3

OpenCL でアルゴリズムを実装しています。C++ で何度もループし、毎回同じ OpenCL カーネルを呼び出します。カーネルは、次の反復の入力データとこれらのデータの数を生成します。現在、次の 2 つの用途について、各ループでこの数値を読み返しています。

  1. この数を使用して、次のループに必要な作業項目の数を決定します。と
  2. この数値を使用して、ループを終了するタイミング (数値が 0 の場合) を決定します。

読み取りにはループのほとんどの時間がかかることがわかりました。それを回避する方法はありますか？

一般的に言えば、カーネルを繰り返し呼び出す必要があり、終了条件がカーネルによって生成された結果 (固定数のループではない) に依存している場合、どうすれば効率的に実行できますか? GPU から読み返す代わりに、OpenGL のオクルージョン クエリのようなものはありますか?

4

3 に答える 3

2

GPU カーネルから数値を読み取るには、常に 10 ～ 1000 マイクロ秒以上かかります。

制御数が常に減少している場合は、グローバル メモリに保持し、グローバル ID に対してテストして、カーネルが各反復で機能するかどうかを判断できます。グローバル メモリ バリアを使用して、すべてのスレッドを同期します ...

kernel void x(global int * the_number, constant int max_iterations, ... )
{
    int index = get_global_id(0);
    int count = 0;    // stops an infinite loop

     while( index  < the_number[0] && count < max_iterations )
     {
      count++;
      // loop code follows

      ....

      // Use one thread decide what to do next 
      if ( index  == 0 )
      {
          the_number[0] = ... next value
      }

      barrier( CLK_GLOBAL_MEM_FENCE ); //  Barrier to sync threads
    }
}
于 2012-07-20T20:16:12.257 に答える
1

ここにはいくつかのオプションがあります。

  1. 可能であれば、ループと条件をカーネルに移動できますか? 現在の繰り返しの入力に応じて、追加の作業項目が何もしないスキームを使用します。
  2. 1.が不可能な場合は、「決定」カーネルによって生成されたデータをバッファーに保存し、そのバッファーを使用して他のカーネルを「指示」することをお勧めします。

どちらのオプションでも、復唱をスキップできます。

于 2012-07-20T19:42:59.310 に答える
0

私はちょうどこの正確な問題に取り組まなければならなかったいくつかの研究を終えているところです！

私たちはいくつかのことを発見しました：

  1. 2つ（またはそれ以上）のバッファーを使用してください！カーネルの最初の反復がb1のデータで動作し、次の反復がb2で動作し、次に再びb1で動作するようにします。各カーネル呼び出しの間に、他のバッファーの結果を読み戻し、反復を停止する時間かどうかを確認します。カーネルが読み取りよりも時間がかかる場合に最適に機能します。プロファイリングツールを使用して、読み取りを待機していないことを確認します（待機している場合は、バッファーの数を増やします）。

  2. オーバーシュート！各カーネルに仕上げチェックを追加し、データをコピーして戻す前に数回（数百回）呼び出します。カーネルが低コストの場合、これは非常にうまく機能します。

于 2012-07-23T02:23:08.060 に答える