CUDA学习之一

CUDA 的核心有三个重要抽象概念：线程组层次结构、共享存储器、屏蔽同步（barrier synchronization），可轻松将其作为 C 语言的最小扩展级公开给程序员。

GPU 专用于解决可表示为数据并行计算的问题——在许多数据元素上并行执行的程序，具有极高的计算密度（数学运算与存储器运算的比率）。由于所有数据元素都执行相同的程序，因此对精密流控制的要求不高；由于在许多数据元素上运行，且具有较高的计算密度，因而可通过计算隐藏存储器访问延迟，而不必使用较大的数据缓存。

最新一代的 NVIDIA GPU 基于 Tesla 架构。在 NVIDIA Tesla 架构中，一个线程块最多可以包含 512 个线程。

可以通过调用 _syncthreads()_ 内函数在内核中指定同步点；_syncthreads()_ 起到屏障的作用，块中的所有线程都必须在这里等待处理。

int main()

{

    // Kernel invocation

    dim3 dimBlock(16, 16); //说明每个块里面有16*16个线程

    dim3 dimGrid((N + dimBlock.x – 1) / dimBlock.x,

                   (N + dimBlock.y – 1) / dimBlock.y);  //定义二维块。这种方法经常要用

    matAdd<<<dimGrid, dimBlock>>>(A, B, C);   //内核函数的调用

}

       一个内核可能由多个大小相同的线程块执行，因而线程总数应等于每个块的线程数乘以块的数量。这些块将组织为一个一维或二维线程块网格。

网格的维度由 <<<…>>> 语法的第一个参数指定。网格内的每个块多可由一个一维或二维索引标识，可通过内置的 blockIdx 变量在内核中访问此索引。可以通过内置的 blockDim 变量在内核中访问线程块的维度

 

存储器层次结构：

       CUDA 线程可在执行过程中访问多个存储器空间的数据。每个线程都有一个私有的本地存储器。每个线程块都有一个共享存储器，该存储器对于块内的所有线程都是可见的，并且与块具有相同的生命周期。最终，所有线程都可访问相同的全局存储器。

此外还有两个只读的存储器空间，可由所有线程访问，这两个空间是固定存储器空间和纹理存储器空间。全局、固定和纹理存储器空间经过优化，适于不同的存储器用途。

对于同一个应用程序启动的内核而言，全局、固定和纹理存储器空间都是持久的。

 

主机和设备编程模式如下图所示：

 

CUDA软件栈

CUDA 软件栈包含多个层：设备驱动程序、应用程序编程接口（API）及其运行时、两个较高级别的通用数学库，即 CUFFT 和 CUBLAS。

 

支持CUDA的显卡及其计算能力：

	多处理器数量	计算能力
GeForce GTX 280	30	1.3
GeForce GTX 260	24	1.3
GeForce 9800 GX2	2×16	1.1
GeForce 9800 GTX	16	1.1
GeForce 8800 Ultra, 8800 GTX	16	1.0
GeForce 8800 GT	14	1.1
GeForce 9600 GSO, 8800 GS, 8800M GTX	12	1.1
GeForce 8800 GTS	12	1.0
GeForce 9600 GT, 8800M GTS	8	1.1
GeForce 9500 GT, 8600 GTS, 8600 GT, 8700M GT, 8600M GT, 8600M GS	4	1.1
GeForce 8500 GT, 8400 GS, 8400M GT, 8400M GS	2	1.1
GeForce 8400M G	1	1.1
Tesla S1070	4×30	1.3
Tesla C1060	30	1.3
Tesla S870	4×16	1.0
Tesla D870	2×16	1.0
Tesla C870	16	1.0
Quadro Plex 1000 Model S4	4×16	1.0
Quadro Plex 1000 Model IV	2×16	1.0
Quadro FX 5600	16	1.0
Quadro FX 3700	14	1.1
Quadro FX 3600M	12	1.1
Quadro FX 4600	12	1.0
Quadro FX 1700, FX 570, NVS 320M, FX 1600M, FX 570M	4	1.1
Quadro FX 370, NVS 290, NVS 140M, NVS 135M, FX 360M	2	1.1
Quadro NVS 130M	1	1.1