在 CUDA 内核中使用 Eigen 3.3

Question

自 2016 年 11 月起，可以编译引用 Eigen3.3 的 CUDA 代码 - 请参阅 this answer

This answer 不是我要找的东西，现在可能已经“过时”了，因为现在可能有更简单的方法，因为以下内容写在docs

从 Eigen 3.3 开始，现在可以使用 Eigen 的对象和 CUDA 内核中的算法。然而，只有一部分特征是 支持以确保不会触发动态分配 一个 CUDA 内核。

另见here。不幸的是，我找不到任何例子来说明这种情况。

我的问题

现在是否可以编写如下的内核，它应该简单地计算一堆点积？

__global__ void cu_dot(Eigen::Vector3d *v1, Eigen::Vector3d *v2, double *out, size_t N)
{
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if(idx < N)
    {
        out[idx] = v1[idx].dot(v2[idx]);
    }
    return;
}

我可以编译这个，但它似乎不起作用。当我尝试将数据复制到主机时，我得到illegal memory access。请注意，我最初将 Vector3d 存储为 `std::vector，然后分别使用

cudaMalloc((void **)&p_device_v1, sizeof(Eigen::Vector3d)*n);
cudaMemcpy(p_v1_device, v1.data(), sizeof(Eigen::Vector3d)*n, cudaMemcpyHostToDevice);

我已经使用 CMake 在https://github.com/GPMueller/eigen-cuda

建立了一个MWE 项目

Answer 1

在 github 上的 MWE 项目中，你写道：

double dot(std::vector<Eigen::Vector3d> v1, std::vector<Eigen::Vector3d> v2)
{   
    ...     
    // Dot product
    cu_dot<<<(n+1023)/1024, 1024>>>(v1.data(), v2.data(), dev_ret, n);

v1.data() 和 v2.data() 指针位于 CPU 内存中。您需要使用 GPU 内存中的指针，即

// Dot product
cu_dot<<<(n+1023)/1024, 1024>>>(dev_v1, dev_v2, dev_ret, n);

CPU 与 GPU 的结果不相同，但这是代码的问题，即您没有对多个点积执行归约。