c++ - cudaMemcpyToSymbol : zero value in __constant__ variable

Question

我在同一个问题上看到了很多答案，但从未找到解决方案。只有一些建议在 cudaMemcpyToSymbol(...) 等中使用 char simbol。

我使用来自 cudaMemcpyToSymbol using or not using string 的 100% 工作代码创建了新的 CUDA 项目 (VS2012+CUDA 6.0)

//file: main.cu
#include <stdio.h>
#include <stdlib.h>
#include <cuda.h>

__constant__ float constData[256];
__device__ float devData;
__device__ float* devPointer;

int main(int argc, char **argv)
{
  cudaFree(0);

  float data[256];

  /**>>>>>>>**/   data[0] = 1.0f; 

  cudaError_t err = cudaMemcpyToSymbol(constData, data, sizeof(data));
  printf("Err id: %d, str: %s\n", err, cudaGetErrorString(err));

  /**>>>>>>>**/   printf("constData[0]: %f ", constData[0]);

  float value = 3.14f;
  err = cudaMemcpyToSymbol(devData, &value, sizeof(float));
  printf("Err id: %d, str: %s\n", err, cudaGetErrorString(err));

  /**>>>>>>>**/   printf("devData: %f ", devData);

  float* ptr;
  cudaMalloc(&ptr, 256 * sizeof(float));
  err = cudaMemcpyToSymbol(devPointer, &ptr, sizeof(ptr));
  printf("Err id: %d, str: %s\n", err, cudaGetErrorString(err));
  cudaFree(ptr);

  return EXIT_SUCCESS;
}

还有两个 Unresolved 问题:

1. 运行时错误“无效设备符号”如果不是项目属性中的“compute_10,sm_10”。
2. 调试器和 printf 显示，constData[0] = 0.000000 和 devData = 0.000000

请给我一些意见，有什么问题吗？

Answer 1

当我编译这段代码时，我收到了 2 条警告:

t481.cu(20): warning: a __constant__ variable "constData" cannot be directly read in a host function
t481.cu(26): warning: a __device__ variable "devData" cannot be directly read in a host function

您不应忽略这些警告。尽管它们是“警告”，但它们代表代码中的重大问题。第 20 行看起来像这样:

/**>>>>>>>**/   printf("constData[0]: %f ", constData[0]);

constData 是一个存在于设备上的变量。您不能在普通主机代码中访问这样的变量(例如将它传递给 printf)。这是 CUDA 中的一般规则:设备变量不能在普通主机代码中使用，主机变量也不能在普通设备代码中使用。

第 26 行类似。无论如何，这些行对我们来说并没有多大用处，所以让我们摆脱它们。如果要打印第 20 和 26 行的值，则需要在打印之前将这些值复制回主机(例如使用 cudaMemcpyFromSymbol())。

除此之外，如果我针对正确的 GPU 架构进行编译，您的程序可以毫无错误地运行。您看到的无效设备符号错误是由于您的设备是 cc1.0 设备，它需要 compute_10, sm_10 代码生成。如果您为另一个(更高)架构编译，您的内核将不会加载，因为设备代码与您的架构不匹配。由于没有匹配的设备代码，静态设备变量，包括您的 __constant__ 变量，不会被实例化。因为它们没有实例化，所以符号无效，这是您的程序中出现问题的第一个迹象。

因此您需要针对与您的设备相匹配的 GPU 架构进行编译。