Slide 1
|
|
|
- ありさ じゅふく
- 6 years ago
- Views:
Transcription
1 電子情報通信学会研究会組込みシステム研究会 (IPSJ-EMB) 2010 年 1 月 28 日 超並列マルチコア GPU を用いた高速演算処理の実用化 NVIDIA Solution Architect 馬路徹
2 目次 なぜ今 GPU コンピューティングか? CPUの性能向上速度が減速 性能向上 = 並列処理 にGPUコンピューティングが応える CUDAシステムアーキテクチャによる超並列処理の実現 普及 NVIDIA GPU アーキテクチャの進歩 3Dグラフィックスの進歩及び統合化シェーダエンジンへの変遷統合化シェーダエンジンへのCUDAの組込み 次世代 GPU コンピューティング Fermi アーキテクチャ まとめ プロセッサ / メモリーアーキテクチャ GigaThreadスレッド制御機構による更なる高性能化統合化プログラム開発環境
3 CUDA は Performance = Parallelism を実現 Performance = Parallelismへの期待このParallelismを実現するために必要なことは? 相当な数の高効率プロセッサ (4 個 8 個のオーダではない ) 並列処理を抽象化できるプログラミング システム CUDA (Compute Unified Device Architecture GPU は数百個の多機能 高速プロセッサを内蔵 マルチスレッド アーキテクチャがこの超並列マルチコアを高効率に活用 CUDA は普及している C 言語の拡張からスタートし 現在は Fortran OpenCL や Direct Compute まで拡張されている CUDA は並列プログラミングの詳細な記述からプログラマーを解放
4 最初に C with CUDA Extension を開発 標準 C コード CPU からコールし GPU で実行する関数 CUDA コード ク ローハ ル スレット ID をそのままインテ ックスに活用 void saxpy_serial(int n, float alpha, float *x, float *y) { for (int i=0; i<n; ++i) y[i] = alpha * x[i] + y[i]; } // Invoke serial saxpy() kernel saxpy_serial(n, 2.0, x, y); global void saxpy_parallel(int n, float alpha, float *x, float *y) { int i = blockidx.x * blockdim.x + threadidx.x; if (i < n) y[i] = alpha * x[i] + y[i]; } // Invoke parallel saxpy() kernel (256 threads per block) int nblocks = (n + 255)/256; saxpy_parallel <<< nblocks, 256 >>>(n, 2.0, x, y); 階層的スレッド分割数 ( ブロック数 / グリッド ) ( スレッド数 / ブロック )
5 今日の GPU コンピューティングの普及 OpenCL is trademark of Apple Inc. used under license to the Khronos Group Inc. 市場展開モーメンタム CUDA 実装の GPU が 市場に 1 億個以上出荷 60,000 以上の GPU コンピューティング開発者 Windows, Linux 及び MacOS プラットフォームのサポート 200 以上の大学で GPU コンピューティングの講座 C + CUDA Extension GPU コンピューティング アプリ OpenCL Khronos Apple Direct Compute Microsoft FORTRAN NVIDIA GPU with the CUDA Parallel Computing Architecture Java and Python planned
6 CUDA Not 2x or 3x, Speed-ups are 20x to 150x Results with Telsa 8 Series relative to compute performance using CPU exclusively 146X 36X 19X 17X 100X Interactive visualization of volumetric white matter connectivity Ionic placement for molecular dynamics simulation on GPU Transcoding HD video stream to H.264 Simulation in Matlab using.mex file CUDA function Astrophysics N- body simulation 149X 47X 20X 24X 30X Financial simulation of LIBOR model with swaptions GLAME@lab: An M-script API for linear Algebra operations on GPU Ultrasound medical imaging for cancer diagnostics Highly optimized object oriented molecular dynamics Cmatch exact string matching to find similar proteins and gene sequences
7 並みの性能向上ではない 2-3 倍の性能向上は単なる 高性能化 顧客の基本的なワークフローに変化は無い 5-10 倍の性能向上は 画期的 装置のアップグレードの価値は十分にある ( 一部または大部分の ) アプリ ソフトを書き換える意味はある 100 倍以上の性能向上は 世界観を変える! プラットフォームの取替えの価値はあるアプリケーションのアーキテクチャまでを見直す意味がある今まで実用的に不可能であったアプリの開発が可能になる科学技術において 新発見までの時間 を短縮する画期的な変化をもたらす
8 GPU コンピューティングが導入された OS 身近になる GPU コンピューティング Mac OS X Snow Leopard
9 目次 なぜ今 GPU コンピューティングか? CPUの性能向上速度が減速 性能向上 = 並列処理 にGPUコンピューティングが応える CUDAシステムアーキテクチャによる超並列処理の実現 普及 NVIDIA GPU アーキテクチャの進歩 3Dグラフィックスの進歩及び統合化シェーダエンジンへの変遷統合化シェーダエンジンへのCUDAの組込み 次世代 GPU コンピューティング Fermi アーキテクチャ まとめ プロセッサ / メモリーアーキテクチャ GigaThreadスレッド制御機構による更なる高性能化統合化プログラム開発環境
10 NVIDIA GPU による 3D グラフィックスの進歩 フ ロク ラマヒ リティーの大幅向上 ( 統合化シェータ ーエンシ ン ) 汎用演算器 1995 NV1 1 Million Transistors 1999 GeForce 256 DirectX 7 22 Million Transistors 2002 GeForce4 Direct X 8 63 Million Transistors 2003 GeForce FX DirectX Million Transistors 2004 GeForce 6 DirectX 9c 222 Million Transistors 2005 GeForce 7 DirectX 9c 302 Million Transistors 2006 GeForce 8 DirectX 10 (Vista) 681 Million Transistors 2008 GeForce GTX Billion Transistors フ ロク ラマヒ リティーの向上 Vertex(FP) +Pixel Shader Vertex(FP) +Pixel(FP) Shader 統合化 Shader CUDA Pixel Shade
11 20 年間続いたグラフィックス アーキテクチャ各機能に固定された用途のハードウエア Vertex Triangle 座標変換及び光源処理 DirectX8 よりプログラマブル頂点シェーダーとなる 三角形 点 線セットアップ Pixel Raster OPeration フラット シェーディング テクスチャーマッピング等 DirectX8 よりプログラマブル ピクセルシェーダーとなる ブレンディング Z バッファー アンチエイリアシング Memory メモリ
12 DirectX 10 (Vista) 以降の GPU 統合型プログラマブル シェーダエンジン GPU はマルチスレッドを処理する超並列マルチコアプロセッサとなった 頂点シェーダ ジオメトリ シェーダ及びピクセル シェーダは上記により実行される Host 頂点シェーダ ジオメトリシェーダ ピクセルシェーダ SP Input Assembler Thread Execution Manager Thread Processors Thread Processors Thread Processors Thread Processors Thread Processors Thread Processors Thread Processors Thread Processors Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache Parallel Data Cache SM SP: Stream Processor SM: Stream Multiprocessor Load/store Global Memory
13 第 1 世代の SM (Streaming Multiprocessor) Streaming Multiprocessor (SM) 8 Streaming Processors (SP) 各 SP は 32b 単精度浮動小数点 整数演算サポート 2 Super Function Units (SFU) 三角関数 平方根等サポート 全ての SP は同一命令を実行開始 ( 途中分岐あり ) SIMT (Single instruction Multiple Thread) マルチスレッド命令ディスパッチ 1 から768 スレッドがアクティブ 32スレッド単位 (warp) でSIMT 命令実行 巨大なローカル レジスタファイルRF 8,192 Registers / SM ハードウエア コンテキストスイッチを容易化 16 KB シェアード メモリ S F U SP 0 RF 0 SP 1 RF 1 SP 2 RF 2 SP 3 RF 3 SM 命令フェッチ I 命令キャッシュ マルチスレッド命令ディスパッチ シェアド メモリ RF 4 RF 5 RF 6 RF 7 定数キャッシュ SP 4 SP 5 SP 6 SP 7 S F U テキスチャーフェッチ
14 1024 NVIDIA GPU コア数の変遷 Number of programmable shaders / die Ti GeForce 3XXX Vertex Shader + Pixel Shader Ti Ti GeForce 4XXX Programmable Shader FX FX FX GeForce 5XXX 6800GT XT GeForce 6XXX 7900GT GT GT GT SE SE 1+4 GeForce 7XXX 8800GTX GS GTS GT GT 8 GeForce 8XXX 9800GTX GT GT GT 16 GeForce 9XXX Unified Shader Tesla Architecture GTX GTX GTX GTX GeForce GT2XX C C Tesla Unified Shader CUDA, OpenCL, DirectX Compute etc Fermi Architecture FERMI Vertical axis shows the number of cores, but not the performance - Only major products are shown Fermi
15 GPU-CPU 性能及びメモリーバンド幅の差は拡大 メモリーバンド幅ピーク値 GB/sec TF Single Precision 4GB Memory 8x double precision ECC L1, L2 Caches 100 NVIDIA GPU X86 CPU
16 目次 なぜ今 GPU コンピューティングか? CPUの性能向上速度が減速 性能向上 = 並列処理 にGPUコンピューティングが応える CUDAシステムアーキテクチャによる超並列処理の実現 普及 NVIDIA GPU アーキテクチャの進歩 3Dグラフィックスの進歩及び統合化シェーダエンジンへの変遷統合化シェーダエンジンへのCUDAの組込み 次世代 GPU コンピューティング Fermi アーキテクチャ まとめ プロセッサ / メモリーアーキテクチャ GigaThreadスレッド制御機構による更なる高性能化統合化プログラム開発環境
17 Fermi アーキテクチャのご紹介スーパコンピュータの魂を持つ GPU DRAM I/F DRAM I/F 30 億個のトランジスターコア数を倍増 (512 コア ) HOST I/F DRAM I/F 倍精度浮動小数点演算ピーク性能が8 倍 GPUとして初めてECCを導入 Giga Thread L2 DRAM I/F L1 及び L2 キャッシュを内蔵約 2 倍のメモリーバンド幅 (GDDR5) DRAM I/F DRAM I/F 最大 1 Terabyte の GPU メモリ複数 Kernel 同時実行 C++ サポート
18 SM (Streaming Multiprocessor) アーキテクチャ Instruction Cache Scheduler Scheduler SM 当たり 32 CUDA コア ( 総数 512 個 ) Dispatch Dispatch Register File 倍精度浮動小数点演算ピーク性能は 8 倍 単精度演算ピーク性能の 50% 2 個の Thread Scheduler 2 ワープを 2 組の 16 個の CUDA, 16 個の Load/Store Unit, 4 個の SFU に同時にディスパッチする シェアードメモリ L1 キャッシュとして使用する 64 KB RAM ( 構成可変 ) Load/Store Units x 16 Special Func Units x 4 Interconnect Network 64K Configurable Cache/Shared Mem Uniform Cache
19 CUDA コア アーキテクチャ Instruction Cache Scheduler Scheduler 最新の浮動小数点演算規格 IEEE 準拠 ( 最新 CPU をも凌駕 ) Dispatch Dispatch Register File Fused multiply-add (FMA) 命令を 倍精度 単精度でサポート ( 積和の最終段でラウンディングするため CUDA Dispatch Port 各段で各々行うより精度が向上 ) Operand Collector 新規設計の全命令 32b サポート整数 ALU 64-bit 及びそれ以上の精度に対しても最適化設計 FP Unit INT Unit Result Queue Load/Store Units x 16 Special Func Units x 4 Interconnect Network 64K Configurable Cache/Shared Mem Uniform Cache
20 統合化された 64 ビットメモリー空間 C/C++ ポインターのフルサポート Non-unified Address Space Local *p_local Shared C 言語やC++ 言語のポインタはターゲットとするアドレス空間が必ずしもコンパイル時に確定せず ランタイムに動的に決まるケースがあるため このポインタを完全に実装することが困難な状況となっていました しかし 統合化されたアドレス空間をもつFermiではこの問題もなくなりました *p_shared Global 0 32-bit *p_global Unified Address Space Local Shared Global 0 40-bit *p
21 キャッシュメモリ階層構造 オンチップ シェアードメモリとともに本格的な キャッシュ階層構造を有する初めての GPU Fermi Memory Hierarchy Thread 各 SM (32 cores) に L1 キャッシュ (48KB または 16KB) メモリバンド幅の改善及びアクセス遅延の低減 Shared Memory L1 Cache ユニファイド L2 キャッシュ (768 KB) GPU 上の全コアにわたり高速に コーヒレンント データをシェア L2 Cache M A R D
22 拡張され 高速化されたメモリ インタフェース GDDR5 メモリーインタフェース GDDR3 の 2 倍の速度 最大 1 Terabyte の GPU メモリ 大きなデータセットを取り扱うことが可能現在 40bアドレス 命令は64bをサポート 384 ビットのメモリーインタフェース 64b partitioning x 6 DRAM I/F Giga Thread HOST I/F DRAM I/F L2 DRAM I/F DRAM I/F DRAM I/F DRAM I/F
23 ECC (Error Checking and Correction) DRAM の ECC によるデータエラーの除去 GDDR5 メモリーの ECC サポート 内部の主要な記憶素子も ECC で保護 レジスターファイル L1 キャッシュ L2 キャッシュ Single-Error Correct Double-Error Detect (SECDED) サポート 2 ビットの誤りはソフト処理 ( 再度実行等 )
24 IEEE 規格準拠の高精度演算 IEEE results 64-bit double precision 32-bit single precision full-speed denormal operands & results NaNs, +/- Infinity IEEE rounding nearest even, zero, +inf, -inf IEEE Fused Multiply-Add (FMA) D = A*B + C; No loss of precision IEEE divide & sqrt use FMA Multiply-Add (MAD): D = A*B + C; A B = Product + C = D (truncate digits) Fused Multiply-Add (FMA): D = A*B + C; A B = (retain all digits) Product + C = D (no loss of precision)
25 GigaThread TM Hardware Thread Scheduler (HTS) 階層的に何千ものアクティブなスレッドを管理 コンテキストスイッチが 10 倍高速 HTS 複数 kernel の同時実行
26 GigaThread Hardware Thread Scheduler 複数 Kernel 同時実行 + 高速コンテキストスイッチ Kernel 1 Kernel 1 Kernel 2 Kernel 2 Kernel 2 Kernel 3 Ker 4 Time Kernel 2 Kernel 3 nel Kernel 5 Kernel 4 Kernel 5 Serial Kernel Execution Parallel Kernel Execution
27 GigaThread Streaming Data Transfer Engine デュアル DMA エンジン CPU GPU 及び GPU CPU データ転送の同時実行 CPU と GPU 演算と完全にオーバラップ可能 処理の流れ : SDT Kernel 0 CPU SDT0 GPU SDT1 Kernel 1 CPU SDT0 GPU SDT1 Kernel 2 CPU SDT0 GPU SDT1 Kernel 3 CPU SDT0 GPU SDT1
28 各世代の比較 GPU G80 GT200 Fermi 集積トランジスタ数 6 億 8100 万個 14 億個 30 億個 CUDAコア数 倍精度浮動小数点演算能力 30 FMA 演算 / クロック 256 FMA 演算 / クロック 単精度浮動小数点演算能力 128 MAD 演算 / クロック 240 MAD 演算 / クロック 512 MAD 演算 / クロック ワープスケジューラ (SMあたり搭載数) 特殊関数ユニット (SFU) /SM 共有メモリ /SM 16KB 16KB 48KB/16KB ( 構成可能 ) L1キャッシュ /SM 16KB/48KB ( 構成可能 ) L2キャッシュ /SM 768KB ECCメモリのサポート 平行実行カーネル数 最大 16 ロード / ストアのアドレス幅 32ビット 32ビット 64ビット
29 NVIDIA Nexus IDE( 統合化開発環境 ) 業界初の超並列アプリ開発用 IDE (Integrated Development Environment) C, C++, OpenCL, DirectCompute 及び DirectX と OpenGL の両グラフィックス API をサポート 完全にVisual Studioに組込まれた開発環境 (CPU + GPU) コプロセッシング アプリの開発を効率化 両プロセッサにまたがるソース デバッグ 性能解析 両プロセッサにわたるイベント データのキャプチャ
30 NVIDIA Nexus IDE( 統合化開発環境 ) 画面
31 まとめ CPUのIPL 性能向上は減速している Performance = Parallelism が今後の性能確保の要統合化シェーダエンジンによりGPUは汎用の超並列プロセッサとなる CUDA GPU コンピューティングはGPUの並列性能を最大限に引き出し 並列プログラミングの抽象化によりプログラミングも容易にする次世代 Fermiアーキテクチャは更にコンピューティング機能 性能を向上した
32 Thank you for your attention
Slides: TimeGraph: GPU Scheduling for Real-Time Multi-Tasking Environments
計算機アーキテクチャ第 11 回 マルチプロセッサ 本資料は授業用です 無断で転載することを禁じます 名古屋大学 大学院情報科学研究科 准教授加藤真平 デスクトップ ジョブレベル並列性 スーパーコンピュータ 並列処理プログラム プログラムの並列化 for (i = 0; i < N; i++) { x[i] = a[i] + b[i]; } プログラムの並列化 x[0] = a[0] + b[0];
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 コンピューティング環境
07-二村幸孝・出口大輔.indd
GPU Graphics Processing Units HPC High Performance Computing GPU GPGPU General-Purpose computation on GPU CPU GPU GPU *1 Intel Quad-Core Xeon E5472 3.0 GHz 2 6 MB L2 cache 1600 MHz FSB 80 GFlops 1 nvidia
GPU GPU CPU CPU CPU GPU GPU N N CPU ( ) 1 GPU CPU GPU 2D 3D CPU GPU GPU GPGPU GPGPU 2 nvidia GPU CUDA 3 GPU 3.1 GPU Core 1
GPU 4 2010 8 28 1 GPU CPU CPU CPU GPU GPU N N CPU ( ) 1 GPU CPU GPU 2D 3D CPU GPU GPU GPGPU GPGPU 2 nvidia GPU CUDA 3 GPU 3.1 GPU Core 1 Register & Shared Memory ( ) CPU CPU(Intel Core i7 965) GPU(Tesla
Images per Second Images per Second VOLTA: ディープラーニングにおける大きな飛躍 ResNet-50 トレーニング 2.4x faster ResNet-50 推論 TensorRT - 7ms レイテンシ 3.7x faster P100 V100 P10
NVIDIA TESLA V100 CUDA 9 のご紹介 森野慎也, シニアソリューションアーキテクト (GPU-Computing) NVIDIA Images per Second Images per Second VOLTA: ディープラーニングにおける大きな飛躍 ResNet-50 トレーニング 2.4x faster ResNet-50 推論 TensorRT - 7ms レイテンシ
熊本大学学術リポジトリ Kumamoto University Repositor Title GPGPU による高速演算について Author(s) 榎本, 昌一 Citation Issue date Type URL Presentation
熊本大学学術リポジトリ Kumamoto University Repositor Title GPGPU による高速演算について Author(s) 榎本, 昌一 Citation Issue date 2011-03-17 Type URL Presentation http://hdl.handle.net/2298/23539 Right GPGPU による高速演算について 榎本昌一 東京大学大学院工学系研究科システム創成学専攻
Microsoft PowerPoint - GPU_computing_2013_01.pptx
GPU コンピューティン No.1 導入 東京工業大学 学術国際情報センター 青木尊之 1 GPU とは 2 GPGPU (General-purpose computing on graphics processing units) GPU を画像処理以外の一般的計算に使う GPU の魅力 高性能 : ハイエンド GPU はピーク 4 TFLOPS 超 手軽さ : 普通の PC にも装着できる 低価格
GPGPU
GPGPU 2013 1008 2015 1 23 Abstract In recent years, with the advance of microscope technology, the alive cells have been able to observe. On the other hand, from the standpoint of image processing, the
GPUコンピューティング講習会パート１
GPU コンピューティング (CUDA) 講習会 GPU と GPU を用いた計算の概要 丸山直也 スケジュール 13:20-13:50 GPU を用いた計算の概要 担当丸山 13:50-14:30 GPU コンピューティングによる HPC アプリケーションの高速化の事例紹介 担当青木 14:30-14:40 休憩 14:40-17:00 CUDA プログラミングの基礎 担当丸山 TSUBAME の
Microsoft PowerPoint - GPGPU実践基礎工学（web）.pptx
GPU のプログラム構造 長岡技術科学大学電気電子情報工学専攻出川智啓 今回の内容 GPU プログラミング環境 (CUDA) GPU プログラムの実行の流れ CUDA によるプログラムの記述 カーネル (GPU で処理する関数 ) の構造 記述方法とその理由 GPU 固有のパラメータの確認 405 GPU(Graphics Processing Unit) とは 画像処理専用のハードウェア 具体的には画像処理用のチップ
23 Fig. 2: hwmodulev2 3. Reconfigurable HPC 3.1 hw/sw hw/sw hw/sw FPGA PC FPGA PC FPGA HPC FPGA FPGA hw/sw hw/sw hw- Module FPGA hwmodule hw/sw FPGA h
23 FPGA CUDA Performance Comparison of FPGA Array with CUDA on Poisson Equation ([email protected]), ([email protected]), ([email protected]), ([email protected]),
GPU のアーキテクチャとプログラム構造 長岡技術科学大学電気電子情報工学専攻出川智啓
GPU のアーキテクチャとプログラム構造 長岡技術科学大学電気電子情報工学専攻出川智啓 今回の内容 GPU のアーキテクチャ CUDA CUDA によるプログラミング 58 GPU(Graphics Processing Unit) とは 画像処理専用のハードウェア 具体的には画像処理用のチップ チップ単体では販売されていない PCI Ex カードで販売 ( チップ単体と区別せずに GPU と呼ぶことも多い
MATLAB® における並列・分散コンピューティング ~ Parallel Computing Toolbox™ & MATLAB Distributed Computing Server™ ~
MATLAB における並列 分散コンピューティング ~ Parallel Computing Toolbox & MATLAB Distributed Computing Server ~ MathWorks Japan Application Engineering Group Takashi Yoshida 2016 The MathWorks, Inc. 1 System Configuration
1 GPU GPGPU GPU CPU 2 GPU 2007 NVIDIA GPGPU CUDA[3] GPGPU CUDA GPGPU CUDA GPGPU GPU GPU GPU Graphics Processing Unit LSI LSI CPU ( ) DRAM GPU LSI GPU
![1 GPU GPGPU GPU CPU 2 GPU 2007 NVIDIA GPGPU CUDA[3] GPGPU CUDA GPGPU CUDA GPGPU GPU GPU GPU Graphics Processing Unit LSI LSI CPU ( ) DRAM GPU LSI GPU](/thumbs/89/99132402.jpg "1 GPU GPGPU GPU CPU 2 GPU 2007 NVIDIA GPGPU CUDA[3] GPGPU CUDA GPGPU CUDA GPGPU GPU GPU GPU Graphics Processing Unit LSI LSI CPU ( ) DRAM GPU LSI GPU")
GPGPU (I) GPU GPGPU 1 GPU(Graphics Processing Unit) GPU GPGPU(General-Purpose computing on GPUs) GPU GPGPU GPU ( PC ) PC PC GPU PC PC GPU GPU 2008 TSUBAME NVIDIA GPU(Tesla S1070) TOP500 29 [1] 2009 AMD
３次多項式パラメタ推定計算の CUDAを用いた実装 （CUDAプログラミングの練習として） Implementation of the Estimation of the parameters of 3rd-order-Polynomial with CUDA
3 次多項式パラメタ推定計算の CUDA を用いた実装 (CUDA プログラミングの練習として ) Estimating the Parameters of 3rd-order-Polynomial with CUDA ISS 09/11/12 問題の選択 目的 CUDA プログラミングを経験 ( 試行錯誤と習得 ) 実際に CPU のみの場合と比べて高速化されることを体験 問題 ( インプリメントする内容
ストリームを用いたコンカレントカーネルプログラミングと最適化 エヌビディアジャパン CUDAエンジニア森野慎也 GTC Japan 2014
ストリームを用いたコンカレントカーネルプログラミングと最適化 エヌビディアジャパン CUDAエンジニア森野慎也 GTC Japan 2014 コンカレントな処理の実行 システム内部の複数の処理を 平行に実行する CPU GPU メモリ転送 カーネル実行 複数のカーネル間 ストリーム GPU 上の処理キュー カーネル実行 メモリ転送の並列性 実行順序 DEFAULT STREAM Stream : GPU
iphone GPGPU GPU OpenCL Mac OS X Snow LeopardOpenCL iphone OpenCL OpenCL NVIDIA GPU CUDA GPU GPU GPU 15 GPU GPU CPU GPU iii OpenMP MPI CPU OpenCL CUDA OpenCL CPU OpenCL GPU NVIDIA Fermi GPU Fermi GPU GPU
TSUBAME2.0 における GPU の 活用方法 東京工業大学学術国際情報センター丸山直也第 10 回 GPU コンピューティング講習会 2011 年 9 月 28 日
TSUBAME2.0 における GPU の 活用方法 東京工業大学学術国際情報センター丸山直也第 10 回 GPU コンピューティング講習会 2011 年 9 月 28 日 目次 1. TSUBAMEのGPU 環境 2. プログラム作成 3. プログラム実行 4. 性能解析 デバッグ サンプルコードは /work0/gsic/seminars/gpu- 2011-09- 28 からコピー可能です 1.
スライド 1
GPU クラスタによる格子 QCD 計算 広大理尾崎裕介 石川健一 1.1 Introduction Graphic Processing Units 1 チップに数百個の演算器 多数の演算器による並列計算 ~TFLOPS ( 単精度 ) CPU 数十 GFLOPS バンド幅 ~100GB/s コストパフォーマンス ~$400 GPU の開発環境 NVIDIA CUDA http://www.nvidia.co.jp/object/cuda_home_new_jp.html
CUDA を用いた画像処理 画像処理を CUDA で並列化 基本的な並列化の考え方 目標 : 妥当な Naïve コードが書ける 最適化の初歩がわかる ブロックサイズ メモリアクセスパターン
CUDA 画像処理入門 エヌビディアジャパン CUDA エンジニア森野慎也 GTC Japan 2014 CUDA を用いた画像処理 画像処理を CUDA で並列化 基本的な並列化の考え方 目標 : 妥当な Naïve コードが書ける 最適化の初歩がわかる ブロックサイズ メモリアクセスパターン RGB Y( 輝度 ) 変換 カラー画像から グレイスケールへの変換 Y = 0.299 R + 0.587
DO 時間積分 START 反変速度の計算 contravariant_velocity 移流項の計算 advection_adams_bashforth_2nd DO implicit loop( 陰解法 ) 速度勾配, 温度勾配の計算 gradient_cell_center_surface 速
1 1, 2 1, 2 3 2, 3 4 GP LES ASUCA LES NVIDIA CUDA LES 1. Graphics Processing Unit GP General-Purpose SIMT Single Instruction Multiple Threads 1 2 3 4 1),2) LES Large Eddy Simulation 3) ASUCA 4) LES LES
64bit SSE2 SSE2 FPU Visual C++ 64bit Inline Assembler 4 FPU SSE2 4.1 FPU Control Word FPU 16bit R R R IC RC(2) PC(2) R R PM UM OM ZM DM IM R: reserved
(Version: 2013/5/16) Intel CPU ([email protected]) 1 Intel CPU( AMD CPU) 64bit SIMD Inline Assemler Windows Visual C++ Linux gcc 2 FPU SSE2 Intel CPU double 8087 FPU (floating point number processing unit)
Microsoft PowerPoint - suda.pptx
GPU の HWアーキテクチャと高性能化手法 須田礼仁 ( 東京大学 ) 2011/03/22 GPU 高性能プログラミング GPU のハードウェアを理解する CUDA のソフトウェアを理解する CUDA でプログラムを書くのは難しくないが, CUDA で高速なプログラムを書くのは難しい どうすれば遅くなるかを理解する! 効果が大きいものから順に説明します 1 高性能プログラミングの手順 1. 現在のコードの,
Slides: TimeGraph: GPU Scheduling for Real-Time Multi-Tasking Environments
加藤真平計算機アーキテクチャ特論 計算機アーキテクチャ特論後半第 1 回最先端アーキテクチャのトレンド 本資料は授業用です 無断で転載することを禁じます 講師加藤真平 前半の趣旨 : 並列化プログラミング for (i = 0; i < N; i++) { x[i] = a[i] + b[i]; } シングルプロセッサ マルチプロセッサ x[0]=a[0]+b[0]; x[1]=a[1]+b[1];
System Requirements for Geomagic
GEOMAGIC 動作環境 32-bit 版 64-bit 版 OS CPU RAM ハードディスクディスプレイ GPU - Windows XP (32-bitまたは64-bit SP2 以上 ) - Windows XP (64-bit SP2 以上 ) - Windows Vista (32-bitまたは64-bit SP1 - Windows Vista (64-bit SP1 以上 ) 以上
! 行行 CPUDSP PPESPECell/B.E. CPUGPU 行行 SIMD [SSE, AltiVec] 用 HPC CPUDSP PPESPE (Cell/B.E.) SPE CPUGPU GPU CPU DSP DSP PPE SPE SPE CPU DSP SPE 2
![! 行行 CPUDSP PPESPECell/B.E. CPUGPU 行行 SIMD [SSE, AltiVec] 用 HPC CPUDSP PPESPE (Cell/B.E.) SPE CPUGPU GPU CPU DSP DSP PPE SPE SPE CPU DSP SPE 2](/thumbs/88/115445009.jpg "! 行行 CPUDSP PPESPECell/B.E. CPUGPU 行行 SIMD [SSE, AltiVec] 用 HPC CPUDSP PPESPE (Cell/B.E.) SPE CPUGPU GPU CPU DSP DSP PPE SPE SPE CPU DSP SPE 2")
! OpenCL [Open Computing Language] 言 [OpenCL C 言 ] CPU, GPU, Cell/B.E.,DSP 言 行行 [OpenCL Runtime] OpenCL C 言 API Khronos OpenCL Working Group AMD Broadcom Blizzard Apple ARM Codeplay Electronic Arts Freescale
GPGPUクラスタの性能評価
2008 年度理研 HPC シンポジウム第 3 世代 PC クラスタ GPGPU クラスタの性能評価 2009 年 3 月 12 日 富士通研究所成瀬彰 発表の概要 背景 GPGPU による高速化 CUDA の概要 GPU のメモリアクセス特性調査 姫野 BMT の高速化 GPGPU クラスタによる高速化 GPU Host 間のデータ転送 GPU-to-GPU の通信性能 GPGPU クラスタ上での姫野
Slide 1
CUDA プログラミングの基本 パート II - カーネル CUDA の基本の概要 パート I CUDAのソフトウェアスタックとコンパイル GPUのメモリ管理 パート II カーネルの起動 GPUコードの具体像 注 : 取り上げているのは基本事項のみです そのほか多数の API 関数についてはプログラミングガイドを ご覧ください GPU 上でのコードの実行 カーネルは C 関数 + 多少の制約 ホストメモリはアクセスできない戻り値型は
211 年ハイパフォーマンスコンピューティングと計算科学シンポジウム Computing Symposium 211 HPCS /1/18 a a 1 a 2 a 3 a a GPU Graphics Processing Unit GPU CPU GPU GPGPU G
211 年ハイパフォーマンスコンピューティングと計算科学シンポジウム Computing Symposium 211 HPCS211 211/1/18 GPU 4 8 BLAS 4 8 BLAS Basic Linear Algebra Subprograms GPU Graphics Processing Unit 4 8 double 2 4 double-double DD 4 4 8 quad-double
hotspot の特定と最適化
1 1? 1 1 2 1. hotspot : hotspot hotspot Parallel Amplifier 1? 2. hotspot : (1 ) Parallel Composer 1 Microsoft* Ticker Tape Smoke 1.0 PiSolver 66 / 64 / 2.76 ** 84 / 27% ** 75 / 17% ** 1.46 89% Microsoft*
Microsoft PowerPoint - GPGPU実践基礎工学（web）.pptx
GPU のメモリ階層 長岡技術科学大学電気電子情報工学専攻出川智啓 今回の内容 GPU のメモリ階層 グローバルメモリ 共有メモリ モザイク処理への適用 コンスタントメモリ 空間フィルタへの適用 577 GPU の主要部品 基盤 GPU( チップ )+ 冷却部品 画面出力端子 電源入力端子 メモリ 特性の把握が重要 電源入力端子 画面出力端子 メモリ チップ PCI Ex 端子 http://www.geforce.com/whats
GPU チュートリアル :OpenACC 篇 Himeno benchmark を例題として 高エネルギー加速器研究機構 (KEK) 松古栄夫 (Hideo Matsufuru) 1 December 2018 HPC-Phys 理化学研究所 共通コードプロジェクト
GPU チュートリアル :OpenACC 篇 Himeno benchmark を例題として 高エネルギー加速器研究機構 (KEK) 松古栄夫 (Hideo Matsufuru) 1 December 2018 HPC-Phys 勉強会 @ 理化学研究所 共通コードプロジェクト Contents Hands On 環境について Introduction to GPU computing Introduction
NVIDIA Tesla K20/K20X GPU アクセラレータ アプリケーション パフォーマンス テクニカル ブリーフ
NVIDIA Tesla K20/K20X GPU アクセラレータ アプリケーション パフォーマンス テクニカル ブリーフ K20 GPU2 個に対するスピードアップ NVIDIA は Fermi アーキテクチャ GPU の発表により パフォーマンス エネルギー効率の両面で飛躍的な性能向上を実現し ハイパフォーマンスコンピューティング (HPC) の世界に変革をもたらしました また 実際に GPU
GPUコンピューティング講習会パート１
GPU コンピューティング (CUDA) 講習会 GPU と GPU を用いた計算の概要 丸山直也 スケジュール 13:20-13:50 GPU を用いた計算の概要 担当丸山 13:50-14:30 GPU コンピューティングによる HPC アプリケーションの高速化の事例紹介 担当青木 14:30-14:40 休憩 14:40-17:00 CUDA プログラミングの基礎 担当丸山 TSUBAME の
2.1 インテル マイクロアーキテクチャー Haswell インテル マイクロアーキテクチャー Haswell は インテル マイクロアーキテクチャー Sandy Bridge とインテル マイクロアーキテクチャー Ivy Bridge の成功を受けて開発された この新しいマイクロアーキテクチャーの
2 章インテル 64 プロセッサー アーキテクチャーと IA-32 プロセッサー アーキテクチャー 本章では 最新世代のインテル 64 プロセッサーと IA-32 プロセッサー ( インテル マイクロアーキテクチャー Haswell インテル マイクロアーキテクチャー Ivy Bridge インテル マイクロアーキテクチャー Sandy Bridge ベースのプロセッサーと インテル Core マイクロアーキテクチャー
概要 目的 CUDA Fortran の利用に関する基本的なノウハウを提供する 本チュートリアル受講後は Web 上で公開されている資料等を参照しながら独力で CUDA Fortran が利用できることが目標 対象 CUDA Fortran の利用に興味を抱いている方 前提とする知識 Fortran
CUDA Fortran チュートリアル 2010 年 9 月 29 日 NEC 概要 目的 CUDA Fortran の利用に関する基本的なノウハウを提供する 本チュートリアル受講後は Web 上で公開されている資料等を参照しながら独力で CUDA Fortran が利用できることが目標 対象 CUDA Fortran の利用に興味を抱いている方 前提とする知識 Fortran を用いた Linux
WebGL OpenGL GLSL Kageyama (Kobe Univ.) Visualization / 57
WebGL 2014.04.15 X021 2014 3 1F Kageyama (Kobe Univ.) Visualization 2014.04.15 1 / 57 WebGL OpenGL GLSL Kageyama (Kobe Univ.) Visualization 2014.04.15 2 / 57 WebGL Kageyama (Kobe Univ.) Visualization 2014.04.15
スパコンに通じる並列プログラミングの基礎
2016.06.06 2016.06.06 1 / 60 2016.06.06 2 / 60 Windows, Mac Unix 0444-J 2016.06.06 3 / 60 Part I Unix GUI CUI: Unix, Windows, Mac OS Part II 0444-J 2016.06.06 4 / 60 ( : ) 6 6 ( ) 6 10 6 16 SX-ACE 6 17
( CUDA CUDA CUDA CUDA ( NVIDIA CUDA I
GPGPU (II) GPGPU CUDA 1 GPGPU CUDA(CUDA Unified Device Architecture) CUDA NVIDIA GPU *1 C/C++ (nvcc) CUDA NVIDIA GPU GPU CUDA CUDA 1 CUDA CUDA 2 CUDA NVIDIA GPU PC Windows Linux MaxOSX CUDA GPU CUDA NVIDIA
スパコンに通じる並列プログラミングの基礎
2018.06.04 2018.06.04 1 / 62 2018.06.04 2 / 62 Windows, Mac Unix 0444-J 2018.06.04 3 / 62 Part I Unix GUI CUI: Unix, Windows, Mac OS Part II 2018.06.04 4 / 62 0444-J ( : ) 6 4 ( ) 6 5 * 6 19 SX-ACE * 6
1 4 1.1........................................... 4 1.2.................................. 4 1.3................................... 4 2 5 2.1 GPU.....
CPU GPU N Q07-065 2011 2 17 1 1 4 1.1........................................... 4 1.2.................................. 4 1.3................................... 4 2 5 2.1 GPU...........................................
IPSJ SIG Technical Report Vol.2013-ARC-203 No /2/1 SMYLE OpenCL (NEDO) IT FPGA SMYLEref SMYLE OpenCL SMYLE OpenCL FPGA 1
SMYLE OpenCL 128 1 1 1 1 1 2 2 3 3 3 (NEDO) IT FPGA SMYLEref SMYLE OpenCL SMYLE OpenCL FPGA 128 SMYLEref SMYLE OpenCL SMYLE OpenCL Implementation and Evaluations on 128 Cores Takuji Hieda 1 Noriko Etani
VOLTA ARCHITECTURE DEEP DIVE 成瀬彰, シニアデベロッパーテクノロジーエンジニア, 2017/12/12
VOLTA ARCHITECTURE DEEP DIVE 成瀬彰, シニアデベロッパーテクノロジーエンジニア, 2017/12/12 TESLA V100 の概要 Volta Architecture Improved NVLink & HBM2 Volta MPS Improved SIMT Model Tensor Core Most Productive GPU Efficient Bandwidth
GPUを用いたN体計算
単精度 190Tflops GPU クラスタ ( 長崎大 ) の紹介 長崎大学工学部超高速メニーコアコンピューティングセンターテニュアトラック助教濱田剛 1 概要 GPU (Graphics Processing Unit) について簡単に説明します. GPU クラスタが得意とする応用問題を議論し 長崎大学での GPU クラスタによる 取組方針 N 体計算の高速化に関する研究内容 を紹介します. まとめ
main.dvi
PC 1 1 [1][2] [3][4] ( ) GPU(Graphics Processing Unit) GPU PC GPU PC ( 2 GPU ) GPU Harris Corner Detector[5] CPU ( ) ( ) CPU GPU 2 3 GPU 4 5 6 7 1 [email protected] 45 2 ( ) CPU ( ) ( ) () 2.1
スパコンに通じる並列プログラミングの基礎
2018.09.10 [email protected] ( ) 2018.09.10 1 / 59 [email protected] ( ) 2018.09.10 2 / 59 Windows, Mac Unix 0444-J [email protected] ( ) 2018.09.10 3 / 59 Part I Unix GUI CUI:
1 3DCG [2] 3DCG CG 3DCG [3] 3DCG 3 3 API 2 3DCG 3 (1) Saito [4] (a) 1920x1080 (b) 1280x720 (c) 640x360 (d) 320x G-Buffer Decaudin[5] G-Buffer D
![1 3DCG [2] 3DCG CG 3DCG [3] 3DCG 3 3 API 2 3DCG 3 (1) Saito [4] (a) 1920x1080 (b) 1280x720 (c) 640x360 (d) 320x G-Buffer Decaudin[5] G-Buffer D](/thumbs/91/105248386.jpg "1 3DCG [2] 3DCG CG 3DCG [3] 3DCG 3 3 API 2 3DCG 3 (1) Saito [4] (a) 1920x1080 (b) 1280x720 (c) 640x360 (d) 320x G-Buffer Decaudin[5] G-Buffer D")
3DCG 1) ( ) 2) 2) 1) 2) Real-Time Line Drawing Using Image Processing and Deforming Process Together in 3DCG Takeshi Okuya 1) Katsuaki Tanaka 2) Shigekazu Sakai 2) 1) Department of Intermedia Art and Science,
Insert your Title here
マルチコア マルチスレッド環境での静的解析ツールの応用 米 GrammaTech 社 CodeSonar によるスレッド間のデータ競合の検出 2013 GrammaTech, Inc. All rights reserved Agenda 並列実行に起因する不具合の摘出 なぜ 並列実行されるプログラミングは難しいのか データの競合 デッドロック どのようにして静的解析ツールで並列実行の問題を見つけるのか?
PowerPoint Presentation
VME Embedded System ユーザーズマニュアル ~ Slim VME Embedded ~ Tecstar Page: 1 Agenda 1. VME Embedded System 概要 2. VME Embedded の特徴 3. Embedded Overview 4. VMEファイルとHEXファイルについて 5. Slim VME について 6. Deployment Toolの起動方法について
Nios II カスタム・インストラクションによるキャスト（型変換）の高速化
ver. 9.1 2009 年 12 月 1. はじめに Nios II にオプションで実装できる浮動小数演算カスタム インストラクションは 浮動小数四則演算はサポートしているものの 整数から浮動小数にキャスト ( 型変換 ) する機能やその逆の機能は備えていません この資料では 単精度浮動小数型と整数型の変換を簡単に Nios II のカスタム インストラクションに実装する方法を紹介しています なお
Microsoft PowerPoint - NxLec-2010-11-01.ppt
2010 年 後 学 期 レポート 問 題 計 算 機 アーキテクチャ 第 二 (O) 4. シングルサイクルプロセッサの 実 装 とパイプライン 処 理 大 学 院 情 報 理 工 学 研 究 科 計 算 工 学 専 攻 吉 瀬 謙 二 kise _at_ cs.titech.ac.jp S321 講 義 室 月 曜 日 5,6 時 限 13:20-14:50 1 1. 1から100までの 加 算
PowerPoint Presentation
ヘテロジニアスな環境におけるソフトウェア開発 Agenda 今日の概要 ヘテロジニアスな環境の登場 ホモジニアスからヘテロジニアスへ ヘテロジニアスなアーキテクチャ GPU CUDA OpenACC, XeonPhi 自分のプログラムを理解するために デバッガ 共通の操作体験 TotalView 続きはブースで より速く ホモジーニアスな並列 HPC 銀河生成 金融のリスク計算 車の衝突解析 製薬
Microsoft PowerPoint - sales2.ppt
最適化とは何? CPU アーキテクチャに沿った形で最適な性能を抽出できるようにする技法 ( 性能向上技法 ) コンパイラによるプログラム最適化 コンパイラメーカの技量 経験量に依存 最適化ツールによるプログラム最適化 KAP (Kuck & Associates, Inc. ) 人によるプログラム最適化 アーキテクチャのボトルネックを知ること 3 使用コンパイラによる性能の違い MFLOPS 90
64bit SSE2 SSE2 FPU Visual C++ 64bit Inline Assembler 4 FPU SSE2 4.1 FPU Control Word FPU 16bit R R R IC RC(2) PC(2) R R PM UM OM ZM DM IM R: reserved
(Version: 2013/7/10) Intel CPU ([email protected]) 1 Intel CPU( AMD CPU) 64bit SIMD Inline Assemler Windows Visual C++ Linux gcc 2 FPU SSE2 Intel CPU double 8087 FPU (floating point number processing unit)
修士論文
AVX を用いた倍々精度疎行列ベクトル積の高速化 菱沼利彰 1 藤井昭宏 1 田中輝雄 1 長谷川秀彦 2 1 工学院大学 2 筑波大学 1 目次 1. 研究背景 目的 2. 実装, 実験環境 3. 実験 - 倍々精度ベクトル演算 - 4. 実験 - 倍々精度疎行列ベクトル積 - 5. まとめ 多倍長精度計算フォーラム 2 目次 1. 研究背景 目的 2. 実装, 実験環境 3. 実験 - 倍々精度ベクトル演算
PowerPoint プレゼンテーション
vsmp Foundation スケーラブル SMP システム スケーラブル SMP システム 製品コンセプト 2U サイズの 8 ソケット SMP サーバ コンパクトな筐体に多くのコアとメモリを実装し SMP システムとして利用可能 スイッチなし構成でのシステム構築によりラックスペースを無駄にしない構成 将来的な拡張性を保証 8 ソケット以上への拡張も可能 2 システム構成例 ベースシステム 2U
名称未設定
Parallels Desktop 4.0 for Mac ( 4.0.3810 ) Read Me : 1. Parallels Desktop 2. 3. 4. 5. 6. Parallels Desktop 7. Parallels Desktop 4.0 for Mac 8. Parallels Desktop 9. Parallels Desktop 10. 11. 12. 1. Parallels
DRAM SRAM SDRAM (Synchronous DRAM) DDR SDRAM (Double Data Rate SDRAM) DRAM 4 C Wikipedia 1.8 SRAM DRAM DRAM SRAM DRAM SRAM (256M 1G bit) (32 64M bit)
2016.4.1 II ( ) 1 1.1 DRAM RAM DRAM DRAM SRAM RAM SRAM SRAM SRAM SRAM DRAM SRAM SRAM DRAM SRAM 1.2 (DRAM, Dynamic RAM) (SRAM, Static RAM) (RAM Random Access Memory ) DRAM 1 1 1 1 SRAM 4 1 2 DRAM 4 DRAM
strtok-count.eps
IoT FPGA 2016/12/1 IoT FPGA 200MHz 32 ASCII PCI Express FPGA OpenCL (Volvox) Volvox CPU 10 1 IoT (Internet of Things) 2020 208 [1] IoT IoT HTTP JSON ( Python Ruby) IoT IoT IoT (Hadoop [2] ) AI (Artificial
特集新世代マイクロプロセッサアーキテクチャ ( 後編 ) 3. 実例 3 ユビキタス コンピューティング時代の組み込みマイクロコンピュータ, SuperH と M32R 清水徹 * 1 長谷川淳 * 2 服部俊洋 * 3 近藤弘郁 * 4 ( 株 ) ルネサステクノロジシステムソリューション統括本部
3. 実例 3 ユビキタス コンピューティング時代の組み込みマイクロコンピュータ, SuperH と M32R 清水徹 * 1 長谷川淳 * 2 服部俊洋 * 3 近藤弘郁 * 4 ( 株 ) ルネサステクノロジシステムソリューション統括本部システムコア技術統括部 * 1 [email protected] * 2 [email protected] * 3 [email protected]
システムソリューションのご紹介
HP 2 C 製品 :VXPRO/VXSMP サーバ 製品アップデート 製品アップデート VXPRO と VXSMP での製品オプションの追加 8 ポート InfiniBand スイッチ Netlist HyperCloud メモリ VXPRO R2284 GPU サーバ 製品アップデート 8 ポート InfiniBand スイッチ IS5022 8 ポート 40G InfiniBand スイッチ
最新 Visual Studio と DirectX 9.0 Ex で戦う 3D プログラミング
次 第 1 章 はじめに 1 1.1 なぜ今 DirectX 9.0Ex なのか........................ 1 1.2 3D 数学について................................ 2 1.3 動作環境について............................... 2 1.4 DirectX の歴史................................
工学院大学建築系学科近藤研究室2000年度卒業論文梗概
耐災害性の高い通信システムにおけるサーバ計算機の性能と消費電力に関する考察 耐障害性, 消費電力, 低消費電力サーバ 山口実靖 *. はじめに 性能と表皮電力の関係について調査し, 考察を行う 災害においては, 減災活動が極めて重要である すなわち 災害が発生した後に適切に災害に対処することにより, その被害を大きく軽減できる. 適切な災害対策を行うには災害対策を行う拠点が正常に運営されていることが必要不可欠であり,
Presentation Title
コード生成製品の普及と最新の技術動向 MathWorks Japan パイロットエンジニアリング部 東達也 2014 The MathWorks, Inc. 1 MBD 概要 MATLABおよびSimulinkを使用したモデルベース デザイン ( モデルベース開発 ) 紹介ビデオ 2 MBD による制御開発フローとコード生成製品の活用 制御設計の最適化で性能改善 設計図ですぐに挙動確認 MILS: