CUDA 2D Grid Matrix Scale
Problem Statement
Scale a 2D matrix by a runtime scalar using a 2D launch grid: out[i, j] = a[i, j] * scale.
Background
cuda.grid(2) returns a pair (i, j) — the row and column this thread owns. Launch with 2D threadsperblock (e.g. (16, 16)) and a 2D blockspergrid covering both dimensions.
Your Task
Implement scale2d_kernel and run(M=64, N=48, scale=3.0) returning whether the result matches a * scale.
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:
M = 64, N = 48, scale = 3.0
True
- The input values are M=64, N=48, and scale=3.0, representing the dimensions of the 2D matrix and the scaling factor.
- A 2D launch grid is created with a suitable number of blocks and threads per block to cover the entire matrix, allowing each thread to calculate out[i,j]=a[i,j]⋅scale.
- The
scale2d_kernelfunction is launched on the GPU, performing the scaling operation on each element of the input matrix a using the given scale factor: out[i,j]=a[i,j]⋅3.0. - The resulting scaled matrix is compared to the reference solution using
np.allclose, checking if the GPU result matches the expected result within a tolerance, yielding the outputTrueif the results match.
Constraints:
- i, j = cuda.grid(2)
- Guard with if i < a.shape[0] and j < a.shape[1]
- Launch with a 2D block and 2D grid
Background Knowledge
The problem involves using CUDA, a parallel computing platform developed by NVIDIA, to scale a 2D matrix by a runtime scalar. To understand this problem, it's essential to have a basic knowledge of parallel computing, GPU architecture, and CUDA programming model. In CUDA, a kernel is a function that runs on the GPU, and it's launched with a specified number of threads. These threads are organized into a grid, which is a two-dimensional array of blocks. Each block contains a specified number of threads.
In this problem, we're using a 2D launch grid, where each thread owns a specific row and column of the matrix. The cuda.grid(2) function returns a pair (i, j) representing the row and column of the thread. This allows us to access the corresponding element in the matrix. To launch a kernel with a 2D grid, we need to specify the number of threads per block and the number of blocks per grid. The number of blocks per grid should cover both dimensions of the matrix.
The problem also involves memory management, as we need to allocate memory on the GPU, copy data from the host (CPU) to the device (GPU), and copy the result back to the host. We'll use Numba, a library that provides a high-level interface to CUDA, to simplify the process of launching kernels and managing memory. The @cuda.jit decorator is used to compile a Python function into a CUDA kernel.
Algorithm/Approach
The general approach to solve this problem involves the following steps:
- Allocate memory on the GPU for the input matrix, output matrix, and any temporary data structures.
- Copy the input matrix from the host to the device.
- Launch the scale2d_kernel with a 2D grid, where each thread scales the corresponding element in the matrix.
- Copy the result from the device back to the host.
- Compare the result with the reference solution using np.allclose.
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.