我有如下操作的工作流:CUDA 2D陣列 - 在分配之前考慮音高?
- 負荷初始值
- 過程值的中間結果甲
- 方法A中間體結果乙
- 方法B中間體結果Ç
- 過程C和B至中間結果D和E
- 總和部分D至最終結果F
我所有中間結果的自然結構都是2D數組,我用cudaMallocPitch()分配它。
不幸的是,我的算法要求我一次將D,E,C和B保存在內存中,並且單獨地,內存比B大4倍。由於我的處理中有另一個限制(迭代在存儲器中的圖形結構上),A或B的尺寸受D和E的最大尺寸限制,而D和E的最大尺寸又由初始值+ B的內存消耗+ C的內存消耗的內存使用量決定。這種依賴性是因爲我正在從主機中「分頁」到/從設備內存的中間結果部分(以容納非常大的問題集),並且我無法開始第4步,直到完成了整個步驟1-3問題集。
一旦我有B中所有的問題集,我可以刪除A.
我目前確定d + E的最大尺寸用下面的函數:
int gpuCalculateSimulPatterns(int lines, int patterns) {
// get free memory
size_t free_mem, total_mem;
int allowed_patterns;
cudaMemGetInfo(&free_mem, &total_mem);
allowed_patterns = (free_mem - (lines*sizeof(int)))/(lines*(sizeof(int)*2.5) + lines*sizeof(char)*1.5);
return min(patterns, allowed_patterns -(allowed_patterns % 32));
}
它「作品」 ,但僅僅是因爲我高估了D或E(它們的尺寸和內存使用率相同)的大小25%,並且使B的預期大小加倍。即使如此,我仍遇到邊緣情況,因爲它的內存分配失敗內存不足。我想更有效地利用卡上的內存和保持對齊,因爲我的內核對全局內存進行多次讀寫。
不,使用共享內存不是一個選項,因爲我在多個塊之間使用多個內核,並且塊內的線程根本不相互作用。
我發現cudaMallocPitch()只返回已成功分配的內存使用的音調。有沒有辦法給司機一個二維內存分配請求,只是要求它分配的音調?
我想搭建一個試驗/錯誤優化程序,但A,B,D和E之間維度的鏈接依賴性(CI計算先驗,因爲它沒有分配點距線性)使得這個解決方案很糟糕,它需要重新計算每個問題集。
有沒有人有更好的方法,可以讓我確定適合任意數量的設備內存的適當大小的中間數據集?
編輯:
爲中間體A的存儲器被重複使用,我的邊界的計算使得假設C + d + E + B >>初始+ A + B(它憑藉這樣的事實爲真甲& B是相同尺寸的1字節的字符,而C,d,E是整數)和這樣我只需要保證有一個爲B + C + d + E.
我只使用足夠的空間計算能力2.x卡可以用(Quadro 2000,Tesla C2075,GTX460)進行測試。
是間距分配浪費了多少內存?你有多少例子?是否足夠擔心?如果是這樣,你可能會考慮做線性內存分配,並將你的索引線性化。 (另外,從你的依賴鏈看起來中間A的內存可以被重用)。 – harrism 2012-02-29 01:20:11
我使用音調線性內存的主要原因是利用了合併,並且因爲一個進程中位置(X,Y)的結果在另一個進程中再次用於(X,Y)。 音高分配浪費了可變數量的內存。我認爲它保持到最接近的512字節,至少對於我正在測試的其中一張卡片。 – 2012-02-29 15:06:43
你不需要間距線性來合併。您只需要(在Fermi上)查找經線中每個負載的地址,並將其放入單個對齊的128字節高速緩存行中。你永遠不會跨進程獲得緩存重用,所以我不確定爲什麼這是相關的。也許我誤解了。如果經線內的局部性模式是2D(並且具有良好的空間局部性),那麼使用紋理訪問會更好,因爲硬件紋理緩存和紋理內存佈局針對2D局部性進行了優化,而間距線性仍然是,線性的。 – harrism 2012-02-29 23:16:19