CUDA Course 2010 at MSU

  • Published on
    13-Dec-2014

  • View
    1.895

  • Download
    0

Embed Size (px)

DESCRIPTION

CUDA 2010 , .16.02.2010

Transcript

<ul><li> 1. - <ul><li>: </li></ul><ul><li><ul><li> .. ( ) </li></ul></li></ul><ul><li><ul><li> . ( NVidia ) </li></ul></li></ul></li> <li> 2. <ul><li> CPU : </li></ul><ul><li><ul><li>2004 . - Pentium 4, 3.46 GHz </li></ul></li></ul><ul><li><ul><li>2005 . - Pentium 4, 3.8 GHz </li></ul></li></ul><ul><li><ul><li>2006 . - Core Duo T2700, 2333 MHz </li></ul></li></ul><ul><li><ul><li>2007 . - Core 2 Duo E6700, 2.66 GHz </li></ul></li></ul><ul><li><ul><li>2007 . - Core 2 Duo E6800, 3 GHz </li></ul></li></ul><ul><li><ul><li>2008 . - Core 2 Duo E8600, 3.33 Ghz </li></ul></li></ul><ul><li><ul><li>2009 . - Core i7 950, 3.06 GHz </li></ul></li></ul></li> <li> 3. <ul><li> , </li></ul><ul><li><ul><li> ~ </li></ul></li></ul></li> <li> 4. <ul><li> , . </li></ul><ul><li> CPU </li></ul><ul><li><ul><li>Multithreading </li></ul></li></ul><ul><li><ul><li>SSE </li></ul></li></ul></li> <li> 5. Intel Core 2 Duo <ul><li>32 L1 </li></ul><ul><li>2/4 L2 </li></ul><ul><li> - </li></ul></li> <li> 6. Intel Core 2 Quad </li> <li> 7. Intel Core i7 </li> <li> 8. Symmetric Multiprocessor Architecture (SMP) </li> <li> 9. Symmetric Multiprocessor Architecture (SMP) <ul><li> L1 L2 </li></ul><ul><li> (, , ) </li></ul></li> <li> 10. Cell </li> <li> 11. Cell <ul><li>Dual-threaded 64-bit PowerPC </li></ul><ul><li>8 Synergistic Processing Elements (SPE) </li></ul><ul><li>256 Kb on-chip SPE </li></ul></li> <li> 12. BlueGene/L </li> <li> 13. BlueGene/L <ul><li>65536 dual-core nodes </li></ul><ul><li>node </li></ul><ul><li><ul><li>770 Mhz PowerPC </li></ul></li></ul><ul><li><ul><li>Double Hammer FPU (4 Flop/cycle) </li></ul></li></ul><ul><li><ul><li>4 Mb on-chip L3 </li></ul></li></ul><ul><li><ul><li>512 Mb off-chip RAM </li></ul></li></ul><ul><li><ul><li>6 3D- </li></ul></li></ul><ul><li><ul><li>3 collective network </li></ul></li></ul><ul><li><ul><li>4 barrier/interrupt </li></ul></li></ul></li> <li> 14. BlueGene/L </li> <li> 15. G80 </li> <li> 16. G80 </li> <li> 17. </li> <li> 18. <ul><li>CPU SISD </li></ul><ul><li><ul><li>Multithreading: ( MIMD) (SIMD) </li></ul></li></ul><ul><li><ul><li>SSE : 128 </li></ul></li></ul><ul><li><ul><li><ul><li> 4 32 (SIMD) </li></ul></li></ul></li></ul><ul><li>GPU SIMD* </li></ul></li> <li> 19. MultiThreading Hello World <ul><li>#include </li></ul><ul><li>#include </li></ul><ul><li>#include // beginthread( ) </li></ul><ul><li>void mtPrintf( void * pArg); </li></ul><ul><li>int main() </li></ul><ul><li>{ </li></ul><ul><li>int t0 = 0; int t1 = 1; </li></ul><ul><li>_beginthread(mtPrintf, 0, ( void *)&amp;t0 ); </li></ul><ul><li>mtPrintf( ( void *)&amp;t1); </li></ul><ul><li>Sleep( 100 ); </li></ul><ul><li>return 0; </li></ul><ul><li>} </li></ul><ul><li>void mtPrintf( void * pArg ) </li></ul><ul><li>{ </li></ul><ul><li>int * pIntArg = ( int *) pArg; </li></ul><ul><li>printf( "The function was passed %d ", (*pIntArg) ); </li></ul><ul><li>} </li></ul></li> <li> 20. MultiThreading Hello World <ul><li>// </li></ul><ul><li>// : </li></ul><ul><li>// entry point , </li></ul><ul><li>// , 0 OS </li></ul><ul><li>// (void *) </li></ul><ul><li>_beginthread(mtPrintf, 0, ( void *)&amp;t1 ); </li></ul><ul><li>// </li></ul><ul><li>mtPrintf( ( void *)&amp;t0); </li></ul><ul><li>// 100 </li></ul><ul><li>// windows </li></ul><ul><li>// </li></ul><ul><li>// </li></ul><ul><li>Sleep( 100 ); </li></ul></li> <li> 21. SSE Hello World <ul><li>#include </li></ul><ul><li>#include </li></ul><ul><li>struct vec4 </li></ul><ul><li>{ </li></ul><ul><li>union </li></ul><ul><li>{ </li></ul><ul><li>float v[4]; </li></ul><ul><li>__m128 v4; </li></ul><ul><li>}; </li></ul><ul><li>}; </li></ul><ul><li>int main() </li></ul><ul><li>{ </li></ul><ul><li>vec4 a = {5.0f, 2.0f, 1.0f, 3.0f}; </li></ul><ul><li>vec4 b = {5.0f, 3.0f, 9.0f, 7.0f}; </li></ul><ul><li>vec4 c; </li></ul><ul><li>c.v4 = _mm_add_ps(a.v4, b.v4); </li></ul><ul><li>printf( "c = {%.3f, %.3f, %.3f, %.3f} " , c.v[0], c.v[1], c.v[2], c.v[3]); </li></ul><ul><li>return 0; </li></ul><ul><li>} </li></ul></li> <li> 22. SIMD <ul><li> , </li></ul><ul><li> ( kernel ) </li></ul></li> <li> 23. SIMD <ul><li> , </li></ul></li> <li> 24. GPU <ul><li>Voodoo - , </li></ul><ul><li> CPU </li></ul></li> <li> 25. GPU <ul><li><ul><li> ( RivaTNT2 ) </li></ul></li></ul><ul><li><ul><li>T&amp;L </li></ul></li></ul><ul><li><ul><li> () </li></ul></li></ul><ul><li><ul><li> ( GeForceFX ) </li></ul></li></ul><ul><li><ul><li> floating point - </li></ul></li></ul></li> <li> 26. GPU: <ul><li> 4D float- </li></ul><ul><li><ul><li> vendor- </li></ul></li></ul></li> <li> 27. GPU: <ul><li> ( Cg, GLSL, HLSL ) </li></ul><ul><li> ( GeForce 6xxx ) </li></ul><ul><li> GPU , CPU 10 Flop </li></ul></li> <li> 28. CPU GPU <ul><li> , </li></ul></li> <li> 29. GPGPU <ul><li> GPU </li></ul><ul><li> GPU API (OpenGL, D3D) </li></ul><ul><li> (++) </li></ul><ul><li>, API </li></ul></li> <li> 30. CUDA (Compute Unified Device Architecture) <ul><li> - /. </li></ul><ul><li> CPU </li></ul><ul><li><ul><li>Multithreading </li></ul></li></ul><ul><li><ul><li>SSE </li></ul></li></ul><ul><li><ul><li> bottleneck </li></ul></li></ul><ul><li>CUDA - ( C) GPU </li></ul></li> <li> 31. CUDA Hello World <ul><li>#define N (1024*1024) </li></ul><ul><li>__global__ void kernel ( float * data ) </li></ul><ul><li>{ </li></ul><ul><li>int idx = blockIdx.x * blockDim.x + threadIdx.x; </li></ul><ul><li>float x = 2.0f * 3.1415926f * ( float ) idx / ( float ) N; </li></ul><ul><li>data [idx] = sinf ( sqrtf ( x ) ); </li></ul><ul><li>} </li></ul><ul><li>int main ( int argc, char * argv [] ) </li></ul><ul><li>{ </li></ul><ul><li>float a [N]; </li></ul><ul><li>float * dev = NULL; </li></ul><ul><li>cudaMalloc ( ( void **)&amp;dev, N * sizeof ( float ) ); </li></ul><ul><li>kernel ( dev ); </li></ul><ul><li>cudaMemcpy ( a, dev, N * sizeof ( float ), cudaMemcpyDeviceToHost ); </li></ul><ul><li>cudaFree ( dev ); </li></ul><ul><li>for ( int idx = 0; idx &lt; N; idx++) printf( "a[%d] = %.5f " , idx, a[idx]); </li></ul><ul><li>return 0; </li></ul><ul><li>} </li></ul></li> <li> 32. CUDA Hello World <ul><li>__global__ void kernel ( float * data ) </li></ul><ul><li>{ </li></ul><ul><li>int idx = blockIdx.x * blockDim.x + threadIdx.x; // </li></ul><ul><li>float x = 2.0f * 3.1415926f * ( float ) idx / ( float ) N; // </li></ul><ul><li>data [idx] = sinf ( sqrtf ( x ) ); // </li></ul><ul><li>} // </li></ul><ul><li> ( N) , . </li></ul><ul><li> id </li></ul></li> <li> 33. CUDA Hello World float a [N]; float * dev = NULL; // GPU N cudaMalloc ( ( void **)&amp;dev, N * sizeof ( float ) ); // N 512 // - kernel // - dev kernel ( dev ); // GPU (DRAM) // CPU (N ) cudaMemcpy ( a, dev, N * sizeof ( float ), cudaMemcpyDeviceToHost ); // GPU cudaFree ( dev ); </li> <li> 34. Tesla 10 Interconnection Network TPC TPC TPC TPC TPC TPC TPC TPC TPC TPC ROP L2 ROP L2 ROP L2 ROP L2 ROP L2 ROP L2 ROP L2 ROP L2 DRAM DRAM DRAM DRAM DRAM DRAM DRAM DRAM CPU Bridge Host Memory Work Distribution </li> <li> 35. Tesla : Tesla 8 TEX SM SM Texture Processing Cluster Streaming Multiprocessor Instruction $ Constant $ Instruction Fetch Shared Memory SFU SP SP SP SP SFU SP SP SP SP Register File </li> <li> 36. Tesla Tesla 10 TEX SM SM Texture Processing Cluster SM Streaming Multiprocessor Instruction $ Constant $ Instruction Fetch Shared Memory SFU SP SP SP SP SFU SP SP SP SP Double Precision Register File </li> <li> 37. <ul><li>: </li></ul><ul><li><ul><li>[+][-] SM TPC </li></ul></li></ul><ul><li><ul><li>[+][-] TPC </li></ul></li></ul><ul><li><ul><li>[+][-] DRAM </li></ul></li></ul><ul><li> : </li></ul><ul><li><ul><li>Tesla 8: 8800 GTX </li></ul></li></ul><ul><li><ul><li>Tesla 10: GTX 280 </li></ul></li></ul></li> <li> 38. SIMT (Single Instruction, Multiple Threads) <ul><li> SM ( threads ) </li></ul><ul><li> warp ( 32 ) SM warp </li></ul><ul><li> warp </li></ul><ul><li> warp </li></ul></li> <li> 39. Fermi <ul><li>Unified L2 cache (768 Kb) </li></ul><ul><li>Up to 1 Tb of memory (64 bit addressing) </li></ul><ul><li>Unified address space </li></ul><ul><li>ECC protection (DRAM, registers, shared, cache) </li></ul><ul><li>Simultaneous CPU-&gt;GPU, GPU-&gt;CPU, kernel execution </li></ul><ul><li>10x faster context switching, concurrent kernel execution (up to 16 kernels) </li></ul></li> <li> 40. Fermi. Instruction Cache Uniform Cache Warp Scheduler 64 Kb Shared Memory/ L1 Cache SFU Core Interconnection network Register File (32768 32-bit words Dispatch Unit Warp Scheduler Dispatch Unit Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core Core SFU SFU SFU Uniform Cache LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST LD/ST </li> <li> 41. Fermi <ul><li>32 cores per SM </li></ul><ul><li>Dual Thread schedule simultaneous execution of 2 warps </li></ul><ul><li>48 Kb shared + 16 Kb cache 16 Kb shared + 48 Kb cache </li></ul><ul><li>Cheap atomics </li></ul></li> <li> 42. <ul><li>CUDA.CS.MSU.SU </li></ul><ul><li><ul><li> ! </li></ul></li></ul><ul><li><ul><li><ul><li> - ! </li></ul></li></ul></li></ul><ul><li><ul><li><ul><li> , ! </li></ul></li></ul></li></ul><ul><li><ul><li><ul><li> CUDA ! </li></ul></li></ul></li></ul><ul><li>s teps3d. narod . ru </li></ul><ul><li>www. nvidia . ru </li></ul></li> <li> 43. <ul><li> : </li></ul><ul><li><ul><li>CUDA / MT / SSE hello world </li></ul></li></ul><ul><li><ul><li>CUDA / MT / SSE Hello World </li></ul></li></ul><ul><li><ul><li><ul><li> hello world </li></ul></li></ul></li></ul><ul><li><ul><li>SVN ? </li></ul></li></ul></li> <li> 44. <ul><li>11 </li></ul><ul><li>5 </li></ul><ul><li><ul><li> : </li></ul></li></ul><ul><li><ul><li><ul><li> CUDA </li></ul></li></ul></li></ul><ul><li>5 </li></ul></li> <li> 45. <ul><li><ul><li>5 </li></ul></li></ul><ul><li><ul><li><ul><li><ul><li> , </li></ul></li></ul></li></ul></li></ul></li> <li> 46. </li> </ul>