koie blog

- 日時: 2016-08-31 10:00〜16:40

• 場所: 東工大キャンパスイノベーションセンター2F (田町)
• 主催: JST CREST「ポストペタスケール時代のメモリ階層の深化に対応するソフトウェア技術」研究チーム
• 案内
• 資料

初回2014年につづいて2度目の開催。

館内の自動販売機でジュースを100円で売ってて安い。

10:00

オープニング / 遠藤敏夫 (東工大)

10:06

10:07

インテルXeon Phiプロセッサのアーキテクチャとメモリ階層 / 池井満（Intel)

• インテルのマーケティング
  • スケーラブル・システム・フレームワーク
  • インテル・クラスタ・レディ
• Xeonは4IPC(inst/clk)だがPhiは2IPC。そのかわり省電力。
• TSX HLE(hardware lock elision): ロックプレフィックスがついていても無視して楽観的排他制御するらしい(競合したらやりなおし)。
• VMMを経由せずにVMに直接割り込みが渡されるようになった(SR-IOV,Direct I/O)。
• リソース・ディレクタ・テクノロジで使えるキャッシュ容量の制御ができるようになった。
• パッケージ内にMCDRAM(積層DRAM)が載っている。
  • 帯域は広いが途中にバッファが入る分、レイテンシが1clk程度遅くなる。
  • 16GB, 450GB/s。
  • パッケージ内にメモリを載せることでメモリ階層が1段ひきあげられたかんじになる。
• MCDRAMのメモリモード: 3種類ある。BIOSで切り替える必要がある。numactl -Hでみられる。
  • Cache mode: MCDAMは外付けのDDRメモリのキャッシュとして動作する。ダイレクトマップなので注意。
  • Flat mode: アドレス空間にMCDRAMとDDRを並べて配置する。アプリが使い分ける。
  • Hybrid mode: MCDRAMの一部をcacheして残りをflatにする混在モード。
• Intel Parallel Studio EXでいろいろ調べられる。
• AVX-512CD: 左辺値にインデックスがコリジョンしたものがあっても再実行してくれる。
• マスクレジスタでAVXレジスタの書き換えたくないところを指定できる。
• クラスタモード: コアとタグメモリとMCDRAMの関係
  • all-to-all mode: 論理アドレスのハッシュ値からタグメモリを持っているタイルにききにいってMDRAMをR/W。
  • quadrant mode: タイルを4分割することでall-to-allに比べてタイル-タグ-MDRAMの距離を短くする。
  • sub NUMA mode: 完全に4分割して4CPUにみせる。
• 姫野ベンチのgosa1が自動でリダクション命令を使うようにできないので、pragmaを入れる必要がある。
• sub NUMAモードにしてMPIを動かしてhbw_malloc()でMDRAMをアロケートするようにすると性能がでる。

11:00

• Q: インテル的には機械学習はPhiでやるつもりか?
• A: トレーニングはPhi、スコアリングはXeon。
• Q: (メモリまで含めて)電力測定はできるか?
• A: おそらくできる。ターボがあるので管理する必要があるため。メモリについては分からない。
• Q: 半精度演算(16bit float)は入るか?
• A: 今はない。
• Q: どういうツールがあるとうれしいか?
• A: 機械学習系ではCUDAが先行しているのでコミュニティ系のものをやれるとうれしい。

11:08

11:09

Pascal世代以降のGPUメモリシステム / 成瀬彰（NVIDIA）

• Tesla P100 (コードネーム GP100)
• HBM2: ピーク720GB/s、実効で7割は出る。4層で16GB積んでいる。スペースは8層分あるので放熱のためスペーサをかましてある。
• GPUはメモリが少ないとの批判がある→Unified Memory
• Unified Memory:
  • CPUメモリとGPUメモリでポインタを書き分けなくてもよくなる。
  • cudaMallocManaged()したときに物理的にメモリを確保していたのが遅延割り当てになる。
  • GPUメモリをCPUが更新するとGPUカーネルで使うかどうかわからないけど全部コピーしていたのがon-demandコピーになる。リンクリストのようなデータ構造で有効。
  • ページ単位でコピーするがページサイズは非公開らしい。
  • GPUとCPUが競合しないようにページを保護する仕組みは用意してない(簡単なんだからユーザが自分で用意しろというスタンスらしい)。
  • Q: readのためにGPU→CPUにコピーしたページもGPUに書き戻される?
  • A: …
  • GPUメモリがあふれたらpage outできる。
  • ページフォルトのレイテンシは10usくらい。同時に1000程度のページフォルトをあつかえるのでレイテンシは隠せる(との主張だがプログラムはブロックされちゃうよね?との質問もあった)。
• out-of-core
  • GPUメモリに載りきらないデータをCPUメモリに載せて演算。
  • PFNのChainerはよくできている。NumPyのAPIでCuPyというのをつくっている。
• NVLINK
  • CPU-GPU間はPCIeでは遅すぎる。
  • POEWR8ではNVLINKが載ってくる。40GB/s。 (PCIeは16GB/s)
  • それ以外のCPUではGPU-GPU間のみNVLINKということになる。

12:02

• Q: Chainerがunified memoryに対応したら速くなる?
• A: 速くなるのではなくて、いままで動かなかった規模のものが動くようになる。
• Q: 遠藤先生のところのswap in/outがいらなくなる?
• A: 動きのわかっているアプリならon-demandよりも明示的に記述した方がオーバヘッドが少なくなるので棲み分け。

12:06

昼休み

牛タン圭助へ。混んでて15分くらい待ち。日替わりランチが売り切れたので「牛ハラミ定食」に。野菜少ない。ちょっと味付け濃いのでごはんおかわり。

13:11

New 3D Flash Technologies Offer Both Low Cost and Low Power Solutions / 大島成夫（東芝）

• 2020年の予想: 生成されるデータが44[ZB]なのに対して、供給されるストレージ容量は6.6[ZB]しかない。
• SSDはパフォーマンス重視・容量重視・価格重視に分かれていくと予想される。
• メモリ階層のアクセスタイムを光速で距離に換算して、CPUをシェフに例えるとわかりやすい。
• BiCS(ビックス)
  • 三次元スタック構造。
  • 従来の平面のNAMDを縦に立てて横に並べた感じになる(ビットライン・ワードライン的には)。
  • 2017年内には1Tbit品が出る予定。
  • ワイヤ・ボンディングからTSV(through-silicon via)にすることでワイヤーのLCRがなくなるので、信号のアイパターンが開いて1.2Gbpsでもクリア。また消費電力はチップレベルで1/2に、コンポーネントレベルで30%減。
  • 2D→BiCSにすることで溜め込める電子数が6倍になるので、2Dでは8値(TLC)が限界だったが16値(QLC)までいけそう(試作あり)。
• SSDはHDDの2〜3倍程度の価格になればデータセンターなどで置き換えが一気に進むと思われる。

13:52

• Q: コントローラの話がなかったが?
• A: 重要。電力やスピード(レイテンシ)の問題(トレードオフ)があるのでむつかしい。誤り訂正をがんばると遅くなるなど。IPを開発中。
• Q: TSVについて。
• A: DRAMのTSVは高価だが、BiCSのTSVは低コスト。
• Q: どこで儲けていく?
• A: intelは3D XpointでCPU寄りをねらっている。東芝は容量寄りで。
• Q: ストレージシステムへの戦略は?
• A: 部品供給側よりもPure Storageのような上位の方が儲かるが、1段ずつfoot chainを上がっていくしかない。
• Q: コントローラは進化しているか?
• A: 少しずつかわっている。詳しくはNDAむすんで。
• Q: 生Flashメモリの標準化はしないのか?
• A: NAND flashは標準化しなかったので(各社自由にやれて)伸びた。違いを楽しんでください!
• Q: BiCSで高さを伸ばすとセルは小さくなる?
• A: 積層数を増やしてもセルは小さくならない。配線遅延はそもそもNANDの動作は遅いので問題にならない。
• Q: endurance重視の製品の予定は?
• A: コントローラである程度なんとかなる。実はエンターブライズ向けとコンシュマ向けとでメモリチップはほぼ同じでコントローラが違うだけ。それで寿命ば倍もちがったりする。寿命をのばしたかったかったら、シーケンシャルアクセスして書き込みはページサイズに合わせるべし。データセンターの人たちならやってくれるはずだ。ドライブレベルでならアプリケーションからヒント情報を与えて寿命をのばすことはできるかもしれない。実際、appleやハイエンド向けではファームウェアのチューニングをしている。Host Managed(コントローラをホスト側にもってくる)という流れもあるがSSDメーカとしては付加価値的に歓迎できない。

14:14

(休憩)

14:24

GDDR・DDR・Flashの多階層メモリを利用するランタイムライブラリと大規模ステンシルへの応用 / 遠藤敏夫 (東工大)

• エクサ時代に向けての目標: 100PB/s, 100PB
• 既存のHPCアプリに手を入れて out-of-core 実行を実現した。
• GPUクラスタといえば CUDA + MPI
• ステンシル計算: 流体シミュレーションで重要な計算カーネル。メモリ・インテンシブなのでGPU向けだがメモリが足りないことも。
• HHRTライブラリ: CUDAとMPIのラッパを提供。 https://github.com/toshioendo/hhrt
  • cudaMallocをラップしてメモリ割り当てを掌握、MPIが待ちに入るときにメモリのswapを実行することでGPUメモリ以上のサイズの計算を実行可能に。
  • ただ、それだけだと性能が1/30に落ちてしまう。
• テンポラル・ブロッキング(時間方向のブロック化): 時間発展の計算を1空間ブロックに対してまとめて実行することで局所性向上を狙う。
  • 55%くらいの性能低下におさえられた。

15:04

• Q: 頻繁にswapしたら意味ない。unified memoryとの統合は?
• A: unified memoryだけでは、まだCPUのメインメモリを越える規模の計算はできないので、いちおう棲み分けはできている。統合はできたらしたい。
• Q: テンポラルブロッキングについて
• A: コード1万行のうち500行くらい。時間ステップは50 stepくらいをやっている。(これは計算による)
• BLASでもテンポラルブロッキングはやっている

15:12

15:13

動的バイナリ変換によるメモリ階層性能プロファイリングと透過的メモリ階層チューニング / 佐藤幸紀 (東工大)

• openMPやMPIを使うだけでは性能は出ない。ノード単体のメモリチューニングが必須。チューニングは職人芸。
• 開発:
  • Exana: プロファイリング
  • ExanaDBT: 透過的チューニング
• Exana:
  • ループ構造の階層を表示
  • メモリ帯域やアクセス回数を表示
  • キャッシュの挙動はシミュレーションに基づく。(100倍程度の速度低下)
  • アクセスパターン・ワーキングセットを調べられる
• ExanaDBT:
  • バイナリをmcsemaでLLVM IRに変換。
  • Pollyでループ最適化。
  • Pin tool setでバイナリ書き換え。

15:42

• Q: データ依存解析の出力は表示が不十分なのでは?
• A: 出力をみてパイプライン構造が見えてくるのはチューニング屋さんにはメリットがあるが、アプリを作った人には自明。
• Q: ワーキングセットツールの位置付けは?
• A: ブロッキングサイズを調べるのにつかえる。
• Q: DBTとPGOの比較は?
• A: -O3ならやや効果があるとおもっている。
• Q: HPCではホットスポット最適化の要望が多い?
• A: ソースコード再コンパイルでよいという声も多い。
• Q: x86だけなのはHPC的によくないのでは?
• A: LLVMのコミュニティでPollyが対応していくと期待。
• そもそもPollyはきれいなループ構造でないと最適化してくれないとかループの中に関数呼び出しがあるとダメとか、いろいろヘボい。
• キャッシュのシミュレーションがFIFOなのはよくないのでは?

15:56

15:56

Flash利用によるout-of-coreステンシルアルゴリズムとブロックサイズ自動チューニングシステム / 緑川博子 (成蹊大)　

• ioDrive高すぎ
• メモリに載りきらない問題サイズをSSDに逃す
  • DRAMとSSDの速度差1000倍くらいある。
  • スペーシャルブロッキングだけだと64倍遅くなる。
  • テンポラルブロッキングもやると1.3倍におさまる。(実用的)
• ブロックサイズの決定は、問題サイズとデバイス情報を与えることで計算により算出する。いくつか試行して最適なものを選ぶ、というのではない。
• swapやmmapではコピーする時間を隠しきれなかったが、kernel版aioをつかうことでほぼ隠しきれた。内部では細かくaioを出すのではなく、まずブロックの半分をaioでとってきて計算開始、裏で残り半分をaioでとってくる構成におちついた(細かくやっても手間のわりに効果なし)。

16:31

• Q: SSDではなくHDDでも効果があるか?
• A: シーケンシャルアクセスにならない(ベージバウンダリにするためにいろいろ詰め物してる)のでダメだとおもう。SSDとHDDでは性能が1000倍くらい違うし。
• linuxのI/Oサブシステムが昔はキューが1個しかなかったが今はコア毎に用意されている。
• この手法は大きな問題を解くためにメモリが欲くてノード数を増やすことで総メモリ数を稼いでいるユーザにはメリットがある。

16:39

• MemoryPlus Workshop -- メモリとファイルストレージとOSと --
• 2014-09-17 10:00-18:45
• JAIST 品川サテライトオフィス (品川インターシティー 19F)
• 主催: JST CREST「ポストペタスケール時代のメモリ階層の深化に対応するソフトウェア技術」研究チーム
• http://aml.hpcc.jp/swopp-announce/msg05082.html

オープニング / 遠藤敏夫 (東工大)

• 別名 緑川勉強会
• 帯域と容量の両立はむつかしい→階層化
  • HMC-NVRAM-Flash

[2014-09-17 10:13]

NAND型フラッシュメモリーとSSD / 菅野伸一(かんのしんいち) (東芝)

• 会場挙手アンケート:
  • SSD利用用途: ファイルシステムが多い・ワークエリアにつかう人ももそこそこいる。
  • 誤り訂正方式: しってる人はほとんどいない。(でもしっている人が数人いることに驚きだ)
  • write ampification、over provisioning: などのキーワードを知っている人もそこそこいる。
• NANDはトランジスタ自体を配線としてつかって高密度化
• MLCは (0,1) (0,0) (1,0) (1,1) になっている(グレイコード)。1レベルの誤りで2ビットの誤りにならないようにする工夫。
• ブロックの大きさ: いまどき8GB〜16GB? (ききまちがえだとおもう)
• ページ: ブロックに126個くらい
• 内部では16KB単位で。4Kでも性能は16KBの性能になってる。
• テンポラリー用途なら(1年もデータを保持しなくてよいので)寿命は伸びる。
• 東芝製品はウェアレベリングは積極的にやらない。逆にインテルは積極的にやってるようだ。
• 最終的に寿命が尽きるときにウェアレベリングできていればよいという考え。
• SSDコントローラのDRAMのほとんどをLUTがほとんどつかっている。キャッシュはOSがやるのでSSDでやる必要なし。
• オーバプロビジョニング: コンシュマ7% エンタープライズ37%
• write apmplification: 1データをかいたときにN書き換え。
• 客が4Kのベンチマークをしてくる。困る。
  • DRAMサイズがおおきくなる。
• Tips
  • TRIMは過信しない
  • デフラグはワークロードがかわったタイミングでやると、GCがよくなって、トータルで特することもある。自分のPCでは3ヶ月に1回くらいやる。

[2014-09-17 10:53]

• Q:デバイスのDRAMキャッシュが効かないのは？
• A:サイズのミスマッチ。小さすぎる。
• Q:pseudo SLCはやってる
• A:やってる製品もある.
• Q:3D-NAND
• A:ブロックサイズが極端に大きくなる。
• Q:IOPSに効くのは？
• A:readはLUTの性能、writeはwrite amplification

[2014-09-17 10:56]

Massive Capacity SMR Disk and Their Impact on Host System Software / Damien Le Moal (HGST)

• 最近は高速HDDはつかわず大容量HDDに直接アクセスすることがおおい。(Facebookなど)
• SMR: トラックがかさなる。trackNを書くとtrackN+1がoverwriteされる→シーケンシャルライトしかできない。その代りに容量が3倍になる。
• SMRのターゲット: アーカイブ(テープとふつうのHDDの中間)
• シーケンシャルライトしかできないがファームウェアでがんばって(indirection,GC)ランダムライト可にできる。
• T10: SCSIの規格では ZBC(zoned-block command).
• T13: SATAの規格では ZAC(zoned-device ata command).
• どちらもコマンドは同じ。
  • model: autonomous(ランダムライトできる), host-aware(ランダムライトできるが最適化はホストでやれ), host-managed(そもそもデバイスタイプがちがう).
• zoneはLBAのレンジで指定される。大きさは自由。規格でも決められていないしzoneごとにサイズが違う可能性もある。先頭と最後のゾーンがconventionalである保証はない→MBRとか置けるとは限らず。
• zone type: conventional zone, sequential write preferred, sqeuential write only.
• ゾーンごとにwrite pointerがある。
• コマンドは2つだけ: REPORT ZONES, RESET WRITE POINTER. (10ページくらい)
• ランダムリードはゾーンの先頭からwrite pointerの前まで。後ろを読むとエラーになる。
• アプリケーションから直接デバイス(SG_IO)をたたくとページキャッシュなどのサポートがなくなる。
• カーネルでサポートするなら
  • 論理ストレージデバイスを設けるか
  • ファイルシステムを設計するか
  • log-structured FSみたいなのにするか
  • block IO schedulerはzone-awareにしないといけない。
• ファイルステムは安定性の面から心配(安定するまで時間がかかる)なのとSMR専用のファイルシステムになってしまうので避けたいところ。
• sg_util, libzbc, device type 14,
• 来年にはデバイスが出回るけど、カーネルの対応はまったく進んでない。
• fsはhost-awareで。ブロックアロケーションをfsからブロックレイヤーにもっていく提案がでてきている。
• 来年でてくるのはhost-awareのもの(autonomousのはしばらくかかる)

[2014-09-17 11:23]

• Q: OSは？
• A: まずはlinux. SMRはコンシューマ向けではないのでwindowsはないだろう(windows serverならあるかも?)。じつはSMRになっているHDDが出まわっている(性能はプチフリがある)。
• Q:テープとしてつかうのは?
• A:アーカイブコストとしてはテープに比べ苦しいところ。facebookの写真などは向いているだろう。writeスループットはふつうのHDDよりも遅い。トラックピッチが狭くなるのでトラックNをライトしたらN-1のトラックをverifyする必要があって遅い。(N-1のトラックを書き潰してしまっている恐れがあるため)
• Q:ディスクアレイだとどう？
• A:LHC(加速器)などは向いているかな。HPCだとあまり向いてない? Host-managedは特殊なモデルでHost-awareがつかわれるだろう。
• Q:ゾーンのサイズは?
• A:ディスク内でバラバラでもよいことになっている規格では。皆の希望は小さめ。

[2014-09-17 11:33]

[2014-09-17 11:34]

Linuxのページ回収処理のよる高性能計算アプリケーションへの影響 / 大山恵弘 (電通大)

• データインテンシブサイエンス： IOウォール問題。世界最高でいまのところ1TB/sくらい。
• NUSA: non unified storage architecture: 5年後に100TB/sにしたい。
• OSジッタ: HPCで悪影響をあたえる。割り込み処理とかデーモンとか。同期がズレて性能が落ちる。古い問題だが今でも論文が出ている。
• 大規模データをIOするとページキャッシュの操作がはいる→OSジッタ
• (今回の話はNVRAMとは関係ない)
• メモリがたりなくなったら： direct reclaim か kswapdを起こすか。pdflushはダーティページを書き出すもので今回の話とはちょっとちがう。
• 共有ページフレームを回収するときはすべてのPTEを一度に削除
• ページ回収という名の闇または沼! 職人の世界。
• ページをたくさん消費しているとkswapdは頻繁に解放処理する。一度に解放するページ数は少なめ。
• high water mark -- low water mark -- min water mark がある。
• kswapdよりも先に新しい回収デーモンを起こして、ページを一気に回収することで、処理コストを抑える。
  • ディスクアクセスが多ければキャッシュを解放するなど。
• WRF
• コア分離: デーモン用にCPUコアを割り当ててしまう解決方法。デーモン用にCPUコアをどれだけ空けておけばよいかわからないので、そもそものジッタを減らす努力が必要。

[2014-09-17 12:02]

• Q:swap deviceは? Flashだとどうなる？ directIOつかえば？
• A:HDD。swapはしてない。ページキャッシュの回収コスト。directioは検討したがread-aheadを自前でやらないと性能が落ちることもあり、アプリ開発者の負担が増える。
• Q:メモリのcgroup。flexsc。
• A:ファイルアクセスだとcgroupは効きにくい。cgroupのオーバヘッドは大きめ。(そんなに重くないという声が会場から)
• Q:drop_caches。madviseのDONTNEEDのようなのがページキャッシュにあればよい？ DONTNEEDはすぐにフリーリストにのるはず。
• Q:まとめて回収することで性能が改善した？
• A:回収したページ量は同じ。こまかく回収するところにコストがあるようだ。

[2014-09-17 12:12]

ひるやすみ

ビル内のつきじ植むらで、日替わりのつめたいうどんと生姜焼きだったかなぁ780円。

[2014-09-17 13:11]

Linuxのメモリ管理 / 吉田雅徳 (日立)

• Page Frame = page-sizeでかつアラインされているもの。
• ACPI: SRAT(static resource affinity table), SLIT(system locality table)
• Zone: DMA(0-16M), DMA32(-4G), NORMAL(4G-), MOVABLE(hot plug memory向け)
• ゾーンに分けているのは低位アドレスを特定デバイス向けに使いたいため(希少資源)。
• dirty ratio: 未書き込みのページが増えすぎないようwrite backする閾値。
• zone list: より高位のゾーンから割り当てるか(zone)、同じノードから選ぶか(node)、zone_normalとzone_dmaのサイズが同じくらいの容量だったらnodeポリシーで、normal>dmaならzoneで。
• kswapdはnumaノードごとにある。low watermarkをしたまわるとkswapdが動く。page allocationは止まらない。minまで少なくなるとallocationが止まってdirect reclaimがうごく。
• buddy system: 連続したページごとにまとめて管理する。4Kのアロケーションが圧倒的に多いので最適化がされていてCPUごとにキャッシュが用意されてホット(cpu cacheに載っているだろう;通常利用のページ)/コールド(載っていなさそう;DMAでつかったとか)が管理である。
• active_listのなかにinactiveがないかどうか調べて、inactive_listをページアウトできるかどうか。
• laptopの場合はactive->inactiveに移動させるときにcleanなページかどうかチェックが入る。ディスクを動かしたくないため。
• KSM(kernel shared memory): たまたま内容が一致したページを共有する機能。stable_nodeに複数のanon_vmaがぶらさがっている。
• VFS(Various FS): 抽象化: superblock, dentry, inode, pagecache
• filep->dentry->inode->page
• block cache = attribute + page cache
  • メタデータのページキャッシュはデバイス(/dev/sda)でつくられている。
  • 最近の速いデバイスではbio ioスケジューラを経由せずデバイスドライバを直接呼んでいる。(Le Moal Damien談)
• NVM
  • Block Device: NVMe(express) いくつかのレイヤをふっとばす
  • Persistent Memory: PMFS(tmpfs似), DAX(direct access block layer)(EXT4をDAXの上にのせるとpage cacheなしにIOする)。
  • DAX: ext2のXIP(eXecution In Place;キャッシュなしにデバイスを直接マップ)をext4に移植したもの

[2014-09-17 13:48]

• Q: NVMの開発は? エラー訂正が必要ならブロック単位にならざるをえないが。
• A: RAMでエミュレーションしながらやっているようだ。
• Q: NVM PCIプロトコルが重くて半分くらいしか性能が出ない。
• Q: レイテンシは?
• A: わからない。
• Q: read/writeの性能の非対称性はどのくらい問題になる？
• Q: NVMはどのあたりで使われる？
• A: しらない。

[2014-09-17 13:56]

不揮発メモリ向けファイルシステムの設計 / 建部修見 (筑波大)

• 計算科学: 初期値の読み込み(100TB-PB)、スナップショット(約1時間間隔)書き出し(100TB-PB)
• データインテンシブサイエンス: 実験データ(10PB-EB)
• 並列ファイルシステムへの性能要求
  • IOバンド幅とIOPSはコア数に比例することが想定される。
  • 100TB/s 1Miopsをめざしている
• HDDの性能はほとんど向上しない。100TB/sの性能を出すには2MWもHDDで使ってしまうことに。それはシステム全体から見て多すぎる。
• フラッシュなら1GB/sで0.1W。(菅野:1桁くらい小さいすぎない?)
• ハイブリッド(NVM-HDD)にしてもネットワークバンド幅制限は解消されない。
• バーストバッファ(IOフォワーディング・IOデレゲーション): 計算ノードに高速ストレージをもってくる。
• 計算ノードにストレージを分散させる。(共有メモリが分散メモリになったように)
  • 分散メモリがうまくいっているのはアプリがMPIをつかているから。でも同じように分散ファイルをユーザにみせてしまうとそれは面倒すぎる。ユーザがファイルの管理(=メタデータ管理)をしないといけない。
• オブジェクトストア
  • ファイルシステムとディスクの中間的なもの。
  • ファイルシステムのメタデータの管理コストが下がる。
• OpenNVM
  • アトミックバッチ:nvm_batch_atomic_operation(IOベクトル)
  • スパースアドレス:flash translation layer(FTL)の提供
  • キーバリュストア: nvm_kv_put/get
• HDDはシーケンシャルアクセスで性能向上 → NVMは並列アクセスで性能向上の違い。
• persistent trim: fusionIO由来、TRIMするとデータがすぐ消える。(通常のtrimはヒントでしかない)
• NVMによるオブジェクトストアの例
  • リージョンを大きくとる。
  • 固定長にするとIDと直接対応。
  • persistent trimとスパースアドレスとでリージョンは使い捨てにできる。

[2014-09-17 14:27]

• Q: fusionIO
• A: そのもの
• Q: IOが速くなると相対的にCPUの負荷があがってくる。
• A: HPCだとメモリバンド幅で上限があるので1コアくらいOSで専有できるだろう。カーネル専用コアとか.
• Q: オブジェクトストアでFTLをとばしてダイレクトアクセスしないのか？
• A: こまかいアロケーションするので必要
• Q: openNVMは1デバイスを仮定しているがスケーラビリティは？
• A: MDSのスケールアウトが鍵
• Q: そもそもファイルにする必要はないのでは？
• A: hadoopなどのインタフェースで考えるのもあり。

[2014-09-17 14:34]

[2014-09-17 14:35]

Linuxにおける不揮発性メインメモリとストレージの融合とその応用 / 及川修一 (筑波大)

• NVM: linuxではパーシステントメモリとよばれている。
• 今はまだCPUに直結はできないようだが、メインメモリとして使えるとして考えてみた。
• ファイルシステムとメインメモリの両方として使ってみたら？
• 既存のメモリファイルシステムはカーネルとくっつきすぎていて使いにくい。(カーネル構造体をそのままつかってたり。バージョン間の互換性がないファイルシステムはいやだ)
• XIPをつかうことに。
• anonメモリをallocateするのをfilesystem経由でNVMをとってくる。
• PRAMFSはいじくりやすい。
• NVMをつかうと、高機能なfsが遅かったり、bitmap操作が遅かったりするのが見えてくる。
• チェックポイント・リスタート
  • メモリをファイルにできる。コピー不要。
  • カーネルの再起動が若干速くなる。
  • リスタートするとファイルが壊れるがCoWにしとけばok。
• kernel rejuvenation(若返り)
  • カーネルだけリスタートしてシステムを安定にするというやつ。
• wikipedia: Everything_is_a_file
• NVMを搭載したマシンがこうなるであろうというコンセンサスがないので論文書きにくい。現実味がないとしてrejectされてしまう..
• NVMの性能を反映できるcycle accurate simulatorはすごく遅い..
• インテルはCPUをいじって特定の領域だけレイテンシをかけることでエミュレートしているらしい。
• Open SSD

[2014-09-17 15:09]

• Q:filesysとmem、どっちを生かすか？
• A:メモリの沼にははまりたくなかったのと、カーネル構造体に依存してしまうので別のカーネルから使えないので、ファイルシステムにした。 gmpfs?
• Q:メモリとファイルの量的切り分け
• A:固定
• Q:メモリアドレスを仮想化するとうれしい?
• A:かんがえてなかった。
• Q:struct pageをファイルとメモリの両方に用意しているがintelは片方だけだが？
• A:いってることはわかる。そこまでNVMがたくさんのは考えてないので問題ない。
• Q:NVMとCPUのキャッシュと電源断 (write-throughになってればいいけど)
• A:H/Wでがんばってはきだすのが正しい。

[2014-09-17 15:18]

休憩

[2014-09-17 15:26]

不揮発性メモリを考慮した大規模なグラフの高速処理 / 佐藤仁 (東工大)

• tsubameのストレージの設計などをしているが、あえてアプリの話を。
• 大規模グラフ処理はスパコンの重要なコアだと考えられるようになった。
  • graph500: スケールフリー・スモールワールド: kronecker graphをBFSする。
• EUでやってる脳シミュレーションだとgraph500 hugeくらいの規模になる。
• BFS: conventionalなtop-downなのと(2012年)bottom-upなのと。
  • ハイブリッド: 探索済の頂点数でtop-downかbottom-upなのかをきりかえる。
• 非同期IOとかmmapとかopenNVMをつかってみたが、性能改善はアルゴリズムの方が効いた...
• 構成はRAIOD0でSSDぶらさげてPCIe x8で。fusionIOには劣るが安い。
• DRAMに載りきらずSSDの60%くらいをつかうようになっても、それほど性能は落ちず。
• DRAMをつかえばつかうほど性能が落ちてSSDをつかった方が性能が出た→ページ回収のコストか?
• HAMAR: GPUとSSDを直接つなぐ

[2014-09-17 16:00]

• Q: 京と今回のハイブリッドシステムのアルゴリズムの違いは？
• A: たいして違わず。以前は1/3くらいの性能しか出てなかったが、ハイブリッドでほぼおなじになった。京だとネットワークのコストがかかってくるのでチューニングちがってくる。
• Q: SSDインタフェースは?
• A: ふつうのファイルIO。直接SSDをたたいてもせいぜい20%くらいしか向上しないと見込んでいる。
• Q: 不揮発性はどうでもよかった？省電力ならよかった？
• A: スパコン側からしてはメモリのコストと量が気になる。不揮発性は耐故障性の面では必要かもしれない。
• Q: スケール30以上のニーズは本当にあるのか？30以下の方がニーズがあるのでは？
• A: たしかにノード数少なくて省電力の方が需要がある。巨大なグラフの実データがない。twitterでやっと2^30くらい。もっと儲かるアプリからカーネルをもってきたほうがいいかも。

[2014-09-17 16:09]

[2014-09-17 16:10]

格子系アルゴリズムの局所性向上とHHRTライブラリの実践 / 遠藤敏夫 (東工大)

• GPUからPCIex経由でホストのメインメモリにアクセスできる。
• テンポラルブロッキングによる局所性向上
  • 時間発展を小さい空間ブロックで進めることで局所性向上。→ アプリケーションが難しくなる。
• プログラミングコスト重視
  • 自動でテンポラルブロッキングに変換するのはムリだが
  • できるところだけ自動でやる → Hybrid Hierarchical RunTime

[2014-09-17 16:24]

[2014-09-17 16:25]

Exanaツールによるメモリアクセスプロファイリング / 佐藤幸紀 (JAIST)

• ダイナミックコンパイラ： ループアンローリングを動的にやる。
• Exana: EXecution-driven application program ANAlysis Tool set
  • ループ構造解析
  • データ依存性解析 (タスク並列性がわかるとか)
  • 関数インライン展開されてもデバッグ情報から対応づけ
  • BytePerFlops ratio: 手動計算に近い値が得られた。
  • LiwkidだとCPUからメモリに出ていくところをカウントするのでBFratioは低くなる。
  • アクセスパターンのパターン化 (fix, sequential, stride, sequential-stride, それらの組み合わせ(再帰的)) して解析
    • アクセストレースは大きくなる→パターン化するとコンパクトに→階層化するとずっとコンパクトに

[2014-09-17 16:42]

[2014-09-17 16:43]

Flash SSD利用による大規模ステンシル計算 / 緑川博子 (成蹊大)

• fusion iodrive2 mlc: HDDよりは1000倍くらい速いがDRAMよりは1000倍くらい遅い → 単にmmapしただけではダメ。
• spatial blocking と temporal blocking と two-level..: メモリに載りきらなくてもけっこう動く。
• aioを効かせるにはアラインしてないとだめだけど、そろっているとTLBとかメモリバンクコンフリクトしてCPU計算が遅くなってた。
• aioは大きなIO単位でアクセスしてるのにiostatでみてると細切れになってる。。

[2014-09-17 17:01]

休憩

[2014-09-17 17:11]

パネルディスカッション

• 緑川・菅野・Le Moal・大山・建部・追川・佐藤・遠藤 (会場:松岡)
• DRAMとflashの間には2つくらい階層が必要なのでは？
• HDD/Flashはスロット的にあまり増やせないのでアーキテクチャに問題がある。
• 計算とストレージを分ける方向性もある。
• デバイスやってる人はDRAMをNVMが置き換えるとはりきっている。そうなるとOSからファイルがなくなってmallocになる。
• SLCとMLC、値段は2倍、寿命は10倍。
• ハイバネーションの調査は社内調査したことがある。結果は十分もつことがわかった。
• MRAMにきりかわるかどうかは値段できまる。パワーの問題もある。
• MRAM: DRAMとflashの間をうめたいという客はあまり声がきこえてこない。
• ReRAM:記録原理がまだよくわかってない。
• 東芝DRAM 25年持つのをつくっていて5年で不良がでるメーカーに負けたので寿命にはセンシティブ
• メモリの消費電力自体はシステム全体では小さいのでは？
• IOPS/powerで考えるべき。ラックの電力や電力密度も考えるべき。
• google big query: 1ノードにHDD10台ぶらさげてインターコネクトも謎。
• HPCはbigdataデータセンターにとりこまれていく。
• facebookは初期からPCI SSDをいれていたのは、そうしないとさばけないため。1要求で10メッセージくらいとぶ。
• GPU/FPGA: azul たまたまハーフラックに入るGPUはいまいちパワーがないため。baiduはGPU。HPCのネットワークをどうやってIDCに低コストでいれていくか。
• 今のHDDは複数プラッタになっていても並列には処理されてないらしい。
• NAND 3D stack になるとビットクロスがおきてHDDが負けるがなくなりはしない。長期保存向け。
• 松岡: MRAMよりもReRAM? MRAMはDRAMとflashのあいだに、ReRAMは大容量にいく？
• メモリの信頼性はどのくらい必要なのか？ 1byteのなかの1bitエラー訂正するのに4bit必要。最小アクセス単位が1KBとかになればコストがちがってくる。
• キャッシュラインサイズ64Bでアクセスできればよい。
• 1つのデバイスに要求される信頼性が異なるデータが混在している。昔はテープは頻繁にコピーしていた。
• エンタープライズでSATAをつかうようになるのに8年くらいかかった。カップリングの問題もある。
• インメモリデータベースでも小さいアクセス単位でしない。データは数バイトでもレコードはKB。しかもベクトルでなめるので。
• インメモリデータベースでも正規化されてないとレコードはでかい。
• バイトアクセスできてもアプリが性能を出すためには、きれいにならべておくひつようがある。
• 2012のSACSISでMRAM数年のうちに出るって産総研の人がいってたけど、、、
• HPの人がthe computerというのをやってる？ メモリ階層は深くなるより浅くなる？
• 素子が同じでも物理的な距離の制約があるので、そんなに単純ではない？
• メモリ圧縮の話は？ メモリ階層をあつかえるガーベッジコレクション・コンパクション。東芝SSDはコンシューマ向けのは書き換え頻度が高い部分を速いブロックに書くようになっている。エンタープライズでは均一なアクセスを要求しているので向かない。エンタープライズの人は自分のコントロールできないレイヤがあるのが嫌い?
• 組込み向けのGCではアクセスパタンによって置き場所を変えるという研究はある。

なんか、東芝の人がいっぱいいたなぁという印象。

会場はWiMAXは電波の入り弱し。GH3のバッテリは2個で十分足りた。1個で3時間くらいはいける。SDカードは32GBを3枚。