Allowed (Recommended) Resources
- NVidia Forums - NVIDIA engineers
answer here. Search it when the compiler, the driver or the hardware does
something the documentation doesn't explain.
- StackOverflow - most CUDA error
messages and C++ pitfalls have already been asked and answered here.
- Algorithms for Competitive Programming - the
number theory and algorithms behind the problems: modular arithmetic,
Barrett and Montgomery reduction, determinants, prefix sums and more.
- Algorithms for Modern Hardware - how
to make code fast on real hardware: latency versus throughput, memory
bandwidth, caches, SIMD and fast modular arithmetic. It is written for CPUs,
but the same reasoning tells you whether a kernel is limited by memory or by
compute.
- Godbolt - shows the PTX and SASS that your code
compiles to, so you can check what actually runs instead of guessing. Every
submission page links to its source there, with the grader's flags.
- CPP Reference - the exact rules of C++: integer
overflow and promotion,
constexpr, templates and the standard library.
- Cuda Programming Guide -
the programming model: threads, blocks, warps, the memory hierarchy, and what
each compute capability supports.
- Cuda C Best Practices Guide -
NVIDIA's own checklist for speed: coalesced memory access, occupancy, shared
memory bank conflicts and how to measure what limits a kernel.
- Cuda ISA -
the PTX reference: what each instruction does, and the intrinsics and inline
assembly (
dp4a, mma, prmt and so on) that C++ alone won't reach.
- Blackwell GPU Wiki - notes
on NVIDIA's Blackwell architecture: what changed from earlier GPUs and how
that changes the way to write fast kernels.
- Modern GPU Programming for MLSYS -
a course that builds fast kernels step by step, especially matrix
multiplication: tiling, shared memory, tensor cores and pipelining.
- TCGEN05 - a walkthrough of
tcgen05, the tensor core instructions of datacenter Blackwell GPUs
(sm_100), for seeing how the fastest matrix kernels are built.
Command Line
Download cli.py into an empty folder, make it
executable and run it:
curl -o cli.py https://zetatron.thinkinrocks.com/cli.py
chmod +x cli.py
./cli.py setup
It downloads every problem's statement and starter code, and saves your
pincode for submitting. To submit a solution and watch its results, give the
problem's name, or just its first letters:
./cli.py s argmax
FAQ
Q: Can I use Autotuners/Bayesian Optimisers/Parameter Search?
A: You, you can, but if you find yourself needing that to speed your solution up, you are optimising the wrong thing.
Q: What hardware are the programs ran on?
A: NVidia RTX 5070, 12 Gb VRAM, Blackwell.
Q: Can I use printf?
A: Yes! It is only visible when a test has failed, but we welcome it. Remember to remove them from the final submission.
Problems
A small set of GPU exercises. Each problem is self-contained: open it, read
the statement, then implement the solution in CUDA.