如何将包含数组的结构传递给CUDA中的内核？

Question

在下面的代码中，我在结构中有一个数组，我需要将它传递给内核函数。我似乎找不到正确的方法。我尝试查看 SO 上的其他帖子，但不太了解他们的方法。

在我的实际代码中，我接收到两个结构，作为指针，作为调用内核的函数的参数。因此，我需要将这些 'argument structs' 的内容复制到 'GPU memory structs' 并传递给内核。

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

typedef struct{
    int *pass;
    int element;
}Pass;

__global__ void hello(int *a, int *b, Pass *p){

    int i = blockIdx.x * blockDim.x + threadIdx.x;
    if(i < *b)
        a[i] = p -> pass[i] + p -> element;
}

int main(){

    int *a_host, b_host = 5;
    int *a_gpu, *b_gpu;
    Pass *p, *p_gpu;

    a_host = (int*)malloc(sizeof(int) * 5); 
    cudaMalloc(&a_gpu, 5 * sizeof(int));

    cudaMalloc(&b_gpu, sizeof(int));
    cudaMemcpy(b_gpu, &b_host, sizeof(int), cudaMemcpyHostToDevice);

    p = (Pass*)malloc(sizeof(Pass));
    p -> pass = (int*)malloc(5 * sizeof(int));

    for(int i = 0;i < 5;i++)
        p -> pass[i] = i;
    p -> element = 5;

    cudaMalloc(&p_gpu, sizeof(Pass));
    cudaMemcpy(p_gpu, p, sizeof(Pass), cudaMemcpyHostToDevice);

    int numBlocks = 1;
    int threadPerBlock = 512;


    hello<<<numBlocks, threadPerBlock>>>(a_gpu, b_gpu, p_gpu);

    cudaMemcpy(a_host, a_gpu, 5 * sizeof(int), cudaMemcpyDeviceToHost);

    int i;
    for(i = 0;i < 5;i++)
        printf("a[%d]: %d\n", i, a_host[i]);

    cudaFree(p_gpu);
    cudaFree(a_gpu);
    cudaFree(b_gpu);

    free(p);
    free(a_host);

    return(0);
}

Answer 1

在该问题存在的最初 24 小时内，我们已经在该问题中看到了三个不同的代码。这个答案解决了最终的演变。

您遇到的根本问题是这种类型的操作：

cudaMalloc(&p_gpu, sizeof(Pass));
cudaMalloc(&p_gpu -> pass, 5 * sizeof(int));

第二个cudaMalloc 是非法的。这是试图从主机取消引用并为设备内存中的指针分配一个值。会导致段错误。

在设备上分配包含指向其他内存分配的指针的结构的正确过程如下：

1. 为结构指针将指向设备的每个数组或对象分配内存
2. 将这些分配分配给主机内存中的结构副本
3. 将主机内存中的完整初始化结构复制到设备内存

在此问题的代码中，可能如下所示：

// p_hgpu is the device structure copy in host memory
// p_gpu is the device structure copy in device memory
Pass *p_gpu, *p_hgpu;
p_hgpu = (Pass*)malloc(sizeof(Pass));
p_hgpu->element = p->element;
cudaMalloc(&(p_hgpu->pass), sizeof(int) * 5);
cudaMemcpy(p_hgpu->pass, p->pass, sizeof(int) * 5, cudaMemcpyHostToDevice);

// copy p_hgpu to the device
cudaMalloc(&p_gpu, sizeof(Pass));
cudaMemcpy(p_gpu, p_hgpu, sizeof(Pass), cudaMemcpyHostToDevice);

虽然这很简单，但它包含一些微妙之处，如果您对指针及其在 C++ 中的使用有充分的了解，这些细节可能不言自明。在这种情况下，这可能是缺失的部分。