博客

1、CUDA线程模型

Grid/Block/Thread之间的关系，如下图所示。

Kernel：在GPU上执行的核心程序，这个kernel函数是运行在某个Grid上的。

2、核函数原型

核函数只能在主机端调用，调用时必须申明执行参数。调用形式如下：

Kernel<<<Dg,Db, Ns, S>>>(param list);

<<<>>>运算符内是核函数的执行参数，告诉编译器运行时如何启动核函数，用于说明内核函数中的线程数量，以及线程是如何组织的。

<<<>>>运算符对kernel函数完整的执行配置参数形式是<<<Dg, Db, Ns, S>>>

• 参数Dg用于定义整个grid的维度和尺寸，即一个grid有多少个block。对应核函数内部实参
  
  gridDim
  
  ，为dim3类型。Dim3 Dg(Dg.x, Dg.y, 1)表示grid中每行有Dg.x个block，每列有Dg.y个block，第三维恒为1(因为如上图Block组织形式最多支持2维，而Thread最多3维)。整个grid中共有Dg.x*Dg.y个block，其中Dg.x和Dg.y最大值为65535。

• 参数Db用于定义一个block的维度和尺寸，即一个block有多少个thread。对应核函数内部实参
  
  blockDim
  
  ，为dim3类型。Dim3 Db(Db.x, Db.y, Db.z)表示整个block中每行有Db.x个thread，每列有Db.y个thread，高度为Db.z。Db.x和Db.y最大值为512，Db.z最大值为62。 一个block中共有Db.x*Db.y*Db.z个thread。（根据硬件配置和计算能力不同，Db.x和Db.y最大值也可能为768、1024等）

• 参数Ns是一个可选参数，用于设置每个block除了静态分配的shared Memory以外，最多能动态分配的shared memory大小，单位为byte。不需要动态分配时该值为0或省略不写。

• 参数S是一个cudaStream_t类型的可选参数，初始值为零，表示该核函数处在哪个流之中。