Skip to content

[Bug]: Scalar kernel arguments are ignored in arithmetic #6

Description

@Perinban

Problem

Scalar kernel arguments work in comparisons and bounds checks, but their values are not correctly applied when used in arithmetic expressions.

This causes kernels to return incorrect results without raising an error.

Reproduction

import numpy as np
from numba import cuda
from metaxuda import GPUMemoryBuffer

@cuda.jit
def scale(a, out, factor, n):
    i = cuda.grid(1)

    if i < n:
        out[i] = a[i] * factor

n = 1024
factor = 2.0

a = np.arange(n, dtype=np.float32)

buf_a = GPUMemoryBuffer(arr=a)
buf_out = GPUMemoryBuffer(length=n, dtype=np.float32)

scale[32, 32](
    buf_a.dev_array,
    buf_out.dev_array,
    factor,
    n,
)

cuda.synchronize()
print(buf_out.download()[:5])

Actual behaviour

The scalar multiplication is not applied.

[0. 1. 2. 3. 4.]

No exception or warning is raised.

Expected behaviour

The scalar argument should participate normally in the arithmetic expression.

[0. 2. 4. 6. 8.]

Workaround

Store the scalar in a single-element device array and access it inside the kernel.

@cuda.jit
def scale(a, out, factor, n):
    i = cuda.grid(1)

    if i < n:
        out[i] = a[i] * factor[0]
buf_factor = GPUMemoryBuffer(
    arr=np.array([2.0], dtype=np.float32)
)

Area

Runtime

macOS version

26.5.2

Apple Silicon model

MacBook Air (Apple M1, 8 GB)

Python, Numba, and MetaXuda versions

Python 3.13.3, Numba 0.66.0, MetaXuda 2.0.1

Checks

  • I searched existing issues before opening this report.
  • I removed credentials, tokens, and personal paths from the report.

Metadata

Metadata

Assignees

Labels

bugSomething isn't working

Projects

No projects

Relationships

None yet

Development

No branches or pull requests

Issue actions