CUDA & GPU Programming test (Intermediate)

Sample questions

5 sample questions from the CUDA & GPU Programming test, with the answers. Samples are taken out of the live test, so you will not meet these when you sit it.

1. What is the most likely reason for poor performance due to misaligned memory access in a CUDA kernel?

   1. The memory access pattern is not coalesced, leading to inefficient global memory transactions.
   2. The threads are not properly synchronized using __syncthreads().
   3. The kernel is not using shared memory correctly for data reuse.
   4. The threads are diverging within a warp, causing performance degradation.

   Answer: A. The memory access pattern is not coalesced, leading to inefficient global memory transactions.

   In CUDA, memory alignment and coalescing are critical for performance. When threads in a warp access consecutive global memory addresses, the accesses are coalesced into a single transaction, reducing latency. However, if the memory access pattern is not coalesced (e.g., threads access non-consecutive addresses), multiple memory transactions are required, increasing latency and reducing performance. This is why struct-of-arrays (SoA) is preferred over array-of-structs (AoS) for memory coalescing.

2. Which type of memory in the GPU memory hierarchy is located on-chip and is faster than global memory but slower than registers?

   1. Local memory
   2. Constant memory
   3. Registers
   4. Shared memory

   Answer: D. Shared memory

   Shared memory is located on-chip, which makes it much faster than global memory, which resides in device DRAM. Registers are the fastest but are limited in number per thread. Constant memory is cached and used for read-only data, but it is slower than shared memory. Local memory is actually off-chip and is used for variables that cannot fit in registers.

3. Which scenario directly causes thread divergence in a CUDA kernel?

   1. Threads in the same block taking different branches
   2. Threads in the same grid taking different branches
   3. Threads in the same warp taking different branches
   4. Threads in different warps taking different branches

   Answer: C. Threads in the same warp taking different branches

   Thread divergence occurs when threads in the same warp take different execution paths due to conditional branches. This causes the warp to execute both paths sequentially, leading to reduced performance as some threads are inactive during parts of the execution. Divergence between threads in different warps does not incur additional costs.

4. Which of the following scenarios would result in high GPU memory bandwidth usage due to inefficient memory access patterns?

   1. The data is stored in constant memory and accessed by multiple threads, leading to efficient memory usage.
   2. The data is stored in shared memory and reused multiple times within the kernel, leading to efficient memory usage.
   3. The data is stored in local memory and accessed by a single thread, leading to efficient memory usage.
   4. The data is stored in global memory and not reused within the kernel, leading to high memory bandwidth usage due to frequent global memory accesses.

   Answer: D. The data is stored in global memory and not reused within the kernel, leading to high memory bandwidth usage due to frequent global memory accesses.

   In this scenario, the data is stored in global memory and not reused within the kernel, which leads to high memory bandwidth usage due to frequent global memory accesses. This is a common mistake because global memory has high latency and low bandwidth compared to other memory types like shared memory. If the data were reused within the kernel, it would be more efficient to load it into shared memory first, reducing the number of global memory accesses and improving performance.

5. Which of the following scenarios ensures memory coalescing in a CUDA kernel when accessing a global memory array?

   1. Each thread in the warp accesses elements with a stride of 4.
   2. Each thread accesses a random element in the array.
   3. Each thread in the warp accesses consecutive elements in the array.
   4. All threads in the warp access the same element in the array.

   Answer: C. Each thread in the warp accesses consecutive elements in the array.

   Memory coalescing occurs when consecutive threads in a warp access consecutive addresses in global memory. This allows the warp's memory transactions to be combined into a single transaction, improving memory bandwidth efficiency. If threads access non-consecutive elements (e.g., due to a stride larger than 1), memory coalescing is not achieved, and performance suffers. The other options are either incorrect or unrelated to memory coalescing.