Big Data Analytics Big Data Analytics A. Parallel Computing / 3. Graphical Processing Units (GPUs) Lars Schmidt-Thieme Information Systems and Machine Learning Lab (ISMLL) Institute for Computer Science University of Hildesheim, Germany Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 1 / 32
Big Data Analytics Syllabus Tue. 9.4. (1) 0. Introduction A. Parallel Computing Tue. 16.4. (2) A.1 Threads Tue. 23.4. (3) A.2 Message Passing Interface (MPI) Tue. 30.4. (4) A.3 Graphical Processing Units (GPUs) B. Distributed Storage Tue. 7.5. (5) B.1 Distributed File Systems Tue. 14.5. (6) B.2 Partioning of Relational Databases Tue. 21.5. (7) B.3 NoSQL Databases C. Distributed Computing Environments Tue. 28.5. (8) C.1 Map-Reduce Tue. 4.6. — — Pentecoste Break — Tue. 11.6. (9) C.2 Resilient Distributed Datasets (Spark) Tue. 18.6. (10) C.3 Computational Graphs (TensorFlow) D. Distributed Machine Learning Algorithms Tue. 25.6. (11) D.1 Distributed Stochastic Gradient Descent Tue. 2.7. (12) D.2 Distributed Matrix Factorization Tue. 9.7. (13) Questions and Answers Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 1 / 32
Big Data Analytics Outline 1. GPUs vs CPUs 2. Basics of GPU Programming 3. Example: Color to Grayscale 4. Example: Matrix Multiplication 5. Block Shared Memory Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 1 / 32
Big Data Analytics 1. GPUs vs CPUs Outline 1. GPUs vs CPUs 2. Basics of GPU Programming 3. Example: Color to Grayscale 4. Example: Matrix Multiplication 5. Block Shared Memory Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 1 / 32
Big Data Analytics 1. GPUs vs CPUs Massively Parallel Computation I mildly parallel computation: I tens of cores of a CPU I massively parallel computation: I thousands of cores I examples: I compute cluster I computing / data center I grid computing I graphical processing units (GPUs) Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 1 / 32
Big Data Analytics 1. GPUs vs CPUs Flynn’s Taxonomy of Computer Architectures instructions single multiple single SISD: MISD: I old single-core I unusual CPUs data multiple SIMD: MIMD: I GPUs I multi-core CPUs I SIMD operations in I multiple CPUs ordinary CPUs I Terminology goes back to Flynn [1972]. Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 2 / 32
Big Data Analytics 1. GPUs vs CPUs Streaming Multiprocessors (SMs) I multiple cores executing the same instruction in parallel on different data I instruction level parallelism I shared resources: I shared memory I much simpler than a full-fledged CPU: I slower clockrate I less cache I no branch prediction, no speculative execution Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 3 / 32
Big Data Analytics 1. GPUs vs CPUs NVIDIA GPU Generations / Streaming Multiprocessors architecture release cores tensor- shared mem. comp. /SM cores/SM /SM [kB] cap. Turing 2018 64 8 64–96 Volta 2017 64 8 64–96 7.0 Pascal 2016 64 — 64–96 6.x Maxwell 2014 128 — 96 5.x Kepler 2012 192 — 16–48 3.x Fermi 2010 32 — 64 2.x Tesla 2006 8 — 1.x “compute capability” describes available features for the GPU. see http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#compute-capabilities Note: Some pascal GPUs have 128 cores/SM. Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 4 / 32
Big Data Analytics 1. GPUs vs CPUs NVIDIA GPUs name generation GPUs SMs cores/SM cores comp. cap. Tesla V100 Volta 1 80 64 5120 7.0 Tesla P100 Pascal 1 56 64 3584 6.0 Tesla M60 Maxwell 1 24 128 3072 5.2 Tesla K80 Kepler 2 13 192 4992 3.7 Quadro GV100 Volta 1 80 64 5120 7.0 Quadro P6000 Pascal 1 60 64 3840 6.1 Quadro M6000 Maxwell 1 24 128 3072 5.2 Quadro K6000 Kepler 1 15 192 2880 3.5 GeForce TITAN RTX Turing 1 72 64 4608 GeForce TITAN Xp Pascal 1 60 64 3840 6.1 GeForce TITAN X Maxwell 1 24 128 3072 6.1 GeForce TITAN Kepler 1 14 192 2688 3.5 GeForce RTX 2080 Ti Turing 1 68 64 4352 GeForce GTX 1080 Ti Pascal 1 56 64 3584 6.1 GeForce GTX 980 Ti Maxwell 1 22 128 2816 5.2 GeForce GTX 780 Ti Kepler 1 15 192 2880 3.5 Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 5 / 32
Big Data Analytics 1. GPUs vs CPUs Peak Performance name generation cores clock speed peak performance [MHz] [FLOPS] Tesla V100 Volta 5376 1200-1455 15.7 T (tensor: 125 T) Tesla P100 Pascal 3584 1328–1480 10.6 T Tesla M60 Maxwell 3072 948–1114 6.8 T Tesla K40 Kepler 2496 745– 875 4.3 T Intel Core i7-7920HQ Kaby Lake-H 4 3174–4198 0.033 T I to compute peak performance: peakPerformance = 2 · numberOfCores · clockSpeed I where does the factor 2 come from? I Modern GPUs and CPUs have fused-multiply-add instructions (FMA), where one instruction basically performs 2 operations: d := round(a · b + c) I since Intels Haswell architecture in 2013. Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 6 / 32
Big Data Analytics 2. Basics of GPU Programming Outline 1. GPUs vs CPUs 2. Basics of GPU Programming 3. Example: Color to Grayscale 4. Example: Matrix Multiplication 5. Block Shared Memory Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 7 / 32
Big Data Analytics 2. Basics of GPU Programming GPU Programming I Compute Unified Device Architecture (CUDA): I C/C++ language extension I dedicated preprocessor: nvcc I development tools, e.g., profiler, debugger I runtime API I for C, C++, Fortran I proprietary by Nvidia I PyCUDA: language binding for Python I Open Computing Language (OpenCL): I interface for parallel computing across heterogeneous hardware I including GPUs I C/C++ like language I open standard managed by Khronos Compute Working Group I Apple, IBM, AMD, Intel, Qualcomm, Nvidia I PyOpenCL: a language binding for Python Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 7 / 32
Big Data Analytics 2. Basics of GPU Programming GPU program abstraction I execute a procedure (kernel) over a cartesian product / grid of 1 to 3 integer ranges I ranges are called x, y, z. I each range starts at 0. I example: all (x,y) pixel coordinates of an image. I elements are grouped into blocks / tiles of the grid I fixed block size for each range x, y, z I elements are usually coordinates / indices of some data. I used to compute memory address. I used to make control decisions. Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 8 / 32
Big Data Analytics 2. Basics of GPU Programming GPU program abstraction I elements are loaded into GPU registers and accessible through symbolic names in kernels: I number of blocks: gridDim I block size: blockDim I block index: blockIdx I relative index in the block: threadIdx I each variable is of type dim3, having components x, y and z. I element can be computed via: elem.x :=blockDim.x · blockIdx.x + threadIdx.x elem.y :=blockDim.y · blockIdx.y + threadIdx.y elem.z :=blockDim.z · blockIdx.z + threadIdx.z I the total grid size is size.x :=blockDim.x · gridDim.x size.y :=blockDim.y · gridDim.y size.z :=blockDim.z · gridDim.z Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 9 / 32
Big Data Analytics 2. Basics of GPU Programming CPU/GPU Cooperation 1. allocate unified memory on GPU/CPU 1 cudaMallocManaged(&ptr, size); 2. start GPU computation 1 kernel <<<numberOfBlocks, sizeOfBlocks>>>(kernelParams...); 3. wait for GPU program to complete 1 cudaDeviceSynchronize(); 4. free memory 1 cudaFree(ptr ); Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 10 / 32
Big Data Analytics 2. Basics of GPU Programming Kernel Calls code: 1 kernel <<<numberOfBlocks, sizeOfBlocks>>>(kernelParams...); executes on GPU as (for one-dimensional indices): 1 gridDim.x = numberOfBlocks; blockDim.x = sizeOfBlocks; 2 for blockIdx .x = 0 ... numberOfBlocks.x−1 in parallel : 3 for threadIdx .x = 0 ... sizeOfBlocks .x−1 in parallel : 4 kernel (kernelParams ...); executes on GPU as (for three-dimensional indices): 1 gridDim = numberOfBlocks; blockDim = sizeOfBlocks; 2 for blockIdx .x = 0 ... numberOfBlocks.x−1 in parallel : 3 for blockIdx .y = 0 ... numberOfBlocks.y−1 in parallel : 4 for blockIdx . z = 0 ... numberOfBlocks.z−1 in parallel : 5 for threadIdx .x = 0 ... sizeOfBlocks .x−1 in parallel : 6 for threadIdx .y = 0 ... sizeOfBlocks .y−1 in parallel : 7 for threadIdx . z = 0 ... sizeOfBlocks . z−1 in parallel : 8 kernel (kernelParams ...); Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 11 / 32
Big Data Analytics 2. Basics of GPU Programming Example: Add Two Vectors 1 #include <iostream> 2 #include <math.h> 3 #define CEIL(x,y) (x+y−1)/y 4 5 __global__ void add(float ∗x, float ∗y) { 6 int n = blockIdx.x ∗ blockDim.x + threadIdx.x; 7 y[n] = x[n] + y[n ]; 8 } 9 10 int main(void) { 11 int N = 1<<20; 12 float ∗x, ∗y; 13 cudaMallocManaged(&x, N∗sizeof(float)); 14 cudaMallocManaged(&y, N∗sizeof(float)); 15 for ( int n = 0; n < N; n++) { 16 x[n] = 1.0f ; 17 y[n] = 2.0f ; 18 } 19 20 add<<<CEIL(N, 256), 256>>>(x, y); 21 cudaDeviceSynchronize(); 22 23 float maxError = 0.0f; 24 for ( int n = 0; n < N; n++) 25 maxError = fmax(maxError, fabs(y[n]−3.0f)); 26 std :: cout << "Max error: " << maxError << std::endl; 27 28 cudaFree(x ); 29 cudaFree(y ); 30 return 0; 31 } Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 12 / 32
Big Data Analytics 2. Basics of GPU Programming Compiling CUDA Code I code in file ex-add.cu. I code is run through a preprocessor nvcc 1 nvcc −o ex−add ex−add.cu I nvcc separates host code and device code (e.g., kernels), I compiles host code by default c++ compiler, e.g., gcc. I kernel calls are substituted by code that – loads the device code into the GPU and – starts the computation. I compiles device code by device code compiler. Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 13 / 32
Big Data Analytics 2. Basics of GPU Programming If Overall Range is not a Multiple of the Blocksize I then the range has to be extended to the next multiple. I use the ceiling of the rangesize/blocksize ratio. I If the range is extended, the kernel needs to be guarded not to access out-of-index locations. I the correct sizes have to be passed as parameters. Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany 14 / 32
Big Data Analytics 2. Basics of GPU Programming Example: Add Two Vectors (Guarded) 1 #include <iostream> 2 #include <math.h> 3 #define CEIL(x,y) (x+y−1)/y 4 5 __global__ void add(float ∗x, float ∗y, int N) { 6 int n = blockIdx.x ∗ blockDim.x + threadIdx.x; 7 if (n < N) 8 y[n] = x[n] + y[n ]; 9 } 10 11 int main(void) { 12 int N = 1<<20; 13 float ∗x, ∗y; 14 cudaMallocManaged(&x, N∗sizeof(float)); 15 cudaMallocManaged(&y, N∗sizeof(float)); 16 for ( int n = 0; n < N; n++) { 17 x[n] = 1.0f ; 18 y[n] = 2.0f ; 19 } 20 21 add<<<CEIL(N, 198), 198>>>(x, y, N); 22 cudaDeviceSynchronize(); 23 24 float maxError = 0.0f; 25 for ( int n = 0; n < N; n++) 26 maxError = fmax(maxError, fabs(y[n]−3.0f)); 27 std :: cout << "Max error: " << maxError << std::endl; 28 29 cudaFree(x ); 30 cudaFree(y ); 31 return 0; 32 Lars Schmidt-Thieme, Information Systems and Machine Learning Lab (ISMLL), University of Hildesheim, Germany} 15 / 32
Hiển thị 20 trang đầu. Tải tài liệu để xem đầy đủ.
Chưa có bình luận nào. Hãy là người đầu tiên!
Bình luận (0)
Chưa có bình luận nào. Hãy là người đầu tiên!