cudaMalloc的结构和相同结构的元素

Question

我想知道当我分配一个结构然后分配(？)并复制相同结构的指针元素时,设备上发生了什么(内存方面).

我是否还需要cudaMalloc元素*a？

示例代码:

typedef struct {
  int *a;
  ...
} StructA;

int main() 
{
  int row, col, numS = 10; // defined at runtime

  StructA *d_A = (StructA*)malloc(numS * sizeof(StructA));
  int *h_A = d_a->a;

  cudaMalloc( (void**)&(d_A), numS * sizeof(StructA) );

  cudaMalloc( &(d_A->a), row*col*sizeof(int) ); // no (void**) needed?
  cudaMemcpy( d_A->a, h_A, row*col*sizeof(int), cudaMemcpyHostToDevice );

  kernel<<<grid, block>>>(d_A); // Passing pointer to StructA in device
  ...
}

内核定义:

__global__ kernel(StructA *d_A)
{
  d_A->a = ...;
  ...
}

这个问题的另一部分这一问题,并涉及到这个问题.

Answer 1

我建议您通过适当的cuda错误检查来编译和运行代码.学习解释编译器输出和运行时输出将使您成为更好,更智能,更高效的编码器.我还建议我回顾一下我之前在这里指出的文章.它处理这个确切的主题,并包括链接的工作示例.这个问题与那个问题重复.

有各种错误:

StructA *d_A = (StructA*)malloc(numS * sizeof(StructA));

上面的代码行在主机内存中为大小结构创建一个分配StructA,并将指针设置为指向d_A该分配的开始.这一刻没有错.

cudaMalloc( (void**)&(d_A), numS * sizeof(StructA) );

上面的代码行在设备内存中创建大小为的分配StructA,并将指针设置为指向d_A该分配的开头.这有效地消除了先前的指针和分配.(之前的主机分配仍在某处,但您无法访问它.它基本上已丢失.)当然这不是您的意图.

int *h_A = d_a->a;

现在d_A(我假设你的意思d_A,而不是d_a)已被指定为设备内存指针,该->操作将取消引用该指针以定位该元素a.这在主机代码中是非法的,并且会抛出错误(seg fault).

cudaMalloc( &(d_A->a), row*col*sizeof(int) );

这行代码有类似的问题.我们不能cudaMalloc指针存在于设备内存中. cudaMalloc创建存在于主机内存中的指针,但引用设备内存中的位置.此操作&(d_A->a)取消引用设备指针,这在主机代码中是非法的.

一个合适的代码是这样的:

$ cat t363.cu
#include <stdio.h>

typedef struct {
  int *a;
  int foo;
} StructA;

__global__ void kernel(StructA *data){

  printf("The value is %d\n", *(data->a + 2));
}

int main()
{
  int  numS = 1; // defined at runtime

  //allocate host memory for the structure storage
  StructA *h_A = (StructA*)malloc(numS * sizeof(StructA));
  //allocate host memory for the storage pointed to by the embedded pointer
  h_A->a = (int *)malloc(10*sizeof(int));
  // initialize data pointed to by the embedded pointer
  for (int i = 0; i <10; i++) *(h_A->a+i) = i;
  StructA *d_A;  // pointer for device structure storage
  //allocate device memory for the structure storage
  cudaMalloc( (void**)&(d_A), numS * sizeof(StructA) );
  // create a pointer for cudaMalloc to use for embedded pointer device storage
  int *temp;
  //allocate device storage for the embedded pointer storage
  cudaMalloc((void **)&temp, 10*sizeof(int));
  //copy this newly created *pointer* to it's proper location in the device copy of the structure
  cudaMemcpy(&(d_A->a), &temp, sizeof(int *), cudaMemcpyHostToDevice);
  //copy the data pointed to by the embedded pointer from the host to the device
  cudaMemcpy(temp, h_A->a, 10*sizeof(int), cudaMemcpyHostToDevice);

  kernel<<<1, 1>>>(d_A); // Passing pointer to StructA in device
  cudaDeviceSynchronize();
}
$ nvcc -arch=sm_20 -o t363 t363.cu
$ cuda-memcheck ./t363
========= CUDA-MEMCHECK
The value is 2
========= ERROR SUMMARY: 0 errors
$

你会注意到我没有解决你正在处理一个StructA(即numS> 1)数组的情况,这将需要一个循环.我将把它留给你来完成我在这里提出的逻辑以及之前的链接答案,看看你是否可以弄清楚那个循环的细节.此外,为了清楚/简洁起见,我省略了通常的cuda错误检查,但请在您的代码中使用它.最后,如果您还没有得出结论,这个过程(有时称为"深度复制操作")在普通CUDA中有点单调乏味.沿着这些线路上一个建议是"扁平化"这样的结构(所以他们不contiain指针),但你也可以探索cudaMallocManaged,即在6个CUDA通用存储器.