将带有指向数组的指针的类从 C++ 传递到 CUDA

Posted 2023-02-21

【中文标题】将带有指向数组的指针的类从 C++ 传递到 CUDA

【英文标题】：Passing class with pointer to array from C++ to CUDA

【发布时间】：2012-07-19 03:26:29

【问题描述】：

我在 C++ 中有以下类：

template<typename T>
class dynArray 

 public:
    T *elements;
    int size;
    int capacity;
    int initCapacity;

有没有办法使用cudaMemcpy()复制此类的对象以在CUDA内核中使用，而不必逐个元素地复制其内容？

提前致谢。

【参考方案1】：

我认为指针element 会是个问题，因为你必须单独复制elements 数组的内容，然后指针就会被弄乱（即它不会指向element 数组在 GPU 上）。我建议分别复制元素数组和大小/容量值。

【参考方案2】：

最初的想法

在我看来，您似乎希望在 GPU 上拥有类似 std::vector<> 的东西。如果您只需要 GPU 全局内存中的数据或向量的大小，我会给出真正考虑的建议。恕我直言，GPU 上的代码实际上应该只修改数组的数据，而不是调整数组本身的大小。这是应该在主机上完成的事情。

有一个名为AGILE 的开源库，它实现了GPUVector，它基本上类似于GPU 上的std::vector<>。 GPUVector 存储容量、大小和指向 GPU 内存的指针。在GPUVector 上运行的内核获取指向内存区域的指针和大小作为参数，即内核调用看起来像这样：

GPUVector v;
[... initialize v...]
computationKernel<<<blockDim, gridDim>>>(v.data(), v.size());

将此转换为您的类，GPUVector::data() 将只返回 dynArray::elements（指向 GPU 内存），GPUVector::size() 返回 dynArray::size。 dynArray::size 应该保留在 CPU 端，因为您很可能不想从 GPU 代码修改它（例如，因为您不能从 GPU 调用 cudaMalloc）。如果不修改，也可以作为参数传递。

您可能想要查看的另一个库是Thrust，它还在 GPU 上提供了一个类似 STL 的向量。

dynArray的复制方法

由于仍然需要复制整个数组，我建议采用以下方法：

template<typename T>
class dynArray 

  public:
    //! Copies this dynArray to the GPU and returns a pointer to the copy.
    void* copyToDevice()
    
        // Copy the dynArray to the device.
        void* deviceArray;
        cudaMalloc(&deviceArray, sizeof(dynArray<T>));
        cudaMemcpy(deviceArray, this, sizeof(dynArray<T>), 
                   cudaMemcpyHostToDevice);
    
        // Copy the elements array to the device.
        void* deviceElements;
        cudaMalloc(&deviceElements, sizeof(T) * capacity);
        cudaMemcpy(deviceElements, elements, sizeof(T) * capacity, 
                   cudaMemcpyHostToDevice);
    
        // On the device, the elements pointer has to point to deviceElements.
        cudaMemcpy(deviceArray, deviceElements, sizeof(T*),
                   cudaMemcpyHostToDevice);

        return deviceArray;
    
    
    T *elements;
    int size;
    int capacity;
    int initCapacity;

【讨论】：

我看过 Thrust，但我不知道 AGILE。我的问题是我想传递给 CUDA 的代码包括创建一个添加元素的向量，上述结构的目的是避免使用向量...

@dx_mrt：所以你的意思是，你会在GPU上执行的代码中动态改变dynArray的大小？如果size 增长到capacity 以上，我预计您会遇到问题。当然，您可以为所有情况创建一个足够大的数组，但这可能会浪费 非常宝贵的 GPU 内存。

好的，如果使用类似这样的东西，看起来可以在内核中执行 free() 和 malloc() ：cudaDeviceSetLimit (cudaLimitMallocHeapSize, 128*1024*1024)

@dx_mrt：太好了！到目前为止，我还不知道您现在还可以从设备函数中调用 malloc() 和 free()。对于reference。