GPU Architecture and Introduction to GPU Programming

Disclaimer: These are notes for CSE 599K "LLM Serving Systems" at the University of Washington, Spring 2025 instructed by both Prof. Baris Kasikci and TA Kan Zhu

Training vs Inference

• Training: Learning from existing data
• Inference: Applying capability to new data

Serving

ML model serving is about building a system to efficiently and scalably perform inference with:

• High throughput
• Low latency
• Compliance with diverse Service Level Objectives (SLOs)

LLM Applications and Market Context

Applications Enabled by LLMs

• AI Assistants (ChatGPT, Google Bard)
• Text-to-Image (DALLcdotE, MidJourney)
• Creative Writing (Jasper, Copy.ai)
• AI Coding Tools (GitHub Copilot, Replit)
• Text-to-Speech & Audio (Descript, Synthesia)

Key principle: Batching user requests and aiming for optimal throughput are key

Market Demand

• ChatGPT monthly visits grew from ~500M (Dec 2022) to ~2000M (Jan 2024)
• Users frequently encounter "We're experiencing exceptionally high demand" messages

Infrastructure Costs

Large-scale H100 investments in 2024:

• Meta: 300K units
• Google: 150K units
• Microsoft: 150K units
• X: 85K units

NVIDIA H200 HGX Server specs:

• Cost: ~$250,000
• High operating cost: Up to ~10,000W
• Long lead time

GPU Fundamentals

What is a GPU?

• Graphics Processing Unit
• Originally designed for accelerated graphics rendering
• Now handles scientific computing and machine learning
• Come with software stacks: CUDA (NVIDIA), ROCm (AMD)

CPU vs GPU Architecture

Example Specifications

Key difference:

• CPU DRAM: Low latency random access
• GPU HBM: Higher bandwidth, structured batch access

GPU Hardware Architecture

Data Center Context

• GPUs are deployed in server clusters
• Connected via high-speed networks (NVLink: 900 GB/s)
• Network connectivity: 200 Gb/s = 25 GB/s to data center network

GPU Memory Hierarchy

• Global Memory (HBM): 80 GB, 3TB/s bandwidth
• L2 Cache: 50MB, ~10TB/s bandwidth
• Shared Memory ("Smem"): 228 KB per SM, ~20TB/s bandwidth
• Registers: 64K imes 32 Bit per SM, ~600TB/s bandwidth

Streaming Multiprocessors (SMs)

Components:

• CUDA Cores: Scalar computation
• Tensor Cores: Matrix (dense) computation
• Shared/Constant Memory: High bandwidth temp buffer

GPU Programming Model

Hierarchy of Execution Units

Key Concepts

• 32 threads form a warp
• Threads in a warp run in parallel with:
  • Same instructions
  • Same pace
  • Different data at register level
• 4 warps run on one SM simultaneously
• Scheduler swaps warps on and off SM
• Blocks operate independently
• Block-Block communication via L2/Global memory

GPU Programming Approaches

1. PyTorch (Easiest)

import torch

def add_tensors(a, b):
    return a + b

if __name__ == "__main__":
    num_elements = 10**9

    # Create tensors on CPU
    tensor1 = torch.rand(num_elements, device='cpu')
    tensor2 = torch.rand(num_elements, device='cpu')

    # Move to GPU
    tensor1 = tensor1.to('cuda')
    tensor2 = tensor2.to('cuda')

    # Compute addition
    for i in range(10):
        result = add_tensors(tensor1, tensor2)

    # Move back to CPU
    result = result.cpu()
    print("Result of addition:", result)

2. Triton (Intermediate)

• Compiler framework from OpenAI
• Python interface with automated thread management
• Higher performance than PyTorch for complex kernels
• Operates at block level

@triton.jit
def add_kernel(x_ptr, y_ptr, output_ptr, n_elements, BLOCK_SIZE: tl.constexpr):
    pid = tl.program_id(axis=0)
    block_start = pid * BLOCK_SIZE
    offsets = block_start + tl.arange(0, BLOCK_SIZE)
    mask = offsets < n_elements

    x = tl.load(x_ptr + offsets, mask=mask)
    y = tl.load(y_ptr + offsets, mask=mask)
    output = x + y

    tl.store(output_ptr + offsets, output, mask=mask)

3. CUDA (Most Control)

• Bare-bone, one-to-one mapping to hardware
• Highest performance
• Heavy implementation burden

CUDA Memory Management

// Memory allocation
cudaMalloc          // device memory allocation  
cudaMallocHost      // pinned host memory allocation
cudaFree            // free memory

// Memory operations
cudaMemcpy          // synchronous copy
cudaMemcpyAsync     // asynchronous copy
cudaMemset          // synchronous set
cudaMemsetAsync     // asynchronous set

CUDA Kernel Structure

// Kernel declaration
__global__ void kernel_name(args...)

// Device helper function  
__device__ T helper_name(args...)

// Example addition kernel
__global__ void add(int *a, int *b, int *c, size_t num) {
    int block_start = blockIdx.x * blockDim.x;
    int thread_id = threadIdx.x;
    int index = block_start + thread_id;
    if (index < num) {
        c[index] = a[index] + b[index];
    }
}

CUDA Kernel Launch

// Define block and thread dimensions
dim3 block(x, y, z);
dim3 thread(x, y, z);

// Launch kernel
kernel_name<<<block, thread>>>(args);

CUDA Synchronization

__syncthreads()           // Thread synchronization (device function)
cudaDeviceSynchronize()   // Device synchronization (host function)

// Error checking
cudaGetLastError()        // Get last error
cudaGetErrorString()      // Get error description

Performance Analysis

Timing Considerations

• PyTorch dispatches kernels non-blocking
• CPU continues execution before GPU finishes
• Must use CUDA events for accurate GPU timing

Profiling Tools

• Torch.profiler: Good at showing CPU activity, slow processing
• Nsight Systems (nsys): High performance system-level profiling
• Nsight Compute (ncu): Tailored for intra-kernel profiling

CUDA Streams

• Multiple streams may execute in parallel
• Depends on kernels and schedulers
• Use cudaEvents to synchronize between streams
• Events act as "flags" between kernels
• cudaStreamWaitEvent for synchronization

Modern GPU Features

Advanced CUDA Features

• Unified memory address (P100+)
• NvLink (P100+)
• Clusters (H100+)
• TMA (Tensor Memory Accelerator) (H100+)
• NVSHARP (H100+)
• FP4 and FP6 precision (B100+)

Summary

Key Takeaways

1. GPU Architecture Understanding

   • Parallel processing focus
   • SMs, blocks, threads hierarchy

2. Programming Approaches

   • PyTorch: Easiest, high-level
   • Triton: Balance of control and ease
   • CUDA: Maximum control and performance

3. Performance Analysis

   • Proper timing with CUDA events
   • Profiling tools for optimization
   • Understanding memory hierarchy impact

Core Principle: Batching user requests and aiming for optimal throughput are key to effective LLM serving systems.