HPC:nVidia、AMD、Sony Cell、x86

nVidia CUDAAMD BrookIBM / Sony Cellなど、神話的な言葉について多くの人が聞いています。噂を聞いただけで、プログラムの実行が何百倍も速くなります。 それらが何であるか、この魔法の高性能プログラミングが一般的にどのように見えるか、そして古き良きx86プロセッサーと比較してどのようなゲインを提供できるかを理解してみましょう。





私は最近、高性能コンピューティングに夢中になりました(言語を破ることができます、HPCは高性能コンピューティングの略です)、さまざまなハードウェアプラットフォーム、つまりnVidia CUDA、AMD Brook +、Sony / Toshiba / IBM Cell (PS3の主力)、比較のために-Core2Quad / i7、AMD Phenom / Phenom IIに代表される老人x86。 パフォーマンスに影響を与える多くのニュアンスがあります-さまざまなレベルのキャッシュのサイズとレイテンシー、さまざまなタイプのチームの互換性、そして1つの記事ではうまくいかないことをすぐに注意したいので、今私は一般的な用語で状況を示します。



会話はすべて約32ビットの数字です。 64ビットの数値を評価すると、GTX285のパフォーマンスは約10〜12倍、AMD 4870は4倍、x86とCellは2倍低下します。



結果をすぐにシードに表示します。



... nVidia GTX285 AMD 4870 ソニーセル X86 i7 3.2Ghz
リアルスピード、GFLOP * 500 600 153 153
結晶面積 470mm 2 256mm 2 120mm 2 263mm 2
プロセス技術 55nm 55nm 65nm 45nm
* GFLOP-1秒あたり数十億回の操作



PS:45nmのセルはまだあるはずです

起動しなかった、とソーニャでは使用されていません

さようなら。



nVidia CUDA



nVidiaは、おそらくその技術を宣伝することに最も力を入れている人です。 魔法の言葉CUDAを聞いたのは、それが何であるかを試すことに決めたずっと前です。 CUDA-これは、ビデオカード内のSIMDプロセッサによる実行のために特別に記述されたプログラムをコンパイルするC ++コンパイラ(かなり単純で安定した)です。 たとえば、GTX285には240個のプロセッサがあり、1476Mhzで1クロックあたり1〜3(乗算+加算+特殊、グラフィックシェーダ2〜3は実際ですが、通常の計算では1.5に焦点を当てる方がよい) (1秒あたり1兆回の操作)、約500 GFLOPの実際に達成可能な速度。



プログラムは次のようになります。



__device__ unsigned int * data_d;

__global__ void gpu_thing(int somevalue)

{

const

int ix = blockDim.x * blockIdx.x + threadIdx.x;



data_d [ix]

* = somevalue;

}



次のようにGPUで実行されます。



dim3

スレッド(256);

dim3グリッド(256);

int

src [256 * 256]、結果[256 * 256]; //ビデオカードのメモリにデータをコピーします

cudaMemcpyAsync(data_d、src、sizeof(int)* 256 * 256、cudaMemcpyHostToDevice、0); //ビデオカードで実行、256 * 256スレッド

同時に

gpu_thing <<< grid、threads >>>(125); //メモリから結果を取得します

cudaMemcpyAsync(result、data_d、sizeof(int)* 256 * 256、cudaMemcpyDeviceToHost、

NULL);



制限事項:

スレッドは16のグループで実行されます。すべての命令は同じでなければなりません(そうでない場合、パフォーマンスが低下します。コンパイラ/ハードウェアはこれを監視します)。



ストリームの開始時間はもちろん1ミリ秒ではありませんが、顕著であるため、小さなタスクをビデオカードに移行しない方が良いでしょう。 プログラムの実行中

モニターが接続されているビデオカードでは、マウスカーソルが移動しない場合でも、ウィンドウは画面を更新できません。ユーザーが悪くないように、作業量を10〜20ミリ秒選択することで問題が解決します。 カーネルの動作時間が5秒を超える場合、Windowsはドライバーがフリーズして再起動することを決定します-結果はプログラムにとって嘆かわしいものです(これは、モニターがビデオカードに接続されている場合にのみ当てはまります):-) LinuxおよびMacOSがサポートされています。 複数のビデオカードのサポート-SLIモードが無効になっている場合にのみ機能します。 8400GS以降のすべてのビデオカードをサポートします(ただし、通常はプロセッサよりも遅いことが判明しています)。 正式には、さまざまな種類のビデオカードが機能する必要はありませんが、実際には組み合わせて機能します(タスクを手動で分離する必要があります)。



AMD Brook +



AMDは鉄で最善を尽くしました-その4870には160のプロセッサがあり、それぞれが750 Mhzの周波数でクロックサイクルごとに5つの操作(小さな制限付き)を実行できます。 理論的および実用的なパフォーマンス-600GFlop。 AMDファン-次の段落は読んでいないが、気をつけてください-実際のアプリケーション(書くことができる場合)では、4870はGTX285を10-20%にします。 さて、今悪い点:Cブルック+ただの熊手。 彼と仕事をするには、とても落ち着いた人である必要があります。 Cコンパイラ(++なし)、したがって、長い間忘れられていた制限(関数本体の先頭でのみローカル変数を定義するなど)に再び精通します。完全になじみのないものです。ローカル配列はサポートされません。通常、それをコンパイルします(クラッシュ/問題

間違ったコード-これはStream SDK 1.3に当てはまります。 一般に、バージョンが1.3と呼ばれ、0.3alphaのように機能することを準備します:-D VLIWアーキテクチャがあり、ループを展開できると便利ですが(たとえば、Intel C ++-プラグマアンロール(5)など)、これはできません。

はい、プリプロセッサもありません、自分でねじ込みます。 まあ、いくつかのビデオカードを使用できないと、すべて(または、1つのプロセス(スレッドではなく)-1つのビデオカード)がすべて完了します。 低レベルのプログラミング(マシンコードではありませんが、ほとんど)-AMD CALにクレジットを与える必要があります-問題はほとんどありませんが、欠点は明らかです。 プログラムは次のようになります。



カーネルvoid gpu_thing(int input <>、

int somevalue、out int出力<>)

{



出力

= somevalue * input;



unsigned int md5_dim [] = {800 * 256};

:: brook :: Stream <int>入力(1、サイズ);

::小川::ストリーム<int>結果(1、サイズ);



int src [800 * 256]、dst [800 * 256];



input.read(src); //ビデオカードのデータを準備します

hello_brook_check(入力、45、出力); // 800 * 256スレッドを実行

output.write(dst);



制限:起動時間も非常に長く、マウスカーソルを除いてすべてがフリーズします。また、Windowsでの最大実行時間は5秒です。

Ironは、HD 2xxx(制限付き)、HD3xxxおよび4xxxでサポートされています。



ソニーセル



ソニーセルのパフォーマンスには多くの人が興奮しています。 もちろん、x86が進歩したことを忘れると印象的です(グラフィックカードは言うまでもありません)。 セルは、PPEと呼ばれる通常のPowerコア(前述のモンスターと比較してパフォーマンスが低い)と8つのSPEコアで構成され、そのうち1つは潜在的に欠陥があります(拒否されていないチップは、はるかに大きく販売されています

お金)、1はオペレーティングシステムによって予約されています(Linuxでプレフィックスを変更した場合、7にアクセスできますが、これはエンドユーザーにとって困難です;-))、合計6です。周波数ごとに6コア(2 * 4)があります3.2Ghz、総実数を提供します

パフォーマンス153.6 GFLOP。 プログラミングでは、私はあなたに教えません、なぜなら これらの153のGFLOPは使用するのが非常に非実用的であるため、

試してみました(プログラムのリリースに関しては国内消費向けではありません)。



x86



すでにx86を落としましたか? しかし、それはまだ古さによって揺れる可能性があります:i7 / Core2Quadでは、4つのコアのそれぞれがサイクルごとに4つの数で3つのSSE2操作を実行でき、合計でサイクルあたり48回の操作を実行できます、VLIW / RISCへの移行を開始しました。これにより、153.6 GFLOP(ops、Cellとまったく同じ)Phenom / PhenomIIも使用できますが、特殊なSSE2モジュール(1つの乗算、1つの加算、1つのユニバーサル)があります。パフォーマンスは20〜40%低下します。 クロックごとに48の操作を使用することが難しいと思われる場合は、次の例をご覧ください。



intデータ[1024 * 1024 * 12];

const int value1 = 123;

for(int i = 0; i <1024 * 1024 * 12; i ++)data [i] ^ = value1;



これをIntel C ++でコンパイルし、4つのスレッドで実行する場合(ただし、現在では、単独で実行するスレッドを推測することもできます)-操作をSSE2にとって便利なグループに結合します(すべての人にとって常にスムーズに進むとは限らないことを認識する価値があります) #プラグマまたはSSE組み込み関数を使用しますが、これは別の記事のトピックです)



どのような結論を出すことができますか? もちろん、ビデオカードは現在、かなり狭いカテゴリのタスク(特にデュアルチップタスク)で驚異的なパフォーマンスを示しており、通常のx86プロセッサの優れたコードよりも約8倍高速です。 8宣伝されているように30〜100にはほど遠いが、それでも時には不可能を可能にするのに十分である。 さらに、4つのビデオカード用のマザーボードは、4つのプロセッサ用のマザーボードよりもはるかに安価です。 PS3のCellのパフォーマンスに関する多幸感は、シリアルx86プロセッサーがパフォーマンスに追いついたため、すでに落ち着いているはずです。



また、habrパーサーは<code>タグを有効に表示したくないため、削除します。 :(



All Articles