c++ - 如何在cuda中将结构的指针变量从主机复制到设备

我有一个包含一些变量和一些指针变量的结构。我想用两种不同的功能将这个结构从主机复制到设备。在第一个函数中，我必须复制除一个指针变量之外的整个结构，然后在第二个函数中，我必须复制剩余的指针。

我能够复制整个结构，但无法复制第二个函数中剩余的指针变量。

#include<iostream>

#define cudaCheckErrors(msg) \
    do { \
        cudaError_t __err = cudaGetLastError(); \
        if (__err != cudaSuccess) { \
            fprintf(stderr, "Fatal error: %s (%s at %s:%d)\n", \
                msg, cudaGetErrorString(__err), \
                __FILE__, __LINE__); \
            fprintf(stderr, "*** FAILED - ABORTING\n"); \
            exit(1); \
        } \
    } while (0)

struct MultiSGDKernelParam {
  int count;
  size_t sizes;
  float *weights;
  float *mom; 
  float lrs;
};


__global__ void Launch(MultiSGDKernelParam *param, int N, MultiSGDKernelParam *result)
{
  for(int i=0; i<N; i++)
  {
     result[i] =param[i];     
  }
}

MultiSGDKernelParam *fillStructure(float *temp, const int N)
{       
    MultiSGDKernelParam *param;
        param = (MultiSGDKernelParam*) malloc( N * sizeof(MultiSGDKernelParam));
        for( int i=0; i< N ; i++)
        {
            param[i].count = i;
            param[i].sizes =  i*2;
            param[i].lrs =  param[i].sizes - i;
            param[i].weights = &temp[i];
        }

    std::cout<<"Inside the function"<<"\n"; 
        for(int i=0; i< N; i++)
        {
                std::cout<<param[i].sizes<<" ,"<<param[i].lrs<<"\t";
        }

    std::cout<<std::endl;   
        for(int i =0 ; i<N;i++)
        {
          std::cout<<*(param[i].weights)<<"\t";

        }
        std::cout<<std::endl;
    MultiSGDKernelParam *d_param;
        cudaMalloc((void**)&d_param, N  * sizeof(MultiSGDKernelParam));
        cudaMemcpy(d_param,param,N  * sizeof(MultiSGDKernelParam),cudaMemcpyHostToDevice);

    return d_param;

}

MultiSGDKernelParam * fillFullStructure(float *tweight, float *tmom,  const int N )
{
  MultiSGDKernelParam *param = fillStructure( tweight, N );

 /* float *d_mom;

   cudaMalloc((void**)&d_mom,N*sizeof(float));
   cudaCheckErrors("cudaMalloc1 fail");
   cudaMemcpy(d_mom,tmom,N*sizeof(float), cudaMemcpyHostToDevice);
   cudaCheckErrors("cudaMemcpy1 fail");*/
   for( int i=0; i< N ; i++)
        {
          cudaMemcpy(&(param[i].mom),&(tmom[i]),sizeof(float), cudaMemcpyHostToDevice);
      cudaCheckErrors("cudaMempcpy2 fail");
        }

    std::cout<<"Momentum Values copied"<<"\n";
   /*cudaMemcpy(&(param->mom),tmom,N*sizeof(float), cudaMemcpyHostToDevice);
   cudaCheckErrors("cudaMempcpy1fail");*/
   return param;
}



int main()
{
    static const  int N =5;
    float tempweight [N], tempmom[N] ;
    for(int i=0; i< N; i++)
    {
            tempweight[i] = i*3 +1;
        tempmom[i] = i+3;
    }

    MultiSGDKernelParam *result;
    MultiSGDKernelParam *param = fillFullStructure( tempweight,tempmom, N ); 
     const unsigned blocks = 1;
         const unsigned threadsPerBlock = 4;
    cudaMalloc(&result, N  * sizeof(MultiSGDKernelParam));
    Launch<<<blocks,threadsPerBlock>>>(param, N, result);
        cudaDeviceSynchronize();
    MultiSGDKernelParam *paramresult;
    paramresult = (MultiSGDKernelParam*) malloc( N * sizeof(MultiSGDKernelParam));
    cudaMemcpy(paramresult,result, N * sizeof(MultiSGDKernelParam),cudaMemcpyDeviceToHost);
    std::cout<<"Inside Main"<<"\n";
    for(int i=0; i< N; i++)
        {
           std::cout<<paramresult[i].sizes<<" ,"<<paramresult[i].lrs<<"\t";
        }
    std::cout<<std::endl;
    for(int i =0 ; i<N;i++)
    { 
          std::cout<<*(paramresult[i].weights)<<"\t";
          std::cout<<*(paramresult[i].mom)<<"\t";
    }
         std::cout<<std::endl;

    return 0;
}

输出为

Inside the function    
0 ,0    2 ,1    4 ,2    6 ,3    8 ,4    
1   4   7   10  13  
Momentum Values copied
Inside Main
0 ,0    2 ,1    4 ,2    6 ,3    8 ,4    
Segmentation fault (core dumped)

我编译了代码，但在打印值时给出了段错误。复制是否成功如果没有，那是什么问题。

最佳答案

我不建议像这样编写 CUDA 内核:
```
__global__ void Launch(MultiSGDKernelParam *param, int N, MultiSGDKernelParam *result)
{
  for(int i=0; i<N; i++)
  {
     result[i] =param[i];     
  }
}
```
即使它只是为了演示，你也应该做以下两件事之一:要么像那样编写内核(没有针对 CUDA 线程的专门化)并且只启动 1 个线程的 1 个 block (那么很明显这只是为了演示)或使用适当的 CUDA 线程索引(例如 int i = threadIdx.x+blockDim.x*blockIdx.x; )并摆脱 for 循环，并使用多个线程启动您的 block 。就目前而言，您都没有做过。您有一个普通的 for 循环，没有专门化，在多个线程中运行。当然，这也许不是您问题的重点，但是您现在拥有的这种行为意味着线程在尝试写入 result[i] 时将相互踩踏。 .即使您的所有其余代码都是正确的，这也可能会混淆对事情是否正常运行的理解。我们将通过将您的启动配置切换为 <<<1,1>>> 来解决此问题
这个:
```
param[i].weights = &temp[i];
```
不可能是正确的。您正在结构内部设置一个指针以指向主机内存 中的内容。 (此处的 temp 项指向您的 tempweight 主机数组。)这样的指针不能以任何方式在设备代码中使用。这是一个基本的 CUDA 原则。当您将该结构复制到设备时，该指针的数值不会以任何方式改变，这意味着它仍然指向主机内存。如果您打算在设备代码中的任何位置使用此指针，您将必须学习如何通过 CUDA 深度复制操作进行工作。和 this answer逐步完成。碰巧的是，您实际上并没有尝试在设备代码中取消引用该指针——您只是将结构从一个地方复制到另一个地方。因此，我们无需进一步深入研究，即可让您显示的设备代码正常工作。
段错误的近端原因是您没有初始化 mom代码中任何位置的结构成员，但您试图在此处取消引用它:
```
std::cout<<*(paramresult[i].mom)<<"\t";
```
在 C 或 C++ 中，如果您尝试取消引用您尚未初始化的指针，则可能会发生不好的事情。我们可以通过注释掉那行代码来解决这个问题。我们还可以通过从 weights 复制仅数字指针值 来“修复”它mom 的结构成员结构成员，在设备代码中。但是，我们不能直接在设备代码中使用这些指针，因为它们是如上所述的主机指针。

以下代码解决了上面的第一项和第三项。它似乎对我来说运行正确。

$ cat t1529.cu
#include<iostream>
#include <stdio.h>

#define cudaCheckErrors(msg) \
    do { \
        cudaError_t __err = cudaGetLastError(); \
        if (__err != cudaSuccess) { \
            fprintf(stderr, "Fatal error: %s (%s at %s:%d)\n", \
                msg, cudaGetErrorString(__err), \
                __FILE__, __LINE__); \
            fprintf(stderr, "*** FAILED - ABORTING\n"); \
            exit(1); \
        } \
    } while (0)

struct MultiSGDKernelParam {
  int count;
  size_t sizes;
  float *weights;
  float *mom;
  float lrs;
};


__global__ void Launch(MultiSGDKernelParam *param, int N, MultiSGDKernelParam *result)
{
  for(int i=0; i<N; i++)
  {
     result[i] =param[i];
  }
}

MultiSGDKernelParam *fillStructure(float *temp, const int N)
{
    MultiSGDKernelParam *param;
        param = (MultiSGDKernelParam*) malloc( N * sizeof(MultiSGDKernelParam));
        for( int i=0; i< N ; i++)
        {
            param[i].count = i;
            param[i].sizes =  i*2;
            param[i].lrs =  param[i].sizes - i;
            param[i].weights = &temp[i];
        }

    std::cout<<"Inside the function"<<"\n";
        for(int i=0; i< N; i++)
        {
                std::cout<<param[i].sizes<<" ,"<<param[i].lrs<<"\t";
        }

    std::cout<<std::endl;
        for(int i =0 ; i<N;i++)
        {
          std::cout<<*(param[i].weights)<<"\t";

        }
        std::cout<<std::endl;
    MultiSGDKernelParam *d_param;
        cudaMalloc((void**)&d_param, N  * sizeof(MultiSGDKernelParam));
        cudaMemcpy(d_param,param,N  * sizeof(MultiSGDKernelParam),cudaMemcpyHostToDevice);

    return d_param;

}

MultiSGDKernelParam * fillFullStructure(float *tweight, float *tmom,  const int N )
{
  MultiSGDKernelParam *param = fillStructure( tweight, N );

 /* float *d_mom;

   cudaMalloc((void**)&d_mom,N*sizeof(float));
   cudaCheckErrors("cudaMalloc1 fail");
   cudaMemcpy(d_mom,tmom,N*sizeof(float), cudaMemcpyHostToDevice);
   cudaCheckErrors("cudaMemcpy1 fail");*/
   for( int i=0; i< N ; i++)
        {
          cudaMemcpy(&(param[i].mom),&(tmom[i]),sizeof(float), cudaMemcpyHostToDevice);
      cudaCheckErrors("cudaMempcpy2 fail");
        }

    std::cout<<"Momentum Values copied"<<"\n";
   /*cudaMemcpy(&(param->mom),tmom,N*sizeof(float), cudaMemcpyHostToDevice);
   cudaCheckErrors("cudaMempcpy1fail");*/
   return param;
}



int main()
{
    static const  int N =5;
    float tempweight [N], tempmom[N] ;
    for(int i=0; i< N; i++)
    {
            tempweight[i] = i*3 +1;
        tempmom[i] = i+3;
    }

    MultiSGDKernelParam *result;
    MultiSGDKernelParam *param = fillFullStructure( tempweight,tempmom, N );
    const unsigned blocks = 1;
    const unsigned threadsPerBlock = 1;
    cudaMalloc(&result, N  * sizeof(MultiSGDKernelParam));
    Launch<<<blocks,threadsPerBlock>>>(param, N, result);
    cudaDeviceSynchronize();
    MultiSGDKernelParam *paramresult;
    paramresult = (MultiSGDKernelParam*) malloc( N * sizeof(MultiSGDKernelParam));
    cudaMemcpy(paramresult,result, N * sizeof(MultiSGDKernelParam),cudaMemcpyDeviceToHost);
    std::cout<<"Inside Main"<<"\n";
    for(int i=0; i< N; i++)
        {
           std::cout<<paramresult[i].sizes<<" ,"<<paramresult[i].lrs<<"\t";
        }
    std::cout<<std::endl;
    for(int i =0 ; i<N;i++)
    {
          std::cout<<*(paramresult[i].weights)<<"\t";
        //  std::cout<<*(paramresult[i].mom)<<"\t";
    }
         std::cout<<std::endl;

    return 0;
}

$ nvcc -o t1529 t1529.cu
$ cuda-memcheck ./t1529
========= CUDA-MEMCHECK
Inside the function
0 ,0    2 ,1    4 ,2    6 ,3    8 ,4
1       4       7       10      13
Momentum Values copied
Inside Main
0 ,0    2 ,1    4 ,2    6 ,3    8 ,4
1       4       7       10      13
========= ERROR SUMMARY: 0 errors
$

如果你想实际使用 weights和 mom设备代码中的结构成员(指针)，您将需要开始尝试理解 CUDA 中的深度复制操作。我已经为您提供了一个链接，该链接通过一个可行的示例逐步说明了该过程。现在，您的代码中没有任何迹象表明您已经实现了这些，并且为您编写代码超出了我打算在这里回答的范围，因为您没有尝试过。

关于c++ - 如何在cuda中将结构的指针变量从主机复制到设备，我们在Stack Overflow上找到一个类似的问题： https://stackoverflow.com/questions/58361569/

c++ - 如何在cuda中将结构的指针变量从主机复制到设备

上一篇：c++ - 我如何修改此函数，以便如果函数的参数是特定结构，它会返回具有更大整数的结构？

下一篇：c++ - WHERE 列 = 值，仅适用于 INTEGER 值