-1

最後の Xmins を調べて、float[] 内のすべての値の平均を計算するカーネルがある場合、すべてのスレッドが同じコード行を同時に実行していないと、パフォーマンスが低下しますか?

例: @ x=1500 とします。過去 2 時間に 500 個のデータ ポイントがあります。

@ x = 1510、過去 2 時間に 300 個のデータ ポイントがあります。

x = 1500 のスレッドは 500 の場所をさかのぼる必要がありますが、x = 1510 のスレッドは 300 しかさかのぼらないため、後のスレッドは最初のスレッドが終了する前に次の位置に移動します。

これは通常問題ですか?

編集:コード例。申し訳ありませんが、CUDAfy.net を使用する予定だったので C# です。うまくいけば、実行する必要があるプログラミング構造のタイプの大まかなアイデアが得られます (実際のコードはより複雑ですが、同様の構造です)。これが GPU / コプロセッサに適しているのか、それとも単に CPU に適しているのかについてのコメントをいただければ幸いです。

public void PopulateMeanArray(float[] data)
{
    float lookFwdDistance = 108000000000f;
    float lookBkDistance = 12000000000f;
    int counter = thread.blockIdx.x * 1000;    //Ensures unique position in data is written to (assuming i have less than 1000 entries).
    float numberOfTicksInLookBack = 0;
    float sum = 0;    //Stores the sum of difference between two time ticks during x min look back.

    //Note:Time difference between each time tick is not consistent, therefore different value of numberOfTicksInLookBack at each position.
    //Thread 1 could be working here.
    for (float tickPosition = SDS.tick[thread.blockIdx.x]; SDS.tick[tickPosition] < SDS.tick[(tickPosition + lookFwdDistance)]; tickPosition++)
    {
        sum = 0;
        numberOfTicksInLookBack = 0;

        //Thread 2 could be working here. Is this warp divergence?
        for(float pastPosition = tickPosition - 1; SDS.tick[pastPosition] > (SDS.tick[tickPosition - lookBkDistance]); pastPosition--)
        {
            sum += SDS.tick[pastPosition] - SDS.tick[pastPosition + 1];
            numberOfTicksInLookBack++;
        }
        data[counter] = sum/numberOfTicksInLookBack;
        counter++;
    }
}
4

1 に答える 1

2

CUDA は、ワープと呼ばれるグループでスレッドを実行します。これまでに実装されたすべての CUDA アーキテクチャ (計算能力 3.5 まで) では、ワープのサイズは 32 スレッドです。異なるワープのスレッドのみが、コード内の異なる場所に実際に配置できます。ワープ内では、スレッドは常に同じ場所にあります。そのコードが実行されると、特定の場所でコードを実行してはならないスレッドはすべて無効になります。無効化されたスレッドは、ワープ内のスペースを占有するだけであり、対応する処理サイクルが失われます。

あなたのアルゴリズムでは、内側のループの終了条件がワープ内のすべてのスレッドに対して同時に満たされないため、ワープ発散が発生します。GPU は、ワープ内のすべてのスレッドの終了条件が満たされるまで、内側のループを実行し続ける必要があります。ワープ内のより多くのスレッドが終了条件に達すると、それらはマシンによって無効になり、処理サイクルが失われます。

状況によっては、無効なスレッドがメモリ要求を発行しないため、失われた処理サイクルがパフォーマンスに影響を与えないことがあります。これは、アルゴリズムがメモリにバインドされており、無効化されたスレッドが必要とするメモリが、ワープ内の他のスレッドの 1 つによって行われた読み取りに含まれていない場合です。ただし、あなたの場合、データはアクセスが結合されるように配置されているため (これは良いことです)、無効なスレッドでパフォーマンスが低下します。

あなたのアルゴリズムは非常に単純で、現状では、アルゴリズムは GPU にうまく適合しません。ただし、並列削減で使用されるアプローチに似たアプローチを使用する別のアルゴリズムを使用すると、CPU と GPU の両方で同じ計算を劇的に高速化できると思います。それが具体的にどのように行われるかは考えていません。

CPU の速度が大幅に向上する可能性がある簡単な方法は、内側のループが後方ではなく前方に反復するようにアルゴリズムを変更することです。これは、CPU がプリフェッチをキャッシュするためです。これらは、データを繰り返し転送する場合にのみ機能します。

于 2013-06-11T14:39:02.150 に答える