7

假设我们有一个数组int * data，每个线程都会访问这个数组的一个元素。由于该数组将在所有线程之间共享，因此它将保存在全局内存中。

让我们创建一个测试内核：

 __global__ void test(int *data, int a, int b, int c){ ... }

我确定该data数组将位于全局内存中，因为我使用cudaMalloc. 现在至于其他变量，我已经看到了一些示例，它们在不分配内存的情况下立即将整数传递给内核函数。在我的情况下，这些变量是a b和c。

如果我没记错的话，即使我们没有直接调用cudaMalloc为每三个整数分配 4 个字节，CUDA 也会自动为我们做这件事，所以最终变量a b和c将被分配到全局内存中。

现在这些变量只是辅助的，线程只读取它们而没有别的。

我的问题是，将这些变量转移到共享内存不是更好吗？

我想如果我们有例如10带有1024线程的块，我们将需要10*3 = 30读取4字节以便将数字存储在每个块的共享内存中。

如果没有共享内存，并且如果每个线程都必须读取所有这三个变量一次，那么全局内存读取的总量将1024*10*3 = 30720非常低效。

现在这是问题所在，我对 CUDA 有点陌生，我不确定是否可以将变量的内存传输a b到c每个块的共享内存，而无需每个线程从全局内存中读取这些变量并加载它们到共享内存，所以最终全局内存读取的总量将是1024*10*3 = 30720而不是10*3 = 30。

在以下网站上有这个例子：

 __global__ void staticReverse(int *d, int n)
 {
    __shared__ int s[64];
    int t = threadIdx.x;
    int tr = n-t-1;
    s[t] = d[t];
    __syncthreads();
   d[t] = s[tr];
 }

这里每个线程在共享变量中加载不同的数据s。所以每个线程，根据它的索引，在共享内存中加载指定的数据。

就我而言，我只想加载变量a b和c共享内存。这些变量始终是相同的，它们不会改变，因此它们与线程本身没有任何关系，它们是辅助的，并且被每个线程用于运行某些算法。

我应该如何解决这个问题？total_amount_of_blocks*3是否可以通过仅进行全局内存读取来实现这一点？

4

1 回答 1

13

GPU 运行时已经以最佳方式做到了这一点，而您无需做任何事情（并且您对参数传递在 CUDA 中的工作方式的假设是不正确的）。这是目前发生的情况：

  • 在计算能力为 1.0/1.1/1.2/1.3 的设备中，内核参数由运行时在共享内存中传递。
  • 在计算能力为 2.x/3.x/4.x/5.x/6.x 的设备中，内核参数由运行时在保留的常量内存库（具有带广播的专用缓存）中传递。

所以在你假设的内核中

__global__ void test(int *data, int a, int b, int c){ ... }

data, a, b, 和c都通过值自动传递给共享内存或常量内存（取决于 GPU 架构）中的每个块。做你建议的事情没有任何好处。

于 2013-05-26T06:21:35.727 回答