A64FX CPU向けコンパイラとチューニング事例
富士通株式会社

ミッションクリティカルシステム事業本部
ＨＰＣシステム事業部

言語ソフトウェア部

原口正寿© 2022 Fujitsu Limited

○ スーパーコンピュータ京、富岳などの自己紹介

コンパイラ開発を担当○京や富岳において、現在のtradモードのコンパイラ開発をリード○現在は、OSSコンパイラをベースとする
clangモードも含めてコンパイラ全体を担当

2© 2022 Fujitsu Limited

はじめに⚫ 富岳・FX1000のA64FXは、省電力を目指した設計で、x86と特性が異なるため、最高性能を引き出すにはコツがいる⚫ 本日は、コンパイラ機能の活用やアプリチューニングでA64FXの性能を引き出すコツのいくつかを紹介する

アウトライン⚫ A64FXの特徴（電力性能）とアーキテクチャ特性⚫ コンパイラ機能活用による性能改善⚫ clangモードについて⚫ プログラムチューニングによる性能改善⚫ チューニング事例
3© 2022 Fujitsu Limited

A64FXの特徴（電力性能）○ 性能／電力を重視で設計○ 下表は、SPEC CPU2017における性能と電力の測定結果○ ratio値単体では富岳はCascade Lakeの約6割の性能○ 最大電力あたりのratio値（ratio/最大電力）において、富岳はCascade Lakeの約2倍
PRIMERGY RX2530M5 (Intel Xeon
Platinum 8268 Cascade Lake)

富岳
SPEC
CPU2017fp

48コア

24コア×2ソケット

2.0GHz(ノーマルページ）

2.9GHz(Turbo off)

SPEC ratio 最大電力(W)
幾何平均

69.2

118.5

ratio/最大電力

SPEC ratio 最大電力(W)

0.584

118.5

434.2

ratio/最大電力の富岳とXeonの比(1.0以上
で富岳が良い）ratio/最大電力
0.273

2.14

※注: 本電力性能測定データは、弊社で実測した参考値で、SPEC未登録です。

4© 2022 Fujitsu Limited

アプリ性能観点から見たアーキテクチャ特性○ アーキテクチャ特性から言える性能向上のポイント○レイテンシの隠蔽（ソフトウェアパイプライニング、プリフェッチ）○メモリバンド幅の有効活用○SIMD化項目

A64FX

Cascade Lake

周波数2.0GHz2.7/2.2/1.8GHz(int/avx2/avx512)

Turbo周波数

2.2GHz上限4.0/3.8/3.7GHz

L1Dキャッシュ

64KiB/core, 4way

32KiB/core, 8way

load 11load 4～6

fma 9

fma 4

add/logic 4

add/logic 1

メモリバンド幅(ノード)

256GB/s×4

140.78GB/s×２

A64FX有利だが実アプリ
ではより必要なケースもレイテンシ隠蔽手段

OoO※1+コンパイラによるSWP※2等

短レイテンシ+大OoO※1+SMT

SWP有効であれば有利

命令レイテンシ(simd)

性能向上のポイント

レイテンシの隠蔽が重要
スカラ命令のレイテンシ
も長く、SIMD化も重要※1 OoO:Out of Order資源、※2 SWP:ソフトウェアパイプライニング
5© 2022 Fujitsu Limitedコンパイラ機能活用による性能改善⚫レイテンシの隠蔽○ソフトウェアパイプライニング▪ 基本原理、実装、効果▪ 評価方法▪ チューニング機能▪ 促進機能と効果▪ ジレンマ（性能向上しないケース）▪ レジスタ不足の影響対策、ループ分割○キャッシュ活用▪ ソフトウェアプリフェッチ⚫メモリバンド幅ネックの対策○zfillの活用
6© 2022 Fujitsu Limitedソフトウェアパイプライニング：基本原理○ ループの次回転以降を重ね合わせて

最適化後

load
load
load
load
load
load

load
load
load
load
load
load

add
add
add

load
load
load

load
load
load

add
add
add

store
store
store

add
add
add

store

store
store
store
store
store
store

7

© 2022 Fujitsu Limited

命令レベルの並列性を向上
概念説明のためのマシンモデル
load の latency = 3サイクル
add の latency = 3サイクル
store の latency = 1サイクル
演算器の数(load,store用)= 3
commit 数 = 4 (load,storeは3つまで同時発行可能)○ このマシンモデルであれば
右図のようにカーネルは
きっちり埋まるプログラム例
do i=1,na(i) = b(i)+ c(i)
enddo
最適化前

カーネル

load

カーネル

プロローグ

load

add
エピローグ

ソフトウェアパイプライニング：実装（１）○ スケジューラ実装○ データアクセスはオンキャッシュを想定○ カーネルでハードリソースを使い切るように命令スケジュール○ 翻訳時適用判断など○ ソフトウェアパイプライニング時点での最内ループが対象（事前にフルアンローリング等が動作してループでなくなっている場合は親ループが対象）○ ソフトウェアパイプライニング時点で、関数・サブルーチンCALL、分岐があったら抑止○ レジスタ不足と判定されたループは抑止○ ループが巨大（命令数が多い、メモリアクセスが多い）な場合は抑止○ 有限の時間内に良いスケジューリング結果が求まらない、もしくはデータ依存が原因でスケ
ジューリング効果がないと判断した場合は抑止○ 実行時の動き○ ある程度のループ回転数がないと高速ルートを通らない

そのルート以外を通る場合にはかえって遅くなるケースもある
8© 2022 Fujitsu Limitedソフトウェアパイプライニング：実装（２）do i=1,na(i) = b(i) + c(i)*d(i)enddo○ A64FXでの実際の動き■ ■■○ 演算器数、レジスタ数、命令レイテンシ／パイプラインを配慮

■ ■
■ ■

し、カーネルでムダなくリソースを使うスジューリング結果を探し出す
ソフトウェアパイプライニングのマシンモデル

■ ■

■ ■
■ ■
■ ■

■

■ ■
■ ■
■ ■
■
■

■ ■
■ ■
■ ■

■

■
■

■ ■
■ ■
■ ■

■

ソフトウェアパイプライニングのカーネルループ■○ 倍精度浮動小数点レジスタ：３２個○ SIMD loadは２つまで同時実行可、storeとの同時実行は不可○ レイテンシ（下表）

■
■

■ ■
■ ■
■ ■

ld1d(SIMDﾛｰﾄﾞ)■11τ

fmla(FMA)■9τ

st1d(SIMDｽﾄｱ)■

1τ(突き放しのため）

■
■

■
■

■ ■
■ ■
■ ■

■
■

■
■

■ ■
■ ■
■ ■

■
■

■
■
■
■ ■
■

■
■

-Koptmsg=2で出力される最適化メッセージ■■○ このカーネルルートを通るには、ある程度の回転数が必要○ 左例では144回転必要

■

■
■

■
■

■

jwd8205o-i "fma.c", line 3: ループの繰返し数が144回以上の時、ソフト
ウェアパイプライニングを適用したループが実行時に選択されます。

-Nlst=tで生成される *.lst ファイルの最適化出力情報■■

SOFTWARE PIPELINING(IPC: 3.20, ITR: 144, MVE: 4, POL: S)■

9© 2022 Fujitsu Limitedソフトウェアパイプライニング：効果(1)
ソフトウェアパイプライニングの効果(L2)
1.04! n=250000 で sub を 10000回呼出し
subroutine sub(a,b,c,d,n)real(kind=8),dimension(1:n) :: a,b,c,d
do i=1,na(i) = b(i) + c(i)*d(i)enddo
end subroutine

アンロールだけでは効果なし

1.02
1.00
0.98
0.96

ソフトウェアパイプライニングの効果で1.08倍

0.94
0.92
0.90
0.88
-Kfast,nounroll,noswp-Kfast,unroll=8,noswp
10-Kfast© 2022 Fujitsu Limitedソフトウェアパイプライニング：効果(2)! n=500 で sub を 5000000回呼出し
subroutine sub(a,b,c,d,n)real(kind=8),dimension(1:n) :: a,b,c,d
do i=1,na(i) = b(i) + c(i)*d(i)enddo
end subroutine

ソフトウェアパイプライニングの効果(L1D)
1.10
1.00
0.90
0.80
アンロール効果は1.49倍

ソフトウェアパイプライニング効果は1.74倍

0.70
0.60
0.50
-Kfast,nounroll,noswp-Kfast,unroll=8,noswp-Kfast11© 2022 Fujitsu Limitedソフトウェアパイプライニング：評価方法
1
2

subroutine sub(a,b,c,d,n)real(kind=8),dimension(1:n) :: a,b,c,d-Nlst=t指定による翻訳時情報出力<<< Loop-information Start >>>※詳細情報はtrad固有<<< [OPTIMIZATION]<<< SIMD(VL: 8)
IPC（Instruction Per Cycle）<<< SOFTWARE PIPELINING(IPC: 3.20, ITR: 144, MVE: 4, POL: S)
最重要。良いスケジューリング結果かどうかの <<< PREFETCH(HARD) Expected by compiler :<<<a, d, c, b
目安。4.0に近い値で大きい値のほうが良い<<< Loop-information End >>>ITR（ITeRation）
3
12v
do i=1,n
ソフトウェアパイプライニングのカーネルを通412v
a(i) = b(i) + c(i)*d(i)
5
1
2v
enddo
るために必要な1スレッドあたりの回転数
6
end subroutine○○○ MVE（Modulo Variable Extension）

ソフトウェアパイプライニングのカーネルに何回転分が入っているかを示す○ POL（Policy）

ソフトウェアパイプライニングの採用アルゴリ
ズムを示す。詳細後述。
S:small：IMS（Iterated Modulo Scheduling）
L:large：SMS（Swing Modulo Scheduling）
12© 2022 Fujitsu Limitedソフトウェアパイプライニング：チューニング機能○trad固有のチューニング機能を紹介○-Kswp_policy={auto|small|large}○-Kswp_{weak|strong}○-Kswp_{freg|ireg|preg}_rate=N13© 2022 Fujitsu Limitedソフトウェアパイプライニング：アルゴリズム選択○ -Kswp_policy={auto|small|large}
翻訳オプション

マニュアル記載-Kswp_policy=small小さなループ(例えば、
必要レジスタ数が少ない
ループ)に適した命令ス
ケジューリングアルゴリ
ズムを使用します。

※OCL/pragmaもあり採用アルゴリズム詳細

IMS(Iterative Modulo
Scheduling)

京でも採用している従来アルゴリズムをA64FX向けに対応。レジスタ不足とならない(SPILL/FILLが出ない）範囲で解を求めます※clangのソフトウェアパイプライニングは-ffj-swp指定で動作し、実装はIMSのみ。

-Kswp_policy=large大きなループ(例えば、
必要レジスタ数が多い
ループ)に適した命令ス
ケジューリングアルゴリ
ズムを使用します。

SMS(Swing Modulo
Scheduling)

A64FX向けに改良した新アルゴリズム
レジスタプレッシャ削減とクリティカルパス優
先のヒューリスティックを持つモジュロスケ
ジューラ。特徴として、FILLロードはリソースが空いているなら許容します。
レジスタプレッシャが高いループでより良い命
令スケジューリングが期待できます。

-Kswp_policy=auto

ループ毎に命令スケ
ジューリングアルゴリズ
ムを自動で選択します。

IMSとSMSの自動
選択

デフォルト。基本、IMSとSMSの自動選択だが、ヒューリスティックにIMS効果が望めそうと判断したケースはIMSを採用

14© 2022 Fujitsu Limited

SMSの効果○ atan2を含むループで評価（自動でSMSが選択されるケース）
SMS(-Kswp_policy=large)の効果
1.20
1.00

! n=250000 で sub を 10000回呼出し
subroutine sub(a,b,c,n)real(kind=8),dimension(1:n) :: a,b,c
do i=1,na(i) = atan2(b(i),c(i))
enddo
end subroutine

0.80

数学関数のインラインの効果で3.47倍
0.60
0.40

IMSは効果なし

SMSの効果でさらに1.46倍

0.20
0.00
-Kfast,noswp,noilfunc-Kfast,noswp-Kfast,swp_policy=small -Kfast (swp_policy=large)
15© 2022 Fujitsu Limitedソフトウェアパイプライニング：チューニング機能（続）○ ソフトウェアパイプライニングルートを通すためのチューニングオプション
翻訳オプション
（以下OCL/pragma有）-Kswp_weak-Kswp_strong-Kswp_freg_rate=N
他に-Kswp_ireg_rate=N,-Kswp_preg_rate=N
もある

最適化
指示子

マニュアル記載

詳細（ワンポイント）○

ソフトウェアパイプライニングを ITRを小さくしたい
調整し、ループ内の実行文の重な IPCを下げて、ソフトウェアパイプライニングのカーりを小さくすることを指示します。 ネルに入る回転数(ITR)を小さくする。

○

ソフトウェアパイプライニングの条件を緩和し、ソフトウェアパイ
プライニングを促進すうことを指示します。

ソフトウェアパイプライニングの答えが見つからない(jwd8662o-i)
スケジューリングの解が見つからなかった時、もう少し時間をかけて解を探させる。

ソフトウェアパイプライニングを適用する際、浮動小数点レジスタ
ならびにSVEのSIMDレジスタのN%が使用可能であると指示します。

レジスタ不足でソフトウェアパイプライニングできない(jwd8665o-i,jwd8666o-i,jwd8673o-i)
特にデータ依存がないケースで、メモリへの退避復元であるSPILL/FILLが出ても良いからソフトウェア
パイプライニングを実行してみる。
（例）浮動小数点レジスタ不足時、レジスタ数が1.5倍あると思ってソフトウェアパイプライニング実行-Kswp_freg_rate=150○

16© 2022 Fujitsu Limitedソフトウェアパイプライニング：促進機能と効果～逆数近似命令の利用～○ 逆数近似命令の利用(-Kfp_relaxed指定)による性能改善○ 三角関数・指数関数補助命令利用などによる数学関数のインライン(-Kilfunc指定)でも似た傾向
逆数近似命令利用の効果

1.20

fdiv命令のままでソフトウェア
パイプライニングできても効果なし

モード-Kfastとの関係

trad-Kfp_relaxed, -Kilfuncを誘導

clang-ffj-fp-relaxed, -ffj-ilfuncを誘導

1.00! n=250000 で sub を 10000回呼出し
subroutine sub(a,b,c,n)real(kind=8),dimension(1:n) :: a,b,c
do i=1,na(i) = b(i)/c(i)enddo
end subroutine

0.80
0.60
0.40

逆数近似命令利用で18.7倍
ソフトウェアパイプライニング
でさらに1.24倍

0.20
0.00
-Kfast,nofp_relaxed,noswp-Kfast,nofp_relaxed-Kfast,noswp
17-Kfast© 2022 Fujitsu Limitedソフトウェアパイプライニングのジレンマ○ ソフトウェアパイプライニングは万能ではない➡適用されれば必ず性能が上がるというわけではない○ ループ回転数がITR以下の場合は逆効果のケースあり○ データ依存のあるループはIPCが低く性能向上しないケースあり<<<

SOFTWARE PIPELINING(IPC: 3.20, ITR: 144, MVE: 4, POL: S)

：
3
4
5

1
1
1

<<<
：
3
1
4
1
5
1

2v
2v
2v

do i=1,n ! 回転数n＜144a(i) = b(i) + c(i)*d(i)enddo

SOFTWARE PIPELINING(IPC: 0.55, ITR: 6, MVE: 2, POL: S)
2s
2s
2s

do i=1,na(i) = a(i-1) + b(i)enddo

加算のレイテンシ
18© 2022 Fujitsu Limited

レジスタ不足の影響対策○ ループ分割による最適化促進

回転数nで回るループを２つ以上に分け、理想的には分割したループ間の受渡し用
作業配列を最小化し、個々のループに対する最適化促進により、性能向上を狙う
実 do i=1,n
行 a(i) =SPILL/FILL回避
do i=1,n:a(i) =
ソフトウェアパイプライニング適用時p(i)=b(i) =
間 enddo実::p(i) =
q(i) =::
x(i) =endd

行時間

do i=1,n
実 b(i) =行 :時 q(i) =:間 x(i) =endd

上から下への受渡し用作業配列は
最小限

SPILL/FILL回避
ソフトウェアパイプライニング適用

大量のSPILL/FILL
19© 2022 Fujitsu Limitedループ分割：２つのアルゴリズム○ ループ分割による最適化促進（-Kloop_fission,ocl有効時）!OCL LOOP_FISSION_TARGET(CL)!OCL FISSION_THRESHOLD(n)do i=1,m～
enddo!OCL LOOP_FISSION_TARGET(LS)!OCL FISSION_THRESHOLD(n)do i=1,m～enddo

CL：クラスタリングアルゴリズムでループ分割
分割ループ間で利用する作業配列削減を優先（default）

LS：局所探索アルゴリズムでループ分割
ソフトウェアパイプライニング促進を優先

THRESHOLD(n)閾値nは1～100でデフォルトは50。小さい値を指定したほうが分割数が増える

20© 2022 Fujitsu Limitedキャッシュ活用：ソフトウェアプリフェッチ○ ソフトウェアプリフェッチ機能の基本動作（出力確認はリスタ機能で）○ 16ストリームまではハードウェアプリフェッチにお任せ○ 17ストリーム以上でソフトウェアプリフェッチ動作○ インダイレクトプリフェッチ、ストライドプリフェッチは個別オプション指定時に動作

4
5
6

1
1
1

idx,aをハードウェアプリフェッチに任せるのは<<< Loop-information Start >>>プリフェッチのデフォルト動作:<<< PREFETCH(HARD) Expected by compiler :<<<idx, a<<< PREFETCH(SOFT) : 8bに対するソフトウェアプリフェッチ出力は、<<<
INDIRECT : 8-Kprefetch_indirect オプション指定時の<<<
b: 8
オプショナル動作<<< Loop-information End >>>2v
do i=1,n2va(i) = b(idx(i))
2v
enddo

○ 次頁ではインダイレクトプリフェッチの活用事例を示す
21© 2022 Fujitsu Limitedインダイレクトプリフェッチの活用事例○ PA情報のL1D,L2のdemand miss率を確認し、Memory throughputの余裕を見て、インダイ
レクトプリフェッチを活用

インダイレクト
プリフェッチにより1.49倍の効果[秒]!ocl prefetch
do i=1,na(i) = b(d(i)) + scalar * c(e(i))
enddo
浮動小数点
ロードL2
アクセス待ちCache ・・

L1D miss

・・

L1D missdemandrate (%)(/L1D miss)・・

L2 miss

・・

L2 missdemandrate (%)(/L2 miss)

・・

改善前

・・

3.77E+09 ・・

98.23% ・・

4.29E+08 ・・

54.19% ・・

改善後・・

3.82E+09 ・・

2.94% ・・

1.77E+09 ・・

2.02% ・・

Memory throughput(GB/s)

改善前

32.67

改善後

183.71

改善前

22

改善後

© 2022 Fujitsu Limitedメモリバンド幅ネックの対策○ メモリバンド幅ネックのプログラムの傾向○ 命令スケジューリングは効果無し○ 最適化で多少命令を削っても効果無し、最適化がいまいちで多少命令が増えても変化無し○ 性能向上のためには、メモリアクセス量を減

コア store

らすしかない○ キャッシュマシンは、読み書きをキャッシュ

ライン単位で行うため、ストアでもREADする。
しかし、キャッシュライン全体に書き込むことが分かっているなら、そのREADは不要なはず。

L1D
L2
memory

この考えから生まれたのがZFILL
データの中身をREADせず、キャッシュ上の位置だけ確定するプリフェチもどき23© 2022 Fujitsu Limited

zfillの活用（１）○ 最適化指示子 zfill (オプションは-Kzfill)
subroutine foo(a,b,c,d,n)real(kind=8),dimension(1:n) :: a,b,creal(kind=8) ::d!ocl zfill
do i=1,na(i) = b(i) + c(i)*denddo
end subroutine foo⚫ zfill実装はSIMD化を利用⚫ SIMD化できないものはzfillできない⚫ 固定長SIMDのみ⚫ 参照・定義の配列に重なりがないことが条件※clangモードは固定長SIMD指定(-msve-vector-bits=512)が必要
最適化指示行は、「#pragma [fj] loop zfill」○ 無闇に指定しても性能向上せず、逆効果

zfill活用は、キャッシュライン全体を書き替えることを保証しなければならないため、
キャッシュライン全体を書き替えることを保証するコードを出力する必要がある。
また、ハードウェアプリフェッチに任せず、zfillとL1Dへのソフトウェアプリフェッチを出力するため、バンド幅ネックでない場合、これらは全く余分な命令。したがって、ＰＡ情報を見て活用することが重要。
24© 2022 Fujitsu Limited

zfillの活用（２）○ 逆効果を避けるため、zfillはPA結果の[秒]「Memory Throughput(GB/s)」が大きいアプリかどうかを調べて活用

zfillにより1.32倍の効果

MemoryThroughput (GB/s)

改善前211.59

改善後

209.96

浮動小数点
ロードL2
アクセス待ち○ 改善効果は「L2 miss」の削減を確認
L1D miss rate(/Load-store
instruction)

L1D missdemandrate (%)(/L1D miss)

L1D miss
L1D miss
hardwaresoftwareprefetchprefetchrate (%)
rate (%)(/L1D miss) (/L1D miss)L2 miss

メモリ
アクセス待ち・・・・

L1D miss

改善前・・

9.39E+07

0.24 27.78% 72.21%

0.01% 9.38E+07

・・

改善後

・・

9.39E+07

0.21 16.80% 49.98% 33.22% 6.25E+07

・・

メモリ・キャッシュ
ビジー待ち

改善前
25

改善後
© 2022 Fujitsu Limited

clangモードについて⚫tradモードとclangモードの機能比較⚫clangモード優位な機能例⚫strict aliasing

26© 2022 Fujitsu Limited

tradモードとclangモードの機能比較○ 下表にtradモードとclangモードの主要機能比較を示す
項目

備考

tradモードclangモード

OpenMP

○

○

自動

○

×

可変長

○

○(default)

固定長○(default)○

ポインタ等のデータ依存解析はclangモード優位
SIMD化バリエーションはややtradモード優位
固定長は-msve-vector-bits=512で動作

インライン展開△○

C++のインライン展開はclang優位

ソフトウェアプリフェッチ○△

clangモードは一部未サポートあり

数学関数のインライン展開○○

ソフトウェアパイプライニング○△

LTO(LinkTimeOptimization)

×○

ACLE, fp16

×

○

zfill

○

○

セクタキャッシュ○×

スレッド並列化

SIMD化

clangモードは-ffj-swp指定時に動作、tradモードはチューニング
機能が充実⚫ C++翻訳は、インライン展開優位のclang
モードがお勧め⚫ データ依存が明確であれば、SIMD化やソフトウェアパイプライニング性能が安定してるtradモードがお勧め
27© 2022 Fujitsu Limited

clangモード優位な機能例○ インライン展開（例）仮想関数テーブル経由コールのインライン○ 仮想関数テーブル経由コールのインライン○ ラムダ式のインライン○ 不要な関数削除（翻訳時間に影響）など○ LTO(Link Time Optimization)○ -fltoオプションで動作○ リンク時であれば、実行モジュールを構成するファイルはすべて揃う。そこで、オブジェクト
ファイル(*.o)の中にソースプログラム相当を埋め込み、リンク時に再翻訳する。
これにより、ファイルを跨ぐインライン展開などの最適化が可能となる。

○ データ依存解析(以下は一例)// override sample#include <iostream>class Calc{
public: virtual double calc(double a, double b) { return a; }};class Add : public Calc{
public: double calc(double a, double b) { return a+b; }};：int main(void) {Calc *x = new Add;// calcをインラインし定数計算
std::cout << x->calc(2.0,1.0) << std::endl;}○ 型種別を利用したデータ依存解析（詳細次頁）
28© 2022 Fujitsu Limited

strict aliasing○ データ型種別を利用したデータ依存の解消

下記プログラムにおいて、○文法通りだと、a[i]への代入は、b[m]の領域を破壊するかもしれない➡b[m]の不変式不可、SIMD化も不可○しかし、そんなコードは通常書かない
void foo(double *a, long *b, int n, int m) {for (int i=0; i<n; ++i) {a[i] = a[i] + (double)b[m];}}○ trad -Kstrict_aliasing：個別オプション指定時動作○ clang -fstrict_aliasing ：-O2以上で動作（注）上例(C言語)は文法のrestrict指定を推奨。
また、この例では、-Krestp=argでも最適化可能
29b[m]のloadを最内
ループの外に追い出し、
SIMD化© 2022 Fujitsu Limited

プログラムチューニングによる性能改善⚫データ依存の解消⚫コーディングの工夫

30© 2022 Fujitsu Limited

データ依存の解消（C/C++）○ 自動SIMD化(or 自動並列化)できない原因の多くが、データ依存解析の問題○ コンパイラがデータ依存を見抜けない場合、restrict指定、最適化指示行、OpenMPによるデータ依存の解消が必要

void foo(double *a, double *b, int n) {for (i=0; i<n; ++i) {a[i] = a[i] + b[i];}}void foo(double *a, double *b, int n) {#pragma loop novrec// #pragma clang loop vectorize(assume_safety)for (i=0; i<n; ++i) {a[i] = a[i] + b[i];}}void foo(double * restrict a, double * restrict b, int n) {for (i=0; i<n; ++i) {a[i] = b[i] + c[i]*d[i];}}void foo(double *a, double *b, int n) {#pragma omp simdfor (i=0; i<n; ++i) {a[i] = a[i] + b[i];}}31© 2022 Fujitsu Limited

データ依存の解消（Fortran）○ Fortranのデータ依存の解消は、C/C++同様の最適化指示子「!ocl novrec」やOpenMP記述の他に、以下のような文法的対応が可能

subroutine foo(a,b,c,n)real(kind=8),dimension(:),pointer :: a,b,c
do i=1,na(i) = b(i) + c(i)subroutine foo(a,b,c,n)
enddo
real(kind=8),dimension(:),pointer,contiguous :: a,b,c
end subroutine foo
do concurrent(i=1:n)a(i) = b(i) + c(i)enddo
end subroutine fo

do concurrentがお勧めなのですが…
※コンパイラに障害があり、最適化が促進されていない。
今年度末～来年度頭で提供するパッチで対応予定
32© 2022 Fujitsu Limitedコーディングの工夫○ C/C++の２次元以上の配列をFortranと同じように連続領域として確保し、
restrict指定で最適化促進

void boo(int n, double **a, double **b) {for (int j=0; j<n; ++j) {for (int i=0; i<n; ++i) {a[j][i] = a[j][i] + b[j][i];}}a[j]のload + a[j][i]へのstore}

高速化

void foo(int n) {double **a = (double **)malloc(sizeof(double *)*n);double **b = (double **)malloc(sizeof(double *)*n);for (int i=0; i<n; ++i) {a[i] = (double *)malloc(sizeof(double)*n);b[i] = (double *)malloc(sizeof(double)*n);}
calc_a(n,a);nxnの配列生成
calc_b(n,b);boo(n, a, b);}

void boo(int n, double (* restrict a)[n], double (* restrict b)[n]) {for (int j=0; j<n; ++j) {for (int i=0; i<n; ++i) {a[j][i] = a[j][i] + b[j][i];}}a[j][i]へのstore}
void foo(int n) {double (*a)[n] = (double (*)[n])malloc(n*sizeof(double[n]));double (*b)[n] = (double (*)[n])malloc(n*sizeof(double[n]));nxnの配列生成
calc_a(n,a);calc_b(n,b);boo(n, a, b);}33© 2022 Fujitsu Limitedチューニング事例⚫ABINIT_MPアプリのチューニング事例

34© 2022 Fujitsu Limitedチューニング事例○対象アプリケーション○ABINIT-MP➢ フラグメント分子軌道（FMO）計算を高速に行えるソフトウェア➢ 4体フラグメント展開（FMO4）による2次摂動計算も可能➢ MPIおよびOpenMPにより超大規模計算にも対応➢ http://www.cenav.org/abinit-mp-open_ver-2-rev-4/○測定環境○計算機: FX1000 CPU:A64FX(2.0GHz x 48cores)○コンパイラ: Fujitsu Fortran compiler 4.5.0 tcsds-1.2.31○MPI: Fujitsu MPI

35© 2022 Fujitsu Limited

コスト分布○プログラム全体のコスト分布○基本プロファイラによるプロセス0番、スレッド0番のコスト分布○2電子積分処理が全体の約半分を占める
ただし、81種の処理の総和であるため、1種あたりのコストは1%前後と非常に小さい○通信に関連したコストは8%程度と小さい

システム関連
6%

その他
9%

2電子積分
53%

通信関連
8%

2電子積分：81種のサブルーチン(sub_*)のコスト総和
Direct SCF：サブルーチンdirect_scf_gmatのコストリスト作成：3種のサブルーチン(get_tei_rs_fix, get_tei_pq_fix,
リスト作成
get_ixijcs_to_proc_pqfix)のコスト総和
10%
通信関連：通信に関連した処理(putofu_*, opal_*, mca_*)のコスト総和
システム関連：ライブラリやOSなどに関連した処理のコスト総和
その他：上記以外の処理の総和
Direct SCF
14%36© 2022 Fujitsu Limited

【ご参考】コスト分布○81種の2電子積分処理コスト○プログラム全体に対するコスト比率が2%以上の処理は上位4個○上位4個のコストを合わせるとプログラム全体に対して13.2%

sub_pdsd_
sub_pddd_
sub_pdds_
sub_pdpd_
sub_pdsp_
sub_ddpd_
sub_ddsd_
sub_dspd_
sub_dssd_
sub_dddd_
sub_pdss_
sub_ddds_
sub_dpsd_
sub_pddp_
sub_psdd_
sub_ddps_
sub_dsdd_
sub_ddsp_
sub_ssss_ 6.2%
sub_pdpp_
sub_sdpd_
sub_ssdd_
sub_pdps_
sub_pspd_
sub_ddss_
sub_dsds_
sub_dsdp_
sub_dddp_
sub_sddd_
sub_dpdd_
sub_sppd_
sub_ddpp_
sub_dssp_
sub_psdp_
sub_ssps_ 2.4%
sub_sdsd_
sub_spdd_
sub_dpds_
sub_pssd_
sub_dppd_
sub_sddp_
sub_spss_ 2.3%
sub_ppsd_
sub_sdds_
sub_spsd_
sub_dsps_
sub_spdp_
sub_sssp_ 2.2%
sub_sspd_
sub_pppd_
sub_ppdd_
sub_spds_
sub_sspp_
sub_dspp_
sub_dpsp_
sub_dpdp_
sub_psss_
sub_ppds_
sub_psds_
sub_ssdp_
sub_sdss_
sub_sdsp_
sub_sdps_
sub_ssds_
sub_dpss_
sub_dpps_
sub_ppss_
sub_ppdp_
sub_dsss_
sub_spsp_
sub_ppsp_
sub_psps_
sub_dppp_
sub_pspp_
sub_sssd_
sub_pssp_
sub_spps_
sub_sdpp_
sub_ppps_
sub_sppp_
sub_pppp_※数値はプログラム全体に対するコスト比率37© 2022 Fujitsu Limited

2電子積分処理のソースコード分析○ ループ構造分析○ 81種の処理全てで下記の2重ループ構造◼ sub_ssssのループ最内部の実行回数
1,000,000

100.0%

900,000

90.0%

頻度

800,000
700,000

70.0%

600,000

60.0%

500,000

50.0%

400,000

40.0%
頻度300,000○ sub_ssssでは（プロセス0番、スレッド0番）

80.0%

累積頻度

30.0%

200,000

20.0%

100,000

10.0%

0

0.0%
0

○ sub_ssssは4,425,459回実行される○ ngij、ngklは36

累積頻度

do npq=1,ngij
if (abs(dkabm(npq)) > tv) thendo nrs=1,ngkl
if (abs(dkabm(npq)*dkcdm(nrs)) > tv) then
[ループ最内部：2電子積分計算]
end ifend doend ifend do

20

40

60

80

100

120

ループ最内部の実行回数

ループ最内部は最大でも1296回(36*36)の実行
ただし、条件判定により1296回よりも減少する

38• 実行回数40回以下が全実行の90%を占める• 仮に、二重ループを一重化してソフトウェアパイプライ
ニングなどの最適化を促進させようとしても、その効果は限定的と推測される© 2022 Fujitsu Limited

sub_ssssの最適化状況分析（1/2）○ 最適化情報を確認○ 最適化情報
39
40
41
42

sint(1)

1
2

43
44
45
46
47
48
49
50
51
52
53
54
55

3
4
4
4
4
4
4
4
4
4
5
5
5

◼ 最適化メッセージ（一部抜粋）= 0.0_8

do npq=1,ngij
if (abs(dkabm(npq)) > tv) then<<< Loop-information Start >>><<< [OPTIMIZATION]<<< PREFETCH(HARD) Expected by compiler :
<<<
dkcdm
<<< Loop-information End >>>s
do nrs=1,ngkl
mif (abs(dkabm(npq)*dkcdm(nrs)) > tv) thensze = 1.0_8/(zetam(npq)+etam(nrs))sa0 = dkabm(npq)*dkcdm(nrs)*sqrt(ze)srz = etam(nrs)*zesrho = zetam(npq)*rzstt = ((qm(1,nrs)-pm(1,npq))*(qm(1,nrs)-pm(1,npq)) &+(qm(2,nrs)-pm(2,npq))*(qm(2,nrs)-pm(2,npq)) &+(qm(3,nrs)-pm(3,npq))*(qm(3,nrs)-pm(3,npq)))*rhosss• jwd6229s-i "integral/sub_ssss.F90", line 43: IF文が存在す
るため、このDOループのSIMD化を抑止しました。
• jwd8670o-i "integral/sub_ssss.F90", line 43: ループ内に分岐命令があるため、ソフトウェアパイプライニングを適用できません。
• jwd8209o-i "integral/sub_ssss.F90", line 49: 多項式の演算順序を変更しました。
• jwd8209o-i "integral/sub_ssss.F90", line 56: 多項式の演算順序を変更しました。

コンパイラがSIMD化を抑止している

if (tt <= 36.0_8) then ! Tf = 2*m+36 (for the case of m=0)
ts = 0.5_8+tt*fmt_inv_step_size
delta = ts*fmt_step_size-tt⇒ OCL指示子から、SIMD化を促進する

SIMD化されていない
39© 2022 Fujitsu Limited

sub_ssssの最適化状況分析（2/2）○ OCL指示子でSIMD化を促進○ チューニング後の最適化情報
39
43
44
45
46
47

48
49
50
51
52
53
54
55
56
57
58
59
60

sint(1)

1
2

3
4
4
4
4
4
4
4
4
4
5
5
5

= 0.0_8!ocl noswp!ocl eval_concurrent!ocl SIMD
do npq=1,ngij
if (abs(dkabm(npq)) > tv) then<<< Loop-information Start >>><<< [OPTIMIZATION]<<< SIMD(VL: 8)<<< PREFETCH(HARD) Expected by compiler :<<<dkcdm, etam, qm<<< Loop-information End >>>v
do nrs=1,ngkl
v
if (abs(dkabm(npq)*dkcdm(nrs)) > tv) thenvze = 1.0_8/(zetam(npq)+etam(nrs))v
a0 = dkabm(npq)*dkcdm(nrs)*sqrt(ze)vrz = etam(nrs)*zev
rho = zetam(npq)*rzv
tt = ((qm(1,nrs)-pm(1,npq))*(qm(1,nrs)-pm(1,npq)) &+(qm(2,nrs)-pm(2,npq))*(qm(2,nrs)-pm(2,npq)) &+(qm(3,nrs)-pm(3,npq))*(qm(3,nrs)-pm(3,npq)))*rhovvv

OCL指示子

説明

noswp

ソフトウェアパイプライニング機能を無効に

eval_concurrent

tree-height-reduction最適化において演算の並列性を優先することを指示

SIMD

SIMD 化を有効に✓ noswpを指定している理由
・ ngklは36で固定
・ swpは回転数が40以上のループにしか適用されない◼ チューニング前後の性能比較
実行時間[s]asis
9.17

if (tt <= 36.0_8) then ! Tf = 2*m+36 (for the case of m=0)
ts = 0.5_8+tt*fmt_inv_step_size
delta = ts*fmt_step_size-tt

tune
4.43

性能比 対asis
tune
2.07

SIMD化された40© 2022 Fujitsu Limited

sub_sspsの最適化状況分析（1/4）○ sub_ssssと同じチューニングを試みる◼ 最適化情報
40
41
42
43

44
45
…

51
52
53
54
55

f

1
2

sint(1:3)◼ 最適化情報= 0.0_8

3
4

do npq=1,ngij
if (abs(dkabm(npq)) > tv) then<<< Loop-information Start >>><<< [OPTIMIZATION]<<< PREFETCH(HARD) Expected by compiler :
<<<
dkcdm
<<< Loop-information End >>>s
do nrs=1,ngkl
mif (abs(dkabm(npq)*dkcdm(nrs)) > tv) then

5
5
5
5
5

<<< Loop-information Start >>><<< [OPTIMIZATION]<<< FULL UNROLLING<<< Loop-information End >>>fs
do i=1,3
fs
qc(i) = qm(i,nrs)-c(i)fspq(i) = qm(i,nrs)-pm(i,npq)
fswq(i) =-re*pq(i)fsend do

40

44
45
46
47
48

f

56
57
58
59
60

SIMD化されていない

41= 0.0_8

34!ocl noswp!ocl eval_concurrent!ocl SIMD
do npq=1,ngij
if (abs(dkabm(npq)) > tv) then<<< Loop-information Start >>><<< [OPTIMIZATION]<<< PREFETCH(HARD) Expected by compiler :
<<<
dkcdm
<<< Loop-information End >>>s
do nrs=1,ngkl
v
if (abs(dkabm(npq)*dkcdm(nrs)) > tv) then

5
5
5
5
5

<<< Loop-information Start >>><<< [OPTIMIZATION]<<< FULL UNROLLING
配列qc, wqを求める計算が<<< Loop-information End >>>SIMD化されていない
fs
do i=1,3fm
qc(i) = qm(i,nrs)-c(i)fvpq(i) = qm(i,nrs)-pm(i,npq)
fm
wq(i) =-re*pq(i)fv
end do1
2

oclを追加49
50…

sint(1:3)© 2022 Fujitsu Limited

sub_sspsの最適化状況分析（2/4）○ 最適化情報を観察○ 最適化メッセージ（一部抜粋）○ jwd6208s-i "integral/sub_ssps.F90", line 57: 定義引用の順序が分からないため、定義引用順序が逐次実行と変わる可能性があり、このDOループはSIMD化できません。(名前:qc)○ jwd6208s-i "integral/sub_ssps.F90", line 59: 定義引用の順序が分からないため、定義引用順序が逐次実行と変わる可能性があり、このDOループはSIMD化できません。(名前:wq)○ jwd6208s-i "integral/sub_ssps.F90", line 65: 定義引用の順序が分からないため、定義引用順序が逐次実行と変わる可能性があり、このDOループはSIMD化できません。(名前:f)○ 最適化メッセージより、定義引用の順序が分からないことが原因だと判明○ qc, wq, fは配列であるため、コンパイラが依存関係を認識できなかったと推察⇒ 配列をスカラ変数に変更し、依存関係を明示化する

42© 2022 Fujitsu Limited

sub_sspsの最適化状況分析（3/4）○ 配列qc,wq,fをスカラ変数に変更

qc1 = qm(1,nrs)-c(1)qc2 = qm(2,nrs)-c(2)qc3 = qm(3,nrs)-c(3)pq1 = qm(1,nrs)-pm(1,npq)
pq2 = qm(2,nrs)-pm(2,npq)
pq3 = qm(3,nrs)-pm(3,npq)
wq1 =-re*pq1
wq2 =-re*pq2
wq3 =-re*pq3
tt = (pq1*pq1+pq2*pq2+pq3*pq3)*rho※変更した行：赤色
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
…
81
82
83
84

do i=1,3
qc(i) = qm(i,nrs)-c(i)pq(i) = qm(i,nrs)-pm(i,npq)
wq(i) =-re*pq(i)end dott = (pq(1)*pq(1)+pq(2)*pq(2)+pq(3)*pq(3))*rho
if (tt <= 38.0_8) then
! Tf = 2*m+36 (for the case of m=1)
ts = 0.5_8+tt*fmt_inv_step_size
delta = ts*fmt_step_size-tt
f(0) = ((fmt_table(3,ts)*inv6 *delta &+ fmt_table(2,ts)*inv2)*delta &+ fmt_table(1,ts)) *delta &+ fmt_table(0,ts)
f(1) = ((fmt_table(4,ts)*inv6 *delta &+ fmt_table(3,ts)*inv2)*delta &+ fmt_table(2,ts)) *delta &+ fmt_table(1,ts)
else
t_inv = inv2/tt
f(0) = sqrt(pi_over2*t_inv)
f(1) = t_inv * f(0)end iff0 = ((fmt_table(3,ts)*inv6 *delta &f1 = ((fmt_table(4,ts)*inv6 *delta &f0 = sqrt(pi_over2*t_inv)
f1 = t_inv * f0

ssss0=f0*a0
ssss1=f1*a0
sint(1) = sint(1)+qc1*ssss0+wq1*ssss1
sint(2) = sint(2)+qc2*ssss0+wq2*ssss1
sint(3) = sint(3)+qc3*ssss0+wq3*ssss1

ssss(0:1)=f(0:1)*a0
do k=1, 3
sint(k) = sint(k)+qc(k)*ssss(0)+wq(k)*ssss(1)end doオリジナルコード変更後
43© 2022 Fujitsu Limited

sub_sspsの最適化状況分析（4/4）○ チューニング結果の確認○ 最適化情報
53
54
55
56
57
58
59
…
65
66
67
68
69
70
71
72
73
…
75
…
80
81
82
83
…

3
4
4
4
4
4
4

v
v
v
v
v
v
v

4
4
4
4
4
4
4
4
4

v
v
v
v
v
v
v
v
v

qc1 = qm(1,nrs)-c(1)qc2 = qm(2,nrs)-c(2)qc3 = qm(3,nrs)-c(3)pq1 = qm(1,nrs)-pm(1,npq)
pq2 = qm(2,nrs)-pm(2,npq)
pq3 = qm(3,nrs)-pm(3,npq)
wq1 =-re*pq1
wq2 =-re*pq2
wq3 =-re*pq34v

tt = (pq1*pq1+pq2*pq2+pq3*pq3)*rho

5
5
5
5

v

f0 = ((fmt_table(3,ts)*inv6 *delta &+ fmt_table(2,ts)*inv2)*delta &+ fmt_table(1,ts)) *delta &+ fmt_table(0,ts)

…
85
86
87
88
…
93
94
…
105
106
…
110
111
112
…
114
115

do nrs=1,ngkl
if (abs(dkabm(npq)*dkcdm(nrs)) > tv) thenze = 1.0_8/(zetam(npq)+etam(nrs))
a0 = dkabm(npq)*dkcdm(nrs)*sqrt(ze)rz = etam(nrs)*ze
re = zetam(npq)*ze
rho = zetam(npq)*rz

5
5
5
5

v

f1 = ((fmt_table(4,ts)*inv6 *delta &+ fmt_table(3,ts)*inv2)*delta &+ fmt_table(2,ts)) *delta &+ fmt_table(1,ts)

5
5

v
v

f0 = sqrt(pi_over2*t_inv)
f1 = t_inv*f0

4
4

v
v

ssss0=f0*a0
ssss1=f1*a0

4
4
4

v
v
v

sint(1) = sint(1)+qc1*ssss0+wq1*ssss1
sint(2) = sint(2)+qc2*ssss0+wq2*ssss1
sint(3) = sint(3)+qc3*ssss0+wq3*ssss1

4
3

v
v

end ifend do◼ チューニング前後の性能比較
実行時間[s]asis

依存がないことが認識され、SIMD化された

3.56
44

tune

2.38

性能比 対asis
tune

1.49
© 2022 Fujitsu Limited

ABINIT-MP チューニングまとめ○ソースコードの分析○基本プロファイラにより最も実行時間を要している個所を特定する○ソースコード等から処理の特徴を見極め、チューニングの指針を決定する○ループ回転数等の条件によってはソフトウェアパイプライニングを諦める○SIMD化による高速化○ソースリストの最適化出力情報からSIMD化の状況を確認する○コンパイラによる自動SIMD化が阻害されている場合には、最適化メッセージから原因を特定し、指示行の挿入などの対策を実施する○謝辞

本チューニング事例は、JHPCNのjh210036-NAH課題のご協力のもと、
立教大学 望月祐志教授よりABINIT-MPのソースコードをご提供いただき、
SS研「A64FXシステムアプリ性能検討WG」での活動成果を再構成しました。
45© 2022 Fujitsu Limited

その他の事例のご紹介○GitHubでのA64FX向けアプリケーションの公開○https://github.com/fujitsu○A64FX向けOSSパッチ
FrontISTRLAMMPS

MPAS

OpenFOAM

Quantum Espresso

SPECFEM3D○AIフレームワーク
tensorflow

pytorch○HPCAsia2022 Best Paper Awardの受賞○152K-computer-node parallel scalable implicit solver for dynamic

nonlinear earthquake simulation○https://sighpc.ipsj.or.jp/HPCAsia2022/index.html#awards○「富岳」のコデザイン活動を通じて、有限要素法ソルバに対する単体性能およ
び並列性能チューニングで貢献
46© 2022 Fujitsu Limited

Thank you© 2022 Fujitsu Limited