2014年8月26日SS研HPCフォーラム

新ベンチマークプログラム：

HPCGの概要と「京」における性能
理化学研究所 計算科学研究機構&
運用技術部門 ソフトウェア技術チーム&!
南一生

RIKEN&ADVANCED&INSTITUTE&FOR&COMPUTATIONAL&SCIENCE

2
OUTLINE2014年8月26日SS研HPCフォーラム• はじめに• LINPACK• HPCGとは• HPCGベンチマークプログラム• HPCGのチューニングと性能

2014年8月26日SS研HPCフォーラム3

2014年8月26日SS研HPCフォーラムはじめに

2014年8月26日SS研HPCフォーラム

4
スパコン世界一をめぐる競争
2014年8月26日SS研HPCフォーラム

スパコン世界一の推移

A*cjnl&.Sequoia「京」 Titan 天河2

A
A@@@



A@@



A@



A



A@@@



A@@



A@



平均すると1年に約1.9倍の性能向上！

「京」は、世界初のスパコンCRAY-1（1976年）の約7000万倍！


2014年8月26日SS研HPCフォーラム

ちなみに2014年8月26日SS研HPCフォーラム5
世界初のスパコンCRAY-1の演算性能は、約160メガフロップス
一方、iPhone4Sの演算性能は、約140メガフロップス

AEDC

B@AA
b

2014年8月26日SS研HPCフォーラム

6
最近のスパコンランキングの推移
2014年8月26日SS研HPCフォーラム

Jun. 2011>

Nov.2011>

Systemcountry

PFLOPS

Systemcountry

1

K computerJPN> 8.16

K computer
JPN

2

TianheA
CHN

> 2.57

TianheA
CHN

3

JaguarUSA

1.76

JaguarUSA

4

Nebulae
CHN

1.27

NebulaeCHN

5

TSUBAME2.0
JPN

> 1.19TSUBAME2.0
JPN

>
>

Jun. 2012>

PFLOPS

Systemcountry> 10.51

SequoiaUSA> 2.57

K computer
JPN

>

PFLOPS>

SystemcountryTianhe-2(CHN)> 16.32

Titan
USA

> 10.51

Sequoia
USA

8.16

K computer
JPN

Sequoia
USA

8.16

K computer
JPN

1.27

SuperMUC
GER

> 1.19TianheA
CHN

>

17.59

> 10.51>

Jun. 2014
Systemcountry

Titan
USA

Mira
USA

Jun. 2013〜

PFLOPS> 16.32

1.76>

Nov. 2012>> 2.90

Mira
USA

>

> 2.57JUQUEEN
GER

> 4.14PFLOPS> 33.86>

17.59> 17.17> 10.51

>

Mira
USA

8.59

2014年8月26日SS研HPCフォーラム7

2014年8月26日SS研HPCフォーラム

LINPACK2014年8月26日SS研HPCフォーラム

High Performance 2014年8月26日SS研HPCフォーラム
LINPACK(HPL) 8• 世界で最も高速なコンピュータシステムの上位500位までを定期的にランク付
けし、評価するプロジェクトとしてTOP500がある&• TOP500は、米国のテネシー大学のJ.0Dongarra博士によって開発された密行列計
算による連立一次方程式を解くベンチマーク・プログラム：LINPACKが用いられている&• LINPACKを用いたTOP500プロジェクトは、1993年に発足&• LINPACKが浮動小数点演算性能のベンチマーク・テストに重きを置いている

2014年8月26日SS研HPCフォーラム

HPLのプロセス分割
9
2014年8月26日SS研HPCフォーラム「京」の例
N(8-"+l)p(1 vll



)

































N

q(1 vl
l0)nb( v

l1f6.io0m )



nb

2014年8月26日SS研HPCフォーラム

HPLの計算手順10
2014年8月26日SS研HPCフォーラム(1)Lパネル分解

p(2) 分解済みLパネル放送

nb

p

q

DTRSM
(三角行列計算)nbq>G[[XSaYW_>

0>QZ^W[Z\W>
T_]_VUS`a(3)Uブロック行放送とPivot行交換
p

nb
q

>
QZ^W[Z\W>
T_]SVUS`a>G[[XSaYW_

2014年8月26日SS研HPCフォーラム

HPLの計算手順11
2014年8月26日SS研HPCフォーラム(5) 列パネル更新(4) Uブロック分解

p

p

nb

nb
q

q

DTRSM
(三角行列計算)

DGEMM(行列行列積計算)の計算

この更新計算が計算時間の大半を占める2014年8月26日SS研HPCフォーラム

12
TOP500ベンチマークの成功と失敗
2014年8月26日SS研HPCフォーラム

(*)&HPCG WorkShop,25,March,2014,Bethesda, MD開催より• 成功– わかりやすさ– システム最大性能との線形相関• 問題点– 実問題との乖離• 発足して20年以上が経過• 実際のアプリケーションで求められる性能要件との乖離• Stunt machineの開発を助長する– 長すぎる実行時間• 途中からの実行など実行時間短縮案を検討中
2014年8月26日SS研HPCフォーラム13

2014年8月26日SS研HPCフォーラム

HPCGとは2014年8月26日SS研HPCフォーラム

14
HPCGベンチマークの提案
2014年8月26日SS研HPCフォーラム

(*)&HPCG WorkShop,25,March,2014,Bethesda, MD開催より• ベンチマーク自体の紹介– HPCG(High Performance Conjugate Gradient)– Version 2の紹介

• TOP500との関係– 置き換えるものではない• TOP500のリストに列を追加してHPCGスコアも表示するなどの案あり• WS参加者のなかでは将来的に置き換えるべきという意見もあった

2014年8月26日SS研HPCフォーラム

15
連立一次方程式と共役勾配法(CG法)2014年8月26日SS研HPCフォーラム

左図のような2次元の領域で微分方程式を解くとする．
例として5点差分で差分化．

元の微分方程式は，差分化により連立一次方程式を解く事に帰着．

Ax = bf (x) =

1 2
ax − bx2

f(x)の最小値問題

1
f (x) = (x,Ax) − (b,x)
2∂ f (x) 1 n
1 n= ∑ aik xi + ∑ akj x j − bk
∂xk
2 i=1
2 j=1

f '(x) = ax − b = 0

本式を解く事と同値∂ f (x) n= ∑ aki xi − bk = 0∂xk
i=1
2014年8月26日SS研HPCフォーラム16
共役勾配法(CG法)2014年8月26日SS研HPCフォーラム1
f (x) = (x,Ax) − (b,x)
2CG法アルゴリズムの基本形．
この他にアルゴリズムの派生形あり．
f(x)が最小値が得られるようにxiの列を求める．
ri は残差ベクトルであり互いに直交し一次独立である．
残差ベクトルはN元の連立一次方程式にはN個しか存在しない．
N元の連立一次方程式は高々N回の反復で収束する．
またAの固有値が縮重しているか密集していれば収束が速くなる性質を持つ．
2014年8月26日SS研HPCフォーラム17
前処理付き共役勾配法(CG法)2014年8月26日SS研HPCフォーラム

M = UT U

Ux = x!

U b = b!
−T!
U −T AU −1 = A

Aに近い正値対称行列:Mを考えコレスキー
分解する．

!
このようなx,b,Aの置き換えを行う．

!
!
置き換えたAの性質は単位行列に近い．
置き換えた行列Aの固有値は1の周りに密集している! = U −T AU −1 ≈ I
A! x! = b!
A!
置き換えた行列とベクトルに対する連立一次方程式の収束は早いものと期待．

この前処理によりCG法が広く使用されるようになった
2014年8月26日SS研HPCフォーラム18
前処理付き共役勾配法(CG法)2014年8月26日SS研HPCフォーラム

不完全コレスキー分解による前処理CG法アルゴリズムの基本形．
計算は行列ベクトル積・ベクトルスカラ積・ベクトルの和・内積・除算で構成される．
赤線で示した前処理部分は前進代入・後退代入で計算される．
1.ICCG法のアルゴリズム
ステップ1:

αk = (rik •(LL T )−1 rik )/(Apik • p ik )ステップ2:

xi = xi +α pi

ステップ3:

ri k+ 1 = ri k − αk Apik

ステップ4:

β k = (rik+1 •(LLT ) −1 rik+1 ) /(ri k •(LLT ) −1 rik )ステップ5:

pi = (LL ) ri + β pi

k+ 1

k+ 1

k

k

k

T −1 k+1

k k

2014年8月26日SS研HPCフォーラム

19
前処理としてのマルチグリッド法
2014年8月26日SS研HPCフォーラム

連立一次方程式の解法としてのマルチグリッド法の例
Rytv~w
qtw

x|w

2014年8月26日SS研HPCフォーラム

20
前処理としてのマルチグリッド法
2014年8月26日SS研HPCフォーラム

連立一次方程式の解法としてのマルチグリッド法の例(NPB MG)
マルチグリッドも行列ベクトル積の計算が中心

Rytv~w
qtw
;1:

fk = Rk+1→k fk+1
fk−1 = Rk→k−1 fk
fk−2 = Rk−1→k−2 fk−1
f1 = R2→1 f2
x|w

Ak+1uk+1 = rk+1

rk+1 = fk+1 − Ak+1uk+1

uk+1 = Rk→k+1uk

Ak uk = rk 75$5
rk = fk − Ak uk #6.

uk = Rk−1→k uk−1 91:
2014年8月26日SS研HPCフォーラム

21
LINPACKとHPCGの単体性能比較
2014年8月26日SS研HPCフォーラム

High Performance

A Processor Performance

OMPQGINF>
1

0

2uz uw/3

3$H?Jfhsfz4

3$H?Jfdfsf2<

5

3$H?Jf=d

6

3$H?Jf=h{rv}{p'

0

,k

ge/

4!0 l8-"+l )5%>0 0 ,l6.fLQIKF>
4!(0 l8-"+l 5%>(0 v,fLow Performance
2014年8月26日SS研HPCフォーラム22

2014年8月26日SS研HPCフォーラム

HPCGベンチマークプログラム

2014年8月26日SS研HPCフォーラム

HPCGベンチマーク
2014年8月26日SS研HPCフォーラム

23• 連立一次方程式 Ax = b を前処理付CG法で解く際の性能をベ
ンチマーク– 前処理はVer.2.0以降はMulti-Grid法を採用(Ver1.1まではGauss-Seidel法)• 係数行列Aは有限要素法の行列– 疎で対称– 具体的には直方体領域を規則的に分割したもの• 特徴– LINPACKと比して通信・メモリアクセスの比重が高い– メモリアクセスはリストアクセスであり，より高いメモリ性能が必要– C++で書かれているが便利なCとしてしか使っていない
2014年8月26日SS研HPCフォーラム

HPCGベンチマークアプリの概要
2014年8月26日SS研HPCフォーラム
前処理

24

初期化

!
Problem&Setup&Phase

検証

Validation&Testing&Phase
Reference&SpMV+MG&Timing&Phase
Reference&CG&Timing&Phase

測定

Optimized&CG&Setup&Phase
Optimized&CG&Timing&Phase

後処理• 大まかに4つの部分に分けられる

Report&Results■前処理• メモリ確保などの初期化• 問題設定，行列の最適化!
���■検証• ユーザーチューニングの正当性検証• 雑多なデータ測定!
■測定• 実行時間の予測• 本番測定 (1時間)!
■後処理
ここでは各部分の主要な処理に• 結果を整理し提出するデータを作成

ついて紹介

2014年8月26日SS研HPCフォーラム

HPCGベンチマークアプリの概要
2014年8月26日SS研HPCフォーラム

25

前処理

初期化

• XYZ各軸に割り当てるプロセス数の決定
GenerateGeometry()&関数&• 連立方程式Ax=bを作成

Problem&Setup&Phase

GenerateProblem()&関数&•

Validation&Testing&Phase•

行列をプロセス内ローカルなものに変換&
袖通信用の通信相手の情報を作成

検証

SetupHalo()関数&

Reference&SpMV+MG&Timing&Phase•

Multi^Gridのデータ構造を作成
GenerateCoarseProblem()&関数&

Reference&CG&Timing&Phase•

ユーザー定義のデータ最適化
OptimizeProblem()&関数
標準では何もしない

測定

Optimized&CG&Setup&Phase

これを独自に実装して&最適化を行う&所要時間は計測される

後処理

Optimized&CG&Timing&Phase
Report&Results

2014年8月26日SS研HPCフォーラム

GenerateProblem
関数の処理
2014年8月26日SS研HPCフォーラム26

連立一次方程式Ax=bを設定する (この時点ではグローバルな節点番号で作成)
48

P0

42

P1

24

3×3×3&節点&
2プロセスの場合
節点番号はx→y→zの順&
②行の内容
列番号
値

z

25

18

19

6

x

0

1
②と隣接

2

節点②に&

16

9
3

29
23

15

8

35

28
22

14

7

34

27

21

13

41

33

②と隣接

20

12

y

26

47

40

32

53

46

39

38
31

52

45

44

37
30

51

50

43

36

例)&

に動く

49

17

10
4

11
5

相当する&
行に着目

0 1 2 3 4 5 6 7 8 9 10 11 12 13 14 15 16 17 18 19 20 21 22 23 24 25 26 27 28 29 ･･･ 53-1 26 -1
-1 -1 -1
-1 -1 -1
-1 -1 -1

赤字は外部節点への参照

実際の格納形式 ELL形式
・・ 26 27 非ゼロ個数は 27個/行 に固定
非ゼロ番号 0 1 2 3 4 5 6 7 8 9 10 11 12 ・・・・
global_int_t型(実体はint)
列番号
1 2 3 7 8 9 19 20 21 25 26 27・・・
doulbe型
値-1 26 -1 -1 -1 -1 -1 -1 -1 -1 -1 -1

2014年8月26日SS研HPCフォーラム

27

GenerateCoarseProblemの処理
2014年8月26日SS研HPCフォーラム
MG用の粗いメッシュに対応する係数行列を作成
粗い格子は3レベル作成して&初期メッシュ

合計4レベルの格子データ作成&
プロセス①,
Generate,
Geometry

粗いメッシュ

XYZの節点数を1/2にしてデータ作成プロセスを先頭から再度呼ぶ

行列データの構造体

行列データの構造体

格子の情報

格子の情報
①,
Generate,
Geometry• XYZの節点数&• XYZのプロセス形状&
• etc
行列の情報

行列の情報

• 行数，行毎の非ゼロ数&②,
Generate,
Problem②,
Generate,
Problem• 列番号，非ゼロ値&• etc
袖通信の情報③,
SetupHalo• 隣接相手&• 送信節点情報
粗い行列データへのポインタ&
細←→粗間の節点対応表• 行数，行毎の非ゼロ数&• 列番号，非ゼロ値&• etc
袖通信の情報③,
SetupHalo• 隣接数&④,
Generate,
CoarseProblem• XYZの節点数&• XYZのプロセス形状&
• etc

④,
Generate,
CoarseProblem• 隣接数&• 隣接相手&• 送信節点情報
粗い行列データへのポインタ&
細←→粗間の節点対応表

2014年8月26日SS研HPCフォーラム

HPCGベンチマークアプリの概要
2014年8月26日SS研HPCフォーラム
前処理

28

初期化
Problem&Setup&Phase

検証

Validation&Testing&Phase
Reference&SpMV+MG&Timing&Phase
Reference&CG&Timing&Phase

演算関数は&
2種類存在

チューニング版

測定
後処理

Report&Results

ComputeCG_ref()
標準では単にラッパー

ComputeCG()

カーネルのチューニングや行列の最適化が正しいかをチェック!

TestCG()• チューニング版CG関数に，強く対角優位な行列を与えて，規程回数で収束するかを判定(収
束判定値1.0×10-12)

Optimized&CG&Setup&Phase
Optimized&CG&Timing&Phase

基準版!

• 前処理なしなら12回以内• 前処理ありなら2回以内

条件は
決め打ち

TestSymmetry()• OptimizeProblem()で最適化した行列が対称になっているかチェック
2014年8月26日SS研HPCフォーラム

HPCGベンチマークアプリの概要
2014年8月26日SS研HPCフォーラム
前処理

29

初期化

基準版のSPMV(疎行列ベクトル積)とMG(MultiGrid)関数の実行時間を計測

Problem&Setup&Phase!

ComputeSPMV_ref()• 基準版の疎行列ベクトル積��

検証

Validation&Testing&Phase!

ComputeMG_ref()• 基準版のMG

Reference&SpMV+MG&Timing&Phase!

この時のSPMVとMGの1回当りの平均時間を求める

Reference&CG&Timing&Phase!

Optimized&CG&Setup&Phase

測定

Report Result部で� OptimizeProblem()の時間と，この平均
� 時間との比を出力している

Optimized&CG&Timing&Phase

後処理

10回

Report&Results
2014年8月26日SS研HPCフォーラム

HPCGベンチマークアプリの概要
2014年8月26日SS研HPCフォーラム
前処理

30

初期化
Problem&Setup&Phase

検証

Validation&Testing&Phase
チューニング版CGを1回実行し，要した時間を測定!

Reference&SpMV+MG&Timing&Phase

測定

Reference&CG&Timing&Phase
Optimized&CG&Setup&Phase
1回実行時間後処理

Optimized&CG&Timing&PhaseCG()• チューニング版CG• 反復上限は500回
収束判定はCG_ref()を実行して求めた
ref_tolerance(1.0E-3程度)�
• 1回だけ実行!

この時間を元に，次のフェーズでCG()を何回実行するかを決定Report&Results
2014年03月12日(水)HPCGの京における状況説明

2014年8月26日SS研HPCフォーラム

HPCGベンチマークアプリの概要
2014年8月26日SS研HPCフォーラム

31

本番測定
チューニング版CGを複数回実行し測定

Problem&Setup&Phase!

検証

Validation&Testing&Phase
Reference&SpMV+MG&Timing&Phase
Reference&CG&Timing&Phase

測定

CG()
• チューニング版CG• 反復上限は先の1回実行の反復数(収束判定は0なので必ず指定回数反復する)• 合計実行回数は，先に計った1回の実行
時間を元に，入力データで指定した時間となるよう決定

提出は1時間以上必要!

Optimized&CG&Setup&Phase
1回実行時間Optimized&CG&Timing&Phase

後処理

解は実行の度にゼロクリア

複数回

前処理

初期化

TestNorms()• 再現性・安定性のチェック!
!
• CGの実行毎に r50r0 2の値を記録
2
しておき，ブレがないかチェック��

Report&Results

分散が1.0×10^6以下なら合格

2014年8月26日SS研HPCフォーラム

MGの処理

前進Sweep&&&k =&1〜N!
!
z = MG ( A, r )


!
z =0



後退Sweep&&&k=N〜1



!

32

N
N
#
1 &
( ) 2014年8月26日SS研HPCフォーラム(k + 1 )zi= %ri − ∑ aij z j 2 − ∑ aij z (jk )"
aii $
j <i
j >i
!
k + 12zi(k +1) =

!

N
N
#
1 &
(k + 12 )r−az−
aij z (jk +1)"
% i ∑ ij j
∑
aii $
j <i
j >i
!

Relaxation&&&&&&&&&&&&&&&&&&&&&&&by&&Gauss^Seidel&&&&(Pre^smoothing)Az = r

袖通信
!
!
z
最細
Az = A ⋅ z 袖通信
Z=0&Pre^smoothing&
細かい格子での値を!
!
Residual&rcoarse = Restrict (r − Az )粗い格子上に補間
MG& Restrict
再帰!
!
Prolongate
zcoarse = MG ( Acoarse , rcoarse ) (再帰呼び出し)
A
MG& !
補間)&
!
! !
!
再帰 r
粗い格子での値を&z
z = z + Prolongate( zcoarse )格子サイズが1:2の整数
細かい格子上に補間! !
倍なので，補間といってProlongate&
Relaxation&&&&&&&&&&&&&&&&&&&&&&&&by&&Gauss^Seidel&&&&(Post^smoothing)Az = rMG&も代入するだけPost^Smoothing
Restrict

coarsecoarsecorase

袖通信

! !
Relaxation&&&&&&&&&&&&&&&&&&&&&&&by&&Gauss^Seidel
Az = r

再帰

最粗
Relax

袖通信

2014年8月26日SS研HPCフォーラム

HPCGベンチマークアプリの概要
2014年8月26日SS研HPCフォーラム

33

初期化
Problem&Setup&Phase

提出する実行結果を出力(HPCG-Benchmark-2.1-年月日時刻.yaml)!

Optimized&CG&Timing&Phase

並列数・スレッド数
全体問題サイズ
プロセス形状
1ノード当りの問題サイズ
マルチグリッド各階層での行列の行数と非ゼ
ロ数• チューニングの正当性検証結果• 再現性チェック結果• 測定結果(各カーネルと全体)• Elapse時間• 理論的な浮動小数点演算数• GFLOPS値

Report&Results

検証フェーズ★で不合格があればFAILED, 正
常ならVALID＋GFLOPS値が出力される

Validation&Testing&Phase
Reference&SpMV+MG&Timing&Phase
Reference&CG&Timing&Phase
Optimized&CG&Setup&Phase

•
•
•
•
•

!

2014年8月26日SS研HPCフォーラム34

2014年8月26日SS研HPCフォーラム

HPCGのチューニングと性能

2014年8月26日SS研HPCフォーラム

HPCG テスト実行on 京
2014年8月26日SS研HPCフォーラム

35•

コードはAs Isそのまま– ただし大並列への対応として一部の変数の型を変更– Geometry.hpp中のtypedefされている型をintからlong longに変更•

Make方法– QUICKSTART.TXTに従って実行

•

テスト問題
– デフォルトの104×104×104=112万節点/プロセスを利用– 測定時のCG繰り返し回数は5回 (本来は1時間分で約50回)

2014年8月26日SS研HPCフォーラム

ウィークスケール測定結果
2014年8月26日SS研HPCフォーラム36

反復上限は50回
CG実行回数は5回に固定大並列で動かすよう一部変数の型を変更

15222GFLOPS&(0.36%)
9835GFLOPS&(0.23%)

8並列→32768並列の4096倍でGFLOPS値は並列化効率
4054倍 (ベンチ結果)
4040倍 (プロファイラ)
98%

8スレッド/プロセス• 8並列から32768並列まで並列数を換え測定• ベンチ結果��HPCGコードが自身で算出した(提出する) GFLOPS値• プロファイラ プロファイラを利用して算出した GFLOPS値• 32768並列までの範囲ではウィークスケール性能は良好

3回平均

ベンチ結果とプロファイラとで1.5倍程度の差がある↓
浮動小数点演算数カウント方式の違いが原因
2014年8月26日SS研HPCフォーラム

37

測定値のプロファイラとの違いについて
2014年8月26日SS研HPCフォーラム
プロファイラでの浮動小数点演算数•

測定区間で発生する浮動小数点演算全てがカウント対象•

除算等は複数の浮動小数点演算としてカウント

HPCGでの浮動小数点演算数•

理論的に必要な演算のうち，主要なもののみをカウントしており，無視しているものもある! !
!
x = CG( A, b , x0 , ε , max)

! !
!
x = SYMGS( A, b , x )

SPMV()
 !
WAXPBY()  !
※V2.4ではカウントするようになった&
DotProductfor(k=1; k<max && !conv; k++){MG()
DotProduct() !
WAXPBY()
SPMV()
}

for(i=0; i< ; i++){for(j=0; j<nz[i]; j++){col = A.mtxIndL[i][j] sum -= val[j] * x[col];}sum += x[i] * Diag[i];sum = sum / Diag[i];}  

2014年8月26日SS研HPCフォーラム

通信の比率変化2014年8月26日SS研HPCフォーラム

38

4x4x4
8x8x8

通信はおおむね6%程度

2x2x2

実行時のプロセス数指定はGenerateGeometryで決定した3次元形状に合わせた
3回平均• プロファイラで計測したMPIのコストの，対全コスト比率• 通信のコストの割合は，512並列以上ではサチってきており，合計6%程度

2014年8月26日SS研HPCフォーラム

HPCGコスト分析
2014年8月26日SS研HPCフォーラム

39

ComputeDotProduct_ref&&46^79

ExchangeHalo_ref&&59^70

ComputeSPMV_ref&&59^70

ComputeSYMGS_ref&&54^103

8192並列のもの(他もほぼ同じ)• プロファイラで採取したCG内で呼ぶ手続きのコスト分布• 上位4つで98%以上を占める(Allreduce, Isend, Irecv, Waitを含む)• チューニングはまず単体性能の向上
ComputeSYMGSとComputeSPMVが主眼に2014年8月26日SS研HPCフォーラム

単体性能チューニング
2014年8月26日SS研HPCフォーラム40

メモリ領域の連続化
メモリ確保部ソース

行列の各行の情報を保持する配列は

for(int i=0; i<nrow; ++i){
matrixValues[i] = new double[27];mtxIndL[i] = newint[27];}

で不連続になっている

メモリ空間

A.matrixValues[0]行毎にnewされているので，メモリ上

A.mtxIndL[0]A.mtxIndL[nrow^1]

…

…

…

A.matrixValues[nrow^1]
A.mtxIndL[1]…

…

A.matrixValues[1]…

2014年8月26日SS研HPCフォーラム

単体性能チューニング
2014年8月26日SS研HPCフォーラム41

メモリ領域の連続化
メモリ確保部ソース
double* tmp1 = new double[nrow * 27];
double* tmp2 = newint[nrow * 27];!

for(int i=0; i<nrow; ++i){
matrixValues[i] = &(tmp1[i*27]);
mtxIndL[i] = &(tmp2[i*27]);}

メモリ確保を一括して行い，各行の情報が連続するように変更&!

ムダな領域もなくなった

メモリ空間
A.matrixValues[0] A.matrixValues[1]…
…

………………

A.matrixValues[nrow^1]…

A.mtxIndL[0] A.mtxIndL[1]A.mtxIndL[nrow^1]…
…
………………

2014年8月26日SS研HPCフォーラム

単体性能チューニング
2014年8月26日SS研HPCフォーラム

42

ケース


実行時間



連続化

 


メモリスル スループ
ープット

ット



 
ミス率ミス率ロード・ストア ロード・ストア  


数

数

 


 
 

CG()から直接呼ばれるSPMVについて，
実行時間と，キャッシュミス率，スループッ
ト等を調べた※ MG()から呼ばれる分は除外!
×2.3高速化

メモリ領域の連続化の効果によりキャッシュミス率減少，スループット向上!
実行時間は1/2.3に短縮
8プロセス，8スレッド/プロセス
2014年8月26日SS研HPCフォーラム

その他のチューニングを実施
2014年8月26日SS研HPCフォーラム

43※CGから呼ばれるもののみ※MG以下のSPMV,SYMGS等の合計

16.2GFLOPS&
1.58%

2.3GFLOPS&
0.22%

X2.4

X3.2
39.1GFLOPS&
3.82%

7.2GFLOPS&
0.7%• HPCG自身が測定した提出するデータから抜粋したもの

2.6GFLOPS&
0.25%

X3.2
8.4GFLOPS&
0.82%

!

• SPMVの時間CG()から呼ばれる分のみの時間• 前処理MGはSPMV,SYMGS等々を含む時間• 2点の簡単なチューニングだけで3.2倍の高速化が得られた• 行列のメモリ領域の連続化• SYMGSのBAKループでのループ方向の反転!
!

2014年8月26日SS研HPCフォーラム

前処理のマルチスレッド化
2014年8月26日SS研HPCフォーラム

44

種別

関数名

行番号

ComputeSYMGS_ref&&93^96

ComputeSYMGS_ref
74^77
ComputeSPMV_ref&&67^68• プロファイラで採取した各手続きのループ単位のコスト比率• ComputeSYMGSのループはマルチスレッド化されておらず逐次で動作!
• ここではComputeSYMGSのマルチスレッド化を実施
2014年8月26日SS研HPCフォーラム45
カラーリング2014年8月26日SS研HPCフォーラム!

SYMGSは対称行列についてガウス・ザイデル法を使用したマルチグリッ
ド処理を行う．
i−1
n
(m+1)
(m+1)
)
= − ∑ aij x j− ∑ aij x (m+ kj
ガウス・ザイデル法 aii xi
j
j=1

j=1+1

(m+1)世代の解を得るために(m+1)世代の下三角行列を使用して(リカレンスの発生)．
(a)オリジナル(c)RED-BLACK２色カラーリング

参照関係
計算順序!
(スイープ順序)2014年8月26日SS研HPCフォーラム46
カラーリング2014年8月26日SS研HPCフォーラム

XYZ各方向の隣りと斜めを同時に参照する27点
ステンシルなので，8色必要
for(int i=0; i<nrow; i++){
最内ループ
SYMGSの外側ループ}
カラーを回す外側ループ追加で3重化
for(int ic=0; i<8; ic++){
リカレンスがないのでディレクティブ
挿入で並列化#pragma omp parallel for
for(int i=st; i<=ed ++){
最内ループ
}
8色

z y
x

}

演算順序を変更しカラー毎に処理することで並列化が可能に2014年8月26日SS研HPCフォーラム47
カラーリング2014年8月26日SS研HPCフォーラム

演算順序の変更に伴い，メモリ上での再配置も行い，メモリアクセスを連続化
1行分の情報値…

列番号27個のdouble…

27個のint

メモリ空間

行列の情報

1行

2行

3行

4行

5行

6行

7行

8行

9行

10行

…… N^1行

N行

5行

6行

7行

8行

9行

10行

…… N^1行

N行…… 100&行 101行 102行 103行……

N^2行 N^1行

N行

カラーリング

メモリ空間

行列の情報

1行

2行

3行

4行

並び替えて連続化

メモリ空間

行列の情報

1行

2行

3行

4行

2014年8月26日SS研HPCフォーラム48
カラーリング2014年8月26日SS研HPCフォーラム

1

5

2

6

1

5

9

13 10 149

13 10 14

3

7

3

7

4

8

11 15 12 16

2

4

6

8

11 15 12 16

メッシュ点のグループ

メモリ上配置
1

2

3

4

カラー1

5

6

7

8 ・・・

カラー2

・・・
9 10 11 12 13 14 15 16
カラー3• ここで説明した8色カラーリングは問題の構造的なメッシュ構造を前提としたプログラミングを禁止するレギュレーショ
ンに抵触する．
• そのためもっと一般的なカラーリング
手法に変更．
• メッシュ点を複数まとめてグループを構成する．
• そのグループ毎にカラーリングを実施．
• カラー内ではグループ間でリカレンス
は発生しない．
• そこをスレッド並列に利用．

カラー4
2014年8月26日SS研HPCフォーラム

49
カラーリング2014年8月26日SS研HPCフォーラム※&データサイズ1043→1123に変更

カラー化によりコスト上位のループが全て並列化OK

ComputeSYMGS
144^147
ComputeSPMV
87^88
ComputeSYMGS
114^117※全てのSPMVを含むコスト

MG%SPMV%700.0##

20.0##
18.0##

600.0##

16.0##

500.0##

12.0##

Time%[s]%Time%[s]%14.0##

4.0

10.0##
8.0##
6.0##

400.0##

22.4

300.0##
200.0##

4.0##

100.0##

2.0##

0.0##

0.0##
As#Is#As#Is#

Tune#

Tune#

2014年8月26日SS研HPCフォーラム50
トータル性能2014年8月26日SS研HPCフォーラム

Total%800.0##
700.0##

500.0##

19.2

400.0##

882944

300.0##

524288##200.0##

262144##131072##100.0##

65536##

0.0##

32768##As#Is#

Tune#

16384##GFLOPS'(log)

Time%[s]%600.0##

8192##
4096##
2048##

GFLOPS#

1024##
512##
256##
128##
64##
32##
16##
8#

32#

128#

512#
2048#
(log)

8192#

32768#

2014年8月26日SS研HPCフォーラム

51
ISCʼ142014年8月26日SS研HPCフォーラム
Site

天河２
「京」

ComputerCores

Tianhe-2 NUDT,NSCC / Guangzhou Xeon 12C 2.2GHz + Intel Xeon 3,120,000
Phi 57C + Custom

HPL
Rmax
(Pflops)

HPL
Rank

HPCG
(Pflops)

HPCG/
HPL

33.9

1

.580

1.7%

HPL
HPCG

RIKEN Advanced
Inst for Comp Sci

K computer Fujitsu SPARC64
VIIIfx 8C + Custom

705,024

10.5

4

.427

4.1%

Titan

DOE/OSOak Ridge Nat Lab

Titan, Cray XK7 AMD 16C +Nvidia Kepler GPU 14C +Custom

560,640

17.6

2

.322

1.8%

Mira

DOE/OS
Argonne Nat Lab

Mira BlueGene/Q, Power BQC
16C 1.60GHz + Custom

786,432

8.59

5

.101#

1.2%

Swiss CSCS

Piz Daint, Cray XC30, Xeon 8C+ Nvidia Kepler 14C + Custom

115,984

6.27

6

.099

1.6%

Leibniz
Rechenzentrum

SuperMUC, Intel 8C + IB

147,456

2.90

12

.0833

CEA/TGCC-GENCI

Curie tine nodes Bullx B510
Intel Xeon 8C 2.7 GHz + IB

79,504

1.36

26

.0491

2.9% * scaled to reflect the same
number of cores# unoptimized implementation
3.6%

62,640

3.00

11

.0489

1.6%

132,8401.65

18.0439 #

2.7%

78,848

.881*

7

.0161

1.8%

79

.0110

2.4%

80

.00998

2.4%

Exploration and
HPC2, Intel Xeon 10C 2.8 GHz
Production+ Nvidia Kepler 14C + IB
Eni S.p.A.
DOE/OS
Edison Cray XC30, Intel Xeon
L Berkeley Nat Lab
12C 2.4GHz + Custom
Texas Advanced
Stampede, Dell Intel (8c) +Computing Center
Intel Xeon Phi (61c) + IB
Beaufix Bullx B710 Intel Xeon
Meteo France
12C 2.7 GHz + IB
Prolix Bullx B710 Intel Xeon
Meteo France
2.7 GHz 12C + IB
CALMIP Bullx DLC Intel Xeon
U of Toulouse
10C 2.8 GHz + IB

24,192
23,760.469(.467*).464(.415*)

12,240.255

184.00725

2.8%

Cambridge U

Wilkes, Intel Xeon 6C 2.6 GHz+ Nvidia Kepler 14C + IB

3584

.240

201

.00385

1.6%

TiTechTUSBAME-KFC Intel Xeon 6C
2.1 GHz + IB

2720

.150

436

.003702.5%
2014年8月26日SS研HPCフォーラム

52
ISCʼ142014年8月26日SS研HPCフォーラム• ISC’14では正式のベンチマークではなく参考のミニリストとして発表• 京はこのミニリストで2位を獲得• 1位の天河2はピーク性能で京の5倍程度• HPCGの性能差は1.3倍• HPCG/HPL性能比：4.1%は群を抜いている• HPCGはSC’14で正式のベンチマークとなる予定

2014年8月26日SS研HPCフォーラム

53
まとめ
2014年8月26日SS研HPCフォーラム• はじめに• LINPACK• HPCGとは• HPCGベンチマークプログラム• HPCGのチューニングと性能

2014年8月26日SS研HPCフォーラム