cuda编程基础
1.基礎(chǔ)知識整理
一個典型的CUDA程序是按這樣的步驟執(zhí)行的：
? ?一般CPU一個核只支持一到兩個硬件線程，而GPU往往在硬件層面上就支持同時成百上千個并發(fā)線程。不過這也要求我們在GPU編程中更加高效地管理這些線程，以達到更高的運行效率。在CUDA編程中，線程是通過線程網(wǎng)格（Grid）、線程塊（Block）、線程束（Warp）、線程（Thread）這幾個層次進行管理的.
??第二點，為了達到更高的效率，在CUDA編程中我們需要格外關(guān)注內(nèi)存的使用。與CPU編程不同，GPU中的各級緩存以及各種內(nèi)存是可以軟件控制的，在編程時我們可以手動指定變量存儲的位置。具體而言，這些內(nèi)存包括寄存器、共享內(nèi)存、常量內(nèi)存、全局內(nèi)存等。這就造成了CUDA編程中有很多內(nèi)存使用的小技巧，比如我們要盡量使用寄存器，盡量將數(shù)據(jù)聲明為局部變量。而當存在著數(shù)據(jù)的重復(fù)利用時，可以把數(shù)據(jù)存放在共享內(nèi)存里。而對于全局內(nèi)存，我們需要注意用一種合理的方式來進行數(shù)據(jù)的合并訪問，以盡量減少設(shè)備對內(nèi)存子系統(tǒng)再次發(fā)出訪問操作的次數(shù)。
線程管理:
?
內(nèi)存管理:?
其中寄存器(Registers)是GPU上運行速度最快的內(nèi)存空間，通常其帶寬為8TB/s左右，延遲為1個時鐘周期。核函數(shù)中聲明的一個沒有其他修飾符的自變量，通常就存儲在寄存器中。最快速也最受偏愛的存儲器就是設(shè)備中的寄存器，屬于具有重要價值有極度缺乏的資源。
接下來是共享內(nèi)存(shared memory)，共享內(nèi)存是GPU上可受用戶控制的一級緩存。共享內(nèi)存類似于CPU的緩存，不過與CPU的緩存不同，GPU的共享內(nèi)存可以有CUDA內(nèi)核直接編程控制。由于共享內(nèi)存是片上內(nèi)存，所以與全局內(nèi)存相比，它具有更高的帶寬與更低的延遲，通常其帶寬為1.5TB/s左右，延遲為1～32個時鐘周期。對于共享內(nèi)存的使用，主要考慮數(shù)據(jù)的重用性。當存在著數(shù)據(jù)的重復(fù)利用時，使用共享內(nèi)存是比較合適的。如果數(shù)據(jù)不被重用，則直接將數(shù)據(jù)從全局內(nèi)存或常量內(nèi)存讀入寄存器即可。
全局內(nèi)存(global memory)是GPU中最大、延遲最高并且最常使用的內(nèi)存。全局內(nèi)存類似于CPU的系統(tǒng)內(nèi)存。在編程中對全局內(nèi)存訪問的優(yōu)化以最大化程度提高全局內(nèi)存的數(shù)據(jù)吞吐量是十分重要的。
?
1.向量相加的cuda實現(xiàn)
(按對應(yīng)位置相加)
#include <iostream> #include <cuda_runtime.h>#define length 10000//__global__ int foo(int a){}表示一個內(nèi)核函數(shù)， //是一組由GPU執(zhí)行的并行計算任務(wù)，以foo<<>>(a)的形式或者driver API的形式調(diào)用。 //目前__global__函數(shù)必須由CPU調(diào)用，并將并行計算任務(wù)發(fā)射到GPU的任務(wù)調(diào)用單元。隨著GPU可編程能力的進一步提高，未來可能可以由GPU調(diào)用。 __global__ void vector_add(int *a,int *b,int *& c) {for(int i=0;i<length;i++){c[i]=a[i]+b[i];} }int main() {int *a , *b, *c,*a_G , *b_G, *c_G;a=new int[length];b=new int[length];c=new int[length];for (int i =0;i<length;i++){a[i]=i+1;b[i]=i+2;}//gpu分配空間//cudaMalloc//函數(shù)原型: cudaError_t cudaMalloc (void **devPtr, size_t size );//此函數(shù)返回值是CUDA中定義的一個錯誤代碼。cudaMalloc((void**)&a_G, sizeof(int) * length);cudaMalloc((void**)&b_G, sizeof(int) * length);cudaMalloc((void**)&c_G, sizeof(int) * length);//CPU變量給GPU賦值cudaMemcpy(a_G,a,sizeof(int )*length,cudaMemcpyHostToDevice );cudaMemcpy(b_G,b,sizeof(int )*length,cudaMemcpyHostToDevice );//函數(shù)名稱<<<block 數(shù)目, thread 數(shù)目, shared memory 大小>>>(參數(shù)...);vector_add<<<1,1>>>(a_G,b_G,c_G);//Gpu結(jié)果取到cpucudaMemcpy(c,c_G, sizeof(int )* length,cudaMemcpyDeviceToHost);for (int i=0;i<length;i++){cout<<c[i]<<endl;}return 0; }核函數(shù):
__global__ int foo(int a){}表示一個內(nèi)核函數(shù)， 是一組由GPU執(zhí)行的并行計算任務(wù)，以foo<<>>(a)的形式或者driver API的形式調(diào)用。 目前__global__函數(shù)必須由CPU調(diào)用，并將并行計算任務(wù)發(fā)射到GPU的任務(wù)調(diào)用單元。隨著GPU可編程能力的進一步提高，未來可能可以由GPU調(diào)用。cudaMalloc函數(shù):
函數(shù)原型:
cudaError_t cudaMalloc (void ** devPtr, size_t size );
返回值是cuda中定義的錯誤代碼.
第一個變量:指向指針變量的指針,參數(shù)傳遞的是存儲在cpu內(nèi)存中的指針變量的地址，cudaMalloc在執(zhí)行完成后，向這個地址中寫入了一個GPU的地址值
第二個變量:大小.
?
cudaMemcpy函數(shù):
cudaMemcpy用于在主機（Host）和設(shè)備（Device）之間往返的傳遞數(shù)據(jù)，用法如下：
主機到設(shè)備：cudaMemcpy(d_A,h_A,nBytes,cudaMemcpyHostToDevice)
設(shè)備到主機：cudaMemcpy(h_A,d_A,nBytes,cudaMemcpyDeviceToHost)
拷貝數(shù)據(jù)時cpu主進程是鎖死的.
?
2.向量相加的并行
cuda變成只單線程與在CPU下差不多,沒有任何意義,因此要進行并行才能體現(xiàn)GPU的優(yōu)勢.
?
?
＞＞＞＞＞＞＞＞＞＞未完待續(xù)
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
?
總結(jié)
- 上一篇： 前端笔试题总结---持续更新
- 下一篇： Querydsl使用fetchCount