-インテルコンパイラ・レシピ-

インテル®MIC アーキテクチャ製品が発売されてから、かなり時間が経ちました。MICの独自のプログラミング手法として、OFFLOADプログラミングと実行について簡単な紹介をします。

本書は既にMPSSパッケージとインテルコンパイラ2015(C++)、及びインテルMPIライブラリ5.0.2がインストール済みであることを前提とします。

OFFLOADプログラミング

OFFLOADプログラミングは、オリジナルのプログラムにOpenMPと同様にディレクティブを挿入してXeon Phi へのデータ転送、プログラムの実行を指示します。 このOFFLOADプログラミングには、非共有メモリーモデルと仮想共有メモリーモデルが存在します。仮想共有メモリーモデルを使用する場合は、インテル®Click™Plus が必要となるため本書では割愛します。

OFFLOADプログラムの実行の流れは以下の用になります。

• CPUでの処理(非OFFLOAD 部分)
• CPUでXeon Phi に転送するデータを準備
• Xeon Phi にデータの転送
• Xeon Phi でプログラムの実行
• Xeon Phi からCPUに結果の転送
• CPUでの処理(非OFFLOAD 部分)

サンプルプログラムの解説

ここで使用するサンプルプログラムは巻末に提示しています。
プログラム内のOFFLOADに関する部分の説明を以下に示します。

__declspec (target (mic)) float fa[SIZE][SIZE] __attribute__((aligned(64)));
__declspec (target (mic)) float fb[SIZE][SIZE] __attribute__((aligned(64)));
__declspec (target (mic)) float fc[SIZE][SIZE] __attribute__((aligned(64)));

__declspec (target (mic)) は作成した関数がXeon Phi でも利用できることを保証するために、特別な関数属性を指定します。

#pragma offload target (mic) inout(fa,fb,fc)
#pragma omp parallel for
for (i=0; i for (j=0; j fa[i][j] = i * 0.1;
fb[i][j] = i * 0.05;
fc[i][j] = 0.0;
}
}

もしXeon Phi が利用可能であれば、#pragma offload target(mic) ディレクティブの次の文を Xeon Phi で実行します。 Xeon Phiが利用できない場合は、CPUで実行します。
上記の例では、OpenMP のディレクティブ部分からXeon Phi上で並列実行されます。

コンパイルと実行

以下の用にしてプログラムをコンパイルします。 OFFLOADのディレクティブを有効にするコンパイルオプションはありません。 デフォルトでOFFLOADが有効になります。

hpcs@server ~]$ icc -O2 -openmp test.c -o test

コンパイルが完了したら、環境変数を設定してからプログラムを実行します。

[hpcs@server ~]$ export MIC_ENV_PREFIX=MIC
[hpcs@server ~]$ export MIC_OMP_NUM_THREADS=30
[hpcs@server ~]$ export MIC_MKP_AFFINITY=scatter
[hpcs@server ~] $ ./test
Sum = 4.396342e+15, Sec = 6.568, Gflops = 40.871

ここで、設定した環境変数の内容は以下の通りです。

MIC_ENV_PREFIX

ホストの環境変数は Xeon Phi に引き継がれますが、この変数で指定した文字列(ここではMIC)が頭についた環境変数は、Xeon Phi にのみ設定されます。

MIC_OMP_NUM_THREADS

MIC_ENV_PREFIXで指定された接頭辞がついていますので、プログラムをXeon Phi で動作させるときの最大スレッド数を指定しています。

MIC_MKP_AFFINITY

MIC_ENV_PREFIXで指定された接頭辞がついていますので、Xeon Phiでプロセシング・ユニットにスレッドをバインドする方法を指定します。

OFFLOADのレポート機能

実際のプログラムがCPU、Xeon Phi のどちらで実行されているのかを知るために、OFFLOADのレポート機能を使用して実際のプログラムの動作を確認することが可能です。

環境変数 OFFLOAD_REPORT を設定することで、レポート機能を使用できます。レポート機能は、1～3の値を指定可能で、大きな値を指定すればより詳細なレポートが表示されます。

[hpcs@server ~]$ export MIC_ENV_PREFIX=MIC
[hpcs@server ~]$ export MIC_OMP_NUM_THREADS=30
[hpcs@server ~]$ export MIC_MKP_AFFINITY=scatter
[hpcs@server ~]$ export OFFLOAD_REPORT=1
[hpcs@server ~]$ ./test
[Offload] [MIC 0] [File] test.c
[Offload] [MIC 0] [Line] 29
[Offload] [MIC 0] [Tag] Tag 0
[Offload] [HOST] [Tag 0] [CPU Time] 0.866272(seconds)
[Offload] [MIC 0] [Tag 0] [MIC Time] 0.052415(seconds)

[Offload] [MIC 0] [File] test.c
[Offload] [MIC 0] [Line] 41
[Offload] [MIC 0] [Tag] Tag 1
[Offload] [HOST] [Tag 1] [CPU Time] 6.501428(seconds)
[Offload] [MIC 0] [Tag 1] [MIC Time] 6.411892(seconds)

Sum = 4.396342e+15, Sec = 6.513, Gflops = 41.216

Xeon Phi 特有の環境変数

OpenMP では、プログラム実行時のスレッド数を環境変数 OMP_NUM_THREADS で指定します。 Xeon Phi では、環境変数KMP_PLACE_THREADSでより詳細な指定が可能です。 環境変数 KMP_PLACE_THREADS を使用することで、使用するコア数と、コアあたりに割り当てるスレッド数も指定できます。

たとえば30コアを使用しコアあたり3スレッドを割り当て、合計 90 スレッドでプログラムを実行する場合は、以下の用に指定します。

[hpcs@server ~]$ export MIC_ENV_PREFIX=MIC
[hpcs@server ~]$ export MIC_KMP_AFFINITY=compact
[hpcs@server ~]$ export MIC_KMP_PLACE_THREADS=30c,3t
[hpcs@server ~]$ ./test
Sum = 4.396342e+15, Sec = 4.098, Gflops = 65.503

(注意)上記環境変数を使用する場合は、環境変数OMP_NUM_THREADS とは一緒に使用しないでください。

サンプルプログラムリスト

本書で使用したサンプルプログラムを以下に示します。

サンプルプログラム: test.c

#include
#include
#include
#include
#include

#define SIZE 5120

__declspec (target (mic)) float fa[SIZE][SIZE] __attribute__((aligned(64)));
__declspec (target (mic)) float fb[SIZE][SIZE] __attribute__((aligned(64)));
__declspec (target (mic)) float fc[SIZE][SIZE] __attribute__((aligned(64)));

double second()
{
double tseconds = 0.0;
struct timeval tv;
gettimeofday(&tv, (struct timezone*)0);
tseconds = (double) (tv.tv_sec + tv.tv_usec*1e-6);
return ( tseconds );
}

int main()
{
int i,j,k;
double tstart, tstop, ttime;
float sum = 0.0;
double gflops = 0.0;

#pragma offload target (mic) inout(fa,fb,fc)
#pragma omp parallel for
for (i=0; i for (j=0; j fa[i][j] = i * 0.1;
fb[i][j] = i * 0.05;
fc[i][j] = 0.0;
}
}

tstart = second();

#pragma offload target (mic) in(fa,fb) inout(fc)
#pragma omp parallel for private (i,j,k)
for (i=0; i for (j=0; j for (k=0; k fc[i][j] += fa[i][k] * fb[k][j] ;
}
}
}

tstop = second();
ttime = tstop – tstart;

for (i=0; i for (j=0; j sum += fc[i][j];
}
}


gflops = (double)2.0*SIZE*SIZE*SIZE*1.0e-9/ttime;

if ((ttime) > 0.0 )
{
printf(“Sum = %10.6e, Sec = %10.3lf, Gflops = %10.3lf\r\n”, sum, ttime, gflops);
}
return (0);
}