全局内核中的CUDA变量

Question

我的问题是:

1)我理解正确,当你在全局内核中声明一个变量时,每个线程都会有不同的变量副本.这允许您为每个线程在此变量中存储一些中间结果.示例:vector c = a + b:

__global__ void addKernel(int *c, const int *a, const int *b)
{
   int i = threadIdx.x;
   int p;
   p = a[i] + b[i];
   c[i] = p;
} 

这里我们声明中间变量p.但实际上这个变量有N个副本,每个副本用于每个线程.

2)是否为真,如果我将声明数组,将为每个线程创建此数组的N个副本？只要全局内核中的所有内容都发生在gpu内存上,对于声明的任何变量,你需要在gpu上多花费n倍的内存,其中N是你的线程数.

3)在我当前的程序中,我有35*48 = 1680个块,每个块包括32*32 = 1024个线程.这是否意味着,在全局内核中声明的任何变量将花费我N = 1024*1680 = 1 720 320倍于内核之外？

4)为了使用共享内存,我需要比每个变量多M倍的内存.这里M是块数.真的吗？

Answer 1

1)是的.每个线程都有一个在函数中声明的非共享变量的私有副本.这些通常会进入GPU register内存,但可能会溢出到local内存中.

2),3)和4)虽然你需要许多私有内存的副本,但这并不意味着你的GPU必须同时为每个线程提供足够的私有内存.这是因为在硬件中,并非所有线程都需要同时执行.例如,如果你启动N个线程,那么在给定时间可能有一半是活动的,而另一半在有空闲资源运行之前不会启动.

线程使用的资源越多,硬件可以同时运行的越少,但这并不限制您可以要求运行的数量,因为GPU没有资源的任何线程将在一些资源释放后运行.

这并不意味着你应该疯狂并宣布大量的本地资源.GPU很快,因为它能够并行运行线程.要并行运行这些线程,它需要在任何给定时间适合大量线程.从一般意义上讲,每个线程使用的资源越多,在给定时刻活动的线程就越少,硬件可以利用的并行性就越少.