c++ - 为什么设备代码可以像这样访问cuda中的主机变量？

Question

在下面的代码中，我使用了推力库，我不明白的是在设备中

code 我可以访问 v，我认为它在主机内存中，因为我直接给它赋值

4 在构造函数中。

struct op{
    float v;
    op(float iv):v(iv){}
    __device__
    void operator()(float v2)
    {
         printf("%f\n", v+v2);
    }
};
int main()
{
    device_vector<float> data(4,1);
    for_each(data.begin(), data.end(), op(4));
    return 0;
}

上面代码的输出是5.0 5.0 5.0 5.0。

那么，你能给出一些关于这种行为的提示吗？我已经通过官方搜索了

手动，但未获得有关此行为的任何信息。

---

抱歉我的表达，就像 Ben Voigt 说的，我的问题是“我不明白的是我可以访问 v 的设备代码，我认为它在主机内存中”

Answer 1

v 不在主机内存中，至少如果仿函数对象在设备代码中实例化时不在。 (而你的结构，其 () 运算符只有一个 __device__ 装饰，只能在设备代码中“工作”。)

iv 从主机内存开始，但是从 iv(主机)复制到 v(设备，至少如果op 对象用于推力设备操作)在 for_each 函数计算之前由推力处理。

op() 初始化参数(在您的情况下为 4)的复制实际上与此处的 1 参数需要发生的情况没有什么不同:

device_vector<float> data(4,1);
                            ^

1 参数存在于主机代码 中。然而，device_vector 的 thrust 构造函数会启动必要的主机->设备内存拷贝(和/或内核调用，具体取决于特定的 thrust 函数)以初始化 device_vector 存储，这存在于设备上。

这是在运行时处理的，而不是编译时。您可以用一个变量替换您的 4 常量，并且仍然获得有效结果，就像您可以用一个变量替换 device_vector 的 1 初始值设定项一样。

这就是我要解释的方式。 talonmies 或 JaredHoberock(或其他人)可能可以纠正我的术语，但我认为要点是正确的。