-1

我有一个奇怪的问题，所以我想我会问一下，看看是否有比我更有经验的人能找到解决方案。

我正在用 CUDA C/C++ 编写一个程序，我有一些常量整数来指定各种东西，比如计算边界的坐标等。目前我只有这些东西在全局设备内存中。它们被每个内核调用中的每个线程访问，所以我认为如果它们在全局内存中，那么它们永远不会被缓存或广播（对吗？）。所以这些小整数占用了很多（相对）开销，并且有很多“读取冗余”。

所以我在标题中声明：

__constant__ int* number;

我包括那个标题，并且，当我做记忆的东西时，我会：

cutilSafeCall( cudaMemcpyToSymbol(number, &(some_host_int), sizeof(int) );

然后我传入number我所有的内核：

__global__ void magical_kernel(int* number, ...){

   //and I access 'number' like this
   int data_thingy = big_array[ *number ];

}

我的代码崩溃了。有了全局内存中的数字，就可以了。我已经确定它在访问内核中的数字时会崩溃。这意味着我访问或分配错误。如果它持有错误的值，它也会导致崩溃，因为它是用来索引到数组中的。

最后，我会问几个问题。首先，我做错了什么？作为奖励：有没有比常量内存更好的方法来完成这个任务 - 我不知道number编译时的值，所以一个简单的 #define 不起作用。恒定内存是否会完全加快代码速度，或者它是否一直被缓存和广播？我能否以某种方式将每个线程块的数据放入共享内存中，并通过多个内核调用将其保留在共享内存中？

4

1 回答 1

1

这里有几个问题：

  1. 您已声明number为指针，但从未为其分配一个值，该值是 GPU 内存中的有效地址
  2. 您有一个变量范围冲突：int * number定义的参数变量与定义为编译单元范围magic_kernel 的变量不同。__constant__ int * variable
  3. 调用的第一个参数cudaMemcpyToSymbol几乎可以肯定是不正确的。

如果您不明白为什么前两点中的任何一点都是正确的，那么您需要对 C++ 中的指针和范围进行一些修改。

根据您对现已删除的答案的回复，我怀疑您实际上想要做的是：

__constant__ int number;

__global__ void magical_kernel(...){

   int data_thingy = big_array[ number ];

}

cudaMemcpyToSymbol("number", &(some_host_int), sizeof(int));

ienumber旨在成为常量内存中的整数，而不是指针，也不是内核参数。

编辑：这是一个例子，它显示了这一点：

#include <cstdio>

__constant__ int number;

__global__ void magical_kernel(int * out)
{
   out[threadIdx.x] = number;
}

int main()
{
    const int value = 314159;
    const size_t sz = size_t(32) * sizeof(int);
    cudaMemcpyToSymbol("number", &value, sizeof(int));

    int * _out, * out;

    out = (int *)malloc(sz);
    cudaMalloc((void **)&_out, sz);

    magical_kernel<<<1,32>>>(_out);

    cudaMemcpy(out, _out, sz, cudaMemcpyDeviceToHost);
    for(int i=0; i<32; i++)
        fprintf(stdout, "%d %d\n", i, out[i]);

    return 0;
}

您应该能够自己运行它并确认它像宣传的那样工作。

于 2012-06-17T06:52:35.557 回答