できるXeonPhi！
はじめに
XeonPhiハッカソンがあまりにも人が集まらないと嘆いていたら、
「そもそもXeonPhiに対する事前情報がなさ過ぎてプログラム書けるかどうかすら疑問」
という話をいろんな所から聞いたので、じゃあチュートリアル的なもの書きますねというのが今回の趣旨です
公式資料
Intelさんからの公式資料は以下にあります
http://software.intel.com/en-us/mic-developer
XeonPhiについて
Intelの開発したメニーコアアクセラレータボードのことです。
正式にはブランド名。現在のは第二世代MIC, Knights Cornerというもの。
PCIExpress Gen 2.0 x16スロットに接続し、PCIe経由でホストCPUとやり取りを行います
XeonPhiはいくつか種類がありますが、5110Pでは内部では60コア / 240スレッドが動作しています
次世代は既にいくつか情報が出ており、名称はKnights Landing。現在のようにPCIe経由での動作するカードに加え、
現在のXeonの代替、つまりLGAソケットに刺さるモデルも出るようです
アーキテクチャ
詳細はPCWatchの後藤さんの記事に詳しいです
http://pc.watch.impress.co.jp/docs/column/kaigai/20120910_558566.html
http://pc.watch.impress.co.jp/docs/column/kaigai/20120911_558738.html
言葉でさらっと言ってしまうと
プログラミングモデル
XeonPhiでコードを実行する基本的なモデルというのは3種類あります
- XeonPhiのみでコードを実行する
- ホストの特定の部位をXeonPhiにオフロードする
- XeonPhiとホストを協調動作させる
XeonPhiではLinuxが走っているため、1の場合はそのままログインしてXeonPhi上でプログラムを実行します
対して2では、CUDAやOpenCLのように特定の場所だけをXeonPhiに渡して実行してもらうことになります
最後の3は、XeonPhiの1スレッドを1プロセスのように扱うMPIのプログラミングモデルです
また、現時点では開発にICCが必要です。
現時点では、開発にICC、またはMPSS(Intel Manycore Platform Software Stack) を導入したgccが必要です。
ただし、gccの場合、XeonPhiの組み込み関数はサポートしていません
0710 修正
この記事では1,2について解説します
XeonPhiでのみコードを実行する
ここで重要なことは、XeonPhi上でのみコードを実行する場合、基本的にはコンパイルし直すだけで動作する、ということにあります
取り出したるは何の変哲もないただのHello world.
// hello_world_mic.cpp #include <iostream> int main() { std::cout << "Hello world" << std::endl; return 0; }
これを、
$ icpc hello_world_mic.cpp -mmic -o hello_world
とコンパイルすると、hello_worldというバイナリが出来るのはいつも通りですが、これはXeonPhi用のバイナリになっています
先に指定した-mmicオプションがXeonPhi用のオプションで、このコードを使って生成されたバイナリはXeonPhiで実行可能なものになります
出来上がったバイナリを
$ scp hello_world mic0:~/
とかやると、XeonPhiに転送されます
したら続いて
$ ssh mic0
でXeonPhiにログインしてからの
$ ./hello_world
でXeonPhiのHello worldが実行できました。はい
これだけだと味気ないんで、並列化して240のHello worldにしましょう
// hello_world_mic.cpp #include <iostream> int main() { #pragma omp parallel { std::cout << "Hello world" << std::endl; } return 0; }
変更点は、並列にしたいところをブロック化してOpenMPプラグマを突っ込んだぐらいです
$ icpc hello_world_mic.cpp -mmic -openmp -o hello_world
同様に、こちらの変更点もOpenMPをEnableにするオプションを追加したぐらいです
先ほどと同様に、hello_worldバイナリをXeonPhiに転送して実行するとアホっぽいぐらい画面が埋め尽くされたかと思います
XeonPhiのみでのコード実行は大体こんな感じ
ホストの特定の部位をXeonPhiにオフロードする
ここで重要なことは、ホストの特定の部位をオフロードする場合は、そのオフロードしたい場所をプラグマでくくってあげるだけでオフロードが可能、ということです
続いて取り出したのはこんな感じのvecAdd.cpp。ただ単に要素分足しているだけのコード。
// vecAdd.cpp #include <iostream> int main() { const size_t num = 1024; float* A = new float[num]; float* B = new float[num]; float* C = new float[num]; for(size_t i = 0; i < num; ++i) { A[i] = static_cast<float>(i); B[i] = static_cast<float>(i); C[i] = 0.0f; } // vector add for(size_t i = 0; i < num; ++i) { C[i] = A[i] + B[i]; } delete []A; delete []B; delete []C; }
vector add部分をXeonPhiに乗せたいと考えます
オフロードしたい箇所をプラグマで指定します
// vector add #pragma offload target(mic) \ in(A:length(num)) \ in(B:length(num)) \ out(C:length(num)) { for(size_t i = 0; i < num; ++i) { C[i] = A[i] + B[i]; } }
このようにすることによって、プラグマ以下のブロックがXeonPhi上で実行されます
また、target(mic)行ですが、target(mic:n)でXeonPhiの番号が指定できます。複数枚積んでいるケースではXeonPhiの指定が出来るのでそれなりに便利
その下にあるin, outは、それぞれデータの入力、出力を制御しています
in, outのみならず様々に制御構文があるのですが、それはIntelさんのドキュメントを参照してください
http://software.intel.com/sites/default/files/Beginning%20Intel%20Xeon%20Phi%20Coprocessor%20Workshop%20Offload%20Compiling%20Part%201.pdf
これです
このときのコンパイルは
$ icpc vecAdd.cpp -o vecAdd
です。先ほどまであった-mmicオプションは、今回はつけません
前節でもありましたが、XeonPhi上では最大240スレッドが動作します
手っ取り早くそのスレッドを扱うには、OpenMPを使用することが挙げられます
つまり、先ほどのコードはさらにOpenMP対応も加えて
// vector add #pragma offload target(mic) \ in(A:length(num)) \ in(B:length(num)) \ out(C:length(num)) { #pragma omp parallel for for(size_t i = 0; i < num; ++i) { C[i] = A[i] + B[i]; } }
のようになります
これでホストからXeonPhiに対して並列処理をさせることが出来るようになりました
