2013-01-22 38 views
0

我想將SM的所有可用共享內存分配給一個塊。我這樣做是因爲我不希望將多個塊分配給同一個SM。 我的GPU卡有64KB(共享+ L1)內存。在我目前的配置中,48KB分配給共享內存,16KB分配給L1。 我編寫了下面的代碼來使用所有可用的共享內存。如何將所有可用的共享內存分配給CUDA中的單個塊?

__global__ void foo() 
{ 

    __shared__ char array[49152]; 
... 

} 

我有兩個問題:

  1. 我怎麼能確保所有的共享存儲空間用完?
  2. 我可以將「48K」增加到更高的值(沒有任何錯誤或警告)。有沒有人可以證明這一點?

由於提前,

伊曼

+2

當我編譯一個(靜態分配的)共享內存太大的大小時,我得到一個編譯錯誤(來自ptxas)。請提供一個完整的可編譯示例,其中包含共享內存分配過大的代碼以及用於編譯的命令行,以便進一步調查問題#2。如果您在Eugene建議的內核啓動時切換到共享內存的動態分配,那麼如果請求的共享內存過大,您應該會遇到運行時錯誤(您是否檢查錯誤?)。 –

回答

2
  1. 你可以通過調用cudaGetDeviceProperties
  2. cudaDeviceProp::sharedMemPerBlock,你可以得到讀取可用的設備共享內存的大小不必指定的大小你的數組。相反,您可以動態傳遞共享內存的大小作爲第三個內核啓動參數。

「clock」CUDA SDK示例說明如何在啓動時指定共享內存大小。

+0

如果您的應用程序不支持併發內核執行,那麼您只需分配1/2 * MAX_SHARED_MEMORY_PER_SM + 1個字節的共享內存,以將每個SM的佔用量限制爲1個塊。 –