2013-03-04 88 views
6

我是OpenCL的新手。但是,我瞭解C/C++基礎知識和OOP。 我的問題如下:是否可以並行運行總計算任務?這在理論上是可能的嗎?下面我將介紹我所試圖做的:是否可以在OpenCL中並行運行求和計算?

的任務,例如:

double* values = new double[1000]; //let's pretend it has some random values inside 
double sum = 0.0; 

for(int i = 0; i < 1000; i++) { 
    sum += values[i]; 
} 

我試過的OpenCL內核做(我覺得這是錯誤的,因爲或許它訪問相同的「總和」變量從不同的線程/任務同時):

__kernel void calculate2dim(__global float* vectors1dim, 
          __global float output, 
          const unsigned int count) { 
    int i = get_global_id(0); 
    output += vectors1dim[i]; 
} 

此代碼是錯誤的。如果有人回答我,如果理論上可以並行運行這些任務,並且如果是這樣的話,我將非常感激。

+5

這是一個典型的下降問題。看看[這裏](http://developer.download.nvidia.com/compute/cuda/1.1-Beta/x86_website/projects/reduction/doc/reduction.pdf)一步一步解釋優化這個過程,核心架構(它是CUDA,但是原理完全一樣,除了關於模板的部分可能)。雖然有關該主題的更多介紹性材料可能會更有幫助,但我會將其留給適當的答案。 – 2013-03-04 11:43:55

+0

非常感謝!現在我知道這是一個普遍的問題,並且會很快解決它! – Vladimir 2013-03-04 12:28:20

回答

0

如果您想以並行方式對數組的值進行求和,則應確保減少爭用並確保跨線程無數據依賴關係。

數據依賴性將導致線程必須等待對方,造成爭用,這是你想避免得到真正的並行化。

你可以做到這一點的一種方法是將你的數組拆分成N個數組,每個數組包含你原始數組的一些子部分,然後用每個不同的數組調用你的OpenCL內核函數。

最後,當所有內核都完成了艱苦的工作後,您可以將每個數組的結果總結爲一個。該操作可以通過CPU輕鬆完成。

關鍵是在每個內核中完成的計算之間沒有任何依賴關係,因此您必須拆分數據並進行相應的處理。

我不知道你的數據是否與你的問題有任何實際的依賴關係,但這是你要弄清楚的。

+0

謝謝你的回答。可能,我應該將我的陣列分成幾個獨立的陣列!你知道任何方式將二維數組(如double [] [])傳遞給內核嗎?因爲pointet-to-pointer不能用作函數參數。 – Vladimir 2013-03-04 12:21:32

+1

你不需要傳遞二維數組,只需傳遞你的緩衝區並像這樣myarr [y * WIDTH + x]就可以了。 – alariq 2013-03-06 15:50:40

0

我提供的供參考的一段代碼應該可以完成這項工作。

E.g.您有N元素,並且您的工作組的大小爲WS = 64。我認爲N的倍數2 * WS(這很重要,一個工作組計算2 * WS元素的總和)。然後,你需要運行內核指定:

globalSizeX = 2*WS*(N/(2*WS)); 

結果總和陣列將擁有2 * WS元素的部分和。 (例如,sum [1] - 將包含索引爲2 * WS4 * WS-1)的元素之和。

如果你的globalSizeX是2 * WS或更少(這意味着你只有一個工作組),那麼你就完成了。只需使用sum [0]作爲結果。 如果不是 - 您需要重複此過程,此時使用總和數組作爲輸入數組並輸出到其他數組(在它們之間創建2個數組和乒乓)。等等,直到你將只有一個工作組。

還搜索Hilli Steele/Blelloch並行算法。 This文章可能還有

下面是實際的例子有用:

__kernel void par_sum(__global unsigned int* input, __global unsigned int* sum) 
{ 
    int li = get_local_id(0); 
    int groupId = get_group_id(0); 

    __local int our_h[2 * get_group_size(0)]; 
    our_h[2*li + 0] = hist[2*get_group_size(0)*blockId + 2*li + 0]; 
    our_h[2*li + 1] = hist[2*get_group_size(0)*blockId + 2*li + 1]; 

    // sweep up 
    int width = 2; 
    int num_el = 2*get_group_size(0)/width; 
    int wby2 = width>>1; 

    for(int i = 2*BLK_SIZ>>1; i>0; i>>=1) 
    { 

     barrier(CLK_LOCL_MEM_FENCE); 

     if(li < num_el) 
     { 
      int idx = width*(li+1) - 1; 
      our_h[idx] = our_h[idx] + our_h[(idx - wby2)]; 
     } 

     width<<=1; 
     wby2 = width>>1; 
     num_el>>=1; 
    } 

     barrier(CLK_LOCL_MEM_FENCE); 

    // down-sweep 
    if(0 == li) 
     sum[groupId] = our_h[2*get_group_size(0)-1]; // save sum 
} 
相關問題