CUDA中的複數/ cuComplex算術

Question

我對CUDA相當陌生，想了解更多關於複數算術及其速度的信息。

我需要解決所有元素下面的複數公式中的「f] []」陣列和存儲在「答[]」的答案：

Ans [0] = (2.0/((20.5*(j[0]*j[0]))+(5.55*j[0])+20)); 
Ans [1] = (2.0/((20.5*(j[1]*j[1]))+(5.55*j[1])+20)); 
... 
... 
... 
Ans [n] = (2.0/((20.5*(j[n]*j[n]))+(5.55*j[n])+20)); 

由於我需要執行相同的計算到'j'的所有元素，我可以並行化這些代碼，並讓每個線程/塊處理每個計算（blockIdx.x = 0 - > Ans [0]等）。 根據我的理解，如果我這樣做平行的很多元素我應該能夠看到速度的增加。但是，可以用一行C++代碼編寫的代碼需要幾行代碼才能在GPU中執行。

我的問題是，所有額外的代碼行意味着更長的處理時間，因爲它涉及到將許多臨時值保存在中間值。如果是這樣，當元素數量少於1000時，在GPU中進行這種計算還是有意義的嗎？ （任意數）

的公式：

C++ -> Ans [0] = (2.0/((20.5*(j[0]*j[0]))+(5.55*j[0])+20)); 

我對它的GPU版本：

int tid = blockIdx.x; 

    temp1[tid] = cuCmul(j[tid], j[tid]); 
    temp2[tid] = cuCmul(temp1[tid], make_cuDoubleComplex(20.5, 0)); 
    temp3[tid] = cuCmul(j[tid], make_cuDoubleComplex(5.55, 0)); 
    temp4[tid] = cuCadd(temp2[tid], temp3[tid]); 
    temp5[tid] = cuCadd(temp4[tid], make_cuDoubleComplex(20, 0)); 
    Ans[tid] = cuCdiv(make_cuDoubleComplex(2.0, 0), temp5[tid]); 

另外，請讓我知道是否有寫這爲GPU更有效的方法

Answer 1

什麼可以寫在一行C++代碼需要在GPU中做幾行。

這可能是不正確的，至少對於你已經顯示的例子。您似乎擔心臨時存儲，但編譯器（主機和GPU）對於確定臨時存儲是否有意義以及是否優化它們都非常好。許多程序員陷入了陷阱，他們認爲他們編寫的C代碼很好地表示了機器將在存儲使用情況和操作順序方面做什麼，但是對於現代編譯器來說，通常情況並非如此。

舉個例子，你說這是你的CPU代碼：

Ans [0] = (2.0/((20.5*(j[0]*j[0]))+(5.55*j[0])+20)); 

的GPU版本可以寫成：

Ans [0] = cuCdiv(make_cuDoubleComplex(2.0, 0), cuCadd(cuCadd(cuCmul(cuCmul(j[tid], j[tid]), make_cuDoubleComplex(20.5, 0)), cuCmul(j[tid], make_cuDoubleComplex(5.55, 0))), make_cuDoubleComplex(20, 0))); 

作出任何明確的臨時存儲的使用。 （然而，代碼當然很難閱讀。）但是，主機（C）或設備（GPU）情況下的「底層」情況可能看起來不同。在編寫如何優化一行或幾行代碼時，編譯器通常比程序員更好。

讓您的代碼先工作。然後基準（時間）它。然後決定是否要仔細研究優化。諸如visual profiler之類的工具可以幫助您發現優化機會。

儘管您的主機C代碼看起來很簡單，但請記住，一個複雜的數字仍然有2個數量與它關聯。儘管這看起來並不明顯，但是在「抽象的」C代碼中，編譯器仍在進行必要的操作，以分別處理各個數字，以適合各種操作：+， - ，*，/

我的問題是，所有額外的代碼行意味着更長的處理時間，因爲它涉及到在許多臨時數據中保存中間值。

不一定，因爲我上面描述的原因。你在做任何實現時都會做同樣的工作，編譯器會觀察這個並可能生成類似的機器代碼。

如果是這樣，當元素數量少於1000時，在GPU中進行這種計算還是有意義的嗎？ （任意數字）

如果您計算的答案總數大約是1000，那麼對於現代GPU而言，您的問題「非常小」。現代GPU可能有8個（或更多）SM，每個SM能夠同時運行1到3個線程（32個線程），並且該機器還需要相當穩定的「準備運行」的warp，以保持所有管線（內存，計算等）忙碌。 1000線程可能是實現GPU的體面利用的最低限度。顯然，這取決於你將在哪個GPU或GPU上運行。例如，筆記本電腦中的小型低端GPU可能能夠以更小的問題實現高利用率。但是，如果你的計算範圍是你在這裏顯示的1000個類型，我無法想象在CPU（主機代碼）上花費很多時間。

Answer 2

CUDA適用於C++的子集。支持的功能之一是超載操作員。

__device__ __host__ cuDoubleComplex operator*(cuDoubleComplex a, cuDoubleComplex b) { return cuCmul(a,b); } 
__device__ __host__ cuDoubleComplex operator+(cuDoubleComplex a, cuDoubleComplex b) { return cuCadd(a,b); } 
__device__ __host__ cuDoubleComplex operator/(cuDoubleComplex a, cuDoubleComplex b) { return cuCdiv(a,b); } 

如果其中一個輸入是double而不是cuDoubleComplex，那麼您也可以類似地重載運算符。

如果你在其他內核中沒有使用相同的操作，那麼最好繼續做你正在做的事情。但是，如果您正在處理大型項目，並且需要在其他內核中繼續使用類似的操作，那麼最好有一個包含所有這些重載操作符的頭文件。

我的問題是，做所有的代碼的其他行意味着更長的 處理時間，因爲它涉及到衆多 臨時工節省中間值。如果是這樣，當元素數量少於1000時，在GPU中進行這種計算 還是有意義的嗎？ （任意數）

編譯器通常應該生成用於同樣的操作相同數目的臨時變量無關的代碼行的。加速來自GPU完成的並行操作數量。在大約1000個元素中，單線程主機端實現應該能夠勝過只執行這些操作的CUDA內核。有參與，從主機的數據複製到設備，啓動內核，讀取和寫入全局內存等開銷

一個CUDA GPU啓用通常能夠同時運行數千個線程的。並且每個線程都應該具有相對較高的計算帶寬比率以最優地使用GPU。