并行编程的通信模式(Communication Patterns)
并行计算:非常多的线程在合作解决一个问题
常见通信模式:
- 映射Map 输入输出关系:一一对应(one-to-one)
- 聚合gather 输入输出关系:多对一(many-to-one)
- 分散scatter 输入输出关系:一对多(one-to-many)
- 模板stencil 输入输出关系:serveral-to-one
- 转换transpose 输入输出关系:一对一(one-to-one)
- 压缩reduce 输入输出关系:多对一(all-to-one)
- 重排scan/sort 输入输出关系:多对多(all-to-all)
GPU硬件模式
包括SM(流处理器),Kernel(核),thread block(线程块),线程。
Thread Block:a group of threads that cooperate to solve a(sub)problem
Thread Blocks:group of thread blocks to solve a function
Kernel核:可以理解为C/C++中的一个函数function
线程索引threadIdx和块索引blockIdx:
(1)CUDA中kernel内的变量和常量
在CUDA中每个块中线程的索引变量threadIdx分为X方向上的索引threadIdx.x以及Y方向上的索引threadIdx.y,块索引变量blockIdx分为X方向上的索引blockIdx.x以及Y方向上的索引block.y,通过这两个索引我们可以准确定位到某个线程。
而kernel中的常量为blockDim以及gridDim,它们分别代表了每个block的大小以及grid中block的分布。
(2)创建内核程序中block和thread
dim3 blocksPerGrid(N1,M1);//在grid中有N1M1个block
dim3 threadsPerBlock(N2,M2);//每个block中有N2M2个thread
(3)创建索引
int threadIndex = threadIdx.x + threadIdx.y * blockDim.x;
int blockIndex = blockIdx.x + blockIdx.y * gridDim.x;
int index = threadIndex + blockIndex * blockDim.x * blockDim.y;
threadIndex为单个block中thread的索引,blockIndex为block在整个grid中的索引,Index即为某个thread在整个grid中的索引。
SM(stream multiprocessor):流处理器
GPU:每个GPU有若干个SM,最少有1个,目前16个算大的,每个SM并行而独立运行