CUDA ReLU Kernel

easyFREECore Compute KernelsCUDA Basics with Numba

Problem Statement

Implement the ReLU activation out = max(x, 0) as a CUDA kernel.

Background

There's no tl.maximum here — write the comparison yourself with a conditional. ReLU clamps negatives to zero and is the most common neural-network activation.

Your Task

Implement relu_kernel and run(n=1024) comparing to np.maximum(x, 0).

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

Input

n = 1024

Output

True

Reasoning

The input n = 1024 determines the size of the input array x to be used for the ReLU activation function.
The relu_kernel function is launched on the GPU, applying the ReLU activation out = \max(x, 0) element-wise to the input array x, effectively clamping all negative values to $0$ .
The result from the GPU is compared to the reference result obtained by applying np.maximum(x, 0) to the input array x.
The run function returns True if the GPU result is close to the reference result, as verified by np.allclose(gpu_result, reference), indicating that the CUDA kernel implementation of the ReLU activation function is correct.

Constraints

out[i] = x[i] if x[i] > 0 else 0.0
Bounds-check the global index

More from CUDA Basics with Numba

CUDA Dot Producthard CUDA Atomic Max Reductionmedium CUDA Two-Stage Reduction with Atomicshard CUDA Atomic Histogrammedium CUDA Atomic Sum Reductionmedium

Concept primer

Background Knowledge

The ReLU (Rectified Linear Unit) activation function is a widely used activation function in neural networks. It maps all negative values to zero and all positive values to the same value, i.e., $out = \text{max}(x, 0)$ . This function is essential in deep learning as it introduces non-linearity into the model, allowing it to learn more complex relationships between inputs and outputs. The ReLU function can be mathematically represented as:

\text{ReLU}(x) = \begin{cases} 0 & \text{if } x < 0 \\ x & \text{if } x \geq 0 \end{cases}

In the context of CUDA (Compute Unified Device Architecture), we are dealing with a parallel computing platform and programming model developed by NVIDIA. CUDA allows developers to harness the power of the GPU (Graphics Processing Unit) to perform general-purpose computing tasks, including scientific simulations, data analysis, and machine learning. To implement the ReLU activation function as a CUDA kernel, we need to understand how to write efficient parallel code that can be executed on the GPU.

The Numba library is a just-in-time compiler that translates Python and NumPy code into fast machine code. It provides a @cuda.jit decorator that allows us to define CUDA kernels, which are functions that can be executed on the GPU. To solve this problem, we need to use Numba's CUDA support to define a kernel function that applies the ReLU activation function to an input array.

Algorithm/Approach

The general approach to solving this problem involves defining a CUDA kernel function that applies the ReLU activation function to each element of the input array. This can be achieved by using a conditional statement to compare each element with zero and assign the maximum value to the corresponding output element. The kernel function will be launched on the GPU, where it will be executed in parallel across multiple threads.

Step-by-Step Strategy

To implement the solution, follow these steps:

Import the necessary libraries, including numba and numpy.
Define the relu_kernel function using the @cuda.jit decorator, which will apply the ReLU activation function to each element of the input array.
Allocate the input and output arrays on the host (CPU) and copy the input array to the device (GPU).
Launch the relu_kernel function on the GPU, specifying the number of threads and blocks.
Copy the output array from the device (GPU) back to the host (CPU).
Compare the result with the reference solution using np.maximum(x, 0) and return a boolean value indicating whether the results are close.

Common Pitfalls

When implementing the solution, watch out for the following common pitfalls:

Incorrectly defining the kernel function or launching it with the wrong number of threads and blocks.
Failing to copy data between the host and device or using incorrect memory management.
Not handling errors or exceptions properly.

Time & Space Complexity

The time complexity of the ReLU activation function is $O(n)$ , where $n$ is the number of elements in the input array, since we need to apply the function to each element. The space complexity is also $O(n)$ , as we need to store the output array. However, since we are using a CUDA kernel, the actual time complexity will depend on the number of threads and blocks used, as well as the memory bandwidth and computational resources of the GPU.

solution.py

Python 3.13

GPU · T4

Test results