2011-03-30 82 views
2

我有問題綁定到紋理內存全局設備內存的一個子部分。CUDA紋理內存綁定全局內存的子部分

我有充滿存儲器的大全球裝置陣列如下:

雙* device_global; ((** **)& device_global,sizeof(double)* N));

cudaMemcpy(device_global,host,sizeof(double)* N,cudaMemcpyHostToDevice));

我在for循環中運行多個內核。

每個內核所需其我綁定到紋理通過device_global一小部分(INT偏移量= 100):

cudaBindTexture(0,texRef的,device_global,channelDesc,的sizeof(雙)* 10) ;

但是,我面臨的問題是,我無法使用指針算術來僅通過循環偏移來綁定循環部分device_global

我想這樣做:

cudaBindTexture(0,將texRef,device_global + offsett * 1,channelDesc,的sizeof(雙)* 10);

應該指出的是,如果偏移量設置爲0,上述方法可以工作,但某種程度上指針算術不起作用。

任何幫助或其他指導方針將不勝感激。

+0

IIRC指針算法是確定的,即使是設備指針。您是否在循環結束時解除了紋理?什麼是錯誤? – LumpN 2011-03-30 17:07:52

回答

2

紋理存儲器的偏移必須對齊。您無法將內存中的任何部分綁定到只有正確對齊的部分,這是因爲內部高性能硬件的工作原理。

一個解決方案是使用傾斜內存,而不是有非常小的紋理 有幾個大的開始在矩陣對齊的行。

我猜在這裏,但我覺得用

sizeof(double)*10 

作爲紋理顯存是命令datasize,需要更多的設置記憶本身,而不是閱讀。

總矩陣有多大?

+0

給出了範圍的指示,假設每個內核的紋理內存大約是2000倍,程序需要啓動100次內核(因此'device_global'的總大小爲100 * 2000)。因此我希望在每次啓動內核時都綁定並取消綁定。這也回答了@talonmies的評論。任何其他見解或提示非常讚賞。 – user683994 2011-03-30 18:25:57

2

我不相信有可能做你想做的事情。我懷疑有一些幕後地址轉換,這意味着如果您傳遞給綁定調用的指針不是運行時內存管理器已知的並且適當地與頁面邊界對齊,那麼它將不允許綁定紋理到地址。

將整個數組綁定到紋理,然後將索引偏移量傳遞到要用於紋理獲取的每個內核中可能會更好。

2

將0或NULL作爲cudaBindTexture的第一個參數傳遞的錯誤做法。 CUDA紋理綁定要求綁定的指針必須對齊。對齊要求可以通過設備屬性cudaDeviceProp::textureAlignment確定。

cudaBindTexture可以綁定到紋理的任何設備指針。如果指針未對齊,則返回第一個參數cudaBindTexture中距離最近的前一個對齊地址的字節偏移量。如果第一個參數是NULL,則函數調用失敗。

綁定應該做的事爲:

size_t texture_offset = 0; 
cudaBindTexture(&texture_offset, texRef, device_global+ offsett * i , channelDesc, sizeof(double)*10);