search

shimomurakazuya / diff_eq

練習用コード
0 stars 0 forks source link

コード並列化 #6

Open shimomurakazuya opened 3 years ago

shimomurakazuya commented 3 years ago

openMPを実装したバージョンをアップしました。 並列化箇所は初期条件の設定関数と、拡散方程式計算関数です。 d269a2fe960d6431abdc39e8996558d2cf17f2b2

デバッグの結果メモ：実行の時dataディレクトリがないとセグフォで落ちる。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea

fn[i] = ... の代入処理が load & store になるという意味ですか？ そうなるのかなと思っていました。言葉足らずですみません

確認頻度については承知しました。

hasegawa-yuta-jaea commented 3 years ago

@shimomurakazuya

そうなるのかなと思っていました。言葉足らずですみません

代入が load&store になるつもりでロード 2 回と置いた、ということは把握しました。

それで、

バージニア大のSTREAM benchのコードを読むと、代入は store のみカウントのようですね？（tri-ad が 3 アクセスとなっている） このカウントで理論バンド幅の 90% 以上出たそうですので、load&store じゃなくて store なんじゃないですかね。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea データ演算数のカウントについては承知しました。教えていただきありがとうございます。

今朝までの進捗を共有します。

stream benchを拡散方程式へと近づけていく中で速度がどのように変化しているかを調べているのですが、二次元拡散方程式まで拡張して、75%程度の性能比が出ていることを確認しています。

シミュレーション設定はthread数=24、コンパイル設定が (gcc -O3 -std=c++11 -fopenmp -Wall -Wextra )です。 格子点数については1次元がnx=10^8、２次元がnx=2000000,ny=100です。

1次元版↓ 3f9bc7ffe1be698ebb84d47d6eaeea64c8493e5e

２次元版↓ 2f92eebb2b61cf8cbbdce75af8a851f5799a9b38

１次元版では1時間ステップの計算いかかる時間が0.0144435sec, でバンド幅がband_width(GB/s)=110.776となりました。 2次元版では1時間ステップの計算にかかる時間が0.0223484sec, でバンド幅がband_width(GB/s)=105.633となりました。 いずれも性能比が75％以上で、十分な性能があると考えています。

　今のところ、元の2次元拡散方程式（演算性能が低いコード）との大きな違いはファイル分割をしたかどうかであるので、stream_benchコードでも同様の処理を実施し、性能を評価する予定です。

hasegawa-yuta-jaea commented 3 years ago

@shimomurakazuya

ご苦労様です。性能出てよかったですね。 以下、コメントです。

拡散方程式ではGB/s と GFLOPS, ルーフラインに対する性能%も出すようにしてください。

混乱させるようですみませんが、STREAM Benchmark ではオリジナルに合わせて GB/s にしただけですので。。 HPC業界としては、原則 ルーフラインモデル で評価すべきです。

２次元がnx=2000000,ny=100

1億格子取れれば何でも良いのですよね？例えば

はどうですか？

元の2次元拡散方程式（演算性能が低いコード）との大きな違いはファイル分割をしたかどうかであるので、

前にも書いた気がしますが、これはたぶん index_xy() がインライン展開されない仕様になっているせいです。 再掲します。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea

拡散方程式ではGB/s と GFLOPS, ルーフラインに対する性能%も出すようにしてください。

承知しました。ルーフラインモデルとの性能比も計算してみます。

1億格子取れれば何でも良いのですよね？

nx=200000にしたのはy方向のアクセス範囲が3(中心差分法で用いた範囲)8（倍精度）2000000(格子点数)=48Mbyte<35Mbyte(キャッシュメモリ容量)となるようにしたかったからです。

ただ、ベクトル長やスレッド数については考慮していなかったので。コメントされた格子点数についても少し調べてみます。

これはたぶん index_xy() がインライン展開されない仕様になっているせいです。 実装するのを忘れていました、すみません。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea

inline関数を導入した結果を報告します。 画像の通り、性能比が77%(23.978)に向上したことを確認しました。 格子点数はnx=2000000, ny=100、演算性能はデータ演算数:4、flop数:7です。

スクリーンショット 2021-06-11 17 05 17

図：演算結果。　elapse_time:出力時間など全ての時間を含めた時間。elapse time loop avetage: １時間ループを演算するのにかかる時間。thread_num: スレッド数。bandwidth：バンド幅。Flops：演算性能。calcuration ratio: 性能比。

※理論演算性能(peak flops)は30.625Gflopsです。

次にnx=ny=6144の結果を載せます。こちらはデータ演算数:2、flop数:7として計算しています。 こちらの場合は演算性能が約60%となりました。ただ試行回数は一回なので、下振れを引いた可能性もあるので、継続して投入してみます。

スクリーンショット 2021-06-11 17 12 56

解析解との整合性はこれから検証します。

idomura commented 3 years ago

下村さん、次はスレッド並列の回し方を変えてみてください。staticのブロック分割をdynamicやcyclicに変えるとy方向のステンシルデータもスレッド間で再利用できるようになるので、性能が変わると思います。

hasegawa-yuta-jaea commented 3 years ago

@shimomurakazuya

inline関数を導入した結果を報告します。 画像の通り、性能比が77%(23.978)に向上したことを確認しました。

了解です。性能出てよかったですね。 画像（std::cout？）も対応していただいてどうもありがとうございます。GB/s と GFLOPS が一度に見れると、だいぶ見通しがいいですね。

次にnx=ny=6144の結果を載せます。こちらはデータ演算数:2、flop数:7として計算しています。 こちらの場合は演算性能が約60%となりました。

y 方向が完全にキャッシュに載る仮定でルーフライン を計算し直したのですね？ GLOPSとかルーフライン比を見た感じ、実際のところはキャッシュ再利用は完璧ではなく、部分的に載ったり載ってなかったりしているのでしょう。 スレッド並列の回し方で改善するといいですね。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea スレッド並列の回し方を変えてみました。

　下の表はnx=ny=6144,スレッド数＝24の時に、チャンク数1としてSchedule指示節の中身ごとの性能比です。(cyclicというオプションはなかったです) defalut（無指定時）と比較してguided以外は性能が低下しました。おそらく割り当て時間の処理+分割が最適ではないことに起因すると思われます。少なくとも、y方向のデータアクセスに関してキャッシュ再利用ができていないままだと考えています。

idomura commented 3 years ago

1.6144だとそもそもオンキャッシュじゃないですか？それだと、チャンクサイズを下げた分だけタスク割り当てのオーバーヘッドがかかります。y方向がメモリアクセスとなるように問題サイズを十分に大きくしましょう。 2.cyclicはstatic,1です。 3.OpenMPのオーバーヘッドを下げるにはチャンクサイズを1以上に設定して、タスク割り当ての頻度を下げます。 4.性能比の定義がよくわかりません。defaultは1ではないのですか？

hasegawa-yuta-jaea commented 3 years ago

コメントが被りました。失礼しました。 井戸村さんのコメントに優先して対応してください。

@shimomurakazuya

承知です。OpenMPよく知らないので特にアドバイスできません。 明日の打ち合わせで井戸村さんか小野寺さんからコメントがあると思うので、それで次の方針を決めてください。

そういえば小野寺さんの言っていたもの：

は試しましたか？

あとすみません、井戸村さんのコメントにあるnx = ny = 6144 は私からお願いした条件ですが、何やらダメらしいので、元の条件に戻していただいて結構です。（ nx ≠ ny の解析解も検討してみてください）

shimomurakazuya commented 3 years ago

@idomura

1.6144だとそもそもオンキャッシュじゃないですか？それだと、チャンクサイズを下げた分だけタスク割り当てのオーバーヘッドがかかります。y方向がメモリアクセスとなるように問題サイズを十分に大きくしましょう。

わかりました。格子点数を変更します。

4.性能比の定義がよくわかりません。defaultは1ではないのですか？

このdefalutの意味は理論値ではないです。schedule 支持節がない時に計測した性能比を乗せています。(schecule(static)に対応していると思います)　紛らわしくてすみません。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea

pragma omp ivdepに関しては試してみましたが、コンパイル時に無視されてしまっています。

aligned_allocについては失念していました。実装します。

shimomurakazuya commented 3 years ago

スレッドの分割についての結果です。 格子点の設定はnx=2000000,ny=96 でシミュレーションを実行しました。結果としては、チャンク数が1ではny方向のデータが再利用できるはずだが、オーバーヘッドに時間がかかり、指定がない時よりも遅い。ただ、チャンク数が４になるについれてオーバーヘッドの時間が短くなり、性能が向上している。一方でチャンク数が8になると、今度は24*8=192>96で、スレッド番号が一周で割り振られるデータ個数の方がnyよりも大きくなり、ny方向のキャッシュデータが再利用ができなくなるため、速度が下がっている。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea

昼におっしゃられたコマンドについて質問なのですが、

pragma omp ivdep　はgccに対応しているとネットでは書かれていたのですが以下のコードでコンパイルしても、次の警告文とともに無視されてしまっているようです。どうすれば読んでもらえるのでしょうか？

「ValuesDiffusion.cpp:62:0: 警告: #pragma omp ivdep を無視します [-Wunknown-pragmas]」 (PS. module load gccを行った後でも同様の警告文が出ました。)

void ValuesDiffusion::
 52 time_integrate(const ValuesDiffusion& valuesDiffusion) {
 53     real* fn = f_;
 54     const real* f = valuesDiffusion.f_;
 55 #pragma omp parallel  
 56     {
 57 #pragma omp for schedule(static,8) 
 58 
 59         for(int j=0; j<defines::ny; j++) {
 60                 const int jm = (j-1 + defines::ny) % defines::ny;
 61                 const int jp = (j+1 + defines::ny) % defines::ny;
 62 #pragma omp ivdep
 63                 for(int i=0; i<defines::nx; i++) {
 64                 const int im = (i-1 + defines::nx) % defines::nx;
 65                 const int ip = (i+1 + defines::nx) % defines::nx;
 66 
 67                 const int ji  = index::index_xy(i,j);
 68                 const int jim = index::index_xy(im,j);
 69                 const int jip = index::index_xy(ip,j);
 70                 const int jim2 = index::index_xy(i,jm);
 71                 const int jip2 = index::index_xy(i,jp);
 72 
 73                 fn[ji] = f[ji]
 74                     + defines::coef_diff * (f[jim] - 4*f[ji] + f[jip] + f[jim2] + f[jip2] );
 75             }
 76         }
 77     }
 78 }
  1. aligned_allocについて、ValueDiffusion.cppのallocateの部分(x, fのメモリ確保を行う関数内)でaligned_allocを用いていますが、一応これでfのアラインメントは揃えられているのでしょうか？(やりたいことはアラインメントを揃えることで計算が内部処理的に速くなる稼働かの検証です)(計算結果については追記します)
 void ValuesDiffusion::
 16 allocate_values() {
 17     x_ = (real*)std::malloc(sizeof(real) * nx_);
 18     y_ = (real*)std::malloc(sizeof(real) * ny_);
 19     f_ = (real*)std::malloc(sizeof(real) * ncell_);
 20  f_ =  (real*)aligned_alloc( defines::alignment, ncell_*sizeof(real));
 21 }

追記：計算結果についてですが、以下のようになりました。(パラメータ設定: nx=2000000,ny=96, スレッド数:24 schedule文 指定無し ) 導入前が性能比 =0.9277なので、若干早くなりました。

スクリーンショット 2021-06-14 17 19 35

hasegawa-yuta-jaea commented 3 years ago

@shimomurakazuya

pragma omp ivdep

#pragma GCC ivdep ではないですか？ https://gcc.gnu.org/onlinedocs/gcc-7.5.0/gcc/Loop-Specific-Pragmas.html#Loop-Specific-Pragmas

20 f_ = (real*)aligned_alloc( defines::alignment, ncell_*sizeof(real));

C言語版の関数が呼ばれているようですね。たぶん問題ないですが。

一応、C++にするなら、

としたほうがお行儀は良い気がします。（あくまでお行儀の話なのでどちらでも良いです）

https://en.cppreference.com/w/cpp/memory/c/aligned_alloc

（追記）あと、 19 f_ = (real*)std::malloc(sizeof(real) * ncell_); はメモリリークなので消してください。 （さらに追記）defines::alignment の数字は幾つですか？まあ私はCPU最適化知らないので聞いても仕方がないのですが。。。それか明日の打ち合わせで小野寺さんに確認すれば最適な数字を教えてくれるかもしれません。

shimomurakazuya commented 3 years ago

@hasegawa-yuta-jaea 　コメントありがとうございます。 defines::alignment は今32です。

hasegawa-yuta-jaea commented 3 years ago

@shimomurakazuya

（追記が多すぎたのでコメントを分割します）

導入前が性能比 =0.9277なので、若干早くなりました。

速くなったのはよかったですね。

ただ、ルーフラインモデルが、y 方向キャッシュに載らない想定のままですね？井戸村さん曰くスレッドの調整でy方向も載るそうですので、 B= (1 load + 1 store) * 8byte = 16 byteのルーフライン でも性能%を出してみてください。