Cuda-OpenCL CPU比OpenCL或CUDA GPU版本快4倍

我一直在用C＃+ Cudafy（C＃ - > CUDA或OpenCL translator）工作的波形模擬器效果很好，除了運行OpenCL CPU版本（Intel驅動，15「MacBook Pro Retina i7 2.7GHz，GeForce 650M（開普勒，384內核））大約是GPU版本的四倍。。後端OpenCL的GPU和CUDA版本執行幾乎相同）Cuda-OpenCL CPU比OpenCL或CUDA GPU版本快4倍

爲了澄清，用於樣品的問題：

OpenCL的CPU 1200赫茲
的OpenCL GPU 320赫茲
CUDA GPU - 〜330赫茲

我茫然地解釋爲什麼CPU版本將更快比GPU。在這種情況下，CPU和GPU上執行的內核代碼（在CL情況下）是相同的。我在初始化期間選擇CPU或GPU設備，但除此之外，所有內容都是相同的。

編輯

下面是啓動內核之一的C＃代碼。（其他是非常相似的。）

public override void UpdateEz(Source source, float Time, float ca, float cb) 
    { 
     var blockSize = new dim3(1); 
     var gridSize = new dim3(_gpuEz.Field.GetLength(0),_gpuEz.Field.GetLength(1)); 

     Gpu.Launch(gridSize, blockSize) 
      .CudaUpdateEz(
       Time 
       , ca 
       , cb 
       , source.Position.X 
       , source.Position.Y 
       , source.Value 
       , _gpuHx.Field 
       , _gpuHy.Field 
       , _gpuEz.Field 
      ); 

    }

而且，這裏是由Cudafy產生的相關CUDA內核函數：

extern "C" __global__ void CudaUpdateEz(float time, float ca, float cb, int sourceX, int sourceY, float sourceValue, float* hx, int hxLen0, int hxLen1, float* hy, int hyLen0, int hyLen1, float* ez, int ezLen0, int ezLen1) 
{ 
    int x = blockIdx.x; 
    int y = blockIdx.y; 
    if (x > 0 && x < ezLen0 - 1 && y > 0 && y < ezLen1 - 1) 
    { 
     ez[(x) * ezLen1 + (y)] = ca * ez[(x) * ezLen1 + (y)] + cb * (hy[(x) * hyLen1 + (y)] - hy[(x - 1) * hyLen1 + (y)]) - cb * (hx[(x) * hxLen1 + (y)] - hx[(x) * hxLen1 + (y - 1)]); 
    } 
    if (x == sourceX && y == sourceY) 
    { 
     ez[(x) * ezLen1 + (y)] += sourceValue; 
    } 
}

只是爲了完整性，這裏是用來生成CUDA的C＃：

[Cudafy] 
    public static void CudaUpdateEz(
     GThread thread 
     , float time 
     , float ca 
     , float cb 
     , int sourceX 
     , int sourceY 
     , float sourceValue 
     , float[,] hx 
     , float[,] hy 
     , float[,] ez 
     ) 
    { 
     var i = thread.blockIdx.x; 
     var j = thread.blockIdx.y; 

     if (i > 0 && i < ez.GetLength(0) - 1 && j > 0 && j < ez.GetLength(1) - 1) 
      ez[i, j] = 
       ca * ez[i, j] 
       + 
       cb * (hy[i, j] - hy[i - 1, j]) 
       - 
       cb * (hx[i, j] - hx[i, j - 1]) 
       ; 

     if (i == sourceX && j == sourceY) 
      ez[i, j] += sourceValue; 
    }

顯然，在這種內核的if是壞的，但即使所產生的流水線停頓不會引起這種極端的性能三角洲。

跳到我身上的唯一另外一件事是我使用了一個不成熟的網格/塊分配方案 - 即網格是要更新的數組的大小，並且每個塊都是一個線程。我確信這會對性能產生一些影響，但我看不到它導致它在CPU上運行的CL代碼速度的四分之一。哎呀！

來源

2013-05-07 David Lively

你有一些你可以共享的代碼示例嗎？ – 2013-05-07 23:13:53

@EricBainville當然 - 你想要C＃，CUDA或CL內核，還是什麼？（這是一個半中型應用程序，我不想將20k行代碼粘貼到SO中） – 2013-05-07 23:16:03

+10

我沒有看到任何跡象表明cuda內核每塊使用多於1個線程（沒有使用'threadIdx.x'或'threadIdx.y'）。此外，啓動每塊指定1個線程。這意味着大約97％的GPU功能未被使用。我對cudafy瞭解不多，所以我不知道你是否可以控制這個，但我並不覺得cuda代碼的運行速度並不快。 – 2013-05-08 00:08:10

回答這個得到它的解答列表。

的代碼發佈指示內核啓動被指定1（活性）螺紋的threadblock。這不是編寫快速GPU代碼的方式，因爲它會讓大部分GPU功能閒置。

典型threadblock尺寸應爲每塊的至少128個線程，以及更高的是往往更好，以32的倍數，高達512或1024每塊的限制，這取決於GPU。

GPU的「喜歡」具有「可用」大量的並行工作來隱藏延遲。爲每個塊指定更多線程有助於實現此目標。（具有在網格中的相當大的數目的threadblocks也可能有幫助。）

此外，該GPU中的32組執行線程。每個塊只指定一個線程或32的非倍數會在執行的每個線程塊中留下一些空閒執行槽。每塊1個線程特別糟糕。

來源

2014-06-27 23:01:58

Cuda-OpenCL CPU比OpenCL或CUDA GPU版本快4倍

回答

相關問題