ML Systems Curriculum: LLMs & Multimodal Models

52 weeks · ~10 hrs/wk · ~520 hrs total Profile: Python/PyTorch practitioner, GPU beginner, full-stack goal (inference + training + kernels)

Year Overview

Dependency Map

flowchart TD
    subgraph P1["① Foundations · Wks 1–12"]
        roofline["Roofline & GPU Architecture"]
        txArith["Transformer Systems Arithmetic"]
        effAttn["Efficient Attention (Flash, GQA, MLA)"]
        inference["Quantization & Serving"]
        dist["Distributed Training (DP / TP / PP / ZeRO)"]
        kernels["Triton Kernels & torch.compile"]
    end

    subgraph P2["② Hardware Mastery · Wks 13–20"]
        advCuda["Advanced CUDA (Warps, Tensor Cores, cp.async)"]
        cutlass["CUTLASS / CuTe"]
        hopper["Hopper ISA (TMA, WGMMA, Ping-Pong)"]
        profiling["Nsight Profiling (Compute + Systems)"]
        interconnects["Interconnects & NCCL Basics"]
    end

    subgraph P3["③ Training at Scale · Wks 21–29"]
        seqPar["Sequence & Context Parallelism"]
        moeTraining["MoE Training (Expert Parallelism)"]
        dataPipe["Data Pipelines & Pretraining Infra"]
        peft["LoRA / QLoRA / PEFT"]
        rlhf["RLHF Systems (PPO, DPO, GRPO)"]
    end

    subgraph P4["④ Advanced Inference · Wks 30–37"]
        longCtx["Long-Context (YaRN, KV Eviction)"]
        advQuant["FP8 & Extreme Quantization"]
        advSpec["Advanced Spec Decoding (EAGLE, Medusa)"]
        disagg["Disaggregated Inference (DistServe)"]
        prodServing["Production Serving & SLO Scheduling"]
    end

    subgraph P5["⑤ Emerging Architectures · Wks 36–46"]
        mamba["Mamba / SSMs & Parallel Scan"]
        moeInf["MoE Inference & Expert Offload"]
        linAttn["Linear Attention & Hybrid Models"]
        diffusion["Diffusion Systems (DiT, Flow Matching)"]
    end

    subgraph P6["⑥ Compilers & Infra · Wks 43–52"]
        mlir["MLIR & Compiler IR Theory"]
        tvm["TVM / XLA / Autotuning"]
        adSys["Automatic Differentiation Systems"]
        memMgmt["CUDA Memory Management"]
        ncclDeep["NCCL Deep Dive & Collectives"]
    end

    %% Part I internal
    roofline --> txArith
    roofline --> kernels
    txArith --> effAttn
    effAttn --> inference
    effAttn --> dist

    %% P1 → P2
    roofline --> advCuda
    kernels --> cutlass
    advCuda --> cutlass
    cutlass --> hopper
    hopper --> profiling
    dist --> interconnects

    %% P1 → P3
    dist --> seqPar
    dist --> moeTraining
    txArith --> dataPipe
    inference --> peft
    peft --> rlhf
    seqPar --> rlhf

    %% P1 → P4
    effAttn --> longCtx
    inference --> advQuant
    inference --> advSpec
    inference --> disagg
    inference --> prodServing

    %% P2 → P4
    hopper --> advQuant

    %% P2 → P5
    hopper --> mamba
    kernels --> mamba

    %% P1 → P5
    effAttn --> linAttn
    txArith --> diffusion

    %% P3 → P5
    moeTraining --> moeInf

    %% P4 → P5
    disagg --> moeInf

    %% P2 → P6
    hopper --> mlir
    kernels --> mlir
    mlir --> tvm
    advCuda --> memMgmt
    interconnects --> ncclDeep
    kernels --> adSys

Part I — Foundations

Goal: Build the hardware intuition, transformer arithmetic, attention theory, inference fundamentals, distributed training basics, and kernel writing skills that underpin everything else.

Week 1 — The Roofline Model

Concepts to understand:

• Arithmetic intensity: what it is and how to compute it for an op
• The roofline model: compute ceiling vs memory bandwidth ceiling
• Ridge point: the arithmetic intensity that separates memory-bound from compute-bound
• HBM vs SRAM: capacity, bandwidth, and latency tradeoffs
• GPU SM, warp, and thread hierarchy at a high level
• How to read a PyTorch profiler trace

Reading:

• Making Deep Learning Go Brrrr — Horace He (1.5 hrs)
• All About Rooflines — JAX Scaling Book (2 hrs)
• Transformer Math 101 — EleutherAI (1.5 hrs)
• NVIDIA Hopper Architecture Whitepaper §1–3 (1.5 hrs)
• PyTorch Profiler Tutorial (1 hr)

Hands-on:

• Run PyTorch profiler on a forward pass of GPT2LMHeadModel. Identify the top 3 ops by CUDA time and classify each as compute- or memory-bound. (1.5 hrs)

Milestone: Given matmul(A, B) with A=(4096,4096), B=(4096,4096) in fp16 on an A100 (312 TFLOP/s, 2 TB/s), compute arithmetic intensity and determine its roofline regime.

Week 2 — CUDA Mental Model & Memory Hierarchy

Concepts to understand:

• Thread/block/grid hierarchy and how it maps to GPU hardware
• Shared memory (SMEM): capacity per SM, latency vs global memory
• Global memory coalescing: why access pattern matters for bandwidth
• Occupancy: how register and SMEM usage limits active warps per SM
• Bank conflicts in shared memory
• L1/L2 cache behavior

Reading:

• Programming Massively Parallel Processors Ch. 1–5 — Hwu & Kirk (4 hrs)
• ECE408/CS483 Lecture Videos Weeks 1–3 — UIUC via NVIDIA (3 hrs)
• An Even Easier Introduction to CUDA — NVIDIA Blog (1 hr)

Hands-on:

• Write a naive CUDA vector-add kernel (in C or via Numba). Measure achieved bandwidth vs theoretical peak. (2 hrs)

Milestone: Explain why a naively written matrix transpose kernel is memory-bandwidth-bound despite O(N²) reads and writes — and describe the shared memory tiling fix.

Week 3 — Transformer Systems Arithmetic

Concepts to understand:

• Parameter count formula for a transformer (embedding, attn projections, MLP, LM head)
• FLOPs ≈ 6ND for training, ≈ 2ND for inference — where this comes from
• KV cache memory: 2 · n_layers · n_heads · d_head · seqlen · batch · dtype_bytes
• Model FLOPs utilization (MFU): definition and how to measure it
• Scaling laws: compute-optimal training (Chinchilla)
• Why seqlen causes quadratic FLOPs but linear KV cache growth

Reading:

• Transformer Math 101 — EleutherAI (re-read carefully, 1.5 hrs)
• Language Models are Few-Shot Learners §2 — GPT-3 (1 hr)
• Scaling Laws for Neural Language Models — Kaplan et al. (2 hrs)
• JAX Scaling Book Ch. 2–3 (2 hrs)

Hands-on:

• For a 7B-parameter LLaMA-2 in bf16, derive: (a) weight memory, (b) KV cache at batch=32 seqlen=2048, (c) FLOPs per forward pass. (3 hrs)

Milestone: Why does doubling sequence length quadratically increase attention FLOPs but only linearly increase KV cache memory? Write out the derivation.

Week 4 — Efficient Attention

Concepts to understand:

• IO complexity of standard attention: why materializing (seqlen × seqlen) is the bottleneck
• FlashAttention tiling: splitting Q/K/V into blocks to avoid writing the full attention matrix
• Online softmax: computing softmax incrementally across tiles
• GQA/MQA: sharing K/V heads across query heads and the memory savings
• MLA: low-rank KV compression in DeepSeek-V2 and the 64× cache reduction
• FlashAttention-2 improvements over v1

Reading:

• FlashAttention — Dao et al. 2022 (2 hrs)
• FlashAttention-2 — Dao 2023 (1.5 hrs)
• GQA: Training Generalized Multi-Query Transformer Models (1.5 hrs)
• DeepSeek-V2 §3 — MLA (1 hr)
• FlashAttention-3 blog (1 hr)

Hands-on:

• Implement FlashAttention forward tiling logic in pure NumPy. Verify against torch.nn.functional.scaled_dot_product_attention. (3 hrs)

Milestone: For a 70B model, 64 heads, d_head=128, seqlen=8192, batch=1 in fp16 — compute the memory footprint of the full attention matrix under standard MHA. Then state FlashAttention’s peak SMEM usage and explain why.

Week 5 — Quantization & Speculative Decoding

Concepts to understand:

• Post-training quantization (PTQ) vs quantization-aware training (QAT)
• Why outlier activations break naive int8 quantization (LLM.int8() mixed-precision)
• Weight-only quantization (GPTQ, AWQ) vs weight+activation quantization
• AWQ: activation-aware scaling to protect salient weights
• Speculative decoding: draft model generates K tokens, target verifies in parallel
• Token acceptance rate and when speculative decoding wins/loses

Reading:

• LLM.int8() — Dettmers et al. (1.5 hrs)
• GPTQ (1.5 hrs)
• AWQ: Activation-aware Weight Quantization (1 hr)
• Fast Inference via Speculative Decoding — Leviathan et al. (1.5 hrs)
• Hugging Face quantization guide (1 hr)

Hands-on:

• Load LLaMA-3-8B in fp16, int8 (bitsandbytes), and GPTQ 4-bit. Measure decode latency, throughput, and perplexity for each. (3 hrs)

Week 6 — Serving Systems

Concepts to understand:

• Prefill vs decode phases: why prefill is compute-bound and decode is memory-bandwidth-bound
• TTFT (time-to-first-token) vs TPOT (time-per-output-token)
• Continuous batching (iteration-level scheduling)
• PagedAttention: virtual memory for KV cache, block table indirection
• KV cache fragmentation and how paging limits it to <4%
• SGLang vs vLLM tradeoffs

Reading:

• Efficient Memory Management for LLM Serving with PagedAttention — Kwon et al. (2 hrs)
• Orca: A Distributed Serving System for Transformer-Based LLMs (1.5 hrs)
• Continuous Batching — 23× LLM Throughput — Anyscale (0.5 hrs)
• vLLM docs — quickstart + architecture (1 hr)
• SGLang (1.5 hrs)

Hands-on:

• Serve LLaMA-3-8B with vLLM. Sweep batch size 1→64. Record throughput and TTFT. At what batch size does the system transition from memory-bound to compute-bound? (3 hrs)

Milestone (Weeks 5–6): For a 7B model on one A100 (80GB), walk through: (1) VRAM available for KV cache after loading int4 weights, (2) concurrent sequences that fit, (3) why continuous batching improves utilization over static batching.

Week 7 — Data Parallelism, Tensor Parallelism, Pipeline Parallelism

Concepts to understand:

• Data parallelism (DDP): gradient all-reduce, overlap with backward pass
• Tensor parallelism: column-parallel and row-parallel linear layers
• Pipeline parallelism: stage assignment, pipeline bubbles, 1F1B schedule
• Micro-batching to fill the pipeline
• 3D parallelism: how TP + PP + DP compose
• Communication primitives: all-reduce, all-gather, reduce-scatter

Reading:

• Megatron-LM — Shoeybi et al. (1.5 hrs)
• Efficient Large-Scale Language Model Training on GPU Clusters — Narayanan et al. (2.5 hrs)
• Everything About Distributed Training and Efficient Finetuning (2 hrs)
• JAX Scaling Book Ch. 4–5 (2 hrs)

Hands-on:

• Using torch.distributed, run DDP on 2 GPUs. Profile all-reduce communication time vs. total step time at varying batch sizes. (2 hrs)

Week 8 — ZeRO, FSDP, Mixed Precision, Gradient Checkpointing

Concepts to understand:

• ZeRO stages 1/2/3: what each stage shards
• ZeRO-3 / FSDP communication: all-gather before forward, reduce-scatter after backward
• Mixed precision: fp16/bf16 forward + fp32 master weights, loss scaling
• Why bf16 is preferred over fp16 for training
• Gradient checkpointing: recomputing activations vs storing them

Reading:

• ZeRO: Memory Optimizations Toward Training Trillion Parameter Models (2 hrs)
• PyTorch FSDP2 blog (1 hr)
• Mixed Precision Training (1.5 hrs)
• NVIDIA Mixed Precision Training Guide (1 hr)
• Reducing Activation Recomputation in Large Transformer Models (1 hr)

Hands-on:

• Train a 1B-parameter toy transformer on 2 GPUs. Record peak memory under: (a) DDP fp32, (b) DDP bf16, (c) FSDP ZeRO-3 bf16. (3 hrs)

Milestone (Weeks 7–8): For a 13B model on 8 GPUs with ZeRO-3: calculate per-GPU memory for parameters, gradients, and optimizer states. Estimate the all-gather + reduce-scatter communication volume per step vs. DDP baseline.

Week 9 — Triton Foundations

Concepts to understand:

• What Triton is: a Python DSL compiling to PTX, sitting between CUDA C and PyTorch ops
• tl.program_id, tl.load, tl.store: the core primitives
• Block tiling in Triton: how BLOCK_SIZE maps to SMEM usage and occupancy
• Masking: handling tensor edges when shape isn’t a multiple of BLOCK_SIZE
• tl.dot: Triton’s matmul primitive
• Autotuning: @triton.autotune decorator and config search
• Fusion: why fusing elementwise ops into a matmul saves memory bandwidth

Reading:

• Triton: An Intermediate Language and Compiler for Tiled Neural Network Computations (1.5 hrs)
• Official Triton tutorials: vector add → fused softmax → matmul (4 hrs)
• Unleashing the Power of Triton (1 hr)

Hands-on:

• Implement fused SiLU (x * sigmoid(x)) as a Triton kernel. Benchmark against PyTorch built-in and two-op naive version. Report bandwidth utilization. (3.5 hrs)

Week 10 — torch.compile & Inductor

Concepts to understand:

• torch.compile pipeline: Dynamo → Inductor → Triton kernels
• Graph capture modes: reduce-overhead vs max-autotune
• Fusion in Inductor: which op patterns get fused automatically
• Graph breaks: what causes them and how to diagnose with TORCH_COMPILE_DEBUG=1
• When torch.compile is and isn’t worth it (dynamic shapes, small models)
• How FlashAttention-3 uses Triton + Hopper TMA/WGMMA for ~75% peak FLOP/s

Reading:

• PyTorch 2 (1.5 hrs)
• torch.compile tutorial (1.5 hrs)
• FlashAttention-3 blog (1.5 hrs)
• Triton blocked matmul tutorial (2 hrs)

Hands-on:

• Apply torch.compile to a full transformer forward pass. Use TORCH_COMPILE_DEBUG=1 to inspect emitted Triton kernels. Identify which ops fused and which caused graph breaks. (3.5 hrs)

Milestone (Weeks 9–10): Write a fused layernorm kernel in Triton that computes mean, variance, and normalization in a single pass. Compare throughput and bandwidth to torch.nn.LayerNorm. Explain why a single-pass implementation is faster.

Week 11 — Vision-Language Model Serving

Concepts to understand:

• VLM anatomy: vision encoder (ViT) + projector + LLM backbone
• How image tokens are injected as “virtual” prefill tokens
• Resolution scaling: how image size determines token count
• Two serving regimes: encoder-heavy vs. LLM-decode-heavy
• Cross-attention injection (Flamingo) vs. token merge/concatenation (LLaVA)
• KV cache implications: image tokens inflate effective context length

Reading:

• CLIP (1.5 hrs)
• LLaVA: Visual Instruction Tuning (1.5 hrs)
• LLaVA-1.5 (1 hr)
• Flamingo §3 (1 hr)
• vLLM multimodal docs (1 hr)

Hands-on:

• Serve LLaVA-1.5-7B with vLLM. Measure: (a) vision encoder latency vs. LLM prefill latency, (b) how doubling image resolution changes latency and KV cache size. (4 hrs)

Milestone: A 336×336 image produces 576 image tokens. For LLaVA-1.5-7B serving 16 concurrent users (one image + 64-token question each), compute effective prefill token count and compare KV cache memory to a text-only baseline.

Week 12 — Part I Capstone

Choose one track:

• Track A — LLM serving optimization: Serve a 7B model on one GPU. Apply quantization, continuous batching, and speculative decoding sequentially. Document measured impact of each.
• Track B — Custom attention kernel: Implement simplified FlashAttention forward in Triton. Benchmark vs. PyTorch SDPA across seqlen 512→8192. Produce a roofline plot.
• Track C — Distributed training audit: Train a 1B model on 4 GPUs. Compare DDP, FSDP ZeRO-2, FSDP ZeRO-3. Identify the dominant communication bottleneck.

Deliverable checklist:

• Baseline profile (PyTorch profiler or Nsight output)
• Each optimization applied with before/after measurements
• Written explanation of why each change helped (roofline reasoning)
• One thing you’d do next with more time or hardware

Part II — Hardware Mastery

Goal: Go from Triton user to someone who can reason about hardware ISA-level behavior, write CUTLASS kernels, profile with Nsight Compute, and understand the full interconnect stack.

Week 13 — Advanced CUDA: Warp Primitives & Tensor Cores

Concepts to understand:

• Warp shuffle functions: __shfl_sync, __shfl_down_sync, __shfl_xor_sync — register-to-register exchange within a warp without SMEM
• Warp vote functions: __ballot_sync, __all_sync, __any_sync
• Warp-level reduce patterns using shuffles
• Tensor Cores via WMMA API: nvcuda::wmma — fragment types, load_matrix_sync, mma_sync, store_matrix_sync
• Supported WMMA shapes (16×16×16, 8×32×16) and data types (FP16, BF16, TF32, INT8, FP8)
• PTX mma.sync for finer control below the WMMA abstraction

Reading:

• NVIDIA Tensor Core Programming — Lei Mao’s Log Book (3 hrs)
• Programming Tensor Cores in CUDA 9 (1.5 hrs)
• Warp Shuffle and Warp Vote Instructions — CSE 599I Slides (1.5 hrs)
• GPU MODE Lectures 14–20 (3 hrs)

Hands-on:

• Implement a fused LayerNorm kernel using warp shuffles for the reduction. Benchmark bandwidth against the naive two-pass version. (3 hrs)

Week 14 — CUDA Concurrency: Streams, Graphs, Async Copy

Concepts to understand:

• cp.async (Ampere+): asynchronous SMEM copy from global memory; commit/wait groups; double-buffering pattern to overlap copy and compute
• CUDA streams: stream semantics, default vs. non-default streams, cudaEventRecord/cudaEventSynchronize
• Overlapping H2D copy, kernel, and D2H copy across streams
• CUDA Graphs: graph capture, instantiation, launch; constant-time ~2.5µs launch overhead; when graphs pay off (many small kernels, inference serving)
• Cooperative Groups: thread-block groups, multi-block cooperative kernels, grid-wide synchronization

Reading:

• Controlling Data Movement on Ampere (cp.async) (2 hrs)
• Getting Started with CUDA Graphs (1.5 hrs)
• Cooperative Groups (2 hrs)
• CUDA Streams and Synchronization (1 hr)

Hands-on:

• Add cp.async double-buffering to your LayerNorm kernel from Week 13. Measure the memory bandwidth improvement vs. the synchronous version. (3.5 hrs)

Milestone (Weeks 13–14): Implement a fused LayerNorm with warp-shuffle reduction and double-buffered cp.async pipelining. Profile in Nsight Compute and identify: (a) achieved bandwidth as % of theoretical, (b) dominant stall reason.

Week 15 — CUTLASS & CuTe

Concepts to understand:

• What CUTLASS is: composable CUDA templates for GEMM and convolution; CUTLASS 2.x vs. 3.x (CuTe-based)
• CuTe layout algebra: Layout = Shape × Stride; composable tiling; how CuTe eliminates hand-coded index arithmetic
• CUTLASS 3.x GEMM pipeline: CollectiveMma, CollectiveEpilogue, GemmUniversal
• Warp specialization: producer warps (TMA-issued loads) vs. consumer warpgroups (WGMMA); PipelineAsync barriers; Ping-Pong scheduling
• CUTLASS vs. Triton: when to use each; most production kernels in vLLM/FlashAttention-3 are CUTLASS-based

Reading:

• CUTLASS: Fast Linear Algebra in CUDA C++ (2 hrs)
• CUTLASS: Principled Abstractions (CuTe overview) (2 hrs)
• Deep Dive on CUTLASS Ping-Pong GEMM (3 hrs)
• CutlassAcademy — curated tutorials (3 hrs)

Hands-on:

• Write a CUTLASS 3.x GEMM using GemmUniversal with a custom epilogue (fused bias + ReLU). Profile against cuBLAS and a hand-rolled Triton kernel. (3 hrs)

Week 16 — Hopper Architecture: TMA & WGMMA

Concepts to understand:

• Tensor Memory Accelerator (TMA): hardware DMA engine that accepts 5-D tensor descriptors; single-thread issue in producer warp; eliminates per-thread address calculation
• wgmma.mma_async: asynchronous warpgroup-level (128-thread) MMA; operand B must be in SMEM; requires wgmma.fence / wgmma.commit_group
• Thread-block clusters: groups of up to 16 CTAs on adjacent SMs sharing Distributed Shared Memory (DSMEM)
• Persistent kernel patterns: one CTA per SM, producer warpgroup feeds TMA, consumer warpgroup executes WGMMA
• Ping-pong warp specialization: alternating consumer warpgroups on back-to-back tiles to hide softmax latency

Reading:

• Benchmarking and Dissecting the NVIDIA Hopper GPU Architecture (4 hrs)
• Dissecting Hopper via Microbenchmarking (3 hrs)
• CUTLASS Tutorial: WGMMA on Hopper — Colfax Research (3 hrs)

Hands-on:

• Implement a double-buffered GEMM on H100 using TMA + WGMMA (without CUTLASS). Read the generated PTX/SASS in Nsight Compute and verify cp.async.bulk and wgmma.mma_async appear in the hot loop. (3 hrs)

Week 17 — FlashAttention-3 as Hopper Case Study

Concepts to understand:

• How FA3 combines: TMA-pipelined Q/K/V loads, ping-pong WGMMA between consumer warpgroups, and FP8 block quantization
• Why FA3 achieves ~740 TFLOPS/s (~75% SOL) on H100 while FA2 achieves ~35%
• Inside NVIDIA GPUs: anatomy of a high-performance matmul kernel
• Blackwell (B200) preview: UMMA, FP4 Tensor Cores, 5th-gen NVLink (1.8 TB/s/GPU)

Reading:

• FlashAttention-3 (4 hrs)
• Anatomy of High-Performance Matmul Kernels — Aleksa Gordić (4 hrs)
• Developing CUDA Kernels for Hopper — Colfax PDF (3 hrs)

Hands-on:

• Read the FlashAttention-3 Triton reference implementation. Annotate each section: which Hopper feature is it using, and what is the expected performance impact? (3 hrs)

Milestone (Weeks 15–17): Write a fused FP8 attention kernel in CUTLASS 3.x with TMA + WGMMA + Ping-Pong scheduling. Profile vs. a Triton FlashAttention-2 implementation. Explain the performance delta using roofline analysis.

Week 18 — Profiling Methodology: Nsight Compute & Nsight Systems

Concepts to understand:

• Nsight Compute workflow: capturing a kernel profile with ncu --set full
• Speed-of-Light (SOL): reading SM throughput vs. memory throughput as % of theoretical peak
• Memory Workload Analysis: L1/L2 hit rates, global load efficiency, bank conflicts
• Warp State Statistics: stall reasons (long scoreboard, memory dependency, no instruction, MIO throttle)
• Scheduler Statistics: issued IPC vs. theoretical IPC; interpreting occupancy
• Nsight Systems timeline: CPU-GPU synchronization bubbles, kernel gaps, multi-stream overlap
• Bottleneck taxonomy: memory-bound (bandwidth wall), compute-bound (FLOP wall), latency-bound (small tiles), launch-overhead-bound

Reading:

• Nsight Compute Profiling Guide (5 hrs)
• Accelerating HPC with Nsight Compute Roofline Analysis (2 hrs)
• Nsight Systems User Guide (3 hrs)

Hands-on:

• Take your WGMMA GEMM from Week 16. Do a full Nsight Compute profile: find the dominant stall reason, interpret the roofline position, implement one optimization, re-profile to verify improvement. (3 hrs)

Week 19 — Alternative Accelerators

Concepts to understand:

• TPU architecture: systolic array, HBM bandwidth, inter-chip interconnect (ICI) mesh
• XLA compilation: HLO IR, fusion passes, layout assignment, SPMD partitioning
• PyTorch/XLA: lazy tensor execution, mark_step(), SPMD mesh partitioning
• AMD ROCm/HIP: near-identical to CUDA; wavefront size = 64 on CDNA; hipify for porting
• CDNA (MI300X): 192 GB unified HBM3, no separate VRAM boundary
• Groq LPU, Cerebras WSE-3, Gaudi: architectural tradeoffs at a conceptual level

Reading:

• TPU Deep Dive — Henry Hmko (2 hrs)
• PyTorch/XLA Overview (2 hrs)
• HIP Programming Model — AMD ROCm Docs (2 hrs)
• AI Accelerators Beyond GPUs (1.5 hrs)

Hands-on:

• Port a Triton matmul kernel to HIP using hipify. Run on ROCm (cloud MI300X instance or ROCm Docker). Document any wavefront-size or memory-layout differences. (2.5 hrs)

Week 20 — Interconnects, NCCL, & Part II Capstone

Concepts to understand:

• NVLink generations: NVLink 4.0 (H100, 900 GB/s bidirectional), NVSwitch fabric (all-to-all within DGX)
• InfiniBand: NDR/XDR generations, fat-tree and dragonfly topologies, IBTA architecture
• RDMA basics: memory registration, QP model, one-sided vs. two-sided operations
• GPUDirect RDMA: NIC reads/writes GPU HBM directly over PCIe — eliminates one bounce copy
• NCCL ring-AllReduce algorithm and bandwidth-latency tradeoff
• Tree-AllReduce for small messages; hierarchical collectives (intra-node NVLink + inter-node IB)
• NVSHMEM: device-initiated one-sided PUT/GET from within CUDA kernels

Reading:

• Scaling Deep Learning Training with NCCL (2 hrs)
• Demystifying NCCL (3 hrs)
• InfiniBand vs. RoCE — Juniper White Paper (2 hrs)
• Inside Multi-Node Training — Together.ai (1.5 hrs)

Hands-on:

• Write a custom AllReduce using NCCL primitives (ReduceScatter + AllGather). Benchmark against dist.all_reduce at varying message sizes. Plot effective bus bandwidth vs. message size. (3 hrs)

Milestone (Part II): End-to-end kernel engineering capstone. Choose one: (a) fused FP8 attention with TMA + WGMMA, (b) INT8/FP8 GEMM with per-block dequantization epilogue vs. cuBLAS, or (c) roofline-guided optimization of an underperforming open-source kernel with 3 distinct improvements verified in Nsight Compute.

Part III — Training at Scale

Goal: Extend beyond the ZeRO/Megatron basics to cover sequence parallelism, MoE training, data infrastructure, fault tolerance, efficient finetuning, and RLHF systems.

Week 21 — Sequence Parallelism & Context Parallelism

Concepts to understand:

• Why standard tensor parallelism doesn’t help with O(N²) attention memory at long contexts
• Megatron-LM sequence parallelism: sharding LayerNorm/Dropout along the sequence dimension, using reduce-scatter/all-gather pairs instead of all-reduce
• Ring Attention: blockwise attention with KV blocks passed ring-wise; communication overlaps with local attention computation
• DeepSpeed Ulysses: all-to-all before QKV projection so each device attends to full sequence but a subset of heads; communication stays constant as sequence length and device count scale proportionally
• Combining sequence parallelism with TP+PP (4D parallelism)
• Context parallelism in Megatron-Core and TorchTitan

Reading:

• Ring Attention with Blockwise Transformers (3 hrs)
• DeepSpeed Ulysses (3 hrs)
• TorchTitan — docs + architecture (3 hrs)

Hands-on:

• Implement Ring Attention for a toy transformer using torch.distributed. Measure effective sequence length per GPU vs. standard DP at equal memory budget. (3 hrs)

Week 22 — MoE Training: Architecture & Routing

Concepts to understand:

• Sparse vs. dense compute: MoE activates K of N experts per token — decouples parameter count from FLOPs
• Top-K routing, gating network, load balancing auxiliary loss
• Switch Transformer: top-1 routing, capacity factor, token dropping under overflow
• GShard: expert parallelism via XLA annotations across 2048 TPUs
• Expert Choice routing: experts select tokens (inverted routing)
• Token dropping vs. capacity buffer padding tradeoffs
• MoE compute efficiency: 4× more compute-efficient than dense at low training budgets

Reading:

• Switch Transformers (4 hrs)
• GShard (3 hrs)
• Mixtral of Experts (2 hrs)
• Cameron Wolfe: MoE LLMs (2 hrs)

Hands-on:

• Implement a sparse MoE FFN layer in PyTorch with top-2 routing and load balancing loss. Train on a toy task and measure expert utilization over time. (3 hrs)

Week 23 — MoE Training: Systems & Communication

Concepts to understand:

• Expert parallelism: sharding experts across devices, all-to-all communication for token dispatch and gather
• Why all-to-all is latency-sensitive at small batch sizes — the key serving challenge
• DeepSeek-MoE: fine-grained expert decomposition (mN sub-experts, mK active); shared expert mechanism
• Megablocks: block-sparse matrix formulation eliminating token dropping; 40% faster than Tutel
• Interaction between expert parallelism and TP/PP
• Expert collapse: diagnosing and preventing degenerate routing

Reading:

• MegaBlocks: Efficient Sparse Training with MoE (3 hrs)
• DeepSeekMoE (3 hrs)
• Survey on MoE Inference Optimization — training sections (3 hrs)

Hands-on:

• Replace naïve token-padded expert dispatch in your MoE from Week 22 with a block-sparse implementation. Measure GPU utilization improvement. (3 hrs)

Milestone (Weeks 22–23): For a 47B MoE model (Mixtral-style, 8 experts, top-2), compute: (a) active parameters per token, (b) VRAM per GPU with EP=8, (c) all-to-all communication volume per forward pass. Compare to a 13B dense model with equivalent per-token FLOPs.

Week 24 — Data Pipelines for Pretraining

Concepts to understand:

• The data loading bottleneck: CPU throughput vs. GPU compute; profiling with DataLoader workers and async prefetch
• WebDataset: tar-based sharded format, sequential streaming without random access
• MosaicML StreamingDataset (MDS): deterministic ordering regardless of GPU count, mid-epoch resumption, multi-cloud support
• NVIDIA DALI: GPU-accelerated preprocessing pipeline; eliminates CPU bottleneck for image/video/audio
• Tokenization at scale: pre-tokenizing and caching, memory-mapped numpy arrays for zero-copy loading
• Dataset mixing and weighting: sampling proportions, upsampling high-quality data
• Data quality filtering: MinHash LSH deduplication, perplexity filtering, rule-based heuristics, toxic content filtering

Reading:

• MosaicML StreamingDataset — docs (3 hrs)
• NVIDIA DALI Documentation (3 hrs)
• Training Compute-Optimal Large Language Models (Chinchilla) (3 hrs)
• LLaMA — data pipeline section (2 hrs)

Hands-on:

• Build a streaming data pipeline using MosaicML StreamingDataset. Profile data loading throughput (samples/sec) and measure how many DataLoader workers are needed to saturate a single GPU. (3 hrs)

Week 25 — Fault Tolerance & Large-Scale Reliability

Concepts to understand:

• The scale problem: at 1000 GPUs, MTBF is hours not days
• Full checkpointing vs. sharded checkpointing (PyTorch DCP): save/load cost comparison
• Async checkpointing: save to host CPU memory in background while training continues
• NCCL error handling: NCCL_ASYNC_ERROR_HANDLING, timeout detection, error propagation to Python
• Elastic training: torch.distributed.elastic (torchrun), job preemption and resumption
• SLURM integration: --signal flag, SIGUSR1 for preemption-aware checkpointing
• Monitoring: per-GPU throughput tracking, NaN/inf gradient detection, loss anomaly detection

Reading:

• BLOOM Training Chronicle (3 hrs)
• OPT-175B Training Logbook (3 hrs)
• TorchElastic / torchrun documentation (2 hrs)
• PyTorch Distributed Checkpoint (DCP) documentation (2 hrs)

Hands-on:

• Implement async sharded checkpointing for a distributed training job. Simulate a mid-training failure. Measure recovery time vs. full checkpoint. (3 hrs)

Week 26 — Efficient Finetuning: LoRA, QLoRA, PEFT

Concepts to understand:

• Full finetuning memory breakdown: 16 bytes/param total (weights + gradients + Adam states)
• LoRA: injecting trainable W = BA into each attention projection; rank-r reduces trainable params by up to 10,000×; weights merge at inference — no latency cost
• QLoRA: NF4 quantization of frozen base + double quantization + paged optimizers; enables 65B finetuning on a single 48GB GPU
• PEFT library internals: get_peft_model(), merge_and_unload(), distributed training with FSDP + PEFT
• IA3: learns three scaling vectors rather than low-rank matrices — even fewer trainable params than LoRA

Reading:

• LoRA: Low-Rank Adaptation of Large Language Models (3 hrs)
• QLoRA: Efficient Finetuning of Quantized LLMs (3 hrs)
• Hugging Face PEFT Library (3 hrs)
• TRL smol-course (4 hrs)

Hands-on:

• Finetune LLaMA-3-8B using: (a) full finetuning with FSDP, (b) LoRA rank-8, (c) QLoRA 4-bit. Record peak memory, training throughput, and eval accuracy for each. (3 hrs)

Milestone (Weeks 24–26): Build a complete finetuning pipeline: streaming data loading → QLoRA training with async checkpointing → PEFT weight merge → evaluation. Document the memory and throughput profile at each stage.

Week 27 — RLHF Systems: SFT, Reward Models & PPO

Concepts to understand:

• Why RLHF changes the training topology: four model instances coexist (Actor, Critic, Reward Model, Reference Model)
• Reference model and KL penalty: frozen initial policy produces per-token log-probs used in KL term r_θ - λ * KL(π || π_ref)
• The rollout bottleneck: generating on-policy samples is ~80% of wall-clock time in PPO
• OpenRLHF architecture: Ray orchestrates model groups; vLLM handles generation; DeepSpeed ZeRO-3 handles training
• SFT infrastructure: sequence packing, efficient attention masking for packed sequences
• Reward model training: Bradley-Terry preference model, ranking loss, process reward models (PRMs)

Reading:

• Illustrating RLHF — Hugging Face (1.5 hrs)
• OpenRLHF — docs + architecture (5 hrs)
• TRL Documentation (4 hrs)

Hands-on:

• Run a full PPO training loop on a 7B model using OpenRLHF. Profile wall-clock time split between rollout generation and model updates. Measure how vLLM integration affects total throughput. (3 hrs)

Week 28 — RLHF Algorithm Variants: DPO, GRPO & Beyond

Concepts to understand:

• DPO: reformulates RLHF as a classification loss; no reward model or RL loop at training time; implicit reward is log(π_θ(y_w|x)/π_ref(y_w|x)) - log(π_θ(y_l|x)/π_ref(y_l|x))
• GRPO: eliminates critic network; normalizes rewards within a group of sampled outputs; reduces memory by removing the value network
• RLOO: leave-one-out baseline within a group of K samples; simpler than GRPO
• KTO: aligns using binary feedback (thumbs up/down); avoids need for paired data
• ORPO: combines SFT + preference loss in a single stage; no reference model needed
• Online vs. offline: DPO/KTO are offline; PPO/GRPO/RLOO are online (generate on-policy rollouts every step)

Reading:

• Direct Preference Optimization — Rafailov et al. (3 hrs)
• DeepSeekMath — GRPO derivation (3 hrs)
• Putting RL back in RLHF — Hugging Face (1.5 hrs)

Hands-on:

• Train LLaMA-3-8B with DPO on a preference dataset. Compare reward margin and eval performance against the PPO run from Week 27. Measure training memory and time per step. (3 hrs)

Milestone (Part III): Full RLHF pipeline: streaming data loading → QLoRA SFT → DPO training with async checkpointing → evaluation. Document GPU memory, throughput, and reward statistics at each stage.

Part IV — Advanced Inference

Goal: Go deep on long-context serving, next-generation quantization, model compression, advanced speculative decoding, disaggregated inference, and production-grade serving infrastructure.

Week 29 — Long-Context Inference: Position Encodings

Concepts to understand:

• RoPE mechanics: encoding relative position as a rotation in the complex plane via e^(imθ)
• Why naive RoPE extrapolation fails beyond training sequence length
• Position interpolation (PI): linearly scaling position indices to fit within training range
• NTK-aware interpolation: applying PI non-uniformly across frequency components
• YaRN: ramp function + temperature correction for improved long-context performance
• LongRoPE: frequency-domain analysis of dimension-wise scaling

Reading:

• YaRN: Efficient Context Window Extension of Large Language Models (3 hrs)
• EleutherAI blog: Extending the RoPE (1.5 hrs)
• Ring Attention for inference — inference section (1.5 hrs)

Hands-on:

• Extend LLaMA-3-8B to 32K context using YaRN. Measure perplexity at 4K, 8K, 16K, 32K and compare to a model without YaRN. (3 hrs)

Week 30 — Long-Context Inference: KV Cache Management

Concepts to understand:

• Chunked prefill: splitting long-prompt prefill into chunks to reduce head-of-line blocking
• Prefix caching: radix-tree KV reuse across requests sharing common prefixes
• H2O (Heavy-Hitter Oracle): evicting KV entries based on attention mass
• SnapKV: query-guided one-shot per-layer token selection before generation
• StreamingLLM: attention sinks + recency window for infinite-length generation with fixed VRAM
• SAGE-KV: one-shot token/head-level top-k eviction

Reading:

• SnapKV (2 hrs)
• vLLM blog: Anatomy of a High-Throughput LLM Inference System (2 hrs)
• StreamingLLM (2 hrs)

Hands-on:

• Implement H2O KV eviction on top of a vLLM-served model. Measure perplexity degradation vs. memory savings at 50% cache budget. (3 hrs)

Milestone (Weeks 29–30): For a LLaMA-3-70B model on 4×A100 with tensor parallelism, compute: (a) baseline KV cache size at seqlen=32K, batch=8, (b) memory savings under 50% H2O eviction, (c) latency impact of chunked prefill at different chunk sizes.

Week 31 — Advanced Quantization: FP8, GGUF, and Extreme Low-Bit

Concepts to understand:

• FP8 floating-point formats: E4M3 (range-optimized) vs. E5M2 (precision-optimized), scaling and saturation behavior
• FP8 inference on Hopper/Ada: NVIDIA TransformerEngine fp8_autocast API
• SmoothQuant: migrating quantization difficulty from activations to weights via per-channel scaling; joint W8A8
• GGUF format: K-quant families (Q2_K through Q6_K), block-scale encoding, CPU/GPU hybrid execution
• QuIP#: Hadamard incoherence processing + E8 lattice codebooks for 2-bit weights
• AQLM: additive quantization with learned codebooks; training cost vs. inference speed
• Mixed-precision quantization: per-layer bit selection via sensitivity analysis
• Calibration dataset design: representativeness, domain shift effects

Reading:

• NVIDIA TransformerEngine FP8 Primer (2 hrs)
• QuIP# (3 hrs)
• AQLM (3 hrs)
• Which Quantization Should I Use? (systematic comparison) (2 hrs)

Hands-on:

• Benchmark LLaMA-3-8B in fp16, int8 (SmoothQuant), int4 (GPTQ), FP8 (TE), and GGUF Q4_K_M. Record throughput (tokens/sec), TTFT, and perplexity. (3 hrs)

Week 32 — Model Compression: Pruning & Distillation

Concepts to understand:

• Unstructured pruning: magnitude pruning, SparseGPT (Hessian-based layer-wise), Wanda (weight × activation magnitude criterion)
• Structured pruning: head pruning, layer pruning, width pruning; hardware-efficiency tradeoffs
• 2:4 structured sparsity: what NVIDIA sparse tensor cores require
• Knowledge distillation at scale: sequence-level vs. token-level losses, forward KL vs. reverse KL (MiniLLM)
• NVIDIA Minitron: structured pruning + short distillation fine-tune as production recipe
• Combined pipelines: prune → distill → quantize stacking

Reading:

• Wanda: A Simple and Effective Pruning Method (2 hrs)
• MiniLLM: Knowledge Distillation of Large Language Models (2.5 hrs)
• ACM Efficient Compressing and Tuning Methods for LLMs — survey (4 hrs)

Hands-on:

• Apply Wanda unstructured pruning to LLaMA-3-8B at 50% sparsity. Measure throughput change (hint: unstructured sparsity doesn’t help on dense hardware) then apply 2:4 structural pattern and remeasure. (3 hrs)

Week 33 — Advanced Speculative Decoding

Concepts to understand:

• Review: draft-model speculative decoding, acceptance rate, expected speedup derivation
• Medusa: multiple decoding heads on frozen backbone; Medusa-1 (frozen LM) vs. Medusa-2 (joint fine-tune); predicts positions +1…+5
• Tree-structured candidate verification: constructing a candidate tree, batching the verify forward pass, token acceptance mask
• EAGLE-1: draft at the feature level (not token level) using a shallow auto-regressive head on frozen LM embeddings
• EAGLE-2: context-aware dynamic draft tree; confidence-score-based acceptance rate; 20–40% faster than EAGLE-1
• Self-speculative decoding (layer skipping): same model with skipped layers as draft — no auxiliary model needed

Reading:

• Medusa (2.5 hrs)
• EAGLE (3 hrs)
• EAGLE-2 (2 hrs)
• vLLM blog: Speculative Decoding up to 2.8× (1 hr)

Hands-on:

• Enable Medusa and EAGLE-2 in vLLM for LLaMA-3-8B. Measure speedup at greedy and temperature=1 decoding. Compare acceptance rates. (3 hrs)

Milestone (Weeks 31–33): Given a 70B model with 50% Wanda pruning + AQLM 2-bit quantization + EAGLE-2 speculative decoding: estimate theoretical tokens/sec vs. the uncompressed baseline. Identify which technique gives the highest throughput-per-quality-point tradeoff.

Week 34 — Disaggregated & Distributed Inference

Concepts to understand:

• Why prefill and decode have fundamentally different compute/memory profiles (roofline)
• Splitwise: phase splitting onto heterogeneous hardware (H100 prefill / A100 decode), KV migration protocol
• DistServe: goodput-optimized disaggregation, independent parallelism strategies per phase
• Tensor parallelism for inference: column/row partition, all-reduce cost
• Multi-GPU vLLM: --tensor-parallel-size, worker topology, NVLink vs. PCIe bandwidth sensitivity
• 2025 landscape: disaggregation as the default (NVIDIA Dynamo, SGLang, LMCache)

Reading:

• DistServe (3 hrs)
• Hao AI Lab: Disaggregated Inference 18 Months Later (1 hr)
• NVIDIA Dynamo announcement (1 hr)

Hands-on:

• Deploy LLaMA-3-70B with tensor parallelism across 2 GPUs using vLLM. Measure TTFT and throughput vs. single-GPU with 4-bit quantization. Identify which strategy gives better throughput/VRAM tradeoff. (3 hrs)

Week 35 — Production Serving Infrastructure

Concepts to understand:

• SLO taxonomy: TTFT, TBT (time-between-tokens), P50/P99 targets
• SLO-aware scheduling: global queue management, preemption, priority classes
• Load balancing across replicas: session affinity for prefix cache hits, least-outstanding-requests
• Autoscaling: request-rate-based vs. queue-depth-based triggers, custom Prometheus metrics, K8s HPA
• Triton Inference Server: model repository, backend API, dynamic batching, ensemble models
• TensorRT-LLM: plugin system, inflight batching, paged KV cache, quantized kernel dispatch
• Cost modeling: $/token, GPU utilization targets, spot instance fault tolerance

Reading:

• SLO-Aware Scheduling for LLM Inferences (2.5 hrs)
• AWS: Multi-node TensorRT-LLM + Triton on EKS (3 hrs)
• A Survey on Inference Engines for LLMs (4 hrs)

Hands-on:

• Deploy LLaMA-3-8B with TensorRT-LLM + Triton Inference Server. Set TTFT P99 < 500ms as an SLO. Measure max throughput (tokens/sec) while holding the SLO. (4 hrs)

Milestone (Part IV): End-to-end serving system: FP8-quantized 70B model on disaggregated prefill/decode infrastructure, with SLO-aware scheduling, autoscaling, and cost monitoring. Document latency, throughput, GPU utilization, and $/token achieved.

Part V — Emerging Architectures

Goal: Understand Mamba/SSMs, MoE inference, linear attention variants, diffusion systems, and advanced multimodal architectures — always through the systems lens of compute, memory, and parallelism.

Week 36 — State Space Models: Mamba

Concepts to understand:

• Selective SSM recurrence: input-dependent (A, B, C) parameters break time-invariance — preventing FFT shortcut
• Parallel associative scan: the core primitive — associativity of state-update operator, tree reduction, work vs. depth analysis
• Mamba’s kernel fusion: why the naïve sequential scan is memory-bandwidth-bound; how kernel fusion eliminates HBM materialization (analogous to FlashAttention tiling)
• Inference memory profile: O(1) KV-cache equivalent — fixed-size recurrent state regardless of sequence length
• Mamba-2 / SSD (Structured State Space Duality): reformulating selective SSM as semiseparable matrix multiplication; enables tensor-core utilization; 2–8× speedup over Mamba-1

Reading:

• Mamba: Linear-Time Sequence Modeling with Selective State Spaces (4 hrs)
• Tri Dao: SSD Blog series (Parts I–III) (3 hrs)
• Princeton PLI: Mamba-2 Algorithms and Systems (2 hrs)

Hands-on:

• Implement the Mamba selective scan in pure PyTorch. Verify output matches the reference implementation. Profile memory usage at seqlen 4K, 16K, 64K — compare to FlashAttention-2. (3 hrs)

Week 37 — Mamba-2, Triton Scan Kernels & Hybrid Architectures

Concepts to understand:

• Mamba-2/SSD: block-decomposable structure enabling sequence parallelism across devices
• Implementing efficient parallel scan in Triton: tile sizes, recomputation vs. checkpointing of states, variable-length batches
• Hybrid architectures (Jamba, Zamba): the 1:7 attention-to-Mamba ratio; interleaved MoE layers; systems implications for serving
• flash-linear-attention library: unified Triton kernels for GLA/Mamba/RWKV

Reading:

• Mamba-2 / SSD (5 hrs)
• Mamba: The Hard Way — Sasha Rush (annotated Triton) (6 hrs)
• Jamba: Hybrid Transformer-Mamba Language Model (2 hrs)

Hands-on:

• Implement a simplified parallel selective scan in Triton. Benchmark against the sequential Mamba reference at seqlen 1K→64K. Measure FLOP/s and memory bandwidth utilization. (4 hrs)

Milestone (Weeks 36–37): Build a Mamba-2 inference server. Measure: (a) per-token VRAM at seqlen 1K vs. 64K vs. 256K (should be constant), (b) throughput vs. an equivalent-parameter transformer at each seqlen, (c) the crossover point where Mamba becomes faster.

Week 38 — MoE Architecture & Inference Serving

Concepts to understand:

• Sparse MoE inference: expert weight storage, all-to-all routing, GPU utilization at small batch sizes
• Expert offloading: streaming expert weights from CPU DRAM; expert prediction/caching to reduce transfer cost
• Serving Mixtral 8×7B and DeepSeek-MoE: how vLLM/SGLang handle expert-parallel routing at batching time
• MoE vs. dense at inference: same FLOPs per token, 4× more parameters — memory bandwidth bottleneck is worse

Reading:

• Efficient Large Scale Language Modeling with MoE — empirical efficiency study (2 hrs)
• Survey on MoE Inference Optimization — inference sections (3 hrs)
• DeepSeekMoE (3 hrs)

Hands-on:

• Serve Mixtral-8×7B with vLLM. Profile GPU memory and expert routing distribution. Measure how batch size affects expert utilization and throughput. (3 hrs)

Week 39 — Linear Attention & Hybrid Architectures

Concepts to understand:

• Kernel-trick reformulation of softmax attention: replacing exp(qᵀk) with φ(q)ᵀφ(k) — the root of O(1)-memory inference
• Why pure linear attention underperforms: missing normalizer term, forgetting over long sequences
• GLA (Gated Linear Attention): chunked-form training kernel in Triton; faster than FlashAttention-2 at 1K seqlen
• RWKV dual-mode: parallel training (WKV operator along sequence) vs. pure RNN inference (O(1) per-token, fixed VRAM)
• RetNet: decay matrix as structured (diagonal + low-rank) operator enabling efficient chunkwise computation
• flash-linear-attention: how subquadratic kernels share the same tiling and recomputation strategy as FlashAttention

Reading:

• GLA: Gated Linear Attention Transformers with Hardware-Efficient Training (3 hrs)
• RWKV: Reinventing RNNs for the Transformer Era (3 hrs)
• flash-linear-attention — code + docs (3 hrs)

Hands-on:

• Benchmark GLA vs. FlashAttention-2 vs. Mamba at seqlen 512, 1K, 2K, 4K, 8K. Plot throughput (tokens/sec) and memory per token. Identify the regime where each is fastest. (3 hrs)

Week 40 — Diffusion Model Systems: Samplers & Architectures

Concepts to understand:

• DDPM reverse process: N forward passes per image — O(N × model_FLOPs) inference cost
• DDIM and DPM-Solver: deterministic ODE-based samplers reducing steps from ~1000 to 10–20 without retraining
• Flow matching: velocity field over straight ODE trajectories — fewer NFEs at inference; numerically better-conditioned
• Consistency models: self-consistency along ODE trajectory → single-step inference
• DiT (Diffusion Transformer) vs. U-Net: patching the latent space with a ViT backbone; uniform FLOP distribution per layer; simpler tensor parallelism

Reading:

• DiT: Scalable Diffusion Models with Transformers (3 hrs)
• DPM-Solver (3 hrs)
• Flow Matching for Generative Modeling (3 hrs)
• Efficient Diffusion Models Survey (4 hrs)

Hands-on:

• Profile DiT-XL/2 inference. Measure latency at 50 steps (DDPM) vs. 10 steps (DPM-Solver) vs. 1 step (consistency). Plot quality (FID) vs. latency tradeoff. (3 hrs)

Week 41 — Diffusion Serving & Distributed Execution

Concepts to understand:

• Batching strategies for diffusion: requests are embarrassingly parallel within a batch; variable CFG guidance scale complicates batching
• DistriFusion: asynchronous parallel denoising over devices — exploiting temporal redundancy across steps
• PipeFusion: pipeline parallelism across transformer layers for DiT
• Long-context sparse attention patterns: BigBird’s random + local + global, Longformer sliding-window + global
• ALiBi: per-head linear bias on attention logits proportional to distance — no positional embedding parameters; generalizes to unseen lengths

Reading:

• BigBird: Transformers for Longer Sequences (3 hrs)
• Hyena Hierarchy (3 hrs)
• Efficient Attention Mechanisms for LLMs Survey (4 hrs)

Hands-on:

• Implement DistriFusion for DiT-XL/2 inference across 2 GPUs. Compare latency to single-GPU at the same step count. Measure communication overhead. (3 hrs)

Milestone (Weeks 40–41): For a DiT-XL/2 model serving at 256 requests/sec, design an architecture: (a) step count vs. quality tradeoff using DPM-Solver, (b) batching strategy with variable CFG, (c) multi-GPU with DistriFusion — estimate total GPU count needed to hit 100ms TTFT P99.

Week 42 — Advanced Multimodal: Video & Audio

Concepts to understand:

• ViT FLOP profile: patch count n = (H × W) / p² grows quadratically with resolution — direct analogy to seqlen scaling
• Image tokenization tradeoffs: continuous patch embeddings vs. discrete VQ-VAE tokens; 32-token tokenization (NeurIPS 2024)
• Video model systems: naive 3D attention is O((T·H·W)²); space-time factored attention; sliding-window temporal attention; memory management for long video sequences
• Frame batching strategies: independent frames (cheap, no temporal coherence) vs. tubelet embeddings (3D patch tokens)
• Audio models (Whisper, EnCodec): spectrogram-to-patch tokenization, streaming inference, causal attention latency constraints
• Cross-modal fusion: Q-Former / cross-attention layers (BLIP-2, Flamingo) vs. simple projection (LLaVA)

Reading:

• An Image is Worth 16×16 Words: ViT (2 hrs)
• An Image is Worth 32 Tokens for Reconstruction and Generation (2 hrs)
• Image and Video Tokenization (ICLR 2025) (2 hrs)
• Vision Transformers on the Edge: Model Compression Survey (3 hrs)

Hands-on:

• Serve a video-language model (e.g., LLaVA-Video or VideoLLaMA). Profile: (a) video tokenization latency vs. LLM prefill vs. decode, (b) how video length and resolution affect KV cache size and TTFT. (4 hrs)

Milestone (Part V): Implement a Mamba-2 parallel scan kernel and benchmark it end-to-end: compare throughput, memory, and quality against FlashAttention-2 at seqlens from 2K to 256K. Write a one-page analysis of which architecture you’d choose for which use case.

Part VI — Compilers & Infrastructure

Goal: Understand the full compiler stack from MLIR down to PTX, the internals of automatic differentiation, memory allocation, and production MLOps infrastructure.

Week 43 — Compiler IR Theory & MLIR

Concepts to understand:

• SSA (Static Single Assignment): basic blocks, CFGs, dominance frontiers, liveness analysis, φ-functions
• MLIR architecture: op, attribute, type, region, block; built-in dialects (func, arith, affine, linalg, memref, scf)
• Dialect conversion framework and the Transform dialect for compiler-controlled transformations
• Lowering chains: MLIR → LLVM dialect → LLVM IR → PTX
• torch.compile internals: Dynamo bytecode interception (PEP 523), FX graph, guard mechanism, graph breaks

Reading:

• SSA-Based Compiler Design (free PDF) — selective chapters (5 hrs)
• MLIR Toy Tutorial Chapters 1–6 (8 hrs)
• depyf: decompiles torch.compile bytecode (3 hrs)
• PyTorch Dynamo Deep-Dive (4 hrs)

Hands-on:

• Complete all 6 chapters of the MLIR Toy tutorial. Then use depyf to inspect the Dynamo FX graph of a transformer forward pass. Identify 3 graph breaks and explain what causes them. (3 hrs)

Week 44 — TVM, XLA & Autotuning

Concepts to understand:

• TVM / TensorIR: first-class schedulable IR; MetaSchedule stochastic search space (block tiling, loop reordering, vectorization); how tuning records are stored
• Ansor: program search space for high-performance tensor programs; cost-model-driven auto-scheduling
• XLA HLO instruction set: algebraic simplification, fusion, layout assignment, buffer assignment
• XLA SPMD partitioner: sharding annotations, per-op partitioning semantics, automatic collective insertion
• GSPMD: generalizing SPMD to arbitrary parallelism strategies

Reading:

• Machine Learning Compilation (MLC) course — TensorIR + MetaSchedule chapters (8 hrs)
• OpenXLA GPU Architecture Overview (3 hrs)
• GSPMD (3 hrs)
• Ansor (2 hrs)

Hands-on:

• Autotune a matrix multiplication using TVM MetaSchedule. Compare achieved FLOP/s to: (a) a naive implementation, (b) cuBLAS, (c) your hand-written Triton kernel. (3 hrs)

Week 45 — Automatic Differentiation Systems

Concepts to understand:

• JVP (forward-mode AD) as Jacobian–vector product; VJP (reverse-mode) as vector–Jacobian product
• Why reverse mode dominates for ML: O(m) VJPs vs. O(n) JVPs for n-input, m-output functions
• JAX transformation model: jax.jvp, jax.vjp, jax.grad; Jaxpr as the internal lambda calculus IR
• vmap as a batch-dimension lifting transformation; how jit + vmap + grad compose
• Custom derivatives in JAX: custom_jvp and custom_vjp for non-differentiable ops and numerical stability fixes
• PyTorch autograd internals: dynamic computation graph, Function.forward/backward, AccumulateGrad nodes, C++ engine thread pool
• AOTAutograd: ahead-of-time joint forward+backward graph capture; why graph breaks hurt backward compilation

Reading:

• JAX Autodiff Cookbook (4 hrs)
• JAX JVP/VJP documentation (2 hrs)
• JAX Custom derivative rules (2 hrs)
• How Computational Graphs are Constructed in PyTorch (2 hrs)

Hands-on:

• Implement a custom vjp in JAX for numerically stable log-softmax. Verify the gradient matches jax.grad on stable inputs but doesn’t NaN on extreme inputs. Then implement the same in PyTorch with torch.autograd.Function. (3 hrs)

Week 46 — Memory Management & Allocators

Concepts to understand:

• PyTorch CUDA caching allocator: block splitting and reuse, per-stream caching, rounding policy, when cudaMalloc/cudaFree are actually called
• Memory fragmentation: external fragmentation from variable-size tensors across streams; max_split_size_mb; PYTORCH_CUDA_ALLOC_CONF knobs
• cudaMallocAsync backend: CUDA 11.4+ virtual memory pools, stream-ordered allocation semantics; when it beats the native allocator
• Memory snapshot tooling: torch.cuda.memory._record_memory_history() + _dump_snapshot() + pytorch.org/memory_viz
• Gradient checkpointing + activation CPU offload: separate CUDA streams for overlap; FSDP + offload combinations

Reading:

• A Guide to PyTorch’s CUDA Caching Allocator — Zach DeVito (2.5 hrs)
• Understanding GPU Memory 1: Visualizing All Allocations over Time (1.5 hrs)
• PyTorch CUDA semantics — PYTORCH_CUDA_ALLOC_CONF reference (1 hr)
• torchtune: Memory Optimization Overview (2 hrs)

Hands-on:

• Capture a memory snapshot for a distributed training run with FSDP ZeRO-3. Use pytorch.org/memory_viz to identify the peak allocation event. Reduce peak memory by 20% through allocator tuning and partial activation offload. (3 hrs)

Week 47 — NCCL Deep Dive

Concepts to understand:

• NCCL ring-AllReduce: bandwidth-optimal for large messages; latency formula O(2(n-1)/n × α + 2(n-1)/n × β × M)
• Double binary tree: logarithmic latency with full bandwidth; when to prefer tree over ring
• NCCL protocols: Simple (bandwidth-optimal), LL (latency-optimal 8-byte writes), LL128 (balanced); dynamic selection based on message size
• NCCL tuning: NCCL_ALGO, NCCL_PROTO, channel count, thread count; benchmarking with nccl-tests
• Compute-communication overlap: async collectives on separate CUDA streams; NCCL 2.28 copy-engine collectives
• SHARP: in-network reduction on InfiniBand switches — eliminating the final merge step

Reading:

• Understanding NCCL Tuning (2 hrs)
• Fast Multi-GPU Collectives with NCCL (1.5 hrs)
• NCCL 2.28 Copy Engine Collectives (1.5 hrs)
• Demystifying NCCL (3 hrs)

Hands-on:

• Use nccl-tests to benchmark AllReduce across 2 and 8 GPUs at message sizes 1KB to 1GB. Plot bus bandwidth vs. message size. Identify the crossover between latency-dominated and bandwidth-dominated regimes. (3 hrs)

Week 48 — Deployment & Production Infrastructure

Concepts to understand:

• Experiment tracking: MLflow (runs, artifacts, model registry) vs. W&B (sweeps, artifact versioning, comparison); when to use which
• ONNX export pipeline: torch.onnx.export dynamo-based path; opset versioning; ONNX Runtime graph optimization levels (basic, extended, all); execution provider selection (CUDA EP, TensorRT EP)
• Kubernetes for ML: resource requests/limits for GPU pods (nvidia.com/gpu), Kubeflow Pipelines, KubeRay, autoscaling GPU node pools
• Cost optimization: spot/preemptible instances, bin-packing, inference-time quantization, request batching

Reading:

• ONNX Runtime — Graph Optimizations (2 hrs)
• ONNX Runtime — Execution Providers (1.5 hrs)
• Ray for ML Infrastructure (4 hrs)
• Full Stack Deep Learning Lecture 6: MLOps (2 hrs)

Hands-on:

• Export LLaMA-3-8B to ONNX. Deploy with ONNX Runtime using the TensorRT EP. Compare latency to vLLM serving. Identify which graph optimizations ONNX Runtime applies automatically. (3 hrs)

Week 49 — Advanced Profiling & Distributed Debugging

Concepts to understand:

• Nsight Compute kernel analysis: SOL section, Memory Workload Analysis chart, Warp State Statistics stall taxonomy, Scheduler Statistics, Source/SASS views for line-level attribution
• Nsight Systems: NVTX range annotations, correlating host-side Python/C++ with device execution, multi-node traces for straggler identification
• Distributed training debugging: NCCL_DEBUG=INFO, rank asymmetry (one rank hangs while others wait), common NCCL error classes (mismatched tensor shapes/dtypes, communicator reuse bugs)
• PyTorch memory profiler: profile_memory=True, memory snapshot workflow, memory_stats counters

Reading:

• Using Nsight Compute to Inspect Your Kernels (2.5 hrs)
• Debugging NCCL Errors in Distributed Training (1.5 hrs)
• Debugging PyTorch Memory Use with Snapshots — Zach DeVito (1.5 hrs)
• PyTorch Mosaic Memory Profiling Tutorial (1.5 hrs)

Hands-on:

• Run a multi-GPU training job where one rank is intentionally slower (add a sleep). Use Nsight Systems + NCCL_DEBUG=INFO to identify the straggler. Then fix the bottleneck and verify the speedup. (3 hrs)

Week 50 — Integrating the Compiler Stack

Concepts to understand:

• Full stack: Python → Dynamo FX graph → AOTAutograd joint graph → Inductor loop IR → Triton / CUTLASS → PTX → SASS
• How each layer in the stack creates or destroys optimization opportunities
• TorchInductor design: define-by-run IR, symbolic shapes, persistent reduction fusions
• When to bypass each layer: hand-written Triton for custom ops, CUTLASS for peak GEMM performance, XLA for TPU/multi-host SPMD

Reading:

• TorchInductor design doc (2 hrs)
• PyTorch 2 paper — re-read for architecture (2 hrs)
• GSPMD paper — re-read for the SPMD big picture (2 hrs)

Hands-on:

• Write a custom PyTorch operator that falls through the entire stack: implement the forward in Triton, register a custom torch.autograd.Function with a custom_vjp, and verify torch.compile can capture and fuse it. (4 hrs)

Week 51 — Reading Week & Integration

A structured review week with no new material. Revisit the notes, exercises, and milestone answers from the hardest weeks.

Suggested review targets:

• Re-read FlashAttention-3, Mamba-2, and DistServe papers with fresh eyes
• Re-derive: ZeRO-3 communication volume, arithmetic intensity for every major op type, speculative decoding speedup formula
• Review your Nsight Compute profiles from Weeks 18 and 49 — do you understand every row now?
• Read Taming the Titans: A Survey of Efficient LLM Inference Serving as a capstone survey (4 hrs)
• Read A Survey on Efficient Inference for Large Language Models (5 hrs)

Week 52 — Year-End Capstone

Goal: Ship something production-quality that integrates at least 3 domains from the curriculum. The deliverable should be something you’d be comfortable presenting as a portfolio piece.

Choose one track:

• Track A — Full-Stack LLM Serving System: Design and deploy a serving system for a 70B model: disaggregated prefill/decode, FP8 quantization, EAGLE-2 speculative decoding, SLO-aware scheduling, autoscaling on K8s. Document latency, throughput, cost, and the roofline analysis behind each design decision.
• Track B — Custom Kernel + Compiler Integration: Write a Mamba-2 parallel scan kernel in CUTLASS using TMA + WGMMA. Register it as a custom PyTorch op, verify torch.compile can capture it, profile end-to-end in Nsight Compute. Compare to the Triton reference implementation and cuBLAS-equivalent GEMM.
• Track C — RLHF System Optimization: Implement a complete RLHF training pipeline (SFT → reward model → GRPO) for a 13B model. Optimize rollout generation with vLLM, training with FSDP + QLoRA, and checkpointing with async DCP. Achieve ≥80% GPU utilization throughout the training loop.
• Track D — Diffusion Systems: Build a production DiT serving stack: DPM-Solver with 10 steps, FP8 inference, DistriFusion across 4 GPUs, TensorRT-LLM backend. Measure samples/sec at various quality levels and plot the quality–latency Pareto frontier.

Deliverable checklist:

• Architecture diagram showing which curriculum concepts were applied and why
• Baseline measurement (before optimizations)
• Each optimization applied with before/after numbers
• Roofline analysis at the bottleneck point
• What you’d do next if you had more hardware or time

Reference Lists

Canonical Papers by Week

Key Blogs & References

Last updated: 2026-03-15. Revisit pacing at Part boundaries.