CUDA C权威编程指南笔记（2）

hello world from CPU!

hello world from GPU<<<1,10>>>();

三重尖括号表示从主线程到设备端代码的调用。

cudaDeviceceReset()；用来显式地释放和清空当前进程中与当前设备有关的所有资源。

cudaMalloc： 执行GPU内存分配

cudaMemcpy:负责主机和设备之间的数据传输

cudaMemset

cudaFree:释放GPU的内存

内存分配成功，函数返回：cudaSuccess 分配失败：cudaErrorMemoryAllocation

可以使用cudaGetErrorString将错误代码转化为可读的错误信息

在GPU内存层次中，最重要的是全局内存和共享内存，全局内存类似CPU的系统内存，共享内存类似CPU的缓存。GPU的共享内存可以由CUDA C的内核直接控制。

同一网格中的所有线程共享相同的全局内存空间。

同一线程协作可以通过同步、共享内存来实现

不同块内线程不能协作。

aKernal<<<Dg,Db>>>(args) 

Dg,网格的维度 Db线程块的维度

girdDim 声明网格中的线程块数目

blockDim声明每个线程块中的线程数目

blockidx给出了这个线程块在网格中的索引

threadIdx给出了这个线程在线程块中的索引

 __syncThreads()可以在一个线程块中进行线程同步 Synchronize 同步

cudaDeviceSynchronize() 可有效同步一个网格中的所有线程

原子操作：atomicAdd()，可防止多线程并发访问一个变量时造成的冲突。

size_t:代表内存大小的专用变量类型

cudaError_t：错误处理的专用变量

API：应用程序接口

图像：宽度W像素，高度H像素，索引值 int i =blockIdx.x*blockDim.x+threadIdx.x；

c 0~W-1 r 0~H-1  列索引 行索引

int c =blockIdx.x*blockDim.x+threadIdx.x；

int r =blockIdy.x*blockDim.y+threadIdx.y；

一维数组中索引i的计算如下：int i=r*w+c；

二维内核启动的语法：

dim 3  blockSize(TX,TY);//z默认为1

计算x,y方向上的线程块数

int bx=(W+blockSize.x-1)/blockSize.x;

int by=(H+blockSize.y-1)/blockSize.y;

指定网格大小（在线程块中）：

dim3 gridSize =dim3(bx,by);

kernelName<<<gridSize,blockSize>>>(args) 启动内核程序

领近线程中的数据通过一个常系数数组进行线性组合-滤波

常系数数组-滤波器/模板

CUDA不支持-计算当前线程中的一个变量的值与相邻网格中其他相同变量值的差 

在SIMT并行程序中，每个线程都可以访问它自己本地的每一个内核变量，但是不能访问其他线程的本地变量，当需要在线程间共享信息时，需要在设备的全局内存中申请数组，每个线程都可以在这个数组中读写它自己的u。这个方法有效但是不高效，当启动一个非常大的线程网络时，可能会有数百万的线程尝试从相同的数组读写数据，这将引起同步以及存储阻塞问题。

数据存储的地方离处理器越远，处理数据所需时间越长。设备存储离多流处理器SM更近，但是使用设备存储需要通过一个相对较慢的通信总线进行数据传输。此外当确实需要进行数据传输时，我们更希望通过尽量少的传输次数来减少总的数据传输时间。

全局内存（所有的设备数组存储在这里，提供通用但是相对较慢的存储访问，比访问主机存储快），寄存处存储（每个线程的本地变量存储在这里，访问速度最快，作用范围仅局限与单个线程内）。

共享内存：作用是缩小存储速度与处理器访问之间的差距。共享内存和SM相邻，并且提供了多达48KB的可被同一线程块中所有线程高效访问的存储空间。可以认为共享内存那就是CUDA        为支持高效的线程间合作所提供的最主要的机制，并且在许多情况下（包括许多模板计算）使用共享内存将获得巨大的性能提升。

模板半径即冗余数据访问

模板半径是指该模板每边所覆盖的相邻元素个数。在本章一维情况下，具有二阶精度和半径r=1的模板的数据访问冗余量为2r+1=3.而在N维情况下冗余量是（2r+1）^N，所以当半径与/或维数增加时，问题将变得更加严重。

第7章 三维数据交互

s stack 0~D-1 D表示图像层的深度

层的索引 int s=blockIdx.z*blockDim.z+threadIdx.z;

每个图像每行有w个体素（二维是像素），每层由w*h个体素，体素的一位数组索引是：

int i=c+r*w+s*w*h;

dim3 blockSize(TX,TY,TZ);

dim3 gridSize(divUp(W,TX),divUp(H,TY),divUp(D,TZ));

函数divUp（）是一个常用技巧，用于确保线程块的数量能够覆盖所有的线程。