在CUDA中，使用结构体指针设置设备内存中的数组。

Question

我正在尝试使用指向内核中创建的结构体的指针在内存中初始化一个数组。以下是到目前为止我的代码，我不知道我做错了什么。如果我尝试对数组中的每个项进行cudaMalloc，则会出现段错误；如果我不这样做，则会出现“未指明的启动失败”错误。

#include <stdio.h>
#include <stdlib.h>
#include <string.h>

typedef struct {
    int status;
    int location;
    double distance;
} Point;

//Macro for checking cuda errors following a cuda launch or api call
#define cudaCheckError() {\
 cudaError_t e=cudaGetLastError();\
 if(e!=cudaSuccess) {\
   printf("\nCuda failure %s:%d: '%s'\n",__FILE__,__LINE__,cudaGetErrorString(e));\
   exit(0); \
 }\
}

__global__ void kernel1(Point** d_memory, int limit){

  int idx = blockIdx.x * blockDim.x * blockDim.y * blockDim.z 
  + threadIdx.z * blockDim.y * blockDim.x 
  + threadIdx.y * blockDim.x + threadIdx.x;

    if(idx < limit) {

        Point* pt = ( Point *) malloc( sizeof(Point) );
        pt->distance = 10;
        pt->location = -1;
        pt->status = -1;

        d_memory[idx] = pt;
    }
}

__global__ void kernel2(Point** d_memory, int limit){
  int i;
  for (i=0; i<limit;i++){
    printf("%f \n",d_memory[i]->distance);
  }
}

int main(int argc, char *argv[])
{
    int totalGrid = 257*193*129;
    size_t size = sizeof(Point) * totalGrid;
    Point ** d_memory;
    cudaMalloc((void **)&d_memory, size);
    /*
    for(int i=0; i<totalGrid; i++){
        printf("%d\n",i);
        cudaMalloc((void **)&d_memory[i], sizeof(Point));
    }*/
    dim3 bs(16,8,8);
    kernel1<<<6249, bs>>>(d_memory, totalGrid);
    cudaCheckError();

    cudaDeviceSynchronize();

    kernel2<<<1,1>>>(d_memory, totalGrid);
    cudaCheckError();

    cudaFree(d_memory);
    return 0;
}

这是我用来编译代码的工具

 nvcc -arch=sm_20 test.cu

- anupshrestha

Answer 1

我相信你的问题是：

Point **d_memory;

它应该是这样的

Point *d_memory;

你不需要将指针转换为void **，你的代码需要这样做是因为传递的指针是Point ***而不是Point **。

请注意，cudaMalloc()将分配连续的内存，Point **表示您想要一个指针数组，为此，我认为您需要类似于以下内容：

Point **d_memory;
cudaMalloc((void **)&d_memory, rows);
for (row = 0 ; row < rows ; ++row)
    cudaMalloc(&d_memory[row], columns * sizeof(Point));

但是，接下来您需要检查将d_memory作为参数的其他对象是否会相应地处理d_memory。

此外，cudaMalloc()在分配成功时返回cudaSuccess，您从未检查过。

- Iharob Al Asimi

---

我希望d_memory是一个指针数组。我尝试了你的代码，其中我设置了rows = 257*193*129和columns = 1，但在for循环的第一次迭代中仍然出现了分段错误。我不确定我漏掉了什么。 - anupshrestha

这取决于您如何访问指针数组，“columns == 1”没有太多意义。 - Iharob Al Asimi

这是您的代码的修改版本，可以“正确”运行。 但是，它只是像@iharob建议的那样将您的数组展平。 （我还缩短了传递给kernel2的长度，以缩短运行时间。让单个线程遍历整个数组将永远花费时间并溢出设备printf缓冲区。）当您在主机和设备之间传输数据时，指针数组更难管理。 我不建议初学者使用它，但如果您想尝试，请搜索“cuda 2d array”并开始阅读。 - Robert Crovella

@talonmies在这里给出的答案涵盖了从主机到设备传输2D数组（指针数组）的细节。 这并不是微不足道的。 如果这就是你想要的，那么你的问题可以说是重复的。 - Robert Crovella

我暂时会像@iharob和你建议的那样压平我的数组，然后阅读一些资料，再尝试使用CUDA 2D数组。谢谢你们的帮助。 - anupshrestha