pointers - 在 CUDA 中，如何将设备指针数组复制到设备内存？

Question

例如，我分配以下指针:

float *data_1, *data_2, *data_3, *data_4;

//Use malloc to allocate memory and fill out some data to these pointers
......
//Filling complete

float *data_d1,*data_d2,*data_d3,*data_d4;

cudaMalloc((void **)&data_d1,size1);
cudaMalloc((void **)&data_d2,size2);
cudaMalloc((void **)&data_d3,size3);
cudaMalloc((void **)&data_d4,size4);

cudaMemcpy(data_d1,data_1,size1,cudaMemcpyHostToDevice);
cudaMemcpy(data_d2,data_2,size2,cudaMemcpyHostToDevice);
cudaMemcpy(data_d3,data_3,size3,cudaMemcpyHostToDevice);
cudaMemcpy(data_d4,data_4,size4,cudaMemcpyHostToDevice);

在此之后，我应该已经获得 4 个设备指针，其中包含主机指针所做的确切数据。现在我想将这些指针存储到一个指针数组中，如下所示，

float *ptrs[4];

ptrs[0] = data_d1;
ptrs[1] = data_d2;
ptrs[2] = data_d3;
ptrs[3] = data_d4;

现在我想将这个指针数组传输到 CUDA 内核。但是，我知道由于 ptrs[4] 实际上在主机内存上，我需要在设备上分配一个新指针。所以我做了这个，

float **ptrs_d;
size_t size = 4 * sizeof(float*);
cudaMalloc((void ***)&ptrs_d,size);
cudaMemcpy(ptrs_d,ptrs,size,cudaMemcpyHostToDevice);

然后调用内核:

kernel_test<<<dimGrid,dimBlock>>>(ptrs_d, ...);
//Declaration should be 
//__global__ void kernel_test(float **ptrs_d, ...);

在 kernel_test 中，使用以下语法加载数据:

if (threadIdx.x < length_of_data_1d)
{
    float element0 = (ptrs[0])[threadIdx.x];
}

编译没问题，但是调试的时候，会报访问冲突的错误。

也许我的代码中有很多错误。但我只想弄清楚为什么我不能以这种方式传递设备指针，以及如果在 CUDA 中允许将设备指针数组传递给内核函数，那么访问它的正确方法是什么。

那么我应该如何解决这个问题？任何建议表示赞赏。提前致谢。

Answer 1

一种可能性是分配一个空指针，就像 CUDA 期望的那样。将其传递到内核时，您可以将其强制转换为 float** .
我是这样做的:

void* ptrs_d = 0;
cudaMalloc(&ptrs_d, 4*sizeof(float*));
cudaMemcpy(ptrs_d, ptrs, 4*sizeof(float*), cudaMemcpyHostToDevice);
kernel_test<<<dimGrid, dimBlock>>>((float**)ptrs_d);