Global Memory
global memory到device memory是經由32,64或128bytes memory transactions進行傳輸。例如,如果為每個thread 4 bytes存取,throughput分成八次的話,可以得到32 bytes的memory transactions。
global memory到device memory是經由32,64或128bytes memory transactions進行傳輸。例如,如果為每個thread 4 bytes存取,throughput分成八次的話,可以得到32 bytes的memory transactions。
Size and Alignment Requirement
若且唯若資料型態大小為1, 2, 4, 8,或16 bytes,而且資料自然地被對齊(address大小為前述的倍數),則對停留在global memory的數據其任何存取(經由variable或pointer)都編譯為單一global memory指令。
對於structure,編譯器可以使用對齊specifiers __align __(8)或 __align __(16)強制執行大小和對齊要求,例如
struct __align__(8) { float x;float y;};
或
struct __align__(16) { float x;float y;
float z;
};
停留在global memory或者是由驅動或runtime API 中的一個memory allocation routines返回的variable其任何address始終對齊至少256個bytes。
讀取非自然對齊的8 byte或16 bytes words會產生不正確的結果,因此必須特別注意保持這些類型的任何值或陣列的起始address的對齊。
以下這個典型情況是當使用一些自定義global memory allocation時,多組陣列的allocation(多次呼叫cudaMalloc()或cuMemAlloc())被單獨大的block memmory分割成多組陣列所取代 ,在這種情況下,可能容易忽視每個的起始地址都已經偏離原本大的block的起始地址。
沒有留言:
張貼留言