c++ - 结构的cudaMalloc和相同结构的元素

我想知道当我分配一个结构然后分配（？）时，在设备上发生了什么（从内存方面看）并复制相同结构的指针元素。
我还需要元素a的cudaMalloc吗？
示例代码：

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 = ...;
  ...
}

这个问题是this question的另一个扩展，与this question有关。

最佳答案

我建议您在编译和运行您的代码时，使用适当的cuda错误检查。学习解释编译器输出和运行时输出将使您成为一个更好、更聪明、更高效的编码器。我还建议回顾一下我之前给你指出的here的文章。它处理这个确切的主题，并包括链接的工作实例。这个问题是那个问题的复本。
有各种错误：

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

上面的代码行为大小为StructA的结构在主机内存中创建一个分配，并设置指向该分配开始的指针目前没什么问题。

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

上面的代码行在设备内存中创建一个d_A大小的分配，并设置指向该分配开始的指针StructA。这实际上消除了以前的指针和分配。（以前的主机分配仍在某个位置，但您无法访问它。它基本上是丢失的）当然那不是你的意图。

int *h_A = d_a->a;

既然d_A（我假设您的意思是d_A，而不是d_A）已被指定为设备内存指针，d_a操作将取消对该指针的引用，以定位元素->。这在主机代码中是非法的，并将抛出一个错误（seg fault）。

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

这行代码也有类似的问题。我们不能在设备内存中使用指针。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
$

您会注意到，我还没有解决您处理cudaMalloc（即cudaMalloc>1）数组的情况，该数组需要循环。我将把它交给你来处理我在这里和我的previous linked answer中提出的逻辑，看看你是否能计算出这个循环的细节。此外，为了清晰/简洁起见，我不再使用通常的cuda error checking，但请在代码中使用它。最后，这个过程（有时称为“深度复制操作”）在普通CUDA中有点乏味，如果您还没有得出结论的话。前面的建议是“展平”这样的结构（这样它们就不会包含指针），但是您也可以探索&(d_A->a)即Unified Memory in CUDA 6。