c++ - 如何将设备内存中分配的结构化数据从设备复制到主机

Question

我是 GPU 和 CUDA 编程的新手。我正在尝试将设备上动态分配的结构化数据从设备复制到主机。我修改了 GPU 编程指南中的一个简单代码。编译代码时我没有收到任何错误，但唯一有问题的是输出错误，即“0”。这是代码:

#include <stdlib.h>
#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <stdio.h>

typedef struct Point
{
    int2 pt;
};

#define NUMOFBLOCKS 1
#define THREDSPERBLOCK  16

__device__ Point* pnt[NUMOFBLOCKS];
Point dataptr_h[NUMOFBLOCKS][THREDSPERBLOCK];

__global__ void allocmem() 
{   
    if (threadIdx.x == 0)       
        pnt[blockIdx.x] = (Point*)malloc(1*blockDim.x * sizeof(Point)); 
    __syncthreads(); 
} 

__global__ void usemem() 
{ 
    Point* ptr = pnt[blockIdx.x]; 
    if (ptr != NULL) 
    {       
        ptr[threadIdx.x].pt.x = threadIdx.x; 
        ptr[threadIdx.x].pt.y = threadIdx.x;
        printf("Ptr = %d\t", ptr[threadIdx.x].pt.x);
    }
}

__global__ void freemem() 
{ 
    Point* ptr = pnt[blockIdx.x]; 
    if (ptr != NULL) 
        printf("Block %d, Thread %d: final value = %d\n", blockIdx.x, threadIdx.x, ptr[threadIdx.x]); 
    if (threadIdx.x == 0) 
        free(ptr); 
}


int main()
{
    Point* d_pt[NUMOFBLOCKS];
    for (int i = 0 ; i < NUMOFBLOCKS; i++)
        cudaMalloc(&d_pt[i], sizeof(Point)*16);  

    // Allocate memory  
    allocmem<<< NUMOFBLOCKS, THREDSPERBLOCK >>>();  
    // Use memory 
    usemem<<< NUMOFBLOCKS, THREDSPERBLOCK >>>(); 
    cudaMemcpyFromSymbol(d_pt, pnt, sizeof(d_pt));
    cudaMemcpy(dataptr_h, d_pt, sizeof(dataptr_h), cudaMemcpyDeviceToHost);

    for (int j = 0 ; j < 1; j++)
        for (int i = 0 ; i < 16; i++)
        {
            printf("\nPtr_h(%d,%d)->X = %d\t", j, i, dataptr_h[j][i].pt.x);
            printf("Ptr_h(%d,%d)->Y = %d", j, i, dataptr_h[j][i].pt.y);
        }

    freemem<<< NUMOFBLOCKS, THREDSPERBLOCK >>>();
    cudaDeviceSynchronize();
    return 0;
}

代码的输出是:

Ptr_h(0,0)->X = 0       Ptr_h(0,0)->Y = 0
Ptr_h(0,1)->X = 0       Ptr_h(0,1)->Y = 0
Ptr_h(0,2)->X = 0       Ptr_h(0,2)->Y = 0
Ptr_h(0,3)->X = 0       Ptr_h(0,3)->Y = 0
Ptr_h(0,4)->X = 0       Ptr_h(0,4)->Y = 0
Ptr_h(0,5)->X = 0       Ptr_h(0,5)->Y = 0
Ptr_h(0,6)->X = 0       Ptr_h(0,6)->Y = 0
Ptr_h(0,7)->X = 0       Ptr_h(0,7)->Y = 0
Ptr_h(0,8)->X = 0       Ptr_h(0,8)->Y = 0
Ptr_h(0,9)->X = 0       Ptr_h(0,9)->Y = 0
Ptr_h(0,10)->X = 0      Ptr_h(0,10)->Y = 0
Ptr_h(0,11)->X = 0      Ptr_h(0,11)->Y = 0
Ptr_h(0,12)->X = 0      Ptr_h(0,12)->Y = 0
Ptr_h(0,13)->X = 0      Ptr_h(0,13)->Y = 0
Ptr_h(0,14)->X = 0      Ptr_h(0,14)->Y = 0
Ptr_h(0,15)->X = 0      Ptr_h(0,15)->Y = 0

我该怎么做才能解决这个问题？

Answer 1

您不能将设备 malloc 操作创建的指针与 CUDA 运行时 API(即 cudaMemcpy)一起使用

所以这行代码是有问题的:

cudaMemcpy(dataptr_h, d_pt, sizeof(dataptr_h), cudaMemcpyDeviceToHost);

d_pt 包含从 pnt 中提取的指针。 pnt 的值由设备 malloc 设置。

相反，您需要使用 cudaMalloc 创建正确分配的区域，然后先将您想要的数据复制到这些区域(从设备上的一个区域到另一个区域)，然后再复制到主机使用 cudaMemcpy。

在我进一步解释你的下一个反对意见之前，让我们明确以上是你的意图(使用在设备 malloc 操作中创建的指针作为 的目标之一cudaMemcpy).这是不合法的。

“但是我使用了 cudaMalloc ??”

d_pt 是驻留在主机内存中的指针数组。您获取了这些指针中的每一个，并使用 cudaMalloc 为其分配了一个值(设备内存中指向的位置)。

然后这行代码:

cudaMemcpyFromSymbol(d_pt, pnt, sizeof(d_pt));

重写您使用从设备内存中其他地方获得的指针设置的所有那些指针，特别是设备 malloc 分配的指针。虽然这在技术上是合法的(该行代码不会引发错误)，但这些指针在主机上无用(无论如何与运行时 API 一起使用)。