def __init__(self, name=None):
if name is not None:
self.x = _internal_types.Data(f'{name}.x', _cuda_types.uint32)
self.y = _internal_types.Data(f'{name}.y', _cuda_types.uint32)
self.z = _internal_types.Data(f'{name}.z', _cuda_types.uint32)
self.__doc__ = f"""dim3 {name}
A namedtuple of three integers represents {name}.
Attributes:
x (uint32): {name}.x
y (uint32): {name}.y
z (uint32): {name}.z
"""
else:
# a dim3 object is created via, e.g., a CUDA API call, in which
# case both the instance name and the attributes are resolved at
# the transpiling time
pass
def z(self, code: str) -> 'Data':
from cupyx.jit import _internal_types # avoid circular import
return _internal_types.Data(f'{code}.z', uint32)
def __init__(self, name):
self.x = _internal_types.Data(f'{name}.x', _cuda_types.uint32)
self.y = _internal_types.Data(f'{name}.y', _cuda_types.uint32)
self.z = _internal_types.Data(f'{name}.z', _cuda_types.uint32)
self.__doc__ = f"""dim3 {name}
self.y = _internal_types.Data(f'{name}.y', _cuda_types.uint32)
self.z = _internal_types.Data(f'{name}.z', _cuda_types.uint32)
self.__doc__ = f"""dim3 {name}
A namedtuple of three integers represents {name}.
Attributes:
x (uint32): {name}.x
y (uint32): {name}.y
z (uint32): {name}.z
"""
threadIdx = _Dim3('threadIdx')
blockDim = _Dim3('blockDim')
blockIdx = _Dim3('blockIdx')
gridDim = _Dim3('gridDim')
warpsize = _internal_types.Data('64' if runtime.is_hip else '32',
_cuda_types.uint32)
warpsize.__doc__ = r"""Returns the number of threads in a warp.
In CUDA this is always 32, and in ROCm/HIP always 64.
.. seealso::
`numba.cuda.warpsize`_
.. _numba.cuda.warpsize:
https://numba.readthedocs.io/en/stable/cuda-reference/kernel.html#numba.cuda.warpsize
"""
warnings.warn(
'The input types of the kernel could not be inferred. '
'Please use `.cached_codes` instead.')
return next(iter(codes.values()))
def rawkernel(*, mode='cuda', device=False):
"""A decorator compiles a Python function into CUDA kernel.
"""
cupy._util.experimental('cupyx.jit.rawkernel')
def wrapper(func):
return functools.update_wrapper(
_JitRawKernel(func, mode, device), func)
return wrapper
threadIdx = _internal_types.Data('threadIdx', _cuda_types.dim3)
blockDim = _internal_types.Data('blockDim', _cuda_types.dim3)
blockIdx = _internal_types.Data('blockIdx', _cuda_types.dim3)
gridDim = _internal_types.Data('gridDim', _cuda_types.dim3)
warpsize = _internal_types.Data(
'64' if runtime.is_hip else '32', _cuda_types.uint32)
warpsize.__doc__ = r"""Returns the number of threads in a warp.
In CUDA this is always 32, and in ROCm/HIP always 64.
.. seealso:: :obj:`numba.cuda.warpsize`
"""
