NCA Performance Optimization & Debugging 4 — Questions and Answers
Question 1: What is warp divergence and how does it impact GPU performance?
- Threads in a warp take different code paths, forcing serial execution of each path (Correct answer)
- Warps from different blocks compete for the same SM resources
- The warp scheduler cannot find a ready warp to execute
- Threads exit early causing other threads to idle
Correct answer: Threads in a warp take different code paths, forcing serial execution of each path
When threads in a warp branch to different code paths due to conditional logic, the GPU must execute each path serially with inactive threads masked off, reducing effective parallelism.
Question 2: A CUDA developer enables Unified Memory (cudaMallocManaged). What is the main performance risk if not managed carefully?
- Page faults trigger on-demand migration, stalling the GPU until pages are transferred (Correct answer)
- Unified Memory has lower peak bandwidth than explicit device memory
- Kernel launches are serialized when Unified Memory is in use
- The CUDA driver limits kernel occupancy for Unified Memory allocations
Correct answer: Page faults trigger on-demand migration, stalling the GPU until pages are transferred
Without prefetching hints, Unified Memory causes GPU page faults that stall kernel execution while pages are migrated from CPU to GPU on demand, creating latency spikes.
Question 3: Which CUDA-GDB command pauses execution when a specific CUDA thread encounters a condition?
- break with a conditional expression targeting thread coordinates (Correct answer)
- watch on a device variable
- cuda thread select followed by step
- set cuda break_on_launch application
Correct answer: break with a conditional expression targeting thread coordinates
CUDA-GDB supports conditional breakpoints using thread/block index expressions like `break kernel.cu:42 if threadIdx.x == 0 && blockIdx.x == 1`.
Question 4: In a multi-GPU setup, which API call is used to enable direct peer-to-peer memory access between two CUDA devices?
- cudaDeviceEnablePeerAccess() (Correct answer)
- cudaMemcpyPeer()
- cudaSetDevice() with both device IDs
- cudaDeviceSynchronize() across devices
Correct answer: cudaDeviceEnablePeerAccess()
cudaDeviceEnablePeerAccess() enables one GPU to directly read/write another GPU's memory over NVLink or PCIe without routing through host memory, reducing latency and improving bandwidth.
Question 5: What is the effect of using too much shared memory per block on SM occupancy?
- Fewer blocks can fit on the SM simultaneously, reducing occupancy (Correct answer)
- Warp scheduling becomes non-deterministic
- Global memory accesses automatically increase to compensate
- Thread synchronization latency increases proportionally
Correct answer: Fewer blocks can fit on the SM simultaneously, reducing occupancy
Shared memory is partitioned among concurrent blocks on an SM; allocating more per block means fewer blocks fit simultaneously, potentially reducing occupancy and latency-hiding ability.
Question 6: A developer profiles a kernel and finds 80% of time is in memory transactions with low arithmetic intensity. Which optimization should be tried first?
- Restructure data access patterns to improve cache hit rate and coalescing (Correct answer)
- Increase thread block size to 1024
- Use atomic operations for memory updates
- Reduce the number of kernel arguments
Correct answer: Restructure data access patterns to improve cache hit rate and coalescing
Memory-bound kernels benefit most from improving access locality and coalescing, which increases effective bandwidth and reduces the total number of memory transactions issued.
Question 7: Which CUDA stream behavior ensures that kernels in different streams can overlap execution on the GPU?
- Kernels in separate non-default streams can run concurrently if SM resources permit (Correct answer)
- Default stream kernels automatically parallelize with all other streams
- Only streams created with cudaStreamNonBlocking flag allow overlap
- Kernel overlap requires CUDA Graph capture to be active
Correct answer: Kernels in separate non-default streams can run concurrently if SM resources permit
Non-default CUDA streams allow independent kernel launches to execute concurrently when sufficient SM resources are available and there are no data dependencies between them.
NCA Performance Optimization & Debugging 4
What is warp divergence and how does it impact GPU performance?