为什么 __device__、__constant__、__shared__ 不支持动态初始化？答案

【问题标题】：Why am I getting dynamic initialization not supported for __device__, __constant__, __shared__?为什么 __device__、__constant__、__shared__ 不支持动态初始化？
【发布时间】：2015-01-18 10:41:30
【问题描述】：

我不明白为什么在编译我的代码时会收到错误 dynamic initialization is not supported for __device__, __constant__, __shared__ variables。我的代码看起来像

wrapper.cu

#include "../Params.hpp"

__constant__ Params cparams;

void wrapperFunction(uint& a)
{
   Params ab;
   a = 20;
}

Params.hpp

#include "Utils.hpp"

typedef struct Params
{
   vectorTypef a;
} Params;

Utils.hpp

#include "Vec2.hpp"

typedef unsigned int uint;
typedef Vec2<float> vectorTypef;

Vec2.hpp

template <typename T>
class Vec2
{
public:
   Vec2(){ x = 0.0; y = 0.0;}
   T x, y;
};

使用命令使用 cmake 构建

CUDA_ADD_EXECUTABLE(test main.cpp cudasrc/wrapper.cu

【问题讨论】：

标签： cuda cmake

【解决方案1】：

您的Params 结构用于cparams 的__constant__ 内存定义。

您的Params 结构包含vectorTypef 类型的元素a，它是float 的Vec2 类的类型定义。此类有一个默认构造函数，即最终分配 Params 结构的元素。这种将数据分配给__constant__ 区域的方法在设备代码或主机代码中都是不合法的。

在设备代码中，修改__constant__ 值是不合法的。在主机代码中（这是这里的视图），__constant__ 值应使用适当的 API 分配，即cudaMemcpyToSymbol。我建议您在主机代码中明确分配这些，而不是通过构造函数。

因此，解决此问题的一种可能方法是将默认构造函数更改为空的：

public:
   __host__ __device__ Vec2(){ }; // change this line
   T x, y;

（你也可以删除空的默认构造函数行）

并且，在 wrapper.cu（可能在 wrapperFunction）中，初始化你的 Params __constant__ 结构：

Params hparams;
hparams.a.x = 0.0;
hparams.a.y = 0.0;
cudaMemcpyToSymbol(cparams, &hparams, sizeof(Params));

【讨论】：

你能举个例子吗？我想我没有完全理解这个想法。
添加了一个可能的修复作为示例。
我试过了，我认为构造函数中缺少像__host__ __code__ 这样的东西，以便拥有矢量类型的设备版本
已更新。（我在设置 hparams 值时也遇到了另一个错误。）您也可以删除空的默认构造函数。
尝试删除默认构造函数，编译报错找不到。

【解决方案2】：

我遇到了和你一样的问题，我找到了两种解决方法。

用 C 类型定义你的结构，像这样：
```
typedef struct {} ClassName;
```

将构造函数和析构函数都定义为 __device__ 类型，如下所示：

struct ClassName{
public:
    __device__ ClassName(){...}
    __device__ ~ClassName(){...}
};

【讨论】：