PowerPoint Presentation

Size: px

Start display at page:

Download "PowerPoint Presentation"

たかとしえいさか
5 years ago
Views:

1 ヘテロジニアスな環境におけるソフトウェア開発

2 Agenda 今日の概要ヘテロジニアスな環境の登場ホモジニアスからヘテロジニアスへヘテロジニアスなアーキテクチャ GPU CUDA OpenACC, XeonPhi 自分のプログラムを理解するためにデバッガ共通の操作体験 TotalView 続きはブースで

pthreads) CPUクラスタマルチプロセス (MPI) 高い汎用性長年の実績とノウハウ

3 より速くホモジーニアスな並列 HPC 銀河生成金融のリスク計算車の衝突解析製薬大規模複雑なアルゴリズム高速単一のCPUコアはクロック数が頭打ち ~ 3GHz ホモジーニアスな並列マルチスレッド (OpenMP, pthreads) CPUクラスタマルチプロセス (MPI) 高い汎用性長年の実績とノウハウメモリアクセス速度のボトルネック高価なインターコネクト複雑な非同期プログラミング The Free Lunch Is Over By Herb Sutter

4 専用の外付けデバイスより速くシンプルで高速な演算器高い並列化効率費用対効果省電力新しい言語拡張の理解ハードウェアの理解 GPU アクセラレータヘテロジニアスな開発環境 MIC コプロセッサ

5 Top 500, Green 500, より速く

6 GPU と XeonPhi ヘテロジニアスな環境 GPU CUDA と OpenACC Device, SM, CUDA Core, Warp, Lane block, thread, global memory, shared, Xeon Phi offload ディレクティブ symmetric モード native モード 512bit のベクトル

7 ハードウェアホストCPUのアクセラレータとしてあるいはホストをGPU 演算の補助装置として数千コアの並列グリッド ( デバイス ) SM CUDAコア (SP) 複数 CPU x 複数 GPU 製品ライン GeForce, Quadro, Tesla, アーキテクチャ Kepler, Fermi, Maxwell, GPU アーキテクチャ

CUDAコアとregisterの組軽量メモリの種類オフチップ Video memory オンチップ shared memory

8 GPU アーキテクチャハードウェア階層 Grid SM (16 個 / Grid) (Fermi) CUDAコア (SP, 32 個 / SM) (Fermi) Warp( 縦糸, 32スレッド ) 制御単位ベクトル Thread ( 撚り糸, 3D, レーン ) CUDAコアとregisterの組軽量メモリの種類オフチップ Video memory オンチップ shared memory (64kB/SM) register constant cache, texture cache SM (Streaming Multiprocessor) Grid CUDA コア

9 CUDA NVIDIA CUDA GPUベンダーのNVIDIA 公式 ( 信頼ベンダーロック ) 無償の開発環境 cufftやcublasなどの数値計算ライブラリ付属 OpenCV, MATLAB, IMSLなどツール経由でCUDA 利用することもホストはWindows, Linux, Mac レイヤーが低くコードが複雑 CUDA Dynamic Parallelism Cuda 6.0 Unified Memory

10 CUDA の論理階層グリッド ( デバイス ) スレッドブロック (2D/3D) SMへの割付単位スレッド (3D) CUDA アーキテクチャ Kernel 関数ホスト側 ( 呼び出し ) MatMulKernel<<<dimGrid, dimblock>>>(mata, matb, matc); デバイス側 ( 実装 ) global void MatMulKernel(Matrix A, Matrix B, Matrix C) {... } 論理構造を明示的に指定

11 メモリ階層 ( 論理的 ) 開発者が明示的に指定 Memory Scope Locality Global Device External Shared Block Chip Local Thread Chip Constant Device Chip (cache) Texture Device Chip (cache) Register Thread Chip shared 16kB CUDA アーキテクチャ

12 CUDA プログラミング言語 C 言語の拡張コンパイラNVCC Python,Javaなどのバインディングホストコードとデバイスコード同一ファイル内に混在可能 NVCCが振り分けてくれるメモリ修飾子 device, constant, shared 関数修飾子 global, device, host

13 CUDA プログラミング単純なプログラムの例 CUDA サンプル vectoradd 2つの配列の足し算データを計算単位に分割する組み込み変数 blockidx, threadidxごとに分ける何番目のBlock (3D), 何番目のThread (3D) にいるのか?( 論理階層 ) 何番目のdevice, SM, Warp, Laneにいるのか?( 物理階層 ) kernel 関数の定義ホストメモリ確保ホストデータ初期化デバイスメモリ確保メモリ転送 kernel 関数呼び出しデバイスメモリ解放ホストメモリ解放

14 OpenACC OpenACC の特徴 Fortran, C/C++ 有償ディレクティブ指向 OpenMPとの類似ホストコードのみ記述通常のCPUコードとしても使えるコンパイラが多くの仕事を担当ループの検出 GPUの側のメモリ管理 CPUとGPUの間のデータ移動 Kernel 関数を作成するかどうかオープン規格 Cray, PGI, NVIDIAがサポート CUDAと組み合わせることができる OpenACC CUDA gang threadblock worker warp vector warp 内のスレッド OpenACC の論理階層

) 独立したOS(Linux) が動作し sshなどによるアクセス柔軟な実行モデル Offloadモデルディレクティブ

15 Xeon Phi Intel Xeon Phi コプロセッサホストCPU にPCI Expressで接続するボード MICアーキテクチャのx86 互換のコプロセッサ汎用的なCPU 用に書かれたコードの移植性が高いコアごとに512bitのSIMD 処理 (16 命令 /clock, ベクトル長 ) 独立したOS(Linux) が動作し sshなどによるアクセス柔軟な実行モデル Offloadモデルディレクティブ必要な部分だけコプロセッサに投げるシンメトリックモデルホストとコプロセッサの間でMPIなどを使ってやりとりする Nativeモデルコプロセッサ上でのみ実行既存のコードをそのまま使える

16 Xeon Phi offload のコード例 host offload void test08() { float pi = 0.0f; const int count = 10000; int i; OpenMP MIC t0 t1 t2 t3 MIC #pragma offload target (mic) #pragma omp parallel for num_threads(4) private(i) reduction(+:pi) for (i = 0; i < count; ++i) { float t = (float)((i + 0.5f) / count); pi += 4.0f / (1.0f + t * t); } pi /= count; } host

自分のコードがどう振る舞うかプログラムは意図したとおりではなく書かれたとおりに動く At Operation.

17 コードを理解する様々なヘテロジニアス環境ハードウェアやプログラミングモデルの特性を理解メモリの制約を理解してエラーを回避する複数言語複数パラダイムクラスタスレッドヘテロジニアス vector 自分のコードがどう振る舞うかプログラムは意図したとおりではなく書かれたとおりに動く At Operation. 仕様書ではなくコードが全て複数人による作業引き継ぎ移植 γνῶθι σεαυτόν ( 汝自身を知れ, Know thyself)

18 コードを理解する自分のコードと仲良くなるために code reading, ペアプログラミングベアプログラミングどの処理がデバイス上でどのように振舞っているか想像するデバッグはコーディングの2 倍大変開発時間の制約手ぶらで立ち向かうのは危険効率的なデバッガ

Phi, OpenACC メモリデバッグ機能 : MemoryScape リバースデバッグ機能 : ReplayEngine

19 TotalView TotalView 幅広いコンパイラプラットフォーム C, C++, Fortran 77 & 90 Unix, Linux, MacOS X, ラップトップからスパコンまで並列デバッグマルチスレッド MPIデバッグ CUDA, Intel Xeon Phi, OpenACC メモリデバッグ機能 : MemoryScape リバースデバッグ機能 : ReplayEngine パワフルで軽量使いやすい GUI パッチ機能 breakpoint: Evaluation Point スクリプト用の CLI, バッチデバッグリモートデバッグ

20 TotalView 世界中の企業や研究機関 HPCwire 影響力のあるミドルウェア 20 (2014/6)

21 TotalView 様々な動作画面 OpenACC CUDA ( デバイスコード ) ホスト XeonPhi コプロセッサ同じインターフェイス breakpoint, ステップ実行, 関数の呼出履歴配列の表示,

22 CUDA のデバッグ画面 TotalView CUDA グリッドとブロックの次元 warp/ レーン warp/sm warp 数 GPU フォーカススレッドの論理座標スタックトレースとインライン関数パラメータレジスタ局所変数と共有変数 warp の PC

23 TotalView CUDA:2 つの座標をマッピング Grid, SM, Warp, Lane Grid, Block, Thread 物理座標論理座標

Summary ヘテロジニアスな環境開発は大変開発者や研究者はGPU, Xeon-Phiなどの新しいデバイスになじまなければならない同時に従来のOpenMPやMPIの手法も知っておく必要があるそれぞれの環境ごとのプログラムの挙動の違いをイメージする

24 Summary ヘテロジニアスな環境開発は大変開発者や研究者はGPU, Xeon-Phiなどの新しいデバイスになじまなければならない同時に従来のOpenMPやMPIの手法も知っておく必要があるそれぞれの環境ごとのプログラムの挙動の違いをイメージするこれらを限られた時間の制約の中で行わなければならないコードを知るにはデバッガが便利! TotalViewは複数のホストとデバイス幅広いアーキテクチャに対し同一の操作体験 Accelerating Great Code 結論 TotalView で定時に帰ろう!!

25 XeonPhi

26 Xeon Phi The spectrum of models CPU 中心マルチコアのホスト offload Intel Xeon Phi 中心シンメトリックメニーコアのホスト汎用的なシリアルおよび並列計算ディレクティブで並列化対等な MPI ネイティブホストコプロセッサ Main( ) Foo( ) MPI_*() Main( ) Foo( ) MPI_*() Foo( ) Main( ) Foo( ) MPI_*() Main( ) Foo( ) MPI_*() Main() Foo( ) MPI_*() PCIe

27 How can TotalView help you? Effective Debugging requires the capability to control and examine specific instances of program execution in detail Threads and/or MPI Deadlocks and hangs Race conditions It provides Asynchronous thread control Powerful group mechanism Fortran and/or C++ Complex data structures Diving and recursive dive STL Collection Classes STLView Rich class hierarchies Powerful type-casting features Memory Analysis Leaks and Bounds Errors Automatic error detection tools Out of Memory Errors Analysis of heap memory usage by file, function and line Numerical errors Extensible data visualization Slicing and filtering of arrays Powerful expression system Conditional watchpoints TotalView provides an answer to the question : What is my program really doing?

28 TotalView NVIDIA CUDA CUDA 4.2, 5.0, 5.5, 6.0 対応 CUDAの統一メモリ (unified memory) CUDAの動的モードのプログラム 1つのセッション内でホストとデバイスのコードをデバッグ TeslaやFermiなどのハードウェア上のCUDA LinuxおよびGPUデバイス上のスレッドを可視化デバイスブロックスレッドメモリの階層構造を完全に可視化デバイスのスレッドを論理座標とデバイス座標の両方で操作可能 CUDAの関数呼び出しホストのピン止めされたメモリ領域 CUDAコンテクストを可視化 CUDAの関数をインラインでもスタック上でもハンドリング使いやすい軽量なGUIと自動化に適したCLI( コマンドライン ) 複数のNVIDIAデバイスを使うアプリケーション CUDAで高速化されたクラスタ上のMPIアプリケーション統一された仮想アドレスとGPUDirect メモリエラーを検知してレポート CUDAの例外をハンドリング

29 CUDA のデバッグ画面 TotalView スレッド (x,y,z) ブロック (x,y,z) ボックスの中にある行番号をクリックして breakpoint を置く GPU フォーカススレッドセレクタで CUDA スレッドのブロック (x,y) やスレッド (x,y,z) インデクスを変更する TotalView のスレッド ID 0 以上 : ホストスレッド 0 より小さい : GPU スレッド

30 TotalView CUDA の例 GPU フォーカススレッドの論理座標 CUDA グリッドとブロックの次元レーンあたりの warp SM あたりのワープワープ数などなどスタックトレースとインライン関数パラメータレジスタ局所変数と共有変数 warp の PC を指す矢印

31 TotalView GPU デバイスの情報を表示デバイス情報論理情報

32 TotalView OpenACC ホストCPU/GPUどちらでもステップ実行コンパイラ Cray CCE 8.x OpenACC

13では未対応 ) ホストとコプロセッサ両方のスレッドを完全に可視化 MPI プログラムの完全サポート

33 TotalView Intel Xeon Phi シンメトリックモデルのデバッグができる初めてのデバッガ付属のMemoryScapeでネイティブおよびシンメトリックモデルのメモリデバッグ ( オフロードは8.13では未対応 ) ホストとコプロセッサ両方のスレッドを完全に可視化 MPI プログラムの完全サポートオフロードコードによる異種混合アプリケーションのシンメトリックデバッグ Xeon Phi ネイティブアプリケーションのリモートデバッグ Xeon および Xeon Phi 両方の非同期スレッド制御マルチホストマルチカードリバースデバッグは未対応

34 TotalView 1 つのデバッグセッションで同じ操作体験ホストコプロセッサ

1. GPU コンピューティング GPU コンピューティング GPUによる汎用コンピューティング GPU = Graphics Processing Unit CUDA Compute Unified Device Architecture NVIDIA の GPU コンピューティング環境 Lin

Windows で始める CUDA 入門 GTC 2013 チュートリアルエヌビディアジャパン CUDA エンジニア森野慎也 1. GPU コンピューティング GPU コンピューティング GPUによる汎用コンピューティング GPU = Graphics Processing Unit CUDA Compute Unified Device Architecture NVIDIA の GPU コンピューティング環境