CUDA Device-to-Device Copy

easyPROMemory & Data MovementCUDA Basics with Numba

Problem Statement

Copy one device array into another with a kernel: dst = src. Both buffers already live on the GPU — no host involvement during the copy.

Background

Allocate the source with cuda.to_device and an empty destination with cuda.device_array. The kernel reads src and writes dst, both device pointers.

Your Task

Implement copy_kernel and run(n=1024) returning whether dst exactly equals src.

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 value n = 1024 determines the size of the source and destination arrays.
The copy_kernel function is launched, which reads from the source array src and writes to the destination array dst on the GPU, effectively performing the operation dst = src.
The run function allocates the source and destination arrays, launches the kernel, and then compares the resulting dst array with the original src array using np.allclose.
The comparison checks for exact equality between dst and src, and since the kernel correctly copies the source array to the destination array, the result is True.

Constraints

src is a device array, dst a separate device array
dst[i] = src[i]
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 primerPREVIEW

Background Knowledge

The problem involves using CUDA, a parallel computing platform developed by NVIDIA, to perform a device-to-device copy of an array. This requires understanding of GPU programming and how to manage memory on the GPU. In CUDA, memory is divided into several types, including global memory, shared memory, and register memory. For this problem, we will be working with global memory, which is the largest memory space on the GPU and is used to store data that is accessed by all threads.

To copy one device array into another, we need to launch a kernel, which is a small program that runs on the GPU. The kernel will read from the source array (src) and write to the destination array (dst). This requires understanding of how to use CUDA threads to parallelize the copy operation. Each thread will be responsible for copying a portion of the array, and we need to ensure that the threads are properly synchronized to avoid any data corruption.

The problem also involves using Numba, a Python library that provides a high-performance interface to CUDA. We will use Numba's @cuda.jit decorator to define the kernel function, and cuda.to_device to allocate the source array on the GPU. We will also use cuda.device_array to allocate the destination array on the GPU. Understanding how to use these functions and decorators is crucial to solving the problem.

Algorithm/Approach

The general approach to solving this problem involves the following steps:

Allocate the source array on the GPU using cuda.to_device.
Allocate the destination array on the GPU using cuda.device_array.
Define a kernel function using @cuda.jit that reads from the source array and writes to the destination array.
Launch the kernel function using the cuda.grid function to specify the number of threads and blocks.
Use thread indexing to determine which element of the array each thread should copy.

PRO

Continue the full explanation

You're reading the free preview. Unlock the complete walkthrough, the code editor, test runner and reference solution with Premium.

PRO

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.

solution.py

Python 3.13

GPU · T4

Test results