CUDA記憶體型別memory

cuda儲存器型別：

每個執行緒擁有自己的register and loacal memory;

每個執行緒塊擁有一塊shared memory;

所有執行緒都可以訪問global memory;

還有，可以被所有執行緒訪問的唯讀儲存器：constant memory and texture memory

1、暫存器register

暫存器是gpu上的快取記憶體器，其基本單元是暫存器檔案，每個暫存器檔案大小為32bit.

kernel中的區域性(簡單型別)變數第一選擇是被分配到register中。

特點：每個執行緒私有，速度快。

2、區域性儲存器 local memory

當register耗盡時，資料將被儲存到local memory。如果每個執行緒中使用了過多的暫存器，或宣告了大型結構體或陣列，或編譯器無法確定陣列大小，執行緒的私有資料就會被分配到local 　　memory中。

特點：每個執行緒私有；沒有快取，慢。

注：在宣告區域性變數時，盡量使變數可以分配到register。如：

unsigned int mt[3];

改為：　unsigned int mt0, mt1, mt2;

3、共享儲存器 shared memory

可以被同一block中的所有執行緒讀寫

特點：block中的執行緒共有；訪問共享儲存器幾乎與register一樣快.

[cpp]view plain

copy

//u(i)= u(i)^2 + u(i-1)

//static

__global__ example(float

* u)

intmain() ;

float

* devu;

size_t

size =

sizeof

(float

)*4;

cudamalloc(&devu, size);

cudamemcpy(devu, hostu, size,

cudamemcpyhosttodevice);

example<<<1,4>>>(devu, devv);

cudamemcpy(hostu, devu, size,

cudamemcpydevicetohost);

cudafree(devu);

return

} //dynamic

extern

__shared__

inttmp;

__global__ example(float

* u)

intmain() ;

float

* devu;

size_t

size =

sizeof

(float

)*4;

cudamalloc(&devu, size);

cudamemcpy(devu, hostu, size, cudamemcpyhosttodevice);

example<<<1,4,size>>>(devu, devv);

cudamemcpy(hostu, devu, size, cudamemcpydevicetohost);

cudafree(devu);

return

}

4、全域性儲存器 global memory

特點：所有執行緒都可以訪問；沒有快取

[cpp]view plain

copy

//dynamic

__global__ add4f(float

* u,

float

* v)

intmain() ;

float

hostv[4] = ;

float

* devu, devv;

size_t

size =

sizeof

(float

)*4;

cudamalloc(&devu, size);

cudamalloc(&devv, size);

cudamemcpy(devu, hostu, size,

cudamemcpyhosttodevice);

cudamemcpy(devv, hostv, size,

cudamemcpyhosttodevice);

add4f<<<1,4>>>(devu, devv);

cudamemcpy(hostu, devu, size,

cudamemcpydevicetohost);

cudafree(devv);

cudafree(devu);

return

} //static

__device__ float

devu[4];

__device__ float

devv[4];

__global__ adduv()

intmain() ;

float

hostv[4] = ;

size_t

size =

sizeof

(float

)*4;

cudamemcpytosymbol(devu, hostu, size, 0, cudamemcpyhosttodevice);

cudamemcpytosymbol(devv, hostv, size, 0, cudamemcpyhosttodevice);

adduv<<<1,4>>>();

cudamemcpyfromsymbol(hostu, devu, size, 0, cudamemcpydevicetohost);

return

}

5、常數儲存器constant memory

用於儲存訪問頻繁的唯讀引數

特點：唯讀；有快取；空間小(64kb)

注：定義常數儲存器時，需要將其定義在所有函式之外，作用於整個檔案

1 __constant__ int
devvar;
2 cudamemcpytosymbol(devvar, hostvar, sizeof(int), 0
, cudamemcpyhosttodevice)
3 cudamemcpyfromsymbol(hostvar, devvar, sizeof(int), 0, cudamemcpydevicetohost)

6、紋理儲存器 texture memory是一種唯讀儲存器，其中的資料以一維、二維或者三維陣列的形式儲存在視訊記憶體中。在通用計算中，其適合實現影象處理和查詢，對大量資料的隨機訪問和非對齊訪問也有良好的加速效果。

特點：具有紋理快取，唯讀。

tne end

CUDA記憶體型別memory

CUDA記憶體型別memory

CUDA記憶體型別memory

3 記憶體 Memory 效能監控

相關推薦