搜索
文章目录1.CUDA是什么2.64位Ubuntu12.04安装CUDA5.53.对CUDAC个人懵懂感觉4.重要概念与名称4.1.主机4.2.设备4.3.线程(Thread)4.4.线程块(Block)4.5.线程格(Grid)4.6.线程束4.7.核函数（Kernel）4.8.dim3构造类型5.函数修饰符6.惯用GPU内存函数6.1.cudaMalloc()6.2.cudaMemcpy()6.3.cudaFree()7.GPU内存分类7.1.全局内存7.2.共享内存7.3.常量内存7.4.纹理内存7.5.固定内存8.原子性9.惯用线程操作函数10.使用事件来测量性能11.流12.技巧CUDA是什么CUDA，ComputeUnifiedDeviceArchitecture简称，是由NVIDIA公司创立基于她们公司生产图形解决器GPUs（GraphicsProcessingUnits,可以通俗理解为显卡）一种并行计算平台和编程模型。通过CUDA，GPUs可以很以便地被用来进行通用计算（有点像在CPU中进行数值计算等等）。在没有CUDA之前，GPUs普通只用来进行图形渲染（如通过OpenGL，DirectX）。开发人员可以通过调用CUDAAPI，来进行并行编程，达到高性能计算目。NVIDIA公司为了吸引更多开发人员，对CUDA进行了编程语言扩展，如CUDAC/C++,CUDAFortran语言。注意CUDAC/C++可以看作一种新编程语言，由于NVIDIA配备了相应编译器nvcc，CUDAFortran同样。更多信息可以参照文献。对CUDAC个人懵懂感觉如果粗暴以为C语言工作对象是CPU和内存条（接下来,称为主机内存），那么CUDAC工作对象就是GPU及GPU上内存（接下来,称为设备内存），且充分运用了GPU多核优势及减少了并行编程难度。普通通过C语言把数据从外界读入，再分派数据，给CUDAC，以便在GPU上计算，然后再把计算成果返回给C语言，以便进一步工作，如进一步解决及显示，或重复此过程。重要概念与名称主机将CPU及系统内存（内存条）称为主机。设备将GPU及GPU自身显示内存称为设备。线程(Thread)普通通过GPU一种核进行解决。（可以表达到一维，二维，三维，详细下面再细说）。线程块(Block)1.由各种线程构成（可以表达到一维，二维，三维，详细下面再细说）。2.各block是并行执行，block间无法通信，也没有执行顺序。3.注意线程块数量限制为不超过65535（硬件限制）。线程格(Grid)/**@file_nameHelloWorld.cu后缀名称.cu*/由各种线程块构成（可以表达到一维，二维，三维，详细下面再细说）。线程束在CUDA架构中，线程束是指一种包括32个线程集合，这个线程集合被“编织在一起”并且“步调一致”形式执行。在程序中每一行，线程束中每个线程都将在不同数据上执行相似命令。核函数（Kernel）1.在GPU上执行函数普通称为核函数。2.普通通过标记符global修饰，调用通过参数1,参数2，用于阐明内核函数中线程数量，以及线程是如何组织。3.以线程格（Grid）形式组织，每个线程格由若干个线程块（block）构成，而每个线程块又由若干个线程（thread）构成。4.是以block为单位执行。5.叧能在主机端代码中调用。6.调用时必要声明内核函数执行参数。7.在编程时，必要先为kernel函数中用到数组或变量分派好足够空间，再调用kernel函数，否则在GPU计算时会发生错误，例如越界或报错，甚至导致蓝屏和死机。dim3构造类型1.dim3是基亍uint3定义矢量类型，相称亍由3个unsignedint型构成构造体。uint3类型有三个数据成员unsignedintx；unsignedinty；unsignedintz;2.可使用亍一维、二维或三维索引来标记线程，构成一维、二维或三维线程块。3.dim3构造类型变量用在核函数调用,中。4.有关几种内置变量4.1.threadIdx，顾名思义获取线程threadID索引；如果线程是一维那么就取threadIdx.x，二维还可以多取到一种值threadIdx.y，以此类推到三维threadIdx.z。4.2.blockIdx，线程块ID索引；同样有blockIdx.x，blockIdx.y，blockIdx.z。4.3.blockDim，线程块维度，同样有blockDim.x，blockDim.y，blockDim.z。4.4.gridDim，线程格维度，同样有gridDim.x，gridDim.y，gridDim.z。5.对于一维block，线程threadID=threadIdx.x。6.对于大小为（blockDim.x，blockDim.y）二维block，线程threadID=threadIdx.x+threadIdx.y*blockDim.x。7.对于大小为（blockDim.x，blockDim.y，blockDim.z）三维block，线程#includestdio.h#includecuda_runtime.h//头文献//核函数声明，前面核心字globalglobalvoidkernel(void){}intmain(void){//核函数调用，注意1,1，第一种1，代表线程格里只有一种线程块；第二个1，代表一种线程块里只有一种线程。kernel1,1();printf(Hello，World!\n);return0;}threadID=threadIdx.x+threadIdx.y*blockDim.x+threadIdx.z*blockDim.x*blockDim.y。8.对于计算线程索引偏移增量为已启动线程总数。如stride=blockDim.x*gridDim.x；threadId+=stride。函数修饰符1.global，表白被修饰函数在设备上执行，但在主机上调用。2.device，表白被修饰函数在设备上执行，但只能在其她device函数或者global函数中调用。惯用GPU内存函数cudaMalloc()1.函数原型：cudaError_tcudaMalloc(void**devPtr，size_tsize)。2.函数用处：与C语言中malloc函数同样，只是此函数在GPU内存为分派内存。3.注意事项：3.1.可以将cudaMalloc()分派指针传递给在设备上执行函数；3.2.可以在设备代码中使用cudaMalloc()分派指针进行设备内存读写操作；3.3.可以将cudaMalloc()分派指针传递给在主机上执行函数；3.4.不可以在主机代码中使用cudaMalloc()分派指针进行主机内存读写操作（即不能进行解引用）。cudaMemcpy()1.函数原型：cudaError_tcudaMemcpy(void*dst，constvoid*src，size_tcount，cudaMemcpyKindkind)。2.函数作用：与c语言中memcpy函数同样，只是此函数可以在主机内存和GPU内存之间互相拷贝数据。3.函数参数：cudaMemcpyKindkind表达数据拷贝方向，如果kind赋值为cudaMemcpyDeviceToHost表达数据从设备内存拷贝到主机内存。4.与C中memcpy()同样，以同步方式执行，即当函数返回时，复制操作就已经完毕了，并且在输出缓冲区中包括了复制进去内容。5.相应有个异步方式执行函数cudaMemcpyAsync()，这个函数详解请看下面流一节关于内容。cudaFree()1.函数原型：cudaError_tcudaFree(void*devPtr)。2.函数作用：与c语言中free()函数同样，只是此函数释放是cudaMalloc()分派内存。下面实例用于解释上面三个函数#includestdio.h#includecuda_runtime.hglobalvoidadd(inta，intb，int*c){*c=a+b;}intmain(void){intc;int*dev_c;//cudaMalloc()cudaMalloc((void**)&dev_c，sizeof(int));//核函数执行add1,1(2，7，dev_c)；//cudaMemcpy()cudaMemcpy(&c，dev_c，sizeof(int),cudaMemcpyDeviceToHost);printf(2+7=%d\n，c);//cudaFree()cudaFree(dev_c);return0;}GPU内存分类全局内存通俗意义上设备内存。共享内存1.位置：设备内存。2.形式：核心字shared添加到变量声明中。如sharedfloatcache[10]。3.目：对于GPU上启动每个线程块，CUDAC编译器都将创立该共享变量一种副本。线程块中每个线程都共享这块内存，但线程却无法看到也不能修改其她线程块变量副本。这样使得一种线程块中各种线程可以在计算上通信和协作。常量内存1.位置：设备内存2.形式：核心字constant添加到变量声明中。如constantfloats[10];。3.目：为了提高性能。常量内存采用了不同于原则全局内存解决方式。在某些状况下，用常量内存替代全局内存能有效地减少内存带宽。4.特点：常量内存用于保存在核函数执行期间不会发生变化数据。变量访问限制为只读。NVIDIA硬件提供了64KB常量内存。不再需要cudaMalloc()或者cudaFree(),而是在编译时，静态地分派空间。5.规定：当咱们需要拷贝数据到常量内存中应当使用cudaMemcpyToSymbol()，而cudaMemcpy()会复制到全局内存。6.性能提高因素：6.1.对常量内存单次读操作可以广播到其她“邻近”线程。这将节约15次读取操作。（为什么是15，由于“邻近”指半个线程束，一种线程束包括32个线程集合。）6.2.常量内存数据将缓存起来，因而对相似地址持续读操作将不会产生额外内存通信量。纹理内存1.位置：设备内存2.目：可以减少对内存祈求并提供高效内存带宽。是专门为那些在内存访问模式中存在大量空间局部性图形应用程序设计，意味着一种线程读取位置也许与邻近线程读取位置“非常接近”。如下图：3.纹理变量（引用）必要声明为文献作用域内全局变量。4.形式：分为一维纹理内存和二维纹理内存。4.1.一维纹理内存4.1.1.用texture类型类型声明，如texturefloattexIn。4.1.2.通过cudaBindTexture()绑定到纹理内存中。4.1.3.通过tex1Dfetch()来读取纹理内存中数据。4.1.4.通过cudaUnbindTexture()取消绑定纹理内存。4.2.二维纹理内存4.2.1.用texture类型,数字类型声明，如texturefloat，2texIn。4.2.2.通过cudaBindTexture2D()绑定到纹理内存中。4.2.3.通过tex2D()来读取纹理内存中数据。4.2.4.通过cudaUnbindTexture()取消绑定纹理内存。固定内存1.位置：主机内存。2.概念：也称为页锁定内存或者不可分页内存，操作系统将不会对这块内存分页并互换到磁盘上，从而保证了该内存始终驻留在物理内存中。因而操作系统可以安全地使某个应用程序访问该内存物理地址，由于这块内存将不会破坏或者重新定位。3.目：提高访问速度。由于GPU懂得主机内存物理地址，因而可以通过“直接内存访问DMA（DirectMemoryAccess)技术来在GPU和主机之间复制数据。由于DMA在执行复制时无需CPU介入。因而DMA复制过程中使用固定内存是非常重要。4.缺陷：使用固定内存，将失去虚拟内存所有功能；系统将更快耗尽内存。5.建议：对cudaMemcpy()函数调用中源内存或者目的内存，才使用固定内存，并