2016-02-05 13 views
7

これは私の最初の投稿です。私はあなたの時間を評価するので、私はそれを短く保つようにしようとします。このコミュニティは私にとって信じられないほどのものでした。この並列処理はOpenCLで実装できます

私はOpenCLを学習しており、以下のアルゴリズムから少しの並列性を抽出したいと考えています。私は私が作業している部分だけを見せてくれます。私はできるだけ単純化しました。

1)入力:長さ(n)の2つの1D配列:A、B、およびnの値。また、値C [0]、D [0]。

2)出力:長さの2つの1Dアレイ(N):C、D.

C[i] = function1(C[i-1]) 
D[i] = function2(C[i-1],D[i-1]) 

したがって、これらは、再帰的定義され、所与のI値しかしながら計算のC & Dがで行うことができ(彼らは明らかにもっと複雑であり、意味を成す)。反復ごとに同期して、一つの初期の比較を行い、その後、それぞれのループを入力します。理想的には

__kernel void test (__global float* A, __global float* B, __global float* C, 
        __global float* D, int n, float C0, float D0) { 
    int i, j=get_global_id(0); 

    if (j==0) { 
     C[0] = C0; 
     for (i=1;i<=n-1;i++) { 
      C[i] = function1(C[i-1]); 
      [WAIT FOR W.I. 1 TO FINISH CALCULATING D[i]]; 
     } 
     return; 
    } 
    else { 
     D[0] = D0; 
     for (i=1;i<=n-1;i++) { 
      D[i] = function2(C[i-1],D[i-1]); 
      [WAIT FOR W.I. 0 TO FINISH CALCULATING C[i]]; 
     } 
     return; 
    } 
} 

2つの作業項目(番号0,1)のそれぞれを:素朴な考えは以下のカーネル用の2つの作業項目を作成することになります。今度はGPUのSIMD実装を考えると、これはうまくいかない(作業項目はすべてのカーネルコードを待っている)と想定していますが、このタイプの作業を2つのCPUコアに割り当てて期待どおりに動作させることは可能ですか?この場合の障壁は何ですか?

+1

CとDのすべての値を保存する必要がありますか、または最終結果のみを気にしますか? – mfa

+1

'function1'と' function2'を定義してください。 –

答えて

0

あなたのケースの依存関係は完全に線形/再帰的です(i-1が必要です)。他の問題(還元、合計、並べ替えなど)と同じようにlogaritmicさえありません。したがって、この問題はSIMDデバイスにはあまり適していません。

あなたができることは、CPUの2スレッドアプローチになります。スレッド1、スレッド2

例えば非常に単純なアプローチのために、データ(C値)「を生成する」であろう:atomic_incatomic_decは、例えばセマフォをカウントして実施することが可能で

Thread 1: 
for(){ 
    ProcessC(i); 
    atomic_inc(counter); //This function should unlock 
} 

Thread 2: 
for(){ 
    atomic_dec(counter); //This function should lock 
    ProcessD(i); 
} 

+0

また、すべてのグローバル作業項目が一緒に実行されるわけではありません(次のセットが完了する前に完全に終了しなければならないブロックで実行される可能性がある)ので、待機はオプションではありません。 – Dithermaster

+0

私のカーネルは基本的にあなたが言及した原子的機能を除いて同じままですか? (私はまだopenCL同期について読んでいるので、これらの関数が何をしているのか把握することはできません)。そして、CPU上で実行すると、コードは期待通りにifステートメントで分岐するでしょうか? – user5887077

1

これはopenclで実装できますが、他の答えのように、最高で2つのスレッドに制限されます。

私のバージョンの関数は、2つの作業項目を持つ1つの作業グループで呼び出す必要があります。

__kernel void test (__global float* A, __global float* B, __global float* C, __global float* D, int n, float C0, float D0) 
{ 
    int i; 
    int gid = get_global_id(0); 

    local float prevC; 
    local float prevD; 

    if (gid == 0) { 
     C[0] = prevC = C0; 
     D[0] = prevD = D0; 
    } 

    barrier(CLK_LOCAL_MEM_FENCE); 

    for (i=1;i<=n-1;i++) { 
     if(gid == 0){ 
      C[i] = function1(prevC); 
     }else if (gid == 1){ 
      D[i] = function2(prevC, prevD); 
     } 

     barrier(CLK_LOCAL_MEM_FENCE); 
     prevC = C[i]; 
     prevD = D[i]; 
    } 
} 

これは任意のopenclハードウェアで実行する必要があります。 CとDの値をすべて保存することに気にしない場合、prevCとprevDをリスト全体ではなく2つのfloatで返すことができます。これはまた、中間値のすべての読出しおよび書込みのために、より低いキャッシュレベル(すなわちローカルメモリ)に固着することにより、これをはるかに速くする。ローカルメモリの増強は、すべてのopenclハードウェアにも適用する必要があります。

これはGPUで実行するのにポイントがありますか?並列性ではありません。あなたは2つのスレッドで立ち往生しています。しかし、CとDのすべての値を返す必要がない場合は、GPUのメモリが非常に高速であるため、おそらく大幅なスピードアップが発生します。

このすべては、function1とfunction2があまりにも複雑ではないことを前提としています。そうであれば、CPUに固執するだけでなく、おそらくOpenMPなどの別のマルチプロセッシング手法を採用しています。

関連する問題