PARALLEL PROGRAMMING
Course Code: BDS701 Semester: VII Academic Year- 2025-26
Module-5
Prepared By-
Bijoy Kumar Das
Assistant Professor
Department Of Computer Science and Engineering- Data Science
1
Bijoy Kumar Das
Text Books:
1. Peter S Pacheco, Matthew Malensek – An Introduction to Parallel Programming, second edition, Morgan
Kauffman.
2. Daniel Michael J Quinn – Parallel Programming in C with MPI and OpenMp, McGrawHill.
Reference Books:
1. Calvin Lin, Lawrence Snyder – Principles of Parallel Programming, Pearson
2. Barbara Chapman – Using OpenMP: Portable Shared Memory Parallel Programming, Scientific and
Engineering Computation
3. William Gropp, Ewing Lusk – Using MPI: Portable Parallel Programming, Third edition, Scientific and
Engineering Computation
Bijoy Kumar Das
Department of CSE-DS 2
GPUs and GPGPU
Evolution of GPU
What are GPUs?
GPU = Graphics Processing Unit
Designed to render detailed images for video games and animations.
Developed in late 1990s and early 2000s
Extremely powerful computational processors
GPGPU: A New Approach
Programmers recognized GPU power for general computing
Early 2000s: Applications beyond graphics emerged
Use cases: searching, sorting, data processing
GPGPU = General Purpose computing on GPUs
Bijoy Kumar Das
Department of CSE-DS 3
GPUs and GPGPU
Evolution of GPU
Bijoy Kumar Das
Department of CSE-DS 4
GPUs and GPGPU
Early Programming Challenges - The Graphics API Problem
The Major Obstacle:
One of the biggest difficulties faced by early developers of GPGPU was a fundamental limitation: the GPUs
of the time could only be programmed using computer graphics APIs (Application Programming
Interfaces).
What are Graphics APIs?
Graphics APIs are specialized programming interfaces designed specifically for graphics operations. The
two dominant graphics APIs at that time were:
•Direct3D (developed by Microsoft for Windows platforms)
•OpenGL (open-source, cross-platform graphics API)
The Problem:
To use a GPU for general computing, programmers had to:
[Link] their algorithms to work with graphics concepts
[Link] about problems in terms of graphics operations
[Link] computational tasks to graphics primitives
Bijoy Kumar Das
Department of CSE-DS 5
GPUs and GPGPU
The Solution - New APIs
Modern GPU Programming APIs
CUDA OpenCL
Developed by NVIDIA for NVIDIA GPUs only Portable across GPUs, FPGAs, and DSPs
CUDA vs OpenCL Comparison
CUDA vs OpenCL
Feature CUDA OpenCL
Setup Requires modest setup Requires more setup code
Optimization Optimized for NVIDIA GPUs Works on multiple platforms
Ease of Use Easier to use More complex but portable
Bijoy Kumar Das
Department of CSE-DS 6
GPU Architectures
Fig 1: Simplified block diagram of a GPU
Bijoy Kumar Das
Department of CSE-DS 7
Heterogeneous Computing
Heterogeneous computing refers to computer
systems that use more than one type of
processor with different architectures to
execute a single program or application.
GPU programming is heterogeneous
computing because programs use both:
Host processor: Conventional CPU
Device processor: GPU
These two processors have different
architectures. Fig 2: Simplified block diagram of a CPU and a GPU
Bijoy Kumar Das
Department of CSE-DS 8
Heterogeneous Computing
Types for Heterogeneous Computing
FPGAs (Field Programmable Gate Arrays)
Contain programmable logic blocks and
interconnects
Configured prior to program execution.
DSPs (Digital Signal Processors)
Contain special circuitry for
manipulating signals
Operations: compressing, filtering Fig 2: Simplified block diagram of a CPU and a GPU
signals
Designed for "real-world" analog
signals.
Bijoy Kumar Das
Department of CSE-DS 9
CUDA (Compute Unified Device Architecture)
CUDA is a software platform for writing GPGPU programs for heterogeneous systems with Nvidia
GPUs.
Current status: Nvidia now considers CUDA simply the name of an API, not an acronym.
Language support: CUDA APIs available for C, C++, Fortran, Python, Java.
CUDA Requires Special Compiler
Unlike MPI and Pthreads, CUDA is NOT just a library:
Cannot be compiled with ordinary C compiler (like gcc)
Requires nvcc (CUDA compiler)
CUDA compiler generates machine language for two different processors: host processor (CPU)
and device processor (GPU).
Bijoy Kumar Das
Department of CSE-DS 10
CUDA Hello Program Structure
Header File
Include <cuda.h> for CUDA programs
Kernel Function (Runs on GPU)
__global__ void Hello(void) {
printf("Hello from thread %d!\n", threadIdx.x);
}
Main Function (Runs on CPU)
int main(int argc, char* argv[]) {
thread_count = strtol(argv[1], NULL, 10);
Hello <<<1, thread_count>>>();
cudaDeviceSynchronize();
return 0;
}
Bijoy Kumar Das
Department of CSE-DS 11
CUDA Hello Program Structure
Header File Compiling and Running
Include <cuda.h> for CUDA programs Compile
Kernel Function (Runs on GPU) $ nvcc -o cuda_hello cuda_hello.cu
__global__ void Hello(void) { Run
printf("Hello from thread %d!\n", threadIdx.x);
} $ ./cuda_hello 10
Main Function (Runs on CPU)
int main(int argc, char* argv[]) {
thread_count = strtol(argv[1], NULL, 10);
Hello <<<1, thread_count>>>();
cudaDeviceSynchronize();
return 0;
}
Bijoy Kumar Das
Department of CSE-DS 12
CUDA
Threads, Blocks, and Grids
Basic Organization
Thread Block (Block): Collection of threads running on a single SM.
Grid: Collection of thread blocks started by a kernel.
Kernel Call Syntax
Hello <<<num_blocks, threads_per_block>>>();
Examples:
Hello <<<1, thread_count>>>() → 1 block, thread_count threads (uses 1 SM)
Hello <<<2, thread_count/2>>>() → 2 blocks, thread_count/2 threads each (uses 2 SMs)
Bijoy Kumar Das
Department of CSE-DS 13
Nvidia Compute Capabilities and Device Architectures
Compute capability is a version number (like 6.1, 7.5, or 8.0) that describes the features and limits of
your Nvidia GPU—how many threads, blocks, or certain functions it supports.
The major number (left of the dot) mainly shows the GPU generation, and the minor (right of the dot)
is for small upgrades.
Different Nvidia GPUs (Tesla, Maxwell, Pascal, etc.) have different compute capabilities.
More recent GPUs (higher numbers) can do more in parallel (more threads per block, more blocks,
bigger memory). Older GPUs (lower numbers) may not work with the newest CUDA features.
CUDA only supports GPUs with compute capability 3.0 or higher.
Bijoy Kumar Das
Department of CSE-DS 14
Nvidia Compute Capabilities and Device Architectures
Compute capability: A numbering scheme GPU architectures and names:
([Link], like 8.0 or 6.1) defining what features and
limits an Nvidia GPU has for programming with Tesla: 1.x
CUDA. Fermi: 2.x
Major versions: 1, 2, 3, 5, 6, 7, 8 (no version 4). Kepler: 3.x
Minor: 0–7, depending on major version. Maxwell: 5.x
CUDA only works with capability 3.0 or higher. Pascal: 6.x
Volta/Turing: 7.0/7.5
Limits dictated by compute capability: Ampere: 8.0
Max threads per block: 1024 for most modern
GPUs.
Max threads per streaming multiprocessor (SM):
1536 (capability 2.x), 2048 (>2.x).
Max dimension sizes (for blocks and grids): x/y up
to 1024, z up to 64.
Bijoy Kumar Das
Department of CSE-DS 15
Vector Addition
CUDA vector addition is a simple, highly parallel program that adds two arrays (vectors)
element-wise using the GPU.
It’s a very common example to show how well GPUs handle parallel, data-intensive tasks. You have
two arrays (vectors) x and y of size n, and you want to create another array z so that for each
position i,
z[i]=x[i]+y[i]
How does it work on the CUDA GPU?
The core idea is to use many threads in parallel so each thread handles just one element.
In the CUDA kernel, every thread figures out which array position it should work on using:
my_elt = blockDim.x * blockIdx.x + threadIdx.x
blockDim.x is the number of threads in each block.
blockIdx.x is the block number within the overall grid.
threadIdx.x is the thread number inside its block.
Bijoy Kumar Das
Department of CSE-DS 16
Vector Addition
Parts of the program
1. CUDA Kernel (Vec_add)
This is the special function that runs on the GPU.
Each thread computes its own position my_elt
(using block and thread indices) and checks if it’s
less than the array size n.
If valid, it does the sum:
z[my_elt] = x[my_elt] + y[my_elt];
In this way, all positions in the vectors get their sum each
thread does one addition.
2. main Function (CPU side)
Reads program arguments and sets up sizes and
pointers.
Allocates memory for vectors.
Fills x and y arrays with values (either user input
or random).
Starts the kernel on the GPU using:
Vec_add<<<blk_ct, th_per_blk>>>(x, y, z, n);
Launches blk_ct blocks, each with th_per_blk
threads.
Bijoy Kumar Das
Department
Each thread runs Vec_add for a different position. of CSE-DS 17
Kernel (in CUDA programming)
A kernel is a special function written in CUDA C/C++ that runs on the GPU rather than the CPU.
It describes the computation each GPU thread should perform.
When launched, thousands of threads can execute the kernel simultaneously, allowing for highly
parallel processing.
For vector addition, the kernel receives arrays x, y, and z, and tells each thread how to compute its
part of the result.
Each thread inside the kernel is assigned a unique global index (rank) so it knows which element to
compute.
Bijoy Kumar Das
Department of CSE-DS 18
Returning results from CUDA kernels
Why CUDA Kernels Can’t “Return” Values Like Normal Functions
CUDA kernels always have void return type:
You can’t use return to send values back to the CPU (host) like you would in regular functions.
Why not?
The GPU (device) and CPU (host) have separate memory, so pointers (addresses) on the CPU
usually make no sense on the GPU, and vice versa.
If you try to directly use a host pointer inside a kernel, it can crash or give wrong results.
Bijoy Kumar Das
Department of CSE-DS 19
Returning results from CUDA kernels
How Do You Get Results from a Kernel to the Host?
1. Unified Memory (simple way, 2. Manual (older or portable way): 3. Global Managed Variables (when
modern GPUs only): Allocate space on the device supported):
Use cudaMallocManaged() to (cudaMalloc) AND on the host Use __managed__ int sum; so both
allocate memory that both CPU (malloc). GPU and CPU can see the same
and GPU can access. Kernel writes to device memory. global variable.
After kernel finishes, copy result
Compute on the GPU, then just with cudaMemcpy from device The variable must be synchronized
read on the CPU, and values will back to host. using cudaDeviceSynchronize()
sync automatically. before the host can read updated
Example: values.
Example: Allocate: cudaMalloc(&dsum_p,
Allocate: sizeof(int)); and hsum_p =
cudaMallocManaged(&sum_p, malloc(sizeof(int));
sizeof(int)); Compute on device:
Kernel computes: *sum_p = x + y; Add<<<1,1>>>(2,3,dsum_p);
CPU reads: printf("%d", *sum_p); Copy result to host:
cudaMemcpy(hsum_p,
dsum_p, ...)
Bijoy Kumar Das
Department of CSE-DS 20
CUDA trapezoidal rule I
What is the Trapezoidal Rule?
It's a method used for numerical integration—estimating the area under a curve (integral) by dividing
it into trapezoids.
For a function f(x) over interval [a,b], you split the interval into n small sub-intervals, calculate the
area of each trapezoid, and sum them for the total area.
Bijoy Kumar Das
Department of CSE-DS 21
CUDA trapezoidal rule I
Example CUDA Kernel
__global__ void TrapezoidalKernel(float a, float
b, int n, float *result) {
int i = blockDim.x * blockIdx.x + threadIdx.x;
float h = (b - a) / n;
float local_sum = 0.0;
// Each thread computes its part
if (i < n) {
float x0 = a + i * h;
float x1 = a + (i + 1) * h;
local_sum = (h / 2) * (f(x0) + f(x1));
}
// Typically, sum up all local_sum values using
a reduction
// (details depend on implementation)
}
Bijoy Kumar Das
Department of CSE-DS 22
THANK YOU
Bijoy Kumar Das
Department of CSE-DS 23