2015-02-07 46 views
1

是什麼宣稱本地內存之間的區別如下:OpenCL的本地內存聲明

__kernel void mmul(const int Ndim, const int Mdim, const int Pdim, 
         const __global int* A, 
         const __global int* B, 
        __global char* C, 
        __local int* restrict block_a, 
        __local int* restrict block_b) 

,並宣佈本地內存的內核中

#define a_size 1024 
#define b_size 1024 * 1024 
__kernel void mmul(const int Ndim, const int Mdim, const int Pdim, 
         const __global int* A, 
         const __global int* B, 
        __global char* C) { 

__local int block_a[a_size] 
__local int block_b[b_size] 

... 
} 

在這兩種情況下,所有的線程將更新某一個單元在共享的A和B陣列中

我知道在內核中不可能有「可變」長度的數組(因此第二個內核頂部的#define),b ut有沒有其他區別?關於何時釋放內存有沒有什麼區別?

+0

在內核中定義本地內存意味着它將被分配給每個工作組。有本地內存作爲參數的手段,在分配內存之前的任何地方,也爲每個工作組它是相同的指針。不要考慮何時以及如何釋放內存。 OpenCL就像1.編譯2.調用內核與多少組和工作人員3.現在知道將使用多少內存。 – Aitch 2015-02-07 22:45:10

+0

優化提示:「如果每個線程只是讀取和寫入本地內存的一個位置,爲什麼你甚至需要本地內存?」 – DarkZeros 2015-02-11 12:20:28

回答

2

在這兩種情況下,工作組的整個生命週期都存在本地內存。正如您所指出的,唯一的區別是將本地內存指針作爲參數傳遞可以動態指定緩衝區的大小,而不是編譯時常量。不同的工作組總是使用不同的本地內存分配。

+0

Nit挑選,但如果使用CL_KERNEL_LOCAL_MEM_SIZE標誌調用clGetKernelWorkGroupInfo,它也會有所幫助。在第一種情況下,大小將被假定爲0.在第二種情況下,它將返回正確的大小...... – CaptainObvious 2015-02-17 08:51:01

1

如果您想將代碼移植到CUDA,則第二種方法會更好,因爲CUDA中的__shared__內存(相當於OpenCL中的__local)不支持將其聲明爲第一種情況。