访问块和内核中的线程数?[复制品]

• blockDim.x,y,z 给出块中的线程数,在 特定方向
• gridDim.x,y,z 给出网格中的块数,在 特定方向
• blockDim.x * gridDim.x 给出网格中的线程数(在本例中,在x方向)

块和网格变量可以是1、2或3维。处理一维数据时,通常只创建一维块和网格。

在CUDA文档中,定义了这些变量 here

特别是,当x维中的线程总数(griddim.x*blockdim.x)为 小于 我希望处理的数组的大小,然后创建一个循环并让线程网格在整个数组中移动是常见的做法。在这种情况下,在处理一个循环迭代之后,每个线程必须移动到下一个未处理的位置,该位置由 tid+=blockDim.x*gridDim.x; 实际上,整个线程网格都是在一维数据数组中跳跃的,一次一个网格宽度。这个主题,有时被称为“网格跨步循环”,将在本文中进一步讨论 blog article .

你可以考虑参加一些介绍性的CUDA网络研讨会 NVIDIA webinar page . 例如,这2:

• 使用cuda c–an introduction的gpu计算(2010)introduction 使用cuda c进行gpu计算的基础知识。 以代码示例的演练演示。没有以前的GPU计算 所需经验
• 使用CUDA C Advanced 1(2010)一级的GPU计算 优化技术,如全局内存优化,以及 处理器利用率。概念将用实际代码来说明 例子

如果你想更好地理解这些概念,那就花2个小时吧。

网格跨步循环的一般主题将在一些细节中介绍 here .

转述自 CUDA Programming Guide :

gridDim:此变量包含网格的维度。

blockidx:这个变量包含网格中的块索引。

blockdim:这个变量包含块的尺寸。

threadidx:此变量包含块内的线程索引。

你似乎对cuda的线程层次结构有点困惑;简而言之,对于内核,将有1个网格(我总是将其视为一个三维立方体)。它的每个元素都是一个块,这样一个声明为 dim3 grid(10, 10, 2); 总共有10*10*2个街区。反过来,每个块都是三维的螺纹立方体。

也就是说,只使用块和网格的x维是很常见的,这就是您所讨论的代码所做的事情。如果你使用的是一维数组,这一点尤其重要。那样的话,你的 tid+=blockDim.x * gridDim.x 行实际上是网格中每个线程的唯一索引。这是因为 blockDim.x 是每个街区的大小, gridDim.x 将是块的总数。

所以如果你用参数启动一个内核

dim3 block_dim(128,1,1);
dim3 grid_dim(10,1,1);
kernel<<<grid_dim,block_dim>>>(...);

然后在你的内核里 threadIdx.x + blockIdx.x*blockDim.x 你将有效地:

threadIdx.x range from [0 ~ 128)

blockIdx.x range from [0 ~ 10)

blockDim.x equal to 128

gridDim.x equal to 10

因此在计算中 threadidx.x+blockidx.x*blockdim.x ,您的值将在以下定义的范围内: [0, 128) + 128 * [1, 10) ,这意味着tid值的范围是{0,1,2,…,1279}。 这对于将线程映射到任务时非常有用,因为这为内核中的所有线程提供了唯一标识符。

但是,如果你有

int tid = threadIdx.x + blockIdx.x * blockDim.x;
tid += blockDim.x * gridDim.x;

那么你基本上会有: tid = [0, 128) + 128 * [1, 10) + (128 * 10) ,而tid值的范围是{1280、1281、…、2559} 我不确定这与什么相关,但这完全取决于您的应用程序以及如何将线程映射到数据。这个映射对于任何内核启动都是非常重要的,而您是决定应该如何进行映射的人。当您启动内核时,您需要指定网格和块维度,并且您必须强制映射到内核中的数据。只要不超过硬件限制(对于现代卡,每个块最多有2^10个线程,每个网格最多有2^16-1个块)

在这个源代码中,我们甚至有4个线程,内核函数可以访问所有10个数组。怎么用?

#define N 10 //(33*1024)

__global__ void add(int *c){
    int tid = threadIdx.x + blockIdx.x * gridDim.x;

    if(tid < N)
        c[tid] = 1;

    while( tid < N)
    {
        c[tid] = 1;
        tid += blockDim.x * gridDim.x;
    }
}

int main(void)
{
    int c[N];
    int *dev_c;
    cudaMalloc( (void**)&dev_c, N*sizeof(int) );

    for(int i=0; i<N; ++i)
    {
        c[i] = -1;
    }

    cudaMemcpy(dev_c, c, N*sizeof(int), cudaMemcpyHostToDevice);

    add<<< 2, 2>>>(dev_c);
    cudaMemcpy(c, dev_c, N*sizeof(int), cudaMemcpyDeviceToHost );

    for(int i=0; i< N; ++i)
    {
        printf("c[%d] = %d \n" ,i, c[i] );
    }

    cudaFree( dev_c );
}

为什么我们不创建10个线程(例如) add<<<2,5>>> or add<5,2>>> 因为如果n大于10 ex)33*1024,我们必须创建相当少的线程数。

这个源代码就是这个例子。 数组为10,CUDA线程为4。 如何仅通过4个线程访问所有10个数组。

请参阅cuda详细信息中关于threadidx、blockidx、blockdim、griddim含义的页面。

在这个源代码中,

gridDim.x : 2    this means number of block of x

gridDim.y : 1    this means number of block of y

blockDim.x : 2   this means number of thread of x in a block

blockDim.y : 1   this means number of thread of y in a block

我们的线程数是4,因为2*2(块*线程)。

在add内核函数中,我们可以访问线程的0、1、2、3索引

-gt; tid = threadIdx.x + blockIdx.x * blockDim.x

0+0×2=0

1+0×2=1

0+1×2=2

–1+1*2=3英镑

如何访问索引4、5、6、7、8、9的其余部分。 while循环中有一个计算

tid += blockDim.x + gridDim.x in while

**内核的首次调用**

-1回路:0+2*2=4

-2回路:4+2*2=8

-3循环:8+2*2=12(但这个值是假的,而不是!)

**内核第二次调用**

-1回路:1+2*2=5

-2回路:5+2*2=9

-3循环:9+2*2=13(但此值为false,而out!)

**内核第三次调用**

-1回路:2+2*2=6

-2循环:6+2*2=10(但这个值是假的,而不是!)

**内核第四次调用**

-1回路:3+2*2=7

-2循环:7+2*2=11(但这个值是假的,而不是!)

因此,0、1、2、3、4、5、6、7、8、9的所有索引都可以通过tid值访问。

请参阅本页。 http://study.marearts.com/2015/03/to-process-all-arrays-by-reasonably.html 我不能上传图片,因为名声不好。