class Cuda_blockDim(AttributeTemplate): key = types.Module(cuda.blockDim) def resolve_x(self, mod): return types.Macro(Cuda_blockDim_x) def resolve_y(self, mod): return types.Macro(Cuda_blockDim_y) def resolve_z(self, mod): return types.Macro(Cuda_blockDim_z)
class Cuda_threadIdx(AttributeTemplate): key = types.Module(cuda.threadIdx) def resolve_x(self, mod): return types.Macro(Cuda_threadIdx_x) def resolve_y(self, mod): return types.Macro(Cuda_threadIdx_y) def resolve_z(self, mod): return types.Macro(Cuda_threadIdx_z)
class CudaAtomicTemplate(AttributeTemplate): key = types.Module(cuda.atomic) def resolve_add(self, mod): return types.Function(Cuda_atomic_add) def resolve_max(self, mod): return types.Function(Cuda_atomic_max) def resolve_min(self, mod): return types.Function(Cuda_atomic_min) def resolve_compare_and_swap(self, mod): return types.Function(Cuda_atomic_compare_and_swap)
class HsaModuleTemplate(AttributeTemplate): key = types.Module(roc) def resolve_get_global_id(self, mod): return types.Function(Hsa_get_global_id) def resolve_get_local_id(self, mod): return types.Function(Hsa_get_local_id) def resolve_get_global_size(self, mod): return types.Function(Hsa_get_global_size) def resolve_get_local_size(self, mod): return types.Function(Hsa_get_local_size) def resolve_get_num_groups(self, mod): return types.Function(Hsa_get_num_groups) def resolve_get_work_dim(self, mod): return types.Function(Hsa_get_work_dim) def resolve_get_group_id(self, mod): return types.Function(Hsa_get_group_id) def resolve_barrier(self, mod): return types.Function(Hsa_barrier) def resolve_mem_fence(self, mod): return types.Function(Hsa_mem_fence) def resolve_wavebarrier(self, mod): return types.Function(Hsa_wavebarrier) def resolve_activelanepermute_wavewidth(self, mod): return types.Function(Hsa_activelanepermute_wavewidth) def resolve_ds_permute(self, mod): return types.Function(Hsa_ds_permute) def resolve_ds_bpermute(self, mod): return types.Function(Hsa_ds_bpermute) def resolve_shared(self, mod): return types.Module(roc.shared) def resolve_atomic(self, mod): return types.Module(roc.atomic)
class CudaModuleTemplate(AttributeTemplate): key = types.Module(cuda) def resolve_grid(self, mod): return types.Macro(Cuda_grid) def resolve_gridsize(self, mod): return types.Macro(Cuda_gridsize) def resolve_threadIdx(self, mod): return types.Module(cuda.threadIdx) def resolve_blockIdx(self, mod): return types.Module(cuda.blockIdx) def resolve_blockDim(self, mod): return types.Module(cuda.blockDim) def resolve_gridDim(self, mod): return types.Module(cuda.gridDim) def resolve_shared(self, mod): return types.Module(cuda.shared) def resolve_syncthreads(self, mod): return types.Function(Cuda_syncthreads) def resolve_threadfence(self, mod): return types.Function(Cuda_threadfence_device) def resolve_threadfence_block(self, mod): return types.Function(Cuda_threadfence_block) def resolve_threadfence_system(self, mod): return types.Function(Cuda_threadfence_system) def resolve_selp(self, mod): return types.Function(Cuda_selp) def resolve_atomic(self, mod): return types.Module(cuda.atomic) def resolve_const(self, mod): return types.Module(cuda.const) def resolve_local(self, mod): return types.Module(cuda.local)
class NumpyModuleAttribute(AttributeTemplate): # note: many unary ufuncs are added later on, using setattr key = types.Module(numpy) def resolve_arctan2(self, mod): return types.Function(Numpy_arctan2) def resolve_add(self, mod): return types.Function(Numpy_add) def resolve_subtract(self, mod): return types.Function(Numpy_subtract) def resolve_multiply(self, mod): return types.Function(Numpy_multiply) def resolve_divide(self, mod): return types.Function(Numpy_divide)
class OclModuleTemplate(AttributeTemplate): key = types.Module(dppy) def resolve_get_global_id(self, mod): return types.Function(Ocl_get_global_id) def resolve_get_local_id(self, mod): return types.Function(Ocl_get_local_id) def resolve_get_global_size(self, mod): return types.Function(Ocl_get_global_size) def resolve_get_local_size(self, mod): return types.Function(Ocl_get_local_size) def resolve_get_num_groups(self, mod): return types.Function(Ocl_get_num_groups) def resolve_get_work_dim(self, mod): return types.Function(Ocl_get_work_dim) def resolve_get_group_id(self, mod): return types.Function(Ocl_get_group_id) def resolve_barrier(self, mod): return types.Function(Ocl_barrier) def resolve_mem_fence(self, mod): return types.Function(Ocl_mem_fence) def resolve_sub_group_barrier(self, mod): return types.Function(Ocl_sub_group_barrier) def resolve_atomic(self, mod): return types.Module(dppy.atomic) def resolve_local(self, mod): return types.Module(dppy.local) def resolve_private(self, mod): return types.Module(dppy.private)
class NumbaTypesModuleAttribute(AttributeTemplate): key = types.Module(types) def resolve_int8(self, mod): return types.Function(ToInt8) def resolve_int16(self, mod): return types.Function(ToInt16) def resolve_int32(self, mod): return types.Function(ToInt32) def resolve_int64(self, mod): return types.Function(ToInt64) def resolve_uint8(self, mod): return types.Function(ToUint8) def resolve_uint16(self, mod): return types.Function(ToUint16) def resolve_uint32(self, mod): return types.Function(ToUint32) def resolve_uint64(self, mod): return types.Function(ToUint64) def resolve_float32(self, mod): return types.Function(ToFloat32) def resolve_float64(self, mod): return types.Function(ToFloat64) def resolve_complex64(self, mod): return types.Function(ToComplex64) def resolve_complex128(self, mod): return types.Function(ToComplex128)
def resolve_local(self, mod): return types.Module(cuda.local)
def resolve_gridDim(self, mod): return types.Module(cuda.gridDim)
def resolve_const(self, mod): return types.Module(cuda.const)
def resolve_threadIdx(self, mod): return types.Module(cuda.threadIdx)
def resolve_blockDim(self, mod): return types.Module(cuda.blockDim)
def resolve_local(self, mod): return types.Module(dppy.local)
class OclPrivateTemplate(AttributeTemplate): key = types.Module(dppy.private) def resolve_array(self, mod): return types.Function(OCL_private_array)
def resolve_shared(self, mod): return types.Module(hsa.shared)
class HsaSharedTemplate(AttributeTemplate): key = types.Module(hsa.shared) def resolve_array(self, mod): return types.Macro(Hsa_shared_array)
return types.Function(Hsa_get_num_groups) def resolve_get_work_dim(self, mod): return types.Function(Hsa_get_work_dim) def resolve_get_group_id(self, mod): return types.Function(Hsa_get_group_id) def resolve_barrier(self, mod): return types.Function(Hsa_barrier) def resolve_mem_fence(self, mod): return types.Function(Hsa_mem_fence) def resolve_wavebarrier(self, mod): return types.Function(Hsa_wavebarrier) def resolve_activelanepermute_wavewidth(self, mod): return types.Function(Hsa_activelanepermute_wavewidth) def resolve_shared(self, mod): return types.Module(hsa.shared) def resolve_atomic(self, mod): return types.Module(hsa.atomic) # intrinsic intrinsic_global(hsa, types.Module(hsa))
signature(types.boolean, types.float32), signature(types.boolean, types.float64), ] class Math_isinf(ConcreteTemplate): key = math.isinf cases = [ signature(types.boolean, types.int64), signature(types.boolean, types.uint64), signature(types.boolean, types.float32), signature(types.boolean, types.float64), ] builtin_global(math, types.Module(math)) builtin_global(math.fabs, types.Function(Math_fabs)) builtin_global(math.exp, types.Function(Math_exp)) if utils.PYVERSION > (2, 6): builtin_global(math.expm1, types.Function(Math_expm1)) builtin_global(math.sqrt, types.Function(Math_sqrt)) builtin_global(math.log, types.Function(Math_log)) builtin_global(math.log1p, types.Function(Math_log1p)) builtin_global(math.log10, types.Function(Math_log10)) builtin_global(math.sin, types.Function(Math_sin)) builtin_global(math.cos, types.Function(Math_cos)) builtin_global(math.tan, types.Function(Math_tan)) builtin_global(math.sinh, types.Function(Math_sinh)) builtin_global(math.cosh, types.Function(Math_cosh)) builtin_global(math.tanh, types.Function(Math_tanh)) builtin_global(math.asin, types.Function(Math_asin))
key = types.float32 class ToFloat64(Caster): key = types.float64 class ToComplex64(Caster): key = types.complex64 class ToComplex128(Caster): key = types.complex128 builtin_global(types, types.Module(types)) builtin_global(types.int8, types.Function(ToInt8)) builtin_global(types.int16, types.Function(ToInt16)) builtin_global(types.int32, types.Function(ToInt32)) builtin_global(types.int64, types.Function(ToInt64)) builtin_global(types.uint8, types.Function(ToUint8)) builtin_global(types.uint16, types.Function(ToUint16)) builtin_global(types.uint32, types.Function(ToUint32)) builtin_global(types.uint64, types.Function(ToUint64)) builtin_global(types.float32, types.Function(ToFloat32)) builtin_global(types.float64, types.Function(ToFloat64)) builtin_global(types.complex64, types.Function(ToComplex64)) builtin_global(types.complex128, types.Function(ToComplex128)) #------------------------------------------------------------------------------
@register_attribute class StringModuleAttr(AttributeTemplate): key = ntypes.Module(string) def resolve_capitalize(self, mod): return ntypes.Function(String_capitalize) def resolve_split(self, mod): return ntypes.Function(String_split) class String_capitalize(ConcreteTemplate): key = string.capitalize cases = [signature(StringVal, StringVal)] class String_split(ConcreteTemplate): key = string.split cases = [ signature(ntypes.Array(StringVal, 1, 'C'), StringVal), signature(ntypes.Array(StringVal, 1, 'C'), StringVal, StringVal), signature(ntypes.Array(StringVal, 1, 'C'), StringVal, StringVal, IntVal) ] register_global(string, ntypes.Module(string)) register_global(string.capitalize, ntypes.Function(String_capitalize)) register_global(string.split, ntypes.Function(String_split))
def resolve_private(self, mod): return types.Module(dppy.private)
class MathModuleAttribute(AttributeTemplate): key = types.Module(math) def resolve_fabs(self, mod): return types.Function(Math_fabs) def resolve_exp(self, mod): return types.Function(Math_exp) def resolve_expm1(self, mod): return types.Function(Math_expm1) def resolve_sqrt(self, mod): return types.Function(Math_sqrt) def resolve_log(self, mod): return types.Function(Math_log) def resolve_log1p(self, mod): return types.Function(Math_log1p) def resolve_log10(self, mod): return types.Function(Math_log10) def resolve_sin(self, mod): return types.Function(Math_sin) def resolve_cos(self, mod): return types.Function(Math_cos) def resolve_tan(self, mod): return types.Function(Math_tan) def resolve_sinh(self, mod): return types.Function(Math_sinh) def resolve_cosh(self, mod): return types.Function(Math_cosh) def resolve_tanh(self, mod): return types.Function(Math_tanh) def resolve_asin(self, mod): return types.Function(Math_asin) def resolve_acos(self, mod): return types.Function(Math_acos) def resolve_atan(self, mod): return types.Function(Math_atan) def resolve_atan2(self, mod): return types.Function(Math_atan2) def resolve_asinh(self, mod): return types.Function(Math_asinh) def resolve_acosh(self, mod): return types.Function(Math_acosh) def resolve_atanh(self, mod): return types.Function(Math_atanh) def resolve_pi(self, mod): return types.float64 def resolve_e(self, mod): return types.float64 def resolve_floor(self, mod): return types.Function(Math_floor) def resolve_ceil(self, mod): return types.Function(Math_ceil) def resolve_trunc(self, mod): return types.Function(Math_trunc) def resolve_isnan(self, mod): return types.Function(Math_isnan) def resolve_isinf(self, mod): return types.Function(Math_isinf) def resolve_degrees(self, mod): return types.Function(Math_degrees) def resolve_radians(self, mod): return types.Function(Math_radians) # def resolve_hypot(self, mod): # return types.Function(Math_hypot) def resolve_copysign(self, mod): return types.Function(Math_copysign) def resolve_fmod(self, mod): return types.Function(Math_fmod) def resolve_pow(self, mod): return types.Function(Math_pow) def resolve_erf(self, mod): return types.Function(Math_erf) def resolve_erfc(self, mod): return types.Function(Math_erfc) def resolve_gamma(self, mod): return types.Function(Math_gamma) def resolve_lgamma(self, mod): return types.Function(Math_lgamma)
def resolve_threadfence_block(self, mod): return types.Function(Cuda_threadfence_block) def resolve_threadfence_system(self, mod): return types.Function(Cuda_threadfence_system) def resolve_selp(self, mod): return types.Function(Cuda_selp) def resolve_atomic(self, mod): return types.Module(cuda.atomic) def resolve_const(self, mod): return types.Module(cuda.const) def resolve_local(self, mod): return types.Module(cuda.local) intrinsic_global(cuda, types.Module(cuda)) ## Forces the use of the cuda namespace by not recognizing individual the ## following as globals. # intrinsic_global(cuda.grid, types.Function(Cuda_grid)) # intrinsic_global(cuda.gridsize, types.Function(Cuda_gridsize)) # intrinsic_global(cuda.threadIdx, types.Module(cuda.threadIdx)) # intrinsic_global(cuda.shared, types.Module(cuda.shared)) # intrinsic_global(cuda.shared.array, types.Function(Cuda_shared_array)) # intrinsic_global(cuda.syncthreads, types.Function(Cuda_syncthreads)) # intrinsic_global(cuda.atomic, types.Module(cuda.atomic))
class HsaAtomicTemplate(AttributeTemplate): key = types.Module(hsa.atomic) def resolve_add(self, mod): return types.Function(Hsa_atomic_add)
class CudaSharedModuleTemplate(AttributeTemplate): key = types.Module(cuda.shared) def resolve_array(self, mod): return types.Macro(Cuda_shared_array)
def resolve_atomic(self, mod): return types.Module(hsa.atomic)
class CudaConstModuleTemplate(AttributeTemplate): key = types.Module(cuda.const) def resolve_array_like(self, mod): return types.Macro(Cuda_const_arraylike)
def _get_global_type(self, gv): ty = self._lookup_global(gv) if ty is not None: return ty if isinstance(gv, pytypes.ModuleType): return types.Module(gv)
class CudaLocalModuleTemplate(AttributeTemplate): key = types.Module(cuda.local) def resolve_array(self, mod): return types.Macro(Cuda_local_array)