在 CUDA 内核中使用 Eigen 3.3

Question

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

This answer is not what I'm looking for and may now be "outdated" in the sense that there might now be an easier way, since the following is written in the docs

Starting from Eigen 3.3, it is now possible to use Eigen's objects and algorithms within CUDA kernels. However, only a subset of features are supported to make sure that no dynamic allocation is triggered within a CUDA kernel.

另见 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);

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

使用 CMake 建立了一个 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 结果不相同，但这是代码的问题，即您没有对多点积进行缩减。

在 CUDA 内核中使用 Eigen 3.3

Using Eigen 3.3 in a CUDA kernel

c++

cuda

c++11

eigen3

我的问题