1 / 1

Hadamard Product as a CUDA Kernel

Linear AlgebraGPU Programming

Easy

Implement a CUDA kernel that computes the elementwise (Hadamard) product of two matrices $A$ and $B$ of shape $(M, N)$ , writing the result to $C$ . For each valid index $(i, j)$ :

C[i, j] = A[i, j] \cdot B[i, j]

Both inputs and the output share the same shape $(M, N)$ . All three matrices are stored as row-major flat arrays of length $M \cdot N$ , so element $(i, j)$ lives at linear index $i \cdot N + j$ .

Loading visualization...

Examples

Input: M = 2, N = 2, A = [[1, 2], [3, 4]], B = [[5, 6], [7, 8]]

Output: [[5.0, 12.0], [21.0, 32.0]]

Input: M = 2, N = 3, A = [[1, 2, 3], [4, 5, 6]], B = [[1, 0, 1], [0, 1, 0]]

Output: [[1.0, 0.0, 3.0], [0.0, 5.0, 0.0]]

Hint 1

Map $\texttt{threadIdx.y}$ and $\texttt{blockIdx.y}$ to the row $i$ , and $\texttt{threadIdx.x}$ and $\texttt{blockIdx.x}$ to the column $j$ , so each thread owns one $(i, j)$ .

Hint 2

Each output element depends only on the matching pair $A[i, j]$ and $B[i, j]$ , so there is no reduction and no need for shared memory.

Hint 3

Because the grid is rounded up to cover any shape, some threads fall outside the matrix. An early-return bounds check on $i$ and $j$ keeps those threads from writing past the buffer.

Requirements

Launch one thread per output element across a 2D grid
Respect row-major layout when computing the linear index
Guard against threads where $i \ge M$ or $j \ge N$
The $\texttt{solve}$ signature must not change

Constraints

$1 \le M, N \le 4096$
Input and output buffers are device pointers
Use $16 \times 16$ thread blocks
Values fit in 32-bit float

Try Similar Problems

Outer Product as a CUDA KernelEasy Matrix Addition as a CUDA KernelEasy Matrix Transpose as a CUDA KernelEasy RMS Normalization as a CUDA KernelMedium Layer Normalization as a CUDA KernelMedium Matrix Multiplication as a CUDA KernelMedium Matrix-Vector Multiplication as a CUDA KernelMedium Vector Addition as a CUDA KernelEasy Vector Subtraction as a CUDA KernelEasy L1 Normalization as a CUDA KernelMedium Cosine Similarity as a CUDA KernelMedium CUDA Dot Product: Multi-Block ReductionMedium

Hadamard Product as a CUDA Kernel

Linear AlgebraGPU Programming

Easy

Implement a CUDA kernel that computes the elementwise (Hadamard) product of two matrices $A$ and $B$ of shape $(M, N)$ , writing the result to $C$ . For each valid index $(i, j)$ :

C[i, j] = A[i, j] \cdot B[i, j]

Both inputs and the output share the same shape $(M, N)$ . All three matrices are stored as row-major flat arrays of length $M \cdot N$ , so element $(i, j)$ lives at linear index $i \cdot N + j$ .

Loading visualization...

Examples

Input: M = 2, N = 2, A = [[1, 2], [3, 4]], B = [[5, 6], [7, 8]]

Output: [[5.0, 12.0], [21.0, 32.0]]

Input: M = 2, N = 3, A = [[1, 2, 3], [4, 5, 6]], B = [[1, 0, 1], [0, 1, 0]]

Output: [[1.0, 0.0, 3.0], [0.0, 5.0, 0.0]]

Hint 1

Map $\texttt{threadIdx.y}$ and $\texttt{blockIdx.y}$ to the row $i$ , and $\texttt{threadIdx.x}$ and $\texttt{blockIdx.x}$ to the column $j$ , so each thread owns one $(i, j)$ .

Hint 2

Each output element depends only on the matching pair $A[i, j]$ and $B[i, j]$ , so there is no reduction and no need for shared memory.

Hint 3

Because the grid is rounded up to cover any shape, some threads fall outside the matrix. An early-return bounds check on $i$ and $j$ keeps those threads from writing past the buffer.

Requirements

Launch one thread per output element across a 2D grid
Respect row-major layout when computing the linear index
Guard against threads where $i \ge M$ or $j \ge N$
The $\texttt{solve}$ signature must not change

Constraints

$1 \le M, N \le 4096$
Input and output buffers are device pointers
Use $16 \times 16$ thread blocks
Values fit in 32-bit float

Try Similar Problems

A =

Accepts: array

B =

Accepts: array

Examples

Requirements

Constraints

Kernel performance

Hadamard Product as a CUDA Kernel

Examples

Requirements

Constraints

Kernel performance