CUDA编程模型

一、CUDA编程模型概述

并行计算的三层：

领域层，在编程和算法设计时考虑如何解析数据和函数；

逻辑层，在编程实现时确保线程和计算可以正确解决问题；

硬件层，通过理解线程如何映射到其核心从而提高性能。

1.CUDA编程结构

        1.分配GPU内存

        2.从CPU内存中拷贝数据到GPU内存

        3.调用CUDA的kernel函数完成运算

        4.将数据从GPU拷贝回CPU

        5.释放GPU空间 

2. 内存管理

cudaMalloc：向GPU设备分配一定字节的线性内存，并返回指针。

cudaMemcpy：负责主机和设备直之间的数据传输。此函数以同步方式执行，在cudaMemcpy函数返回以及传输操作完成之前主机应用程序阻塞。

主要需要注意cudaMemcpy函数

cudaError_t cudaMemcpy(void * dst,const void * src,size_t count,
  cudaMemcpyKind kind)

主机到设备：cudaMemcpy(d_A,h_A,nBytes,cudaMemcpyHostToDevice)

设备到主机：cudaMemcpy(h_A,d_A,nBytes,cudaMemcpyDeviceToHost)

主机到主机：cudaMemcpy(d_A,h_A,nBytes,cudaMemcpyHostToHost)

设备到设备：cudaMemcpy(h_A,d_A,nBytes,cudaMemcpyDeviceToDevice)

CUDA的内存主要分为两部分主机端和设备端，而且是具有层次，简单如下图所示，主要由全局内存和共享内存组成。

3. CUDA 执行模型

3.1 CUDA内核由一组线程执行

所有线程运行相同的代码（SPMD）， 每个线程都有一个ID，用于计算内存地址和做出控制决策。

 Block IDs 和Thread IDs。每个线程使用ID确定其作用于哪块数据。Block ID（1D，2D，3D），Thread ID（1D，2D，3D）。在核函数调用时需要 进行初始化i。

此处需要注意的是Block是一个组织线程的方式，并无明确硬件与其对应。

3.2 线程层次结构

Grid：一组由一个kernel启动所产生的线程block，简单说一个kernel会对应一个Grid，在全局内存中共享数据，在运行时动态调度。

Block： 协作线程阵列（CTA），线程之间的同步，在共享内存中共享数据，1D、2D、3D， 最多512个线程（看CUDA版本）。

kernel：在线程上运行的代码部分。

通常一个grid会被组成成二维，block会被组织为三维。 

一个block内的线程可以通过共享内存、原子操作、同步等进行协作，不同block的线程则无法同步。因为不同block的线程可能分布在不同的SM中 。

关于网格和块的大小设置见下一节。

3.3 线程的执行