GPU, GPGPU, GPU ARCHITECTURE
1. GPUs and GPGPUs
• • GPU (Graphics Processing Unit): Originally designed for graphics
processing, GPUs excel at rendering images by performing the same simple
operations on vast amounts of data simultaneously.
• • GPGPU (General-Purpose Computing on Graphics Processing
Units): This term refers to using the GPU's highly parallel architecture, which was
initially meant for graphics, to solve general scienti c and engineering problems.
• • Model: The GPU operates based on a SIMD (Single-Instruction,
Multiple-Data) model, where a single control unit executes the same instruction on
different data streams using many processing units. This contrasts with the MIMD
(Multiple-Instruction, Multiple-Data) model used by CPUs and typical clusters.
2. GPU Architecture
The GPU architecture is characterized by its massive number of simple processing cores,
designed for high throughput rather than low latency:
• • High Thread Count: GPUs deploy a massive number of threads
simultaneously (often thousands) to hide memory latency. If one thread stalls waiting
for data, the processor switches instantly to another thread.
• • Grid and Blocks: Work is typically organized into a hierarchy:
◦ ◦ A Grid is the overall work assignment structure.
◦ ◦ A Grid is composed of multiple Thread Blocks.
• • Streaming Multiprocessors (SMs): A GPU is composed of several
SMs, which are the main processing units.
◦ ◦ Each SM executes the threads belonging to its assigned thread
blocks.
◦ ◦ Threads within a block cooperate and share a fast, local shared
memory.
• • Memory Hierarchy:
◦ ◦ Global Memory: Large, high-latency (slow) device memory
accessible by all threads.
◦ ◦ Shared Memory: Very fast, low-latency memory that is
private to a single thread block. This is often used for data reuse and block-
level communication.
This architecture makes the GPU highly suitable for applications where the same operation
needs to be applied to millions of data elements independently (data parallelism).
HETEROGENOUS COPMUTING, THREADS, BLOCKS, GRIDS
fi
HETEROGENOUS COPMUTING, THREADS, BLOCKS, GRIDS
Heterogeneous Computing
Heterogeneous computing refers to parallel systems that use multiple types of processors
or cores, each specialized for different tasks, within a single machine.
• • De nition: These systems combine traditional CPUs (hosts), which
are good at sequential control and complex tasks, with specialized devices like GPUs
(devices), which excel at massive, data-parallel computations.
• • Need: Standard CPUs have hit a limit on increasing clock speed (the
power wall), necessitating the use of specialized accelerators to achieve modern
performance demands.
• • Example: A standard desktop computer or a large server where the
main program runs on the CPU, but performance-intensive numerical sections are
of oaded to the GPU.
CUDA Execution Model (Threads, Blocks, Grids)
The CUDA model (used for GPGPU programming) organizes the massive parallelism of the
GPU into a hierarchy of work units, which are mapped onto the GPU's streaming
multiprocessors (SMs).
1. Threads
• • De nition: The smallest unit of execution. Threads execute the same
program code (a kernel) but operate on different data.
• • Characteristic: GPUs use thousands of lightweight threads for high
throughput.
2. Thread Blocks
• • De nition: A group of threads (typically hundreds) that execute
together on a single Streaming Multiprocessor (SM).
• • Cooperation: Threads within the same block can cooperate by
synchronizing using barriers and communicating via fast, local shared memory.
3. Grids
• • De nition: The highest level of the execution hierarchy. A Grid is a
collection of all the Thread Blocks launched to execute a single kernel.
• • Arrangement: The Grid is usually arranged as a one-, two-, or three-
dimensional array of thread blocks.
• • Function: The Grid and its blocks de ne the total work assignment
across the entire GPU.
fl
fi
fi
fi
fi
fi
• • Function: The Grid and its blocks de ne the total work assignment
across the entire GPU.
NVIDIA
NVIDIA Compute Capabilities and Device Architecture
This topic relates to the speci c characteristics of different NVIDIA GPU architectures,
which dictates their performance and supported features for general-purpose programming
(GPGPU).
• • Compute Capability (CC): A version number assigned to a GPU
(e.g., 3.0, 5.2, 7.5). It speci es the features and instruction sets the hardware
supports, such as the amount of shared memory per block, the maximum number of
registers per thread, and other architectural limits.
• • Architecture: The design of the chip, often named after a famous
scientist (e.g., Fermi, Kepler, Maxwell, Pascal, Volta). The architecture determines
the basic hardware layout and the organization of the Streaming Multiprocessors
(SMs).
• • Signi cance: Programmers must consider the Compute Capability to
write code that is optimized for, and compatible with, the target GPU. Code compiled
for a lower CC may run on a higher CC card, but code using features from a high CC
will not run on a lower CC card.
1. Vector Addition
• • De nition: A fundamental, highly parallel operation where two
vectors, A and B, are added element-wise to produce a result vector C (C[i]=A[i]
+B[i]).
• • Parallelism: It is an embarrassingly parallel task. Since the
operation on one pair of elements (A[i], B[i]) is completely independent of all other
pairs, the work can be perfectly distributed across thousands of GPU threads.
• • CUDA Implementation: Typically involves writing a kernel where
each thread calculates exactly one element of the result vector C.
2. Returning Results from CUDA Kernels
• • Problem: CUDA kernels (functions that run on the GPU) cannot
directly return values in the way a standard C function does.
• • Solution: Results must be returned via global memory (device
memory).
fi
fi
fi
fi
fi
• • Mechanism:
1 1 The host (CPU) allocates space in device memory for the result
data (e.g., a buffer to hold the partial sums).
2 2 The kernel threads write their results directly into this pre-
allocated device memory buffer.
3 3 After the kernel completes, the host uses a function
(cudaMemcpy in the CUDA API) to copy the result buffer from device
memory back to host memory.
3. CUDA Trapezoidal Rule I: Initial Implementation
• • Approach: A straightforward parallel decomposition of the
trapezoidal rule calculation for the GPU.
• • Decomposition: The total work is divided so that each thread
calculates the area of one or more individual trapezoids.
• • Communication/Gather: Threads store their partial sums into the
global memory result buffer. A serial reduction (usually performed by the host CPU
after the data is copied back) is still needed to sum these partial results into the nal
answer.
4. CUDA Trapezoidal Rule II: Improving Performance
• • Focus: Improving the performance of the reduction step by avoiding
excessive reliance on slow global memory.
• • Optimization: A parallel reduction is performed on the GPU itself,
often within each thread block.
• • Mechanism: Threads within a block cooperatively sum their partial
results using the fast, local shared memory. This signi cantly reduces the number of
times threads must access the slow global memory.
• • Synchronization: Barriers (__syncthreads()) are used to
ensure all threads in a block nish their computation step before proceeding to the
next step of the reduction.
5. CUDA Trapezoidal Rule III: Blocks with More Than One Warp
• • Warp: A warp is the basic unit of execution and scheduling on an
SM, typically 32 threads. All threads in a warp execute the same instruction in lock-
step.
• • Context: This scenario deals with blocks that are large enough to
contain multiple warps (e.g., a block of 256 threads, which is eight warps).
fi
fi
fi
• • Context: This scenario deals with blocks that are large enough to
contain multiple warps (e.g., a block of 256 threads, which is eight warps).
• • Challenge: Implementing the parallel reduction across the multiple
warps in the block requires careful synchronization and code structure to maintain
ef ciency while ensuring correctness across all threads in the large block size. The
performance is highly sensitive to ef cient management of the reduction phase across
warps.
fi
fi