GPGPU の歴史と応用例 長岡技術科学大学電気電子情報工学専攻出川智啓
今回の内容 GPU の進化の歴史 GPU のアーキテクチャ GPU の産業応用例 38 GPGPU 実践基礎工学
GPU(Graphics Processing Unit) とは 画像処理専用のハードウェア 具体的には画像処理用のチップ チップ単体では販売されていない PCI Ex カードで販売 ( チップ単体と区別せずに GPU と呼ぶことも多い ) マザーボードやノート PC に搭載 PCI Ex カードとして販売される GPU には, ビデオメモリと呼ばれる RAM が搭載 39 GPGPU 実践基礎工学
GPU(Graphics Processing Unit) とは 代表的な製品 NVIDIA GeForce AMD Radeon Intel HD Graphics 代表的な用途 3D グラフィックス処理 3D ゲーム,3DCAD,3DCG 作成 エンコード デコード支援 GPU 上に専用回路を搭載していることが多い デスクトップ PC の GUI 処理 Windows Aero が比較的高性能な GPU を要求 40 GPGPU 実践基礎工学
GPU の性能の遷移 ( 理論演算性能 ) Theoretical GFLOP/s Tesla Fermi GeForce GTX 580 GeForce GTX 480 Kepler GeForce GTX 680 Kepler GeForce GTX TITAN GeForce GTX 280 Tesla M2090 GeForce 8800 GTX Tesla C2050 GeForce 7800 GTX Tesla GeForce 6800 Ultra GeForce FX 5800 Woodcrest Harpertown C1060 Sandy Bridge Pentium 4 Bloomfield Westmere Maxwell GeForce 780 Ti Tesla K40 Tesla K20X Ivy Bridge Apr 01 Sep 02 Jan 04 May 05 Oct 06 Feb 08 Jul 09 Nov 10 Apr 12 Aug 13 Dec 14 GeForce ゲーム用 Quadro CG 用 Tesla GPGPU 用 NVIDIA 社が公開している資料を基に作成 41 GPGPU 実践基礎工学
GPU の性能の遷移 ( 理論バンド幅 ) GeForce 780 Ti Maxwell Tesla K40 GeForce ゲーム用 Theoretical GB/s Tesla GeForce GTX 280 Fermi GeForce GTX 480 Tesla K20X Kepler GeForce GTX 680 Tesla M2090 Tesla C2050 Quadro CG 用 Tesla GPGPU 用 GeForce 8800 GTX Tesla C1060 Ivy Bridge GeForce 7800 GTX Sandy Bridge Bloomfield GeForce 6800 GT Woodcrest GeForce FX 5900 Prescott Westmere Harpertown Northwood 2003 2004 2005 2006 2007 2008 2009 2010 2011 2012 2013 NVIDIA 社が公開している資料を基に作成 42 GPGPU 実践基礎工学
GPU(Graphics Processing Unit) の役割 グラフィクスを表示するために様々な処理を行い, 処理の結果をディスプレイに出力 3 次元グラフィクスの発展に伴って役割が大きく変化 過去 現在 描画情報 CPU 3 次元座標変換 ポリゴンとピクセルの対応付け ピクセル色計算テクスチャ参照 フレームバッファ ( ビデオメモリ ) への書き込み CPU が 3D 描画の演算を実行 GPUが出力 描画情報 3 次元座標変換 GPU ポリゴンとピクセルの対応付け ピクセル色計算テクスチャ参照 フレームバッファ ( ビデオメモリ ) への書き込み GPUが演算から出力までの全てを担当 CPUが描画情報の生成とGPUへの引渡,GPUの制御を行う ディスプレイ出力 ディスプレイ出力 GPU 画面出力 画面出力 43 GPGPU 実践基礎工学
GPU の描画の流れ 1. CPUからGPUへ描画情報を送信 2. 頂点処理 ( 頂点シェーダ ) 座標変換 画面上での頂点やポリゴンの位置 大きさの決定 頂点単位での照明の計算 3. 頂点やポリゴンからピクセルを生成 ( ラスタライザ ) 4. ピクセル処理 ( ピクセルシェーダ ) 画面上のピクセルの色 テクスチャの模様 5. 画面出力 ピクセルの色情報をフレームバッファに書き込み 2. 3. 4. 44 GPGPU 実践基礎工学
ビデオカードの利点 CPU で描画のための演算を行うと,CPU にかかる負荷が大きい 3 次元画像処理の専用回路を備えたハードウェアを導入 CPU にかかる負荷を減らすことができる 頂点 ピクセルごとに並列処理が可能なため, ハードウェアによる並列処理が可能 45 GPGPU 実践基礎工学
ビデオカードの欠点 3 次元画像処理の専用回路を備えたハードウェアを導入 新しい描画方法を開発しても,GPU へ実装 製品化されるまで利用できない ユーザが所有している GPU によって利用できる機能にばらつきが生じる ある描画手法用の専用回路を実装しても, その描画方法が常に使われる訳ではないので GPU 全体の利用効率が下がる 46 GPGPU 実践基礎工学
ビデオカードから GPU へ CG の多様化と共に固定機能の実装が困難に 頂点処理とピクセル処理をユーザが書き換えられるプログラマブルシェーダの実装 グラフィックスカード GPU 頂点処理用回路 頂点シェーダユニット ピクセル処理用回路 ピクセルシェーダユニット 47 GPGPU 実践基礎工学
ビデオカードから GPU へ 描画する画像によって頂点処理とピクセル処理の負荷が変化 処理によっては利用効率に差が発生し, 利用効率が低下 頂点処理重視の処理 GPU 頂点シェーダユニット ピクセル処理重視の処理 GPU 頂点シェーダユニット 空きユニット ピクセルシェーダユニット ピクセルシェーダユニット 空きユニット 48 GPGPU 実践基礎工学
ビデオカードから GPU へ 頂点シェーダとピクセルシェーダを統合したユニファイドシェーダへの進化 頂点処理とピクセル処理を切り替えることで利用率を向上 頂点処理重視の処理 GPU ユニファイドシェーダユニット ピクセル処理重視の処理 GPU ユニファイドシェーダユニット 49 GPGPU 実践基礎工学
ビデオカードから GPU へ 各ピクセルに対して並列に処理実行できるように進化 単純な処理を行う演算器を大量に搭載 高い並列度で処理を実行 GPU の誕生と GPGPU の普及 高性能な 3DCG 画像処理への要求 GPU の高性能化 GPU の長所 消費電力あたりの浮動小数点理論演算性能が高い GPU 単体の消費電力は高い ( 相対的に ) 安価 CPU だけで同等の計算能力を達成するより安価 50 GPGPU 実践基礎工学
GPU の進化 72 Volta 1 Watt あたりの単精度行列 - 行列積の回数 60 48 36 24 12 0 情報処理センター GPGPU システムに搭載 Tesla Fermi Kepler Maxwell Pascal 2008 2010 2012 2014 2016 2018 年 GPU は発展途上 (2~3 年で世代交代 ) NVIDIA 社プレゼンテーションを基に作成 今プログラムを作っておくと, 勝手に速くなってくれる! 51 GPGPU 実践基礎工学
GPU のハードウェア構造 CUDA Core( 旧 Streaming Processor, SP) と呼ばれる演算器を多数搭載 Streaming Multiprocessor(SM, SMX) が複数の CUDA Core と SFU, オンチップメモリをまとめて管理 SFU(Special Function Unit) 数学関数を計算するユニット 複数の SM が集まって GPU を構成 ビデオメモリと呼ばれる大容量のオフチップメモリも管理 52 GPGPU 実践基礎工学
Tesla アーキテクチャ Tesla C1060 の仕様 SM 数 30 CUDA Core 数 240(=8 Core/SM 30 SM) キャッシュを搭載せず 53 GPGPU 実践基礎工学
Tesla アーキテクチャ Tesla C1060 の仕様 CUDA コア数 ( 単精度 ) CUDA コアクロック周波数 240 Cores 1,296 MHz 単精度演算ピーク性能 622* 1 (933* 2 ) GFLOPS 倍精度演算ユニット数 30* 3 Units 倍精度演算ピーク性能 メモリクロック周波数 メモリバス幅 最大メモリバンド幅 * 4 78 GFLOPS 800 MHz 512 bit 102 GB/s * 1 単精度演算ピーク性能 = コアクロック周波数 コア数 命令の同時発行数 (2) * 2 CUDA Core と SFU が同時に命令を発行できれば 1296 MHz 240 3 * 3 一つの SM に倍精度演算器が一つ搭載 * 4 最大メモリバンド幅 = メモリクロック周波数 メモリバス幅 /8 2(Double Data Rate) 54 GPGPU 実践基礎工学
Fermi アーキテクチャ Tesla M2050 の仕様 SM 数 14 CUDA Core 数 448(=32 Core/SM 14 SM) L1/L2 キャッシュを搭載 ECC( 誤り訂正機能 ) を搭載 55 GPGPU 実践基礎工学
Fermi アーキテクチャ Tesla M2050 の仕様 CUDAコア数 ( 単精度 ) CUDAコアクロック周波数単精度演算ピーク性能倍精度演算ユニット数倍精度演算ピーク性能メモリクロック周波数メモリバス幅最大メモリバンド幅 448 Cores 1,150 MHz 1.03 TFLOPS 0* 1 Unit 515 GFLOPS 1.55 GHz 384 bit 148 GB/s * 1 単精度 CUDA Core を 2 基使って倍精度演算を実行 56 GPGPU 実践基礎工学
Kepler アーキテクチャ Tesla K20c/m の仕様 SMX 数 13 Streaming Multiprocessor extreme (?) CUDA Core 数 2,496(=192 Core/SM 13 SMX) 57 GPGPU 実践基礎工学
Kepler アーキテクチャ Tesla K20c/m の仕様 CUDAコア数 ( 単精度 ) 2,496 Cores CUDAコアクロック周波数 706 MHz 単精度演算ピーク性能 3.52 TFLOPS 倍精度演算ユニット数 832* 1 Units 倍精度演算ピーク性能 1.17 TFLOPS メモリクロック周波数 2.6 GHz メモリバス幅 320 bit 最大メモリバンド幅 208 GB/s * 1 64 基 /SMX 13 基 58 GPGPU 実践基礎工学
Maxwell アーキテクチャ GeForce GTX TITAN X の仕様 SM 数 24 CUDA Core 数 3,072(=128 Core/SM 24 SM) http://www.itmedia.co.jp/pcuser/articles/1409/19/news051.html より引用 59 GPGPU 実践基礎工学
Maxwell アーキテクチャ GeForce GTX TITAN X の仕様 * CUDA コア数 ( 単精度 ) CUDA コアクロック周波数 単精度演算ピーク性能 3,072 Cores 1,002 MHz 6.14 TFLOPS 倍精度演算ユニット数 0* 1 Unit 倍精度演算ピーク性能 192 GFLOPS* 2 メモリクロック周波数 3.5 GHz* 3 メモリバス幅 最大メモリバンド幅 *http://ja.wikipedia.org/wiki/flops http://http://www.geforce.com/hardware/desk top gpus/geforce gtx titan x/specifications 384 bit 336.5 GB/s * 1 http://www.4gamer.net/games/121/g012181/20141225075/ * 2 倍精度演算は単精度演算の性能の 1/32 (1/16 Flop/core/clock) * 3 DDR(Double Data Rate) 7GHz 相当と書かれている場合もある 60 GPGPU 実践基礎工学
Pascal アーキテクチャ 2016 年にリリース予定 倍精度演算器を搭載予定 NVLink GPU 同士や GPU と CPU を接続する独自の方式 通信 (CPU メモリ PCI Express メモリ GPU) のボトルネックを解消 (PCI Express3.0 の 5~12 倍 ) 複数の GPU を使って大規模な計算が可能 3D メモリ (High Bandwidth Memory, HBM)* 3 次元積層技術を利用し, メモリの容量と帯域を大幅に増加 最大 32GB, メモリ帯域 1TB/s *http://pc.watch.impress.co.jp/docs/column/kaigai/20150421_698806.html 61 GPGPU 実践基礎工学
Volta アーキテクチャ Pascal の後継 詳しい情報は不明 アメリカの次世代スーパーコンピュータへ採用予定 オークリッジ国立研究所 SUMMIT 150~300PFLOPS ローレンス リバモア研究所 SIERRA 100PFLOPS 以上 地球シミュレータと同等の演算性能を 1 ノードで実現 現在 Top500 2 位のスーパーコンピュータと同じ電力で 5~10 倍高速, サイズは 1/5 *http://www.4gamer.net/games/121/g012181/20141225075/ 62 GPGPU 実践基礎工学
GPU プログラミング環境 Compute Unified Device Architecture NVIDIA 社製 GPU 向け開発環境 (Windows,Linux,Mac OS X) 2007 年頃発表 C/C++ 言語 + 独自のGPU 向け拡張 専用コンパイラ (nvcc) とランタイムライブラリ いくつかの数値計算ライブラリ ( 線形代数計算,FFTなど) CUDA 登場以前 グラフィクスプログラミングを利用 足し算と色を混ぜる処理を対応させる等 汎用計算のためには多大な労力が必要 63 GPGPU 実践基礎工学
プログラマブルシェーダを用いた汎用計算 グラフィックス API(DirectX, OpenGL) による描画処理 + シェーダ言語 (HLSL, GLSL) による演算 void gpumain(){ vec4 ColorA = vec4(0.0, 0.0, 0.0, 0.0); vec4 ColorB = vec4(0.0, 0.0, 0.0, 0.0); vec2 TexA = vec2(0.0, 0.0); vec2 TexB = vec2(0.0, 0.0); TexA.x = gl_fragcoord.x; TexA.y = gl_fragcoord.y; TexB.x = gl_fragcoord.x; TexB.y = gl_fragcoord.y; } ColorA = texrect( texunit0, TexA ); ColorB = texrect( texunit1, TexB ); gl_fragcolor = F_ALPHA*ColorA + F_BETA*ColorB; シェーダ言語を用いた配列加算 (c= *a + *b) の例 void main(){ glutinit( &argc, argv ); glutinitwindowsize(64,64);glutcreatewindow("gpgpuhelloworld"); glgenframebuffersext(1, &g_fb); glbindframebufferext(gl_framebuffer_ext, g_fb); glgentextures(4, g_ntexid); // create (reference to) a new texture glbindtexture(opt1, texid); gltexparameteri(opt1, GL_TEXTURE_MIN_FILTER, GL_NEAREST); gltexparameteri(...); glteximage2d(opt1, 0, opt2, width, height, 0, GL_RGBA, GL_FLOAT, 0); ( 以下省略 ) GPU の処理 (GLSL) 各ピクセルに対して実行 CPU の処理 (OpenGL) 64 GPGPU 実践基礎工学
CUDA による汎用計算 (c= *a + *b) #define N (1024*1024) #define Nbytes (N*sizeof(float)) #define NT 256 #define NB (N/NT) global void init(float *a, float *b, float *c){ int i = blockidx.x*blockdim.x + threadidx.x; a[i] = 1.0; b[i] = 2.0; c[i] = 0.0; } global void add(float *a, float, float *b, float, float *c){ int i = blockidx.x*blockdim.x + threadidx.x; } c[i] = *a[i] + *b[i]; int main(){ float *a,*b,*c; float, ; } cudamalloc((void **)&a, Nbytes); cudamalloc((void **)&b, Nbytes); cudamalloc((void **)&c, Nbytes); =...; =...; init<<< NB, NT >>>(a,b,c); add<<< NB, NT >>>(a,, b,, c); return 0; 65 GPGPU 実践基礎工学
GPGPU の産業応用例
ロボットによる心臓外科手術 ロボットアームと内視鏡を使った心臓外科手術 人工心臓を使わず, 心臓を動かしたまま手術を実施 67 GPGPU 実践基礎工学
ロボットによる心臓外科手術 心臓は複雑な形状で, かつ周期的に脈動 毎秒何十枚と撮られる映像をリアルタイムで処理しながらロボットアームを制御 心臓を 2 次元の画像に変換 その画像を基にロボットが動く道筋を計算 実際に 3 次元の動きに変換 心臓の動きに合わせてロボットアームを制御 https://www.youtube.com/watch?v=g3liyn_oceo 68 GPGPU 実践基礎工学
GPU を利用した研究, 製品開発, 応用例 東京工業大学青木研究室 TSUBAME を活用し, 気象計算や金属凝固のシミュレーションなどの高速化を達成 東工大は CUDA Center of Excellence に認定 ゴードンベル賞 ( その年にスーパーコンピュータを科学的に最も優れた形で応用した成果に対して付与される賞 ) 受賞 CUDA Fellow( 世界で 15 人 ) GPU を利用した計算結果のアニメーション 69 GPGPU 実践基礎工学
住宅設備機器開発 住宅設備機器開発のための混相流シミュレーション 掲示されていた画像を著者が撮影 70 GPGPU 実践基礎工学
GPU を利用した研究, 製品開発, 応用例 Prometech Software 株式会社 粒子法と呼ばれる方法を利用した流体解析ソフトウェアを開発 販売 複雑な流動現象のシミュレーションから映画の CG 製作まで, 幅広く手がける http://www.prometech.co.jp/ 71 GPGPU 実践基礎工学
GPU を利用した研究, 製品開発, 応用例 NVIDIA の最新事例 GPU Technology Conference 2014: Iray VCA (part 7) GTC 先進運転支援システム等に向けた NVIDIA の新画像処理 画像認識プログラム開発環境 エヌビディア最新情報と製造業分野での取り組み 72 GPGPU 実践基礎工学
車載システム Google, NVIDIA, Audi などによる Open Automotive Alliance 自動車への Android プラットフォーム搭載促進を目指す http://microsites.audi.com/marvelsavengers/index.html?locale=en_us 73 GPGPU 実践基礎工学
NVIDIA 社の自動車関連ソリューション NVIDIA DRIVE PX 高機能運転支援システム NVIDIA DRIVE CX 自動車の高度情報化 http://www.nvidia.co.jp/object/drive px jp.html http://www.nvidia.co.jp/object/drive cx jp.html 74 GPGPU 実践基礎工学
Jetson Tegra TK1 世界初の組み込みスーパーコンピューター NVIDIA Tegra K1を採用 Keplarアーキテクチャ 192 CUDA コア NVIDIA 4 Plus 1 クアッドコア ARM Cortex A15 CPU Linux for Tegra が動作 75 GPGPU 実践基礎工学
月面探査 Google Lunar XPRIZE Google による国際宇宙開発レース 日本の民間月面探査チーム HAKUTO が参加 ロケット発射時のシミュレーション 着陸などのシミュレーションや実際の制御に GPU を用いた自動運転技術を利用 http://team hakuto.jp/1549/ https://blogs.nvidia.co.jp/2015/03/18/gpus payloads to moon/ 76 GPGPU 実践基礎工学