1

以下のようなベンチマークを実行しています

CHECK( context = clCreateContext(props, 1, &device, NULL, NULL, &_err); );
CHECK( queue = clCreateCommandQueue(context, device, 0, &_err); );
#define SYNC() clFinish(queue)
#define LAUNCH(glob, loc, kernel) OCL(clEnqueueNDRangeKernel(queue, kernel, 2,\
                                                             NULL, glob, loc,\
                                                             0, NULL, NULL))

/* Build program, set arguments over here */


START;
for (int i = 0; i < iter; i++) {
    LAUNCH(global, local, plus_kernel);
}
SYNC();
STOP;
printf("Time taken (plus) : %lf\n", uSec / iter);

START;
for (int i = 0; i < iter; i++) {
    LAUNCH(global, local, minus_kernel);
}
SYNC();
STOP;
printf("Time taken (minus): %lf\n", uSec / iter);

START;
for (int i = 0; i < iter; i++) {
    LAUNCH(global, local, plus_kernel);
    LAUNCH(global, local, minus_kernel);
}
SYNC();
STOP;
printf("Time taken (both) : %lf\n", uSec / iter);

結果は奇妙に見えます：

Time taken (plus) : 31.450000
Time taken (minus): 28.120000
Time taken (both) : 2256.380000

START、およびSTOPは、タイマーを開始および停止する単なるマクロです。関連するマクロは次のとおりです。

なぜキューイングするのがカーネルの速度を低下させているのかわかりません（そしてAMD GPUでのみ）！

編集私はRadeon7970を使用しています

編集両方のカーネルは独立したメモリで動作しています。また、ここにシステム情報があります。

OS：Ubuntu 11.10

fglrxinfo：

display: :0  screen: 0
OpenGL vendor string: Advanced Micro Devices, Inc.
OpenGL renderer string: AMD Radeon HD 7900 Series 
OpenGL version string: 4.2.11762 Compatibility Profile Context
4

1 に答える 1

1

その答えは、新しい GPU (特に、Graphics Compute Next (GCN)アーキテクチャを使用する Radeon 7970) でのデータのキャッシュに関係していると思います。

このアーキテクチャの利点の 1 つは、キャッシング機能です (現時点では、CPU キャッシングにやや近い)。次のような呼び出しを実行する場合:

PLUS
PLUS 
PLUS
....

次に、GPU の内部キャッシュに常駐するメモリ。一方、次のような呼び出しを行う場合:

PLUS
MINUS
PLUS 
MINUS
...

2 つのカーネルに異なるメモリ オブジェクトが関連付けられている場合、データは各 CU のハードウェア デバイスから追い出されるため、非常に低速なグローバル メモリからデータを取り込む必要があります。

これが当てはまるかどうかをテストする2つの簡単な方法：

  1. さまざまな反復回数で Pluses のみを実行します。反復回数が増えると、最初の実行 (データを取り込む) のコストが償却されるため、平均時間は短くなります。また、最初の呼び出し以降のすべての呼び出しは、比較的均等である必要があることに注意してください。

  2. Plus カーネルと Minus カーネルを同じメモリ オブジェクトで実行します。スローダウンの理由がメモリ オブジェクトのキャッシュにある場合、全体の実行時間は PLUS と MINUS の個々の実行時間の平均になるはずです (おそらく実験 1 による)。

これが事実かどうかわかったら教えてください！

于 2012-09-20T22:11:50.290 に答える