alternative module layout for ext_types

Author: isVoid

numba_cuda/numba/cuda/_internal/cuda_bf16.py

@@ -60,7 +60,7 @@
     uint64,
     void,
 )
-from numba.cuda.ext_types import bfloat16
+from numba.cuda.types.ext_types import bfloat16
 
 float32x2 = vector_types["float32x2"]
 __half = float16

numba_cuda/numba/cuda/cg.py

@@ -6,7 +6,7 @@
 from numba.cuda.typing import signature
 from numba.cuda import nvvmutils
 from numba.cuda.extending import intrinsic
-from numba.cuda.ext_types import grid_group, GridGroup as GridGroupClass
+from numba.cuda.types.ext_types import grid_group, GridGroup as GridGroupClass
 
 
 class GridGroup:

numba_cuda/numba/cuda/cudadecl.py

@@ -15,7 +15,7 @@
     signature,
     Registry,
 )
-from numba.cuda.ext_types import dim3
+from numba.cuda.types.ext_types import dim3
 from numba import cuda
 
 registry = Registry()

numba_cuda/numba/cuda/cudaimpl.py

@@ -19,7 +19,7 @@
 from .cudadrv import nvvm
 from numba import cuda
 from numba.cuda import nvvmutils, stubs
-from numba.cuda.ext_types import dim3, CUDADispatcher
+from numba.cuda.types.ext_types import dim3, CUDADispatcher
 
 registry = Registry("cudaimpl")
 lower = registry.lower

numba_cuda/numba/cuda/debuginfo.py

@@ -10,7 +10,7 @@
 from numba.cuda.core import config
 from numba.cuda import cgutils
 from numba.cuda.datamodel.models import ComplexModel, UnionModel, UniTupleModel
-from numba.cuda.ext_types import GridGroup
+from numba.cuda.types.ext_types import GridGroup
 
 
 @contextmanager

numba_cuda/numba/cuda/dispatcher.py

@@ -24,7 +24,8 @@
 from numba.cuda.typing.templates import fold_arguments
 from numba.cuda.typing.typeof import Purpose, typeof
 
-from numba.cuda import typing, types, ext_types
+from numba.cuda import typing, types
+from numba.cuda.types import ext_types
 from numba.cuda.api import get_current_device
 from numba.cuda.args import wrap_arg
 from numba.cuda.core.bytecode import get_code_object

numba_cuda/numba/cuda/models.py

@@ -10,7 +10,7 @@
 from numba.cuda.datamodel.models import StructModel
 from numba.cuda.extending import core_models as models
 from numba.cuda import types
-from numba.cuda.ext_types import Dim3, GridGroup, CUDADispatcher, Bfloat16
+from numba.cuda.types.ext_types import Dim3, GridGroup, CUDADispatcher, Bfloat16
 
 
 cuda_data_manager = DataModelManager()

numba_cuda/numba/cuda/printimpl.py

@@ -8,7 +8,7 @@
 from numba.cuda.core.errors import NumbaWarning
 from numba.cuda.core.imputils import Registry
 from numba.cuda import nvvmutils
-from numba.cuda.ext_types import Dim3, Bfloat16
+from numba.cuda.types.ext_types import Dim3, Bfloat16
 from warnings import warn
 
 registry = Registry("printimpl")

numba_cuda/numba/cuda/tests/cudapy/test_bfloat16.py

@@ -575,6 +575,9 @@ def kernel(out):
             _bf16_ulp_distance(raw[4:], f8_expected), 2
         )
 
+    def test_bfloat16_type_import(self):
+        self.skip_unsupported()
+
 
 def _bf16_ulp_rank(bits_int16: np.ndarray) -> np.ndarray:
     """

numba_cuda/numba/cuda/tests/cudapy/test_print.py

@@ -110,7 +110,7 @@ def print_too_many(r):
 def print_bfloat16():
     # 0.9375 is a dyadic rational, it's integer significand can expand within 7 digits.
     # printing this should not give any rounding error.
-    a = cuda.ext_types.bfloat16(0.9375)
+    a = cuda.bfloat16(0.9375)
     print(a, a, a)
 
 print_bfloat16[1, 1]()