Introduction To GPU Computing

INTRODUCTION TO GPU COMPUTING

Add GPUs: Accelerate Science ApplicationsCPUGPU2

ACCELERATED COMPUTING IS GROWING RAPIDLY11x GPU Developers450 Applications Accelerated500Available Everywhere 730M485450400350CUDA Enabled GPUs300250615,00020015010045,0005002011 2012 2013 2014 2015 2016 201720122017 2200Universities Teaching CUDA3

SMALL CHANGES, BIG SPEED-UPApplication CodeCompute-Intensive FunctionsGPURest of SequentialCPU Code5% of Code CPU4

3 WAYS TO ACCELERATE y AccelerateApplicationsMaximumFlexibility5

3 WAYS TO ACCELERATE y AccelerateApplicationsMaximumFlexibility6

LIBRARIES: EASY, HIGH-QUALITY ACCELERATIONEASE OF USE Using libraries enables GPU acceleration without in-depthknowledge of GPU programming“DROP-IN” Many GPU-accelerated libraries follow standard APIs, thusenabling acceleration with minimal code changesQUALITY Libraries offer high-quality implementations of functionsencountered in a broad range of applicationsPERFORMANCE NVIDIA libraries are tuned by experts7

GPU ACCELERATED LIBRARIES“Drop-in” Acceleration for Your ApplicationsDEEP LEARNINGcuDNNTensorRTSIGNAL, IMAGE & VIDEODeepStream SDKLINEAR ALGEBRAcuBLAScuSPARSECUDAMath librarycuFFTNVIDIA NPPCODEC SDKPARALLEL ALGORITHMScuSOLVERnvGRAPHNCCLcuRAND8

3 STEPS TO CUDA-ACCELERATED APPLICATIONStep 1: Substitute library calls with equivalent CUDA library callssaxpy ( )cublasSaxpy ( )Step 2: Manage data locality- with CUDA:- with CUBLAS:cudaMalloc(), cudaMemcpy(), etc.cublasAlloc(), cublasSetVector(), etc.Step 3: Rebuild and link the CUDA-accelerated librarygcc myobj.o –l cublas9

DROP-IN ACCELERATION (STEP 1)int N 1 20;// Perform SAXPY on 1M elements: y[] a*x[] y[]saxpy(N, 2.0, d x, 1, d y, 1);10

DROP-IN ACCELERATION (STEP 1)int N 1 20;// Perform SAXPY on 1M elements: d y[] a*d x[] d y[]cublasSaxpy(N, 2.0, d x, 1, d y, 1);Add “cublas” prefixand use device variables11

DROP-IN ACCELERATION (STEP 2)int N 1 20;cublasInit();Initialize cuBLAS// Perform SAXPY on 1M elements: d y[] a*d x[] d y[]cublasSaxpy(N, 2.0, d x, 1, d y, 1);cublasShutdown();Shut down cuBLAS12

DROP-IN ACCELERATION (STEP 3)int N 1 20;cublasInit();cublasAlloc(N, sizeof(float), (void**)&d x);cublasAlloc(N, sizeof(float), (void*)&d y);Allocate device vectors// Perform SAXPY on 1M elements: d y[] a*d x[] d y[]cublasSaxpy(N, 2.0, d x, 1, d y, 1);cublasFree(d x);cublasFree(d y);cublasShutdown();Deallocate device vectors13

DROP-IN ACCELERATION (STEP 4)int N 1 20;cublasInit();cublasAlloc(N, sizeof(float), (void**)&d x);cublasAlloc(N, sizeof(float), (void*)&d y);cublasSetVector(N, sizeof(x[0]), x, 1, d x, 1);cublasSetVector(N, sizeof(y[0]), y, 1, d y, 1);Transfer data to GPU// Perform SAXPY on 1M elements: d y[] a*d x[] d y[]cublasSaxpy(N, 2.0, d x, 1, d y, 1);cublasGetVector(N, sizeof(y[0]), d y, 1, y, 1);Read data back GPUcublasFree(d x);cublasFree(d y);cublasShutdown();14

ACCELERATING OCTAVEScientific Programming LanguageMathematics-oriented syntaxDrop-in compatible with many MATLAB scriptsBuilt-in plotting and visualization toolsRuns on GNU/Linux, macOS, BSD, and WindowsFree SoftwareSource: http://www.gnu.org/software/octave/15

NVBLASDrop-in GPU AccelerationRoutinesTypesOperationgemmS,D,C,Z Multiplication of 2 matricessyrkS,D,C,Z Symmetric rank-k updateherkC,Zsyr2kS,D,C,Z Symmetric rank-2k pdateher2kC,ZtrsmS,D,C,Z Triangular solve, mult right-handtrmmS,D,C,Z Triangular matrix-matrix multsymmS,D,C,Z Symmetric matrix-matrix multhemmC,ZHermitian rank-k updateHemitian rank-2k updateHermitian matrix-matrix mult16

PERFORMANCE COMPARISONCPU (openblas) vs GPU (NVBLAS)Dell C4130 128 GB 36-core, E5-2697 v4 @ 2.30GHz 4x NVIDIA Tesla P100-SXM2 NVLINKSGEMM (GFLOPS)DGEMM 40060030040020020010000N 2048CPUN 4096GPU with NVBLAS libraryN 8192N 2048CPUN 4096N 8192GPU with NVBLAS library17

3 WAYS TO ACCELERATE y AccelerateApplicationsMaximumFlexibility18

OpenACC is a directivesbased programming approachto parallelcomputingdesigned for performanceand portability on CPUsAdd Simple Compiler Directivemain(){ serial code #pragma acc kernels{ parallel code }}and GPUs for HPC.19

TOP HPC APPS ADOPTING OPENACCOpenACC – Performance Portability And Ease of ProgrammingANSYS FluentANSYS Fluent R18.0 Radiation SolverVASPGaussian3 of Top 10 ORB5ORNL55ORNLCAAR CAARCodes5 CSCS CodesCodes5 CSCS CodesCPU: (Haswell EP) Intel(R) Xeon(R) CPU E5-2695 v3 @2.30GHz, 2 sockets, 28 coresGPU: Tesla K80 12 12 GB, Driver 346.4620

CFD12X speedupin 1 weekMedicalImaging10X faster kernels2X faster app40 days to2 hours3X speedupNekCEMComputationalElectromagnetics2.5X speedup60% less DAstrophysics40X speedup3X energy efficiency4X speedupSingle CPU/GPU code4.4X speedup4 weeks effort21

2 BASIC STEPS TO GET STARTEDStep 1: Annotatesource code with directives:! acc data copy(util1,util2,util3) copyin(ip,scp2,scp2i)! acc parallel loop ! acc end parallel! acc end dataStep 2: Compile& run:pgf90 -ta nvidia -Minfo accel file.f22

OpenACC DIRECTIVES EXAMPLE! acc data copy(A,Anew)iter 0do while ( err tol .and. iter iter max )Copy arrays into GPU memorywithin data regioniter iter 1err 0. fp kind! acc kernelsdo j 1,mdo i 1,nAnew(i,j) .25 fp kind *( A(i 1,j ) A(i-1,j ) & A(i ,j-1) A(i ,j 1))err max( err, Anew(i,j)-A(i,j))end doend do! acc end kernelsIF(mod(iter,100) 0 .or. iter 1)A Anewend do! acc end dataParallelize code inside regionClose off parallel regionprint *, iter, errClose off data region,copy data back23

HETEROGENEOUS ARCHITECTURESUnified MemoryGPU 0GPU 1GPU 2GPU 0MEMGPU 1MEMGPU 2MEMCPUSYS MEM24

OPENACC FOR EVERYONENew PGI Community Edition Now AvailableFREEPROGRAMMING MODELSOpenACC, CUDA Fortran, OpenMP,C/C /Fortran Compilers and ToolsPLATFORMSX86, OpenPOWER, NVIDIA GPUUPDATES1-2 times a year6-9 times a year6-9 times a yearSUPPORTUser ForumsPGI SupportPGI ite25

RESOURCESFREE CompilerSuccess storiesGuidesTutorialsVideosCoursesCode SamplesTalksBooks SpecificationTeaching MaterialsSlack&StackOverflowSuccess stories: https://www.openacc.org/success-storiesResources: https://www.openacc.org/resourcesFree Compiler: https://www.pgroup.com/products/community.htm26

CUDA PROGRAMMING LANGUAGES27

GPU PROGRAMMING LANGUAGESNumerical analyticsMATLAB, Mathematica, LabVIEW, OctaveFortranCUDA Fortran, OpenACCC, C CUDA C , OpenACCPythonCUDA Python, PyCUDA, Numba, PyCulibC#OtherAltimesh Hybridizer, Alea GPUR, Julia28

CUDA CStandard C CodeParallel C Code}globalvoid saxpy parallel(int n,float a,float *x,float *y){int i blockIdx.x*blockDim.x threadIdx.x;if (i n) y[i] a*x[i] y[i];}// Perform SAXPY on 1M elementssaxpy serial(4096*256, 2.0, x, y);// Perform SAXPY on 1M elementssaxpy parallel 4096,256 (n,2.0,x,y);void saxpy serial(int n,float a,float *x,float *y){for (int i 0; i n; i)y[i] a*x[i] y[i];http://developer.nvidia.com/cuda-toolkit29

CUDA C : DEVELOP GENERIC PARALLEL CODECUDA C features enablesophisticated and flexibleapplications and middlewareClass hierarchiesdevice methodsTemplatesOperator overloadingFunctors (function objects)Device-side new/deleteMore http://developer.nvidia.com/cuda-toolkittemplate typename T struct Functor {device Functor( a) : a( a) {}device T operator(T x) { return a*x; }T a;}template typename T, typename Oper global void kernel(T *output, int n) {Oper op(3.7);output new T[n]; // dynamic allocationint i blockIdx.x*blockDim.x threadIdx.x;if (i n)output[i] op(i); // apply functor}30

CUDA FORTRAN Program GPU using Fortran Key language for HPC Simple language extensions Kernel functions Thread / block IDs Device & datamanagement Parallel loop directives Familiar syntax Use allocate, deallocate Copy CPU-to-GPU withassignment ( )http://developer.nvidia.com/cuda-fortranmodule mymodule containsattributes(global) subroutine saxpy(n,a,x,y)real :: x(:), y(:), a,integer n, iattributes(value) :: a, ni threadIdx%x (blockIdx%x-1)*blockDim%xif (i n) y(i) a*x(i) y(i);end subroutine saxpyend module mymoduleprogram mainuse cudafor; use mymodulereal, device :: x d(2**20), y d(2**20)x d 1.0; y d 2.0call saxpy 4096,256 (2**20,3.0,x d,y d,)y y dwrite(*,*) 'max error ', maxval(abs(y-5.0))end program main31

PYTHON Numba, a just-in-timecompiler for Pythonfunctions (open-source!)import numpy as npfrom numba import vectorize Numba runs inside thestandard Pythoninterpreter@vectorize(['float32(float32, float32)'], target 'cuda')def Add(a, b):return a b Can compile for GPU orCPU!#NABC Includes PyculibInitialize arrays 100000 np.ones(N, dtype np.float32) np.ones(A.shape, dtype A.dtype) np.empty like(A, dtype A.dtype)# Add arrays on GPUC Add(A, B)NVIDIA CONFIDENTIAL. DO NOT DISTRIBUTE.32

import numpy as np from numba import vectorize @vectorize(['float32(float32, float32)'], target 'cuda') def Add(a, b): return a b # Initialize arrays N 100000 A np.ones(N, dtype np.float32) Bimport numpy as np from numba import vectorize @vectorize(['float32(float32, float32)'], target 'cuda') def Add(a, b): return a b # Initialize arrays N 100000 A np.ones(N, dtype np.float32)PYTHON - PYCULIB Python interface to CUDAlibraries: cuBLAS (dense linearalgebra), cuFFT (FastFourier Transform), andcuRAND (random numbergeneration) Code generates a millionuniformly distributedrandom numbers on theGPU using the “XORWOW”pseudorandom numbergeneratorimport numpy as npfrom pyculib import rand as curandprng curand.PRNG(rndtype curand.PRNG.XORWOW)rand np.empty(100000)prng.uniform(rand)print rand[:10]NVIDIA CONFIDENTIAL. DO NOT DISTRIBUTE.33

JULIA Up and coming scientific language Cross between Python and Matlab Interpreter (like Python) –or- compiled (like C/Fortran) New approach to multi-processing/multi-node Or use MPI Easy to combine it with other languages Works with Jupyter Notebooks!34

import numpy as np from numba import vectorize @vectorize(['float32(float32, float32)'], target 'cuda') def Add(a, b): return a b # Initialize arrays N 100000 A np.ones(N, dtype np.float32) Bimport numpy as np from numba import vectorize @vectorize(['float32(float32, float32)'], target 'cuda') def Add(a, b): return a b # Initialize arrays N 100000 A np.ones(N, dtype np.float32)JULIA – SIMPLE EXAMPLE Simple matrix multiplicationexample (integers) Double precision (Int64) Can also do elementwisemultiplication (just likeMatlab) A .* Bjulia A [1 2 ; 3 4]2x2 Array{Int64,2}:1 23 4julia B [10 11 ; 12 13]2x2 Array{Int64,2}:10 1112 13julia A * B2x2 Array{Int64,2}:34 3778 85NVIDIA CONFIDENTIAL. DO NOT DISTRIBUTE.35

JULIA – GPU EXAMPLEusing CUDAdrv, CUDAnative Options: JuliaGPU (github) Native CUDA (new) GPUarrays Simple native GPU examplefunction kernel vadd(a, b, c)# from CUDAnative: (implicit) CuDeviceArray type,#and thread/block intrinsicsi (blockIdx().x-1) * blockDim().x threadIdx().xc[i] a[i] b[i]return nothingenddev CuDevice(0)ctx CuContext(dev)# generate some datalen 512a rand(Int, len)b rand(Int, len)# allocate & upload on the GPUd a CuArray(a)d b CuArray(b)d c similar(d a)# execute and fetch results@cuda (1,len) kernel vadd(d a, d b, d c)c Array(d c)# from CUDAnative.jlusing Base.Test@test c a bdestroy(ctx)NVIDIA CONFIDENTIAL. DO NOT DISTRIBUTE.36

JULIA – GPU EXAMPLE GPUarrays example Convolutionusing GPUArrays, Colors, FileIO, ImageFilteringusing CLArraysusing GPUArrays: synchronize threadsimport GPUArrays: LocalMemoryusing CLArraysimg pg"));a CLArray(img);out similar(a);k c similar(img)####convolution!(a, out, k);Array(out)outc similar(img)copy!(outc, out)NVIDIA CONFIDENTIAL. DO NOT DISTRIBUTE.37

R Very popular Statistics language Used heavily in Machine Learning gpuR package38

import numpy as np from numba import vectorize @vectorize(['float32(float32, float32)'], target 'cuda') def Add(a, b): return a b # Initialize arrays N 100000 A np.ones(N, dtype np.float32) Bimport numpy as np from numba import vectorize @vectorize(['float32(float32, float32)'], target 'cuda') def Add(a, b): return a b # Initialize arrays N 100000 A np.ones(N, dtype np.float32)R - GPUR gpuR package Simple integer addition oftwo vectors with 1,000valuesA B gpuAgpuBseq.int(from 0, to 999)seq.int(from 1000, to 1) - gpuVector(A) - gpuVector(B)C - A BgpuC - gpuA gpuBall(C gpuC)NVIDIA CONFIDENTIAL. DO NOT DISTRIBUTE.39