cudaMemcpyToSymbol 使用细节

Question

我正在尝试将数据结构从主机移动到 Tesla C1060（计算 1.3）上的常量内存。具有以下功能：

//mem.cu
#include "kernel.cuh"

int InitDCMem(SimuationStruct *sim)
{
  SimParamGPU h_simparam;

  h_simparam.na = sim->det.na;
  h_simparam.nz = sim->det.nz;
  h_simparam.nr = sim->det.nr;

  cudaMemcpyToSymbol(d_simparam, &h_simparam, sizeof(SimParamGPU));
}  

数据结构（在头文件中）：

//kernel.cuh

typedef struct __align__(16)
{
  int na;
  int nz;
  int nr;
} SimParamGPU;

__constant__ SimParamGPU d_simparam;

问题是这些值似乎没有被复制到 GPU 中的常量内存中。

我是否需要像cudaMemcpyToSymbol do not copy data 中所述在\\mem.cu 上重新声明__constant__。
我应该在某处使用extern 吗？

没有错误，值始终设置为 0。

Answer 1

你的内核在哪里？从编译器的角度来看，d_simparam 与您的主机代码在同一个“翻译单元”中 - 您在这里不需要任何“外部”声明。您可能在内核所在的源文件中需要一个。

这对我有用：

device.h - 带有设备符号的文件：

#include <stdio.h>

struct SimpleStruct {
    int a;
    float b;
};

__constant__ SimpleStruct variable = { 10, 0.3f };

__global__ void kernel() {
    printf("%d %f\n", variable.a, variable.b);
}

host.cu - 主机代码：

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

#define CUDA_CHECK_RETURN(value) {                                      \
    cudaError_t _m_cudaStat = value;                                    \
    if (_m_cudaStat != cudaSuccess) {                                   \
        fprintf(stderr, "Error %s at line %d in file %s\n",             \
                cudaGetErrorString(_m_cudaStat), __LINE__, __FILE__);   \
    exit(1);                                                            \
} }

#include "device.h"

int main(void) {
    const SimpleStruct n = { 7, 0.5f };

    CUDA_CHECK_RETURN(cudaMemcpyToSymbol(variable, &n, sizeof(SimpleStruct)));

    kernel<<<1, 1>>>();

    CUDA_CHECK_RETURN(cudaThreadSynchronize()); // Wait for the GPU launched work to complete
    CUDA_CHECK_RETURN(cudaGetLastError());
    CUDA_CHECK_RETURN(cudaDeviceReset());

    return 0;
}

更新： 仅 SM 2.0 和更新的设备支持单独编译。不能对 SM 1.3 代码使用单独编译。