GPU 高性能编程 CUDA : 流

对于大规模数据的并行运算，GPU上的执行性能远高于 CPU 上的性能，除此之外， NVIDAI 图像处理器还支持另一种类型的并行性，类似于 CPU 的多线程应用程序中的任务并行性。任务并行性是指并行执行两个或者多个任务，而不是在大量数据上执行同一个任务。

页锁定主页内存

之前都是使用 cudaMalloc() 在 GPU 上分配内存，以及通过标准 C 库函数 malloc() 在主机上分配内存

除此之外，还可以用 cudaHostAlloc() 来分配主机内存

malloc() 分配的是标准的，可分页的主机内存，cudaHostAlloc() 分配的是页锁定主机内存，页锁定主机内存也叫固定内存，有一个重要的属性：操作系统不会对这块内存分页并交到磁盘上，从而确保改内存始终驻留在物理内存中。操作系统可以安全的使某个应用程序访问该内存的物理地址，因为这块内存不会被破坏或者重新定位。

固定内存是把双刃剑，使用固定内存时，将失去虚拟内存的所有功能。特别是，在应用程序中使用每个页锁定内存时都需要分配物理内存，因为这些内存不能交换到磁盘上，这意味着，系统会更快的耗尽内存。

我们建议，仅对 cudaMemcpy() 调用中的源内存或者目标内存，才使用页锁定内存，在不需要使用它们的时候就立即释放，而不是等程序运行结束的时候

页锁定内存可以提高性能，有一些特殊情况也需要使用页锁定内存

CUDA 流

CUDA 流在加速应用方面起到重要的作用。CUDA 流表示一个 GPU 操作队列，并且该队列中的操作符将以指定的顺序执行，我们可以将每一个流视为 GPU 上的一个任务，并且这些任务是可以并行执行的

使用单个 CUDA 流

假设一个核函数，该函数有两个输入数据缓冲区 a 和 b，核函数将对这些缓冲区中相应位置上的值执行某种计算，将生成的结果保存在输出缓冲区 c 

 

#define N （1024*1024）
__global__ void kernel(int *a ,int *b ,int *c){
    int idx = threadIdx.x + blockDim.x * blockIdx.x;
    if(idx < N){
        c[idx] = a[idx] + b[idx];
    }
}

这个核函数不重要，可以把它看成一个抽象函数，重要的是 main() 中和流相关的代码

首先，要选择一个支持设备重叠（device overlap）功能的设备，支持设备重叠功能的 GPU 能够在执行一个 CUDA C 核函数的同时，还能在设备和主机之间执行复制操作

int main(void){
    cudaDeviceProp prop;
    int whichDevice;
    cudaGetDevice(&whichDevice);
    cudaGetDeviceProperties(&a