将多分支树复制到 GPU 内存

Question

我有一个节点树，我正在尝试将它复制到 GPU 内存。节点看起来像这样：

struct Node
{
   char *Key;
   int ChildCount;
   Node *Children;
}

我的复制功能是这样的：

void CopyTreeToDevice(Node* node_s, Node* node_d)
{


     //allocate node on device and copy host node
     cudaMalloc( (void**)&node_d, sizeof(Node));
     cudaMemcpy(node_d, node_s, sizeof(Node), cudaMemcpyHostToDevice);

     //test
     printf("ChildCount of node_s looks to be : %d\n", node_s->ChildCount);
     printf("Key of node_s looks to be : %s\n", node_s->Key);

     Node *temp;
     temp =(Node *) malloc(sizeof(Node));
     cudaMemcpy(temp, node_d, sizeof(Node), cudaMemcpyDeviceToHost);
     printf("ChildCount of node_d on device is actually : %d\n", temp->ChildCount);
     printf("Key of node_d on device is actually : %s\n", temp->Key);
     free(temp);



     //       continue with child nodes
     if(node_s->ChildCount > 0)
     {
         //problem here
         cudaMalloc( (void**)&(node_d->Children), sizeof(Node)*(node_s->ChildCount));

         cudaMemcpy(node_d->Children, node_s->Children, 
                    sizeof(Node)*node_s->ChildCount, cudaMemcpyHostToDevice);

         for(int i=0;i<node_s->ChildCount;i++)
         {
                 CopyTreeToDevice(&(node_s->Children[i]), &(node_d->Children[i]));
         }
     }

}

但我的线路有问题：

cudaMalloc( (void**)&(node_d->Children), sizeof(Node)*(node_s->ChildCount));

给我访问冲突异常。测试部分工作顺利。初始化字段没有问题。

这是测试部分的输出：

ChildCount of node_s looks to be : 35
Key of node_s looks to be : root
ChildCount of node_d on device is actually : 35
Key of node_d on device is actually : root

这是什么原因？

谢谢。

Answer 1

看起来 node_d 本身在 gpu 上。您不能使用 -> 或 访问 gpu 上的结构。 您需要将node_d复制回主机，分配必要的数据并将其复制回来。

Answer 2

node_d->Children 是驻留在设备代码中的变量。您不能像使用第二个 cudaMalloc 那样直接通过主机代码使用它。此外，将主机指针复制到设备没有多大意义，因为您无法在设备代码中取消引用它们。

更好更快的方法是：

• 为整个树预分配一个大数组。
• 使用数组索引而不是指针。索引的有效性将在与设备之间传输时保留。
• 在设备上一次分配整个阵列。拥有多个memAlloc 可能效率低下（尤其是在 Windows 系统中，当监视器连接到该 GPU 时）。此外，由于memAlloc 返回的地址始终与 512 字节对齐，因此您实际上无法分配较小的内存块。因此，根据您当前的代码，每个子数组至少会消耗 512 个字节，即使里面只有 2 个子数组。
• 将整个阵列从主机复制到设备一次。这比拥有多个 memCopy 指令要快得多，即使您实际上复制了一些未使用的额外内存区域。