CUDA Grid-Stride Loop
Problem Statement
Copy a 1D array using a grid-stride loop so the kernel is correct even when there are fewer total threads than elements: out = x.
Background
A grid-stride loop lets each thread process multiple elements, stepping by the total number of threads (gridDim.x * blockDim.x). This decouples the launch configuration from the data size — the same kernel works for any n.
Your Task
Implement copy_stride_kernel and run(n=4096) (launched with deliberately few blocks) that returns whether the copy is exact.
How it is tested
Your solution must define a top-level function run(...) that allocates the inputs, copies them to the GPU, launches your @cuda.jit kernel, and returns a Python bool from np.allclose(gpu_result, reference). The grader prints run(...); the expected output is True.
Example:
n = 4096 (launched with 8 blocks x 128 threads)
True
- The input size
n = 4096is divided into chunks processed by each thread in a grid-stride loop, with a total of 8⋅128=1024 threads. - Each thread processes multiple elements, stepping by the total number of threads (1024), to ensure all 4096 elements are copied.
- The
copy_stride_kernelfunction copies the input arrayxto the output arrayoutusing the grid-stride loop, resulting in an exact copy of the input array. - The
runfunction compares the GPU result with the reference array usingnp.allclose, returningTrueif the copy is exact, which is the case for the given input.
Constraints:
- start = cuda.grid(1); stride = cuda.gridDim.x * cuda.blockDim.x
- Loop i from start to n stepping by stride
- Must work when total threads < n
Background Knowledge
The CUDA Execution Model is a key concept in parallel computing using NVIDIA GPUs. It defines how threads are organized and executed on the GPU. In CUDA, threads are grouped into blocks, and blocks are grouped into a grid. Each thread has a unique identifier, threadIdx.x, which can be used to identify the thread within a block. The number of threads in a block is defined by blockDim.x, and the number of blocks in a grid is defined by gridDim.x. Understanding these concepts is crucial for writing efficient CUDA kernels.
A grid-stride loop is a technique used in CUDA programming to decouple the launch configuration from the data size. This means that the same kernel can be used to process arrays of different sizes, without having to adjust the launch configuration. In a grid-stride loop, each thread processes multiple elements, stepping by the total number of threads (gridDim.x * blockDim.x). This allows the kernel to handle arrays of any size, even if there are fewer threads than elements.
The CUDA Basics collection provides a foundation for understanding CUDA programming. It covers topics such as memory management, thread synchronization, and kernel optimization. In the context of this problem, understanding how to launch a kernel, manage memory, and use a grid-stride loop are essential skills. The convolution and gradient descent concepts are not directly applicable to this problem, but understanding parallel computing and thread management are crucial.
Algorithm/Approach
The general approach to solving this problem involves using a grid-stride loop to copy a 1D array. The kernel will be launched with a deliberate few blocks, and each thread will process multiple elements using the grid-stride loop technique. The algorithm can be broken down into the following steps:
- Allocate memory for the input and output arrays on the GPU
- Copy the input array to the GPU
- Launch the kernel with the desired launch configuration
- Use a grid-stride loop to copy the input array to the output array
- Copy the output array back to the host and compare it with the reference array
Continue the full explanation
You're reading the free preview. Unlock the complete walkthrough, the code editor, test runner and reference solution with Premium.
Editor locked
The code editor is locked for Pro problems. It is only available for free problems. Please upgrade to gain access to the code editor for all problems.