From 6b8ae2bd34a341cab93253cedcfba87287b7a868 Mon Sep 17 00:00:00 2001
From: Michael Droettboom <mdboom@gmail.com>
Date: Mon, 25 Aug 2025 11:35:13 -0400
Subject: [PATCH 1/4] Move param_packer.cpp -> param_packer.h

---
 .../cuda/bindings/_lib/param_packer.cpp       | 159 -----------------
 .../cuda/bindings/_lib/param_packer.h         | 163 +++++++++++++++++-
 2 files changed, 155 insertions(+), 167 deletions(-)
 delete mode 100644 cuda_bindings/cuda/bindings/_lib/param_packer.cpp

diff --git a/cuda_bindings/cuda/bindings/_lib/param_packer.cpp b/cuda_bindings/cuda/bindings/_lib/param_packer.cpp
deleted file mode 100644
index 6f99c29e3..000000000
--- a/cuda_bindings/cuda/bindings/_lib/param_packer.cpp
+++ /dev/null
@@ -1,159 +0,0 @@
-// SPDX-FileCopyrightText: Copyright (c) 2021-2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
-// SPDX-License-Identifier: LicenseRef-NVIDIA-SOFTWARE-LICENSE
-
-#include <Python.h>
-#include "param_packer.h"
-
-#include <map>
-#include <functional>
-#include <stdexcept>
-#include <string>
-
-PyObject* enum_module = nullptr;
-PyTypeObject* enum_Enum = nullptr;
-
-PyObject* ctypes_module = nullptr;
-PyObject* ctypes_addressof = nullptr;
-PyObject* addressof_param_tuple = nullptr;
-
-PyTypeObject* ctypes_c_char = nullptr;
-PyTypeObject* ctypes_c_bool = nullptr;
-PyTypeObject* ctypes_c_wchar = nullptr;
-PyTypeObject* ctypes_c_byte = nullptr;
-PyTypeObject* ctypes_c_ubyte = nullptr;
-PyTypeObject* ctypes_c_short = nullptr;
-PyTypeObject* ctypes_c_ushort = nullptr;
-PyTypeObject* ctypes_c_int = nullptr;
-PyTypeObject* ctypes_c_uint = nullptr;
-PyTypeObject* ctypes_c_long = nullptr;
-PyTypeObject* ctypes_c_ulong = nullptr;
-PyTypeObject* ctypes_c_longlong = nullptr;
-PyTypeObject* ctypes_c_ulonglong = nullptr;
-PyTypeObject* ctypes_c_size_t = nullptr;
-PyTypeObject* ctypes_c_float = nullptr;
-PyTypeObject* ctypes_c_double = nullptr;
-PyTypeObject* ctypes_c_void_p = nullptr;
-
-PyTypeObject* ctypes_c_ssize_t = nullptr;
-PyTypeObject* ctypes_c_longdouble = nullptr;
-PyTypeObject* ctypes_c_char_p = nullptr;
-PyTypeObject* ctypes_c_wchar_p = nullptr;
-PyTypeObject* ctypes_c_structure = nullptr;
-
-void fetch_ctypes()
-{
-    ctypes_module = PyImport_ImportModule("ctypes");
-    if (ctypes_module == nullptr)
-        throw std::runtime_error("Cannot import ctypes module");
-    // get method addressof
-    PyObject* ctypes_dict = PyModule_GetDict(ctypes_module);
-    if (ctypes_dict == nullptr)
-        throw std::runtime_error(std::string("FAILURE @ ") + std::string(__FILE__) + " : " + std::to_string(__LINE__));
-    // supportedtypes
-    ctypes_c_int = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_int");
-    ctypes_c_char = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_char");
-    ctypes_c_bool = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_bool");
-    ctypes_c_wchar = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_wchar");
-    ctypes_c_byte = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_byte");
-    ctypes_c_ubyte = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ubyte");
-    ctypes_c_short = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_short");
-    ctypes_c_ushort = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ushort");
-    ctypes_c_int = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_int");
-    ctypes_c_uint = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_uint");
-    ctypes_c_long = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_long");
-    ctypes_c_ulong = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ulong");
-    ctypes_c_longlong = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_longlong");
-    ctypes_c_ulonglong = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ulonglong");
-    ctypes_c_size_t = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_size_t");
-    ctypes_c_float = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_float");
-    ctypes_c_double = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_double");
-    ctypes_c_void_p = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_void_p"); // == c_voidp
-}
-
-
-// (target type, source type)
-std::map<std::pair<PyTypeObject*,PyTypeObject*>, std::function<int(void*, PyObject*)>> m_feeders;
-
-void populate_feeders(PyTypeObject* target_t, PyTypeObject* source_t)
-{
-    if (target_t == ctypes_c_int)
-    {
-        if (source_t == &PyLong_Type)
-        {
-            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
-            {
-                *((int*)ptr) = (int)PyLong_AsLong(value);
-                return sizeof(int);
-            };
-            return;
-        }
-    } else if (target_t == ctypes_c_bool) {
-        if (source_t == &PyBool_Type)
-        {
-            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
-            {
-                *((bool*)ptr) = (value == Py_True);
-                return sizeof(bool);
-            };
-            return;
-        }
-    } else if (target_t == ctypes_c_byte) {
-        if (source_t == &PyLong_Type)
-        {
-            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
-            {
-                *((int8_t*)ptr) = (int8_t)PyLong_AsLong(value);
-                return sizeof(int8_t);
-            };
-            return;
-        }
-    } else if (target_t == ctypes_c_double) {
-        if (source_t == &PyFloat_Type)
-        {
-            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
-            {
-                *((double*)ptr) = (double)PyFloat_AsDouble(value);
-                return sizeof(double);
-            };
-            return;
-        }
-    } else if (target_t == ctypes_c_float) {
-        if (source_t == &PyFloat_Type)
-        {
-            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
-            {
-                *((float*)ptr) = (float)PyFloat_AsDouble(value);
-                return sizeof(float);
-            };
-            return;
-        }
-    } else if (target_t == ctypes_c_longlong) {
-        if (source_t == &PyLong_Type)
-        {
-            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
-            {
-                *((long long*)ptr) = (long long)PyLong_AsLongLong(value);
-                return sizeof(long long);
-            };
-            return;
-        }
-    }
-}
-
-int feed(void* ptr, PyObject* value, PyObject* type)
-{
-    PyTypeObject* pto = (PyTypeObject*)type;
-    if (ctypes_c_int == nullptr)
-        fetch_ctypes();
-    auto found = m_feeders.find({pto,value->ob_type});
-    if (found == m_feeders.end())
-    {
-        populate_feeders(pto, value->ob_type);
-        found = m_feeders.find({pto,value->ob_type});
-    }
-    if (found != m_feeders.end())
-    {
-        return found->second(ptr, value);
-    }
-    return 0;
-}
diff --git a/cuda_bindings/cuda/bindings/_lib/param_packer.h b/cuda_bindings/cuda/bindings/_lib/param_packer.h
index c69f47498..6f99c29e3 100644
--- a/cuda_bindings/cuda/bindings/_lib/param_packer.h
+++ b/cuda_bindings/cuda/bindings/_lib/param_packer.h
@@ -1,12 +1,159 @@
 // SPDX-FileCopyrightText: Copyright (c) 2021-2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
 // SPDX-License-Identifier: LicenseRef-NVIDIA-SOFTWARE-LICENSE
-//
-// Please refer to the NVIDIA end user license agreement (EULA) associated
-// with this source code for terms and conditions that govern your use of
-// this software. Any use, reproduction, disclosure, or distribution of
-// this software and related documentation outside the terms of the EULA
-// is strictly prohibited.
-#pragma once
+
 #include <Python.h>
+#include "param_packer.h"
+
+#include <map>
+#include <functional>
+#include <stdexcept>
+#include <string>
+
+PyObject* enum_module = nullptr;
+PyTypeObject* enum_Enum = nullptr;
+
+PyObject* ctypes_module = nullptr;
+PyObject* ctypes_addressof = nullptr;
+PyObject* addressof_param_tuple = nullptr;
+
+PyTypeObject* ctypes_c_char = nullptr;
+PyTypeObject* ctypes_c_bool = nullptr;
+PyTypeObject* ctypes_c_wchar = nullptr;
+PyTypeObject* ctypes_c_byte = nullptr;
+PyTypeObject* ctypes_c_ubyte = nullptr;
+PyTypeObject* ctypes_c_short = nullptr;
+PyTypeObject* ctypes_c_ushort = nullptr;
+PyTypeObject* ctypes_c_int = nullptr;
+PyTypeObject* ctypes_c_uint = nullptr;
+PyTypeObject* ctypes_c_long = nullptr;
+PyTypeObject* ctypes_c_ulong = nullptr;
+PyTypeObject* ctypes_c_longlong = nullptr;
+PyTypeObject* ctypes_c_ulonglong = nullptr;
+PyTypeObject* ctypes_c_size_t = nullptr;
+PyTypeObject* ctypes_c_float = nullptr;
+PyTypeObject* ctypes_c_double = nullptr;
+PyTypeObject* ctypes_c_void_p = nullptr;
+
+PyTypeObject* ctypes_c_ssize_t = nullptr;
+PyTypeObject* ctypes_c_longdouble = nullptr;
+PyTypeObject* ctypes_c_char_p = nullptr;
+PyTypeObject* ctypes_c_wchar_p = nullptr;
+PyTypeObject* ctypes_c_structure = nullptr;
+
+void fetch_ctypes()
+{
+    ctypes_module = PyImport_ImportModule("ctypes");
+    if (ctypes_module == nullptr)
+        throw std::runtime_error("Cannot import ctypes module");
+    // get method addressof
+    PyObject* ctypes_dict = PyModule_GetDict(ctypes_module);
+    if (ctypes_dict == nullptr)
+        throw std::runtime_error(std::string("FAILURE @ ") + std::string(__FILE__) + " : " + std::to_string(__LINE__));
+    // supportedtypes
+    ctypes_c_int = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_int");
+    ctypes_c_char = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_char");
+    ctypes_c_bool = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_bool");
+    ctypes_c_wchar = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_wchar");
+    ctypes_c_byte = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_byte");
+    ctypes_c_ubyte = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ubyte");
+    ctypes_c_short = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_short");
+    ctypes_c_ushort = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ushort");
+    ctypes_c_int = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_int");
+    ctypes_c_uint = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_uint");
+    ctypes_c_long = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_long");
+    ctypes_c_ulong = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ulong");
+    ctypes_c_longlong = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_longlong");
+    ctypes_c_ulonglong = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_ulonglong");
+    ctypes_c_size_t = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_size_t");
+    ctypes_c_float = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_float");
+    ctypes_c_double = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_double");
+    ctypes_c_void_p = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_void_p"); // == c_voidp
+}
+
+
+// (target type, source type)
+std::map<std::pair<PyTypeObject*,PyTypeObject*>, std::function<int(void*, PyObject*)>> m_feeders;
+
+void populate_feeders(PyTypeObject* target_t, PyTypeObject* source_t)
+{
+    if (target_t == ctypes_c_int)
+    {
+        if (source_t == &PyLong_Type)
+        {
+            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
+            {
+                *((int*)ptr) = (int)PyLong_AsLong(value);
+                return sizeof(int);
+            };
+            return;
+        }
+    } else if (target_t == ctypes_c_bool) {
+        if (source_t == &PyBool_Type)
+        {
+            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
+            {
+                *((bool*)ptr) = (value == Py_True);
+                return sizeof(bool);
+            };
+            return;
+        }
+    } else if (target_t == ctypes_c_byte) {
+        if (source_t == &PyLong_Type)
+        {
+            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
+            {
+                *((int8_t*)ptr) = (int8_t)PyLong_AsLong(value);
+                return sizeof(int8_t);
+            };
+            return;
+        }
+    } else if (target_t == ctypes_c_double) {
+        if (source_t == &PyFloat_Type)
+        {
+            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
+            {
+                *((double*)ptr) = (double)PyFloat_AsDouble(value);
+                return sizeof(double);
+            };
+            return;
+        }
+    } else if (target_t == ctypes_c_float) {
+        if (source_t == &PyFloat_Type)
+        {
+            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
+            {
+                *((float*)ptr) = (float)PyFloat_AsDouble(value);
+                return sizeof(float);
+            };
+            return;
+        }
+    } else if (target_t == ctypes_c_longlong) {
+        if (source_t == &PyLong_Type)
+        {
+            m_feeders[{target_t,source_t}] = [](void* ptr, PyObject* value) -> int
+            {
+                *((long long*)ptr) = (long long)PyLong_AsLongLong(value);
+                return sizeof(long long);
+            };
+            return;
+        }
+    }
+}
 
-int feed(void* ptr,  PyObject* value, PyObject* type);
+int feed(void* ptr, PyObject* value, PyObject* type)
+{
+    PyTypeObject* pto = (PyTypeObject*)type;
+    if (ctypes_c_int == nullptr)
+        fetch_ctypes();
+    auto found = m_feeders.find({pto,value->ob_type});
+    if (found == m_feeders.end())
+    {
+        populate_feeders(pto, value->ob_type);
+        found = m_feeders.find({pto,value->ob_type});
+    }
+    if (found != m_feeders.end())
+    {
+        return found->second(ptr, value);
+    }
+    return 0;
+}

From becc33c074904608cd45c99dc6bac3cc9db55be6 Mon Sep 17 00:00:00 2001
From: Michael Droettboom <mdboom@gmail.com>
Date: Mon, 25 Aug 2025 11:48:51 -0400
Subject: [PATCH 2/4] Fix #789: Remove cyclical import between driver and
 _lib.utils

---
 .gitignore                                    |   1 +
 .../cuda/bindings/_lib/param_packer.h         |  63 ++---
 .../cuda/bindings/_lib/param_packer.pxd       |   2 +
 cuda_bindings/cuda/bindings/_lib/utils.pxd.in |  28 +-
 .../_lib/{utils.pyx.in => utils.pxi.in}       | 131 +++++-----
 cuda_bindings/cuda/bindings/driver.pxd.in     |  11 +-
 cuda_bindings/cuda/bindings/driver.pyx.in     | 198 +++++++-------
 cuda_bindings/cuda/bindings/nvrtc.pxd.in      |   3 +-
 cuda_bindings/cuda/bindings/nvrtc.pyx.in      |   6 +-
 cuda_bindings/cuda/bindings/runtime.pxd.in    |   7 +-
 cuda_bindings/cuda/bindings/runtime.pyx.in    | 242 +++++++++---------
 cuda_bindings/setup.py                        |   1 -
 12 files changed, 347 insertions(+), 346 deletions(-)
 rename cuda_bindings/cuda/bindings/_lib/{utils.pyx.in => utils.pxi.in} (91%)

diff --git a/.gitignore b/.gitignore
index 656dd8916..a9e5941f6 100644
--- a/.gitignore
+++ b/.gitignore
@@ -16,6 +16,7 @@ __pycache__/
 cache_driver
 cache_runtime
 cache_nvrtc
+cuda_bindings/cuda/bindings/_lib/utils.pxi
 
 # CUDA Python specific (auto-generated)
 cuda_bindings/cuda/bindings/_bindings/cydriver.pxd
diff --git a/cuda_bindings/cuda/bindings/_lib/param_packer.h b/cuda_bindings/cuda/bindings/_lib/param_packer.h
index 6f99c29e3..96c56b4fe 100644
--- a/cuda_bindings/cuda/bindings/_lib/param_packer.h
+++ b/cuda_bindings/cuda/bindings/_lib/param_packer.h
@@ -1,46 +1,40 @@
 // SPDX-FileCopyrightText: Copyright (c) 2021-2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
 // SPDX-License-Identifier: LicenseRef-NVIDIA-SOFTWARE-LICENSE
 
+// Please refer to the NVIDIA end user license agreement (EULA) associated
+// with this source code for terms and conditions that govern your use of
+// this software. Any use, reproduction, disclosure, or distribution of
+// this software and related documentation outside the terms of the EULA
+// is strictly prohibited.
+
 #include <Python.h>
-#include "param_packer.h"
 
 #include <map>
 #include <functional>
 #include <stdexcept>
 #include <string>
 
-PyObject* enum_module = nullptr;
-PyTypeObject* enum_Enum = nullptr;
-
-PyObject* ctypes_module = nullptr;
-PyObject* ctypes_addressof = nullptr;
-PyObject* addressof_param_tuple = nullptr;
+static PyObject* ctypes_module = nullptr;
 
-PyTypeObject* ctypes_c_char = nullptr;
-PyTypeObject* ctypes_c_bool = nullptr;
-PyTypeObject* ctypes_c_wchar = nullptr;
-PyTypeObject* ctypes_c_byte = nullptr;
-PyTypeObject* ctypes_c_ubyte = nullptr;
-PyTypeObject* ctypes_c_short = nullptr;
-PyTypeObject* ctypes_c_ushort = nullptr;
-PyTypeObject* ctypes_c_int = nullptr;
-PyTypeObject* ctypes_c_uint = nullptr;
-PyTypeObject* ctypes_c_long = nullptr;
-PyTypeObject* ctypes_c_ulong = nullptr;
-PyTypeObject* ctypes_c_longlong = nullptr;
-PyTypeObject* ctypes_c_ulonglong = nullptr;
-PyTypeObject* ctypes_c_size_t = nullptr;
-PyTypeObject* ctypes_c_float = nullptr;
-PyTypeObject* ctypes_c_double = nullptr;
-PyTypeObject* ctypes_c_void_p = nullptr;
+static PyTypeObject* ctypes_c_char = nullptr;
+static PyTypeObject* ctypes_c_bool = nullptr;
+static PyTypeObject* ctypes_c_wchar = nullptr;
+static PyTypeObject* ctypes_c_byte = nullptr;
+static PyTypeObject* ctypes_c_ubyte = nullptr;
+static PyTypeObject* ctypes_c_short = nullptr;
+static PyTypeObject* ctypes_c_ushort = nullptr;
+static PyTypeObject* ctypes_c_int = nullptr;
+static PyTypeObject* ctypes_c_uint = nullptr;
+static PyTypeObject* ctypes_c_long = nullptr;
+static PyTypeObject* ctypes_c_ulong = nullptr;
+static PyTypeObject* ctypes_c_longlong = nullptr;
+static PyTypeObject* ctypes_c_ulonglong = nullptr;
+static PyTypeObject* ctypes_c_size_t = nullptr;
+static PyTypeObject* ctypes_c_float = nullptr;
+static PyTypeObject* ctypes_c_double = nullptr;
+static PyTypeObject* ctypes_c_void_p = nullptr;
 
-PyTypeObject* ctypes_c_ssize_t = nullptr;
-PyTypeObject* ctypes_c_longdouble = nullptr;
-PyTypeObject* ctypes_c_char_p = nullptr;
-PyTypeObject* ctypes_c_wchar_p = nullptr;
-PyTypeObject* ctypes_c_structure = nullptr;
-
-void fetch_ctypes()
+static void fetch_ctypes()
 {
     ctypes_module = PyImport_ImportModule("ctypes");
     if (ctypes_module == nullptr)
@@ -50,7 +44,6 @@ void fetch_ctypes()
     if (ctypes_dict == nullptr)
         throw std::runtime_error(std::string("FAILURE @ ") + std::string(__FILE__) + " : " + std::to_string(__LINE__));
     // supportedtypes
-    ctypes_c_int = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_int");
     ctypes_c_char = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_char");
     ctypes_c_bool = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_bool");
     ctypes_c_wchar = (PyTypeObject*) PyDict_GetItemString(ctypes_dict, "c_wchar");
@@ -72,9 +65,9 @@ void fetch_ctypes()
 
 
 // (target type, source type)
-std::map<std::pair<PyTypeObject*,PyTypeObject*>, std::function<int(void*, PyObject*)>> m_feeders;
+static std::map<std::pair<PyTypeObject*,PyTypeObject*>, std::function<int(void*, PyObject*)>> m_feeders;
 
-void populate_feeders(PyTypeObject* target_t, PyTypeObject* source_t)
+static void populate_feeders(PyTypeObject* target_t, PyTypeObject* source_t)
 {
     if (target_t == ctypes_c_int)
     {
@@ -140,7 +133,7 @@ void populate_feeders(PyTypeObject* target_t, PyTypeObject* source_t)
     }
 }
 
-int feed(void* ptr, PyObject* value, PyObject* type)
+static int feed(void* ptr, PyObject* value, PyObject* type)
 {
     PyTypeObject* pto = (PyTypeObject*)type;
     if (ctypes_c_int == nullptr)
diff --git a/cuda_bindings/cuda/bindings/_lib/param_packer.pxd b/cuda_bindings/cuda/bindings/_lib/param_packer.pxd
index 82b0d9497..ad7fd9566 100644
--- a/cuda_bindings/cuda/bindings/_lib/param_packer.pxd
+++ b/cuda_bindings/cuda/bindings/_lib/param_packer.pxd
@@ -1,5 +1,7 @@
 # SPDX-FileCopyrightText: Copyright (c) 2021-2024 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
 # SPDX-License-Identifier: LicenseRef-NVIDIA-SOFTWARE-LICENSE
 
+# Include "param_packer.h" so its contents get compiled into every
+# Cython extension module that depends on param_packer.pxd.
 cdef extern from "param_packer.h":
     int feed(void* ptr, object o, object ct)
diff --git a/cuda_bindings/cuda/bindings/_lib/utils.pxd.in b/cuda_bindings/cuda/bindings/_lib/utils.pxd.in
index b66ac71f0..d317e69e8 100644
--- a/cuda_bindings/cuda/bindings/_lib/utils.pxd.in
+++ b/cuda_bindings/cuda/bindings/_lib/utils.pxd.in
@@ -6,7 +6,7 @@ cimport cuda.bindings.cydriver as cydriver
 cimport cuda.bindings.cyruntime as cyruntime
 from libcpp.vector cimport vector
 
-cdef class HelperKernelParams:
+cdef class _HelperKernelParams:
     cdef Py_buffer _pybuffer
     cdef bint _pyobj_acquired
     cdef void** _ckernelParams
@@ -14,13 +14,13 @@ cdef class HelperKernelParams:
     cdef int _length
     cdef bint _malloc_list_created
 
-cdef class HelperInputVoidPtr:
+cdef class _HelperInputVoidPtr:
     cdef Py_buffer _pybuffer
     cdef void* _cptr
     cdef bint _pyobj_acquired
 {{if 'CUmemPool_attribute_enum' in found_types}}
 
-cdef class HelperCUmemPool_attribute:
+cdef class _HelperCUmemPool_attribute:
     cdef void* _cptr
     cdef cydriver.CUmemPool_attribute_enum _attr
     cdef bint _is_getter
@@ -31,7 +31,7 @@ cdef class HelperCUmemPool_attribute:
 {{endif}}
 {{if 'CUmem_range_attribute_enum' in found_types}}
 
-cdef class HelperCUmem_range_attribute:
+cdef class _HelperCUmem_range_attribute:
     cdef void* _cptr
     cdef cydriver.CUmem_range_attribute_enum _attr
     cdef size_t _data_size
@@ -42,7 +42,7 @@ cdef class HelperCUmem_range_attribute:
 {{endif}}
 {{if 'CUpointer_attribute_enum' in found_types}}
 
-cdef class HelperCUpointer_attribute:
+cdef class _HelperCUpointer_attribute:
     cdef void* _cptr
     cdef cydriver.CUpointer_attribute_enum _attr
     cdef bint _is_getter
@@ -60,7 +60,7 @@ cdef class HelperCUpointer_attribute:
 {{endif}}
 {{if 'CUgraphMem_attribute_enum' in found_types}}
 
-cdef class HelperCUgraphMem_attribute:
+cdef class _HelperCUgraphMem_attribute:
     cdef void* _cptr
     cdef cydriver.CUgraphMem_attribute_enum _attr
     cdef bint _is_getter
@@ -70,7 +70,7 @@ cdef class HelperCUgraphMem_attribute:
 {{endif}}
 {{if 'CUjit_option_enum' in found_types}}
 
-cdef class HelperCUjit_option:
+cdef class _HelperCUjit_option:
     cdef void* _cptr
     cdef cydriver.CUjit_option_enum _attr
 
@@ -83,11 +83,11 @@ cdef class HelperCUjit_option:
     cdef int _int
     cdef cydriver.CUjit_cacheMode_enum _cacheMode
     cdef vector[char*] _charstarstar # list of names
-    cdef InputVoidPtrPtrHelper _voidstarstar # list of addresses
+    cdef _InputVoidPtrPtrHelper _voidstarstar # list of addresses
 {{endif}}
 {{if 'cudaJitOption' in found_types}}
 
-cdef class HelperCudaJitOption:
+cdef class _HelperCudaJitOption:
     cdef void* _cptr
     cdef cyruntime.cudaJitOption _attr
 
@@ -101,7 +101,7 @@ cdef class HelperCudaJitOption:
 {{endif}}
 {{if 'CUlibraryOption_enum' in found_types}}
 
-cdef class HelperCUlibraryOption:
+cdef class _HelperCUlibraryOption:
     cdef void* _cptr
     cdef cydriver.CUlibraryOption_enum _attr
 
@@ -110,7 +110,7 @@ cdef class HelperCUlibraryOption:
 {{endif}}
 {{if 'cudaLibraryOption' in found_types}}
 
-cdef class HelperCudaLibraryOption:
+cdef class _HelperCudaLibraryOption:
     cdef void* _cptr
     cdef cyruntime.cudaLibraryOption _attr
 
@@ -119,7 +119,7 @@ cdef class HelperCudaLibraryOption:
 {{endif}}
 {{if 'CUmemAllocationHandleType_enum' in found_types}}
 
-cdef class HelperCUmemAllocationHandleType:
+cdef class _HelperCUmemAllocationHandleType:
     cdef void* _cptr
     cdef cydriver.CUmemAllocationHandleType_enum _type
 
@@ -132,12 +132,12 @@ cdef class HelperCUmemAllocationHandleType:
     {{endif}}
 {{endif}}
 
-cdef class InputVoidPtrPtrHelper:
+cdef class _InputVoidPtrPtrHelper:
     cdef void** _cptr
 
 {{if 'CUcoredumpSettings_enum' in found_types}}
 
-cdef class HelperCUcoredumpSettings:
+cdef class _HelperCUcoredumpSettings:
     cdef void* _cptr
     cdef cydriver.CUcoredumpSettings_enum _attrib
     cdef bint _is_getter
diff --git a/cuda_bindings/cuda/bindings/_lib/utils.pyx.in b/cuda_bindings/cuda/bindings/_lib/utils.pxi.in
similarity index 91%
rename from cuda_bindings/cuda/bindings/_lib/utils.pyx.in
rename to cuda_bindings/cuda/bindings/_lib/utils.pxi.in
index da38dd450..0a9f2e4e3 100644
--- a/cuda_bindings/cuda/bindings/_lib/utils.pyx.in
+++ b/cuda_bindings/cuda/bindings/_lib/utils.pxi.in
@@ -6,42 +6,39 @@ from libc.stdlib cimport calloc, free
 from libc.stdint cimport int32_t, uint32_t, int64_t, uint64_t
 from libc.stddef cimport wchar_t
 from libc.string cimport memcpy
-from enum import Enum
-from typing import List, Tuple
-import ctypes
+from enum import Enum as _Enum
+import ctypes as _ctypes
 cimport cuda.bindings.cydriver as cydriver
-import cuda.bindings.driver as driver
+import cuda.bindings.driver as _driver
 cimport cuda.bindings._lib.param_packer as param_packer
 
-ctypedef unsigned long long void_ptr
-
-cdef void* callocWrapper(length, size):
+cdef void* _callocWrapper(length, size):
     cdef void* out = calloc(length, size)
     if out is NULL:
         raise MemoryError('Failed to allocated length x size memory: {}x{}'.format(length, size))
     return out
 
-cdef class HelperKernelParams:
+cdef class _HelperKernelParams:
     supported_types = { # excluding void_p and None, which are handled specially
-        ctypes.c_bool,
-        ctypes.c_char,
-        ctypes.c_wchar,
-        ctypes.c_byte,
-        ctypes.c_ubyte,
-        ctypes.c_short,
-        ctypes.c_ushort,
-        ctypes.c_int,
-        ctypes.c_uint,
-        ctypes.c_long,
-        ctypes.c_ulong,
-        ctypes.c_longlong,
-        ctypes.c_ulonglong,
-        ctypes.c_size_t,
-        ctypes.c_float,
-        ctypes.c_double
+        _ctypes.c_bool,
+        _ctypes.c_char,
+        _ctypes.c_wchar,
+        _ctypes.c_byte,
+        _ctypes.c_ubyte,
+        _ctypes.c_short,
+        _ctypes.c_ushort,
+        _ctypes.c_int,
+        _ctypes.c_uint,
+        _ctypes.c_long,
+        _ctypes.c_ulong,
+        _ctypes.c_longlong,
+        _ctypes.c_ulonglong,
+        _ctypes.c_size_t,
+        _ctypes.c_float,
+        _ctypes.c_double
     }
 
-    max_param_size = max(ctypes.sizeof(max(HelperKernelParams.supported_types, key=lambda t:ctypes.sizeof(t))), sizeof(void_ptr))
+    max_param_size = max(_ctypes.sizeof(max(_HelperKernelParams.supported_types, key=lambda t:_ctypes.sizeof(t))), sizeof(void_ptr))
 
     def __cinit__(self, kernelParams):
         self._pyobj_acquired = False
@@ -58,14 +55,14 @@ cdef class HelperKernelParams:
                 raise RuntimeError("Argument 'kernelParams' failed to retrieve buffer through Buffer Protocol")
             self._pyobj_acquired = True
             self._ckernelParams = <void**><void_ptr>self._pybuffer.buf
-        elif isinstance(kernelParams, (Tuple)) and len(kernelParams) == 2 and isinstance(kernelParams[0], (Tuple)) and isinstance(kernelParams[1], (Tuple)):
+        elif isinstance(kernelParams, (tuple)) and len(kernelParams) == 2 and isinstance(kernelParams[0], (tuple)) and isinstance(kernelParams[1], (tuple)):
             # Hard run, construct and fill out contigues memory using provided kernel values and types based
             if len(kernelParams[0]) != len(kernelParams[1]):
                 raise TypeError("Argument 'kernelParams' has tuples with different length")
             if len(kernelParams[0]) != 0:
                 self._length = len(kernelParams[0])
-                self._ckernelParams = <void**>callocWrapper(len(kernelParams[0]), sizeof(void*))
-                self._ckernelParamsData = <char*>callocWrapper(len(kernelParams[0]), HelperKernelParams.max_param_size)
+                self._ckernelParams = <void**>_callocWrapper(len(kernelParams[0]), sizeof(void*))
+                self._ckernelParamsData = <char*>_callocWrapper(len(kernelParams[0]), _HelperKernelParams.max_param_size)
                 self._malloc_list_created = True
 
             idx = 0
@@ -75,44 +72,44 @@ cdef class HelperKernelParams:
                     # special cases for None
                     if callable(getattr(value, 'getPtr', None)):
                         self._ckernelParams[idx] = <void*><void_ptr>value.getPtr()
-                    elif isinstance(value, (ctypes.Structure)):
-                        self._ckernelParams[idx] = <void*><void_ptr>ctypes.addressof(value)
-                    elif isinstance(value, (Enum)):
+                    elif isinstance(value, (_ctypes.Structure)):
+                        self._ckernelParams[idx] = <void*><void_ptr>_ctypes.addressof(value)
+                    elif isinstance(value, (_Enum)):
                         self._ckernelParams[idx] = &(self._ckernelParamsData[data_idx])
                         (<int*>self._ckernelParams[idx])[0] = value.value
                         data_idx += sizeof(int)
                     else:
-                        raise TypeError("Provided argument is of type {} but expected Type {}, {} or CUDA Binding structure with getPtr() attribute".format(type(value), type(ctypes.Structure), type(ctypes.c_void_p)))
-                elif ctype in HelperKernelParams.supported_types:
+                        raise TypeError("Provided argument is of type {} but expected Type {}, {} or CUDA Binding structure with getPtr() attribute".format(type(value), type(_ctypes.Structure), type(_ctypes.c_void_p)))
+                elif ctype in _HelperKernelParams.supported_types:
                     self._ckernelParams[idx] = &(self._ckernelParamsData[data_idx])
 
                     # handle case where a float is passed as a double
-                    if ctype == ctypes.c_double and isinstance(value, ctypes.c_float):
+                    if ctype == _ctypes.c_double and isinstance(value, _ctypes.c_float):
                         value = ctype(value.value)
                     if not isinstance(value, ctype): # make it a ctype
                         size = param_packer.feed(self._ckernelParams[idx], value, ctype)
                         if size == 0: # feed failed
                             value = ctype(value)
-                            size = ctypes.sizeof(ctype)
-                            addr = <void*>(<void_ptr>ctypes.addressof(value))
+                            size = _ctypes.sizeof(ctype)
+                            addr = <void*>(<void_ptr>_ctypes.addressof(value))
                             memcpy(self._ckernelParams[idx], addr, size)
                     else:
-                        size = ctypes.sizeof(ctype)
-                        addr = <void*>(<void_ptr>ctypes.addressof(value))
+                        size = _ctypes.sizeof(ctype)
+                        addr = <void*>(<void_ptr>_ctypes.addressof(value))
                         memcpy(self._ckernelParams[idx], addr, size)
                     data_idx += size
-                elif ctype == ctypes.c_void_p:
+                elif ctype == _ctypes.c_void_p:
                     # special cases for void_p
-                    if isinstance(value, (int, ctypes.c_void_p)):
+                    if isinstance(value, (int, _ctypes.c_void_p)):
                         self._ckernelParams[idx] = &(self._ckernelParamsData[data_idx])
-                        (<void_ptr*>self._ckernelParams[idx])[0] = value.value if isinstance(value, (ctypes.c_void_p)) else value
+                        (<void_ptr*>self._ckernelParams[idx])[0] = value.value if isinstance(value, (_ctypes.c_void_p)) else value
                         data_idx += sizeof(void_ptr)
                     elif callable(getattr(value, 'getPtr', None)):
                         self._ckernelParams[idx] = &(self._ckernelParamsData[data_idx])
                         (<void_ptr*>self._ckernelParams[idx])[0] = value.getPtr()
                         data_idx += sizeof(void_ptr)
                     else:
-                        raise TypeError("Provided argument is of type {} but expected Type {}, {} or CUDA Binding structure with getPtr() attribute".format(type(value), type(int), type(ctypes.c_void_p)))
+                        raise TypeError("Provided argument is of type {} but expected Type {}, {} or CUDA Binding structure with getPtr() attribute".format(type(value), type(int), type(_ctypes.c_void_p)))
                 else:
                     raise TypeError("Unsupported type: " + str(type(ctype)))
                 idx += 1
@@ -130,7 +127,7 @@ cdef class HelperKernelParams:
     def ckernelParams(self):
         return <void_ptr>self._ckernelParams
 
-cdef class HelperInputVoidPtr:
+cdef class _HelperInputVoidPtr:
     def __cinit__(self, ptr):
         self._pyobj_acquired = False
         if ptr is None:
@@ -138,7 +135,7 @@ cdef class HelperInputVoidPtr:
         elif isinstance(ptr, (int)):
             # Easy run, user gave us an already configured void** address
             self._cptr = <void*><void_ptr>ptr
-        elif isinstance(ptr, (driver.CUdeviceptr)):
+        elif isinstance(ptr, (_driver.CUdeviceptr)):
             self._cptr = <void*><void_ptr>int(ptr)
         elif PyObject_CheckBuffer(ptr):
             # Easy run, get address from Python Buffer Protocol
@@ -160,7 +157,7 @@ cdef class HelperInputVoidPtr:
 
 {{if 'CUmemPool_attribute_enum' in found_types}}
 
-cdef class HelperCUmemPool_attribute:
+cdef class _HelperCUmemPool_attribute:
     def __cinit__(self, attr, init_value, is_getter=False):
         self._is_getter = is_getter
         self._attr = attr.value
@@ -175,7 +172,7 @@ cdef class HelperCUmemPool_attribute:
                             {{if 'CU_MEMPOOL_ATTR_USED_MEM_CURRENT'}}cydriver.CUmemPool_attribute_enum.CU_MEMPOOL_ATTR_USED_MEM_CURRENT,{{endif}}
                             {{if 'CU_MEMPOOL_ATTR_USED_MEM_HIGH'}}cydriver.CUmemPool_attribute_enum.CU_MEMPOOL_ATTR_USED_MEM_HIGH,{{endif}}):
             if self._is_getter:
-                self._cuuint64_t_val = driver.cuuint64_t()
+                self._cuuint64_t_val = _driver.cuuint64_t()
                 self._cptr = <void*><void_ptr>self._cuuint64_t_val.getPtr()
             else:
                 self._cptr = <void*><void_ptr>init_value.getPtr()
@@ -206,7 +203,7 @@ cdef class HelperCUmemPool_attribute:
 {{endif}}
 {{if 'CUmem_range_attribute_enum' in found_types}}
 
-cdef class HelperCUmem_range_attribute:
+cdef class _HelperCUmem_range_attribute:
     def __cinit__(self, attr, data_size):
         self._data_size = data_size
         self._attr = attr.value
@@ -215,7 +212,7 @@ cdef class HelperCUmem_range_attribute:
                           {{if 'CU_MEM_RANGE_ATTRIBUTE_LAST_PREFETCH_LOCATION'}}cydriver.CUmem_range_attribute_enum.CU_MEM_RANGE_ATTRIBUTE_LAST_PREFETCH_LOCATION,{{endif}}):
             self._cptr = <void*>&self._int_val
         elif self._attr in ({{if 'CU_MEM_RANGE_ATTRIBUTE_ACCESSED_BY'}}cydriver.CUmem_range_attribute_enum.CU_MEM_RANGE_ATTRIBUTE_ACCESSED_BY,{{endif}}):
-            self._cptr = callocWrapper(1, self._data_size)
+            self._cptr = _callocWrapper(1, self._data_size)
             self._int_val_list = <int*>self._cptr
         else:
             raise TypeError('Unsupported attribute: {}'.format(attr.name))
@@ -240,13 +237,13 @@ cdef class HelperCUmem_range_attribute:
 {{endif}}
 {{if 'CUpointer_attribute_enum' in found_types}}
 
-cdef class HelperCUpointer_attribute:
+cdef class _HelperCUpointer_attribute:
     def __cinit__(self, attr, init_value, is_getter=False):
         self._is_getter = is_getter
         self._attr = attr.value
         if self._attr in ({{if 'CU_POINTER_ATTRIBUTE_CONTEXT'}}cydriver.CUpointer_attribute_enum.CU_POINTER_ATTRIBUTE_CONTEXT,{{endif}}):
             if self._is_getter:
-                self._ctx = driver.CUcontext()
+                self._ctx = _driver.CUcontext()
                 self._cptr = <void*><void_ptr>self._ctx.getPtr()
             else:
                 self._cptr = <void*><void_ptr>init_value.getPtr()
@@ -260,7 +257,7 @@ cdef class HelperCUpointer_attribute:
         elif self._attr in ({{if 'CU_POINTER_ATTRIBUTE_DEVICE_POINTER'}}cydriver.CUpointer_attribute_enum.CU_POINTER_ATTRIBUTE_DEVICE_POINTER,{{endif}}
                             {{if 'CU_POINTER_ATTRIBUTE_RANGE_START_ADDR'}}cydriver.CUpointer_attribute_enum.CU_POINTER_ATTRIBUTE_RANGE_START_ADDR,{{endif}}):
             if self._is_getter:
-                self._devptr = driver.CUdeviceptr()
+                self._devptr = _driver.CUdeviceptr()
                 self._cptr = <void*><void_ptr>self._devptr.getPtr()
             else:
                 self._cptr = <void*><void_ptr>init_value.getPtr()
@@ -269,7 +266,7 @@ cdef class HelperCUpointer_attribute:
             self._cptr = <void*>&self._void
         elif self._attr in ({{if 'CU_POINTER_ATTRIBUTE_P2P_TOKENS'}}cydriver.CUpointer_attribute_enum.CU_POINTER_ATTRIBUTE_P2P_TOKENS,{{endif}}):
             if self._is_getter:
-                self._token = driver.CUDA_POINTER_ATTRIBUTE_P2P_TOKENS()
+                self._token = _driver.CUDA_POINTER_ATTRIBUTE_P2P_TOKENS()
                 self._cptr = <void*><void_ptr>self._token.getPtr()
             else:
                 self._cptr = <void*><void_ptr>init_value.getPtr()
@@ -287,7 +284,7 @@ cdef class HelperCUpointer_attribute:
             self._cptr = <void*>&self._size
         elif self._attr in ({{if 'CU_POINTER_ATTRIBUTE_MEMPOOL_HANDLE'}}cydriver.CUpointer_attribute_enum.CU_POINTER_ATTRIBUTE_MEMPOOL_HANDLE,{{endif}}):
             if self._is_getter:
-                self._mempool = driver.CUmemoryPool()
+                self._mempool = _driver.CUmemoryPool()
                 self._cptr = <void*><void_ptr>self._mempool.getPtr()
             else:
                 self._cptr = <void*><void_ptr>init_value.getPtr()
@@ -334,7 +331,7 @@ cdef class HelperCUpointer_attribute:
 {{endif}}
 {{if 'CUgraphMem_attribute_enum' in found_types}}
 
-cdef class HelperCUgraphMem_attribute:
+cdef class _HelperCUgraphMem_attribute:
     def __cinit__(self, attr, init_value, is_getter=False):
         self._is_getter = is_getter
         self._attr = attr.value
@@ -343,7 +340,7 @@ cdef class HelperCUgraphMem_attribute:
                           {{if 'CU_GRAPH_MEM_ATTR_RESERVED_MEM_CURRENT' in found_values}}cydriver.CUgraphMem_attribute_enum.CU_GRAPH_MEM_ATTR_RESERVED_MEM_CURRENT,{{endif}}
                           {{if 'CU_GRAPH_MEM_ATTR_RESERVED_MEM_HIGH' in found_values}}cydriver.CUgraphMem_attribute_enum.CU_GRAPH_MEM_ATTR_RESERVED_MEM_HIGH,{{endif}}):
             if self._is_getter:
-                self._cuuint64_t_val = driver.cuuint64_t()
+                self._cuuint64_t_val = _driver.cuuint64_t()
                 self._cptr = <void*><void_ptr>self._cuuint64_t_val.getPtr()
             else:
                 self._cptr = <void*><void_ptr>init_value.getPtr()
@@ -369,7 +366,7 @@ cdef class HelperCUgraphMem_attribute:
 {{endif}}
 {{if 'CUjit_option_enum' in found_types}}
 
-cdef class HelperCUjit_option:
+cdef class _HelperCUjit_option:
     def __cinit__(self, attr, init_value):
         self._attr = attr.value
         if self._attr in ({{if 'CU_JIT_MAX_REGISTERS' in found_values}}cydriver.CUjit_option_enum.CU_JIT_MAX_REGISTERS,{{endif}}
@@ -418,8 +415,8 @@ cdef class HelperCUjit_option:
             self._charstarstar = init_value
             self._cptr = <void*>&self._charstarstar[0]
         elif self._attr in ({{if 'CU_JIT_GLOBAL_SYMBOL_ADDRESSES' in found_values}}cydriver.CUjit_option_enum.CU_JIT_GLOBAL_SYMBOL_ADDRESSES,{{endif}}):
-            pylist = [HelperInputVoidPtr(val) for val in init_value]
-            self._voidstarstar = InputVoidPtrPtrHelper(pylist)
+            pylist = [_HelperInputVoidPtr(val) for val in init_value]
+            self._voidstarstar = _InputVoidPtrPtrHelper(pylist)
             self._cptr = <void*><void_ptr>self._voidstarstar.cptr
         else:
             raise TypeError('Unsupported attribute: {}'.format(attr.name))
@@ -434,7 +431,7 @@ cdef class HelperCUjit_option:
 
 {{if 'cudaJitOption' in found_types}}
 
-cdef class HelperCudaJitOption:
+cdef class _HelperCudaJitOption:
     def __cinit__(self, attr, init_value):
         self._attr = attr.value
         if self._attr in ({{if 'cudaJitMaxRegisters' in found_values}}cyruntime.cudaJitOption.cudaJitMaxRegisters,{{endif}}
@@ -479,7 +476,7 @@ cdef class HelperCudaJitOption:
 
 {{if 'CUlibraryOption_enum' in found_types}}
 
-cdef class HelperCUlibraryOption:
+cdef class _HelperCUlibraryOption:
     def __cinit__(self, attr, init_value):
         self._attr = attr.value
         if False:
@@ -506,7 +503,7 @@ cdef class HelperCUlibraryOption:
 
 {{if 'cudaLibraryOption' in found_types}}
 
-cdef class HelperCudaLibraryOption:
+cdef class _HelperCudaLibraryOption:
     def __cinit__(self, attr, init_value):
         self._attr = attr.value
         if False:
@@ -533,7 +530,7 @@ cdef class HelperCudaLibraryOption:
 
 {{if 'CUmemAllocationHandleType_enum' in found_types}}
 
-cdef class HelperCUmemAllocationHandleType:
+cdef class _HelperCUmemAllocationHandleType:
     def __cinit__(self, attr):
         self._type = attr.value
         if False:
@@ -556,7 +553,7 @@ cdef class HelperCUmemAllocationHandleType:
         {{endif}}
         {{if 'CU_MEM_HANDLE_TYPE_FABRIC' in found_values}}
         elif self._type in (cydriver.CUmemAllocationHandleType_enum.CU_MEM_HANDLE_TYPE_FABRIC,):
-            self._mem_fabric_handle = driver.CUmemFabricHandle()
+            self._mem_fabric_handle = _driver.CUmemFabricHandle()
             self._cptr = <void*><void_ptr>self._mem_fabric_handle.getPtr()
         {{endif}}
         else:
@@ -596,9 +593,9 @@ cdef class HelperCUmemAllocationHandleType:
             raise TypeError('Unsupported attribute: {}'.format(self._type))
 {{endif}}
 
-cdef class InputVoidPtrPtrHelper:
+cdef class _InputVoidPtrPtrHelper:
     def __cinit__(self, lst):
-        self._cptr = <void**>callocWrapper(len(lst), sizeof(void*))
+        self._cptr = <void**>_callocWrapper(len(lst), sizeof(void*))
         for idx in range(len(lst)):
             self._cptr[idx] = <void*><void_ptr>lst[idx].cptr
 
@@ -611,14 +608,14 @@ cdef class InputVoidPtrPtrHelper:
 
 {{if 'CUcoredumpSettings_enum' in found_types}}
 
-cdef class HelperCUcoredumpSettings:
+cdef class _HelperCUcoredumpSettings:
     def __cinit__(self, attr, init_value, is_getter=False):
         self._is_getter = is_getter
         self._attrib = attr.value
         if self._attrib in ({{if 'CU_COREDUMP_FILE' in found_values}}cydriver.CUcoredumpSettings_enum.CU_COREDUMP_FILE,{{endif}}
                           {{if 'CU_COREDUMP_PIPE' in found_values}}cydriver.CUcoredumpSettings_enum.CU_COREDUMP_PIPE,{{endif}}):
             if self._is_getter:
-                self._charstar = <char*>callocWrapper(1024, 1)
+                self._charstar = <char*>_callocWrapper(1024, 1)
                 self._cptr = <void*><void_ptr>self._charstar
                 self._size = 1024
             else:
diff --git a/cuda_bindings/cuda/bindings/driver.pxd.in b/cuda_bindings/cuda/bindings/driver.pxd.in
index 4e2a8bf32..ee01d5b58 100644
--- a/cuda_bindings/cuda/bindings/driver.pxd.in
+++ b/cuda_bindings/cuda/bindings/driver.pxd.in
@@ -3,7 +3,8 @@
 
 # This code was automatically generated with version 13.0.0. Do not modify it directly.
 cimport cuda.bindings.cydriver as cydriver
-cimport cuda.bindings._lib.utils as utils
+
+include "_lib/utils.pxd"
 
 {{if 'CUcontext' in found_types}}
 
@@ -1164,7 +1165,7 @@ cdef class CUDA_KERNEL_NODE_PARAMS_st:
     cdef CUfunction _func
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_st.kernelParams' in found_struct}}
-    cdef utils.HelperKernelParams _cykernelParams
+    cdef _HelperKernelParams _cykernelParams
     {{endif}}
 {{endif}}
 {{if 'CUDA_KERNEL_NODE_PARAMS_v2_st' in found_struct}}
@@ -1237,7 +1238,7 @@ cdef class CUDA_KERNEL_NODE_PARAMS_v2_st:
     cdef CUfunction _func
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_v2_st.kernelParams' in found_struct}}
-    cdef utils.HelperKernelParams _cykernelParams
+    cdef _HelperKernelParams _cykernelParams
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_v2_st.kern' in found_struct}}
     cdef CUkernel _kern
@@ -1316,7 +1317,7 @@ cdef class CUDA_KERNEL_NODE_PARAMS_v3_st:
     cdef CUfunction _func
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_v3_st.kernelParams' in found_struct}}
-    cdef utils.HelperKernelParams _cykernelParams
+    cdef _HelperKernelParams _cykernelParams
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_v3_st.kern' in found_struct}}
     cdef CUkernel _kern
@@ -3201,7 +3202,7 @@ cdef class CUDA_LAUNCH_PARAMS_st:
     cdef CUstream _hStream
     {{endif}}
     {{if 'CUDA_LAUNCH_PARAMS_st.kernelParams' in found_struct}}
-    cdef utils.HelperKernelParams _cykernelParams
+    cdef _HelperKernelParams _cykernelParams
     {{endif}}
 {{endif}}
 {{if 'CUDA_EXTERNAL_MEMORY_HANDLE_DESC_st.handle.win32' in found_struct}}
diff --git a/cuda_bindings/cuda/bindings/driver.pyx.in b/cuda_bindings/cuda/bindings/driver.pyx.in
index 196c5365e..975153c58 100644
--- a/cuda_bindings/cuda/bindings/driver.pyx.in
+++ b/cuda_bindings/cuda/bindings/driver.pyx.in
@@ -17,6 +17,8 @@ from cpython.bytes cimport PyBytes_FromStringAndSize
 import cuda.bindings.driver
 from libcpp.map cimport map
 
+include "_lib/utils.pxi"
+
 ctypedef unsigned long long signed_char_ptr
 ctypedef unsigned long long unsigned_char_ptr
 ctypedef unsigned long long char_ptr
@@ -9546,7 +9548,7 @@ cdef class CUaccessPolicyWindow_st:
         return <void_ptr>self._pvt_ptr[0].base_ptr
     @base_ptr.setter
     def base_ptr(self, base_ptr):
-        _cbase_ptr = utils.HelperInputVoidPtr(base_ptr)
+        _cbase_ptr = _HelperInputVoidPtr(base_ptr)
         self._pvt_ptr[0].base_ptr = <void*><void_ptr>_cbase_ptr.cptr
     {{endif}}
     {{if 'CUaccessPolicyWindow_st.num_bytes' in found_struct}}
@@ -9799,7 +9801,7 @@ cdef class CUDA_KERNEL_NODE_PARAMS_st:
         return <void_ptr>self._pvt_ptr[0].kernelParams
     @kernelParams.setter
     def kernelParams(self, kernelParams):
-        self._cykernelParams = utils.HelperKernelParams(kernelParams)
+        self._cykernelParams = _HelperKernelParams(kernelParams)
         self._pvt_ptr[0].kernelParams = <void**><void_ptr>self._cykernelParams.ckernelParams
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_st.extra' in found_struct}}
@@ -10052,7 +10054,7 @@ cdef class CUDA_KERNEL_NODE_PARAMS_v2_st:
         return <void_ptr>self._pvt_ptr[0].kernelParams
     @kernelParams.setter
     def kernelParams(self, kernelParams):
-        self._cykernelParams = utils.HelperKernelParams(kernelParams)
+        self._cykernelParams = _HelperKernelParams(kernelParams)
         self._pvt_ptr[0].kernelParams = <void**><void_ptr>self._cykernelParams.ckernelParams
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_v2_st.extra' in found_struct}}
@@ -10339,7 +10341,7 @@ cdef class CUDA_KERNEL_NODE_PARAMS_v3_st:
         return <void_ptr>self._pvt_ptr[0].kernelParams
     @kernelParams.setter
     def kernelParams(self, kernelParams):
-        self._cykernelParams = utils.HelperKernelParams(kernelParams)
+        self._cykernelParams = _HelperKernelParams(kernelParams)
         self._pvt_ptr[0].kernelParams = <void**><void_ptr>self._cykernelParams.ckernelParams
     {{endif}}
     {{if 'CUDA_KERNEL_NODE_PARAMS_v3_st.extra' in found_struct}}
@@ -10798,7 +10800,7 @@ cdef class CUDA_HOST_NODE_PARAMS_st:
         return <void_ptr>self._pvt_ptr[0].userData
     @userData.setter
     def userData(self, userData):
-        _cuserData = utils.HelperInputVoidPtr(userData)
+        _cuserData = _HelperInputVoidPtr(userData)
         self._pvt_ptr[0].userData = <void*><void_ptr>_cuserData.cptr
     {{endif}}
 {{endif}}
@@ -10879,7 +10881,7 @@ cdef class CUDA_HOST_NODE_PARAMS_v2_st:
         return <void_ptr>self._pvt_ptr[0].userData
     @userData.setter
     def userData(self, userData):
-        _cuserData = utils.HelperInputVoidPtr(userData)
+        _cuserData = _HelperInputVoidPtr(userData)
         self._pvt_ptr[0].userData = <void*><void_ptr>_cuserData.cptr
     {{endif}}
 {{endif}}
@@ -12758,7 +12760,7 @@ cdef class CUctxCigParam_st:
         return <void_ptr>self._pvt_ptr[0].sharedData
     @sharedData.setter
     def sharedData(self, sharedData):
-        _csharedData = utils.HelperInputVoidPtr(sharedData)
+        _csharedData = _HelperInputVoidPtr(sharedData)
         self._pvt_ptr[0].sharedData = <void*><void_ptr>_csharedData.cptr
     {{endif}}
 {{endif}}
@@ -12962,7 +12964,7 @@ cdef class CUlibraryHostUniversalFunctionAndDataTable_st:
         return <void_ptr>self._pvt_ptr[0].functionTable
     @functionTable.setter
     def functionTable(self, functionTable):
-        _cfunctionTable = utils.HelperInputVoidPtr(functionTable)
+        _cfunctionTable = _HelperInputVoidPtr(functionTable)
         self._pvt_ptr[0].functionTable = <void*><void_ptr>_cfunctionTable.cptr
     {{endif}}
     {{if 'CUlibraryHostUniversalFunctionAndDataTable_st.functionWindowSize' in found_struct}}
@@ -12979,7 +12981,7 @@ cdef class CUlibraryHostUniversalFunctionAndDataTable_st:
         return <void_ptr>self._pvt_ptr[0].dataTable
     @dataTable.setter
     def dataTable(self, dataTable):
-        _cdataTable = utils.HelperInputVoidPtr(dataTable)
+        _cdataTable = _HelperInputVoidPtr(dataTable)
         self._pvt_ptr[0].dataTable = <void*><void_ptr>_cdataTable.cptr
     {{endif}}
     {{if 'CUlibraryHostUniversalFunctionAndDataTable_st.dataWindowSize' in found_struct}}
@@ -13226,7 +13228,7 @@ cdef class CUDA_MEMCPY2D_st:
         return <void_ptr>self._pvt_ptr[0].srcHost
     @srcHost.setter
     def srcHost(self, srcHost):
-        _csrcHost = utils.HelperInputVoidPtr(srcHost)
+        _csrcHost = _HelperInputVoidPtr(srcHost)
         self._pvt_ptr[0].srcHost = <void*><void_ptr>_csrcHost.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY2D_st.srcDevice' in found_struct}}
@@ -13304,7 +13306,7 @@ cdef class CUDA_MEMCPY2D_st:
         return <void_ptr>self._pvt_ptr[0].dstHost
     @dstHost.setter
     def dstHost(self, dstHost):
-        _cdstHost = utils.HelperInputVoidPtr(dstHost)
+        _cdstHost = _HelperInputVoidPtr(dstHost)
         self._pvt_ptr[0].dstHost = <void*><void_ptr>_cdstHost.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY2D_st.dstDevice' in found_struct}}
@@ -13709,7 +13711,7 @@ cdef class CUDA_MEMCPY3D_st:
         return <void_ptr>self._pvt_ptr[0].srcHost
     @srcHost.setter
     def srcHost(self, srcHost):
-        _csrcHost = utils.HelperInputVoidPtr(srcHost)
+        _csrcHost = _HelperInputVoidPtr(srcHost)
         self._pvt_ptr[0].srcHost = <void*><void_ptr>_csrcHost.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY3D_st.srcDevice' in found_struct}}
@@ -13753,7 +13755,7 @@ cdef class CUDA_MEMCPY3D_st:
         return <void_ptr>self._pvt_ptr[0].reserved0
     @reserved0.setter
     def reserved0(self, reserved0):
-        _creserved0 = utils.HelperInputVoidPtr(reserved0)
+        _creserved0 = _HelperInputVoidPtr(reserved0)
         self._pvt_ptr[0].reserved0 = <void*><void_ptr>_creserved0.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY3D_st.srcPitch' in found_struct}}
@@ -13820,7 +13822,7 @@ cdef class CUDA_MEMCPY3D_st:
         return <void_ptr>self._pvt_ptr[0].dstHost
     @dstHost.setter
     def dstHost(self, dstHost):
-        _cdstHost = utils.HelperInputVoidPtr(dstHost)
+        _cdstHost = _HelperInputVoidPtr(dstHost)
         self._pvt_ptr[0].dstHost = <void*><void_ptr>_cdstHost.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY3D_st.dstDevice' in found_struct}}
@@ -13864,7 +13866,7 @@ cdef class CUDA_MEMCPY3D_st:
         return <void_ptr>self._pvt_ptr[0].reserved1
     @reserved1.setter
     def reserved1(self, reserved1):
-        _creserved1 = utils.HelperInputVoidPtr(reserved1)
+        _creserved1 = _HelperInputVoidPtr(reserved1)
         self._pvt_ptr[0].reserved1 = <void*><void_ptr>_creserved1.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY3D_st.dstPitch' in found_struct}}
@@ -14257,7 +14259,7 @@ cdef class CUDA_MEMCPY3D_PEER_st:
         return <void_ptr>self._pvt_ptr[0].srcHost
     @srcHost.setter
     def srcHost(self, srcHost):
-        _csrcHost = utils.HelperInputVoidPtr(srcHost)
+        _csrcHost = _HelperInputVoidPtr(srcHost)
         self._pvt_ptr[0].srcHost = <void*><void_ptr>_csrcHost.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY3D_PEER_st.srcDevice' in found_struct}}
@@ -14376,7 +14378,7 @@ cdef class CUDA_MEMCPY3D_PEER_st:
         return <void_ptr>self._pvt_ptr[0].dstHost
     @dstHost.setter
     def dstHost(self, dstHost):
-        _cdstHost = utils.HelperInputVoidPtr(dstHost)
+        _cdstHost = _HelperInputVoidPtr(dstHost)
         self._pvt_ptr[0].dstHost = <void*><void_ptr>_cdstHost.cptr
     {{endif}}
     {{if 'CUDA_MEMCPY3D_PEER_st.dstDevice' in found_struct}}
@@ -16552,7 +16554,7 @@ cdef class CUDA_LAUNCH_PARAMS_st:
         return <void_ptr>self._pvt_ptr[0].kernelParams
     @kernelParams.setter
     def kernelParams(self, kernelParams):
-        self._cykernelParams = utils.HelperKernelParams(kernelParams)
+        self._cykernelParams = _HelperKernelParams(kernelParams)
         self._pvt_ptr[0].kernelParams = <void**><void_ptr>self._cykernelParams.ckernelParams
     {{endif}}
 {{endif}}
@@ -16609,7 +16611,7 @@ cdef class anon_struct12:
         return <void_ptr>self._pvt_ptr[0].handle.win32.handle
     @handle.setter
     def handle(self, handle):
-        _chandle = utils.HelperInputVoidPtr(handle)
+        _chandle = _HelperInputVoidPtr(handle)
         self._pvt_ptr[0].handle.win32.handle = <void*><void_ptr>_chandle.cptr
     {{endif}}
     {{if 'CUDA_EXTERNAL_MEMORY_HANDLE_DESC_st.handle.win32.name' in found_struct}}
@@ -16618,7 +16620,7 @@ cdef class anon_struct12:
         return <void_ptr>self._pvt_ptr[0].handle.win32.name
     @name.setter
     def name(self, name):
-        _cname = utils.HelperInputVoidPtr(name)
+        _cname = _HelperInputVoidPtr(name)
         self._pvt_ptr[0].handle.win32.name = <void*><void_ptr>_cname.cptr
     {{endif}}
 {{endif}}
@@ -16704,7 +16706,7 @@ cdef class anon_union5:
         return <void_ptr>self._pvt_ptr[0].handle.nvSciBufObject
     @nvSciBufObject.setter
     def nvSciBufObject(self, nvSciBufObject):
-        _cnvSciBufObject = utils.HelperInputVoidPtr(nvSciBufObject)
+        _cnvSciBufObject = _HelperInputVoidPtr(nvSciBufObject)
         self._pvt_ptr[0].handle.nvSciBufObject = <void*><void_ptr>_cnvSciBufObject.cptr
     {{endif}}
 {{endif}}
@@ -17102,7 +17104,7 @@ cdef class anon_struct13:
         return <void_ptr>self._pvt_ptr[0].handle.win32.handle
     @handle.setter
     def handle(self, handle):
-        _chandle = utils.HelperInputVoidPtr(handle)
+        _chandle = _HelperInputVoidPtr(handle)
         self._pvt_ptr[0].handle.win32.handle = <void*><void_ptr>_chandle.cptr
     {{endif}}
     {{if 'CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC_st.handle.win32.name' in found_struct}}
@@ -17111,7 +17113,7 @@ cdef class anon_struct13:
         return <void_ptr>self._pvt_ptr[0].handle.win32.name
     @name.setter
     def name(self, name):
-        _cname = utils.HelperInputVoidPtr(name)
+        _cname = _HelperInputVoidPtr(name)
         self._pvt_ptr[0].handle.win32.name = <void*><void_ptr>_cname.cptr
     {{endif}}
 {{endif}}
@@ -17197,7 +17199,7 @@ cdef class anon_union6:
         return <void_ptr>self._pvt_ptr[0].handle.nvSciSyncObj
     @nvSciSyncObj.setter
     def nvSciSyncObj(self, nvSciSyncObj):
-        _cnvSciSyncObj = utils.HelperInputVoidPtr(nvSciSyncObj)
+        _cnvSciSyncObj = _HelperInputVoidPtr(nvSciSyncObj)
         self._pvt_ptr[0].handle.nvSciSyncObj = <void*><void_ptr>_cnvSciSyncObj.cptr
     {{endif}}
 {{endif}}
@@ -17411,7 +17413,7 @@ cdef class anon_union7:
         return <void_ptr>self._pvt_ptr[0].params.nvSciSync.fence
     @fence.setter
     def fence(self, fence):
-        _cfence = utils.HelperInputVoidPtr(fence)
+        _cfence = _HelperInputVoidPtr(fence)
         self._pvt_ptr[0].params.nvSciSync.fence = <void*><void_ptr>_cfence.cptr
     {{endif}}
     {{if 'CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_st.params.nvSciSync.reserved' in found_struct}}
@@ -17773,7 +17775,7 @@ cdef class anon_union8:
         return <void_ptr>self._pvt_ptr[0].params.nvSciSync.fence
     @fence.setter
     def fence(self, fence):
-        _cfence = utils.HelperInputVoidPtr(fence)
+        _cfence = _HelperInputVoidPtr(fence)
         self._pvt_ptr[0].params.nvSciSync.fence = <void*><void_ptr>_cfence.cptr
     {{endif}}
     {{if 'CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_st.params.nvSciSync.reserved' in found_struct}}
@@ -19573,7 +19575,7 @@ cdef class CUmemAllocationProp_st:
         return <void_ptr>self._pvt_ptr[0].win32HandleMetaData
     @win32HandleMetaData.setter
     def win32HandleMetaData(self, win32HandleMetaData):
-        _cwin32HandleMetaData = utils.HelperInputVoidPtr(win32HandleMetaData)
+        _cwin32HandleMetaData = _HelperInputVoidPtr(win32HandleMetaData)
         self._pvt_ptr[0].win32HandleMetaData = <void*><void_ptr>_cwin32HandleMetaData.cptr
     {{endif}}
     {{if 'CUmemAllocationProp_st.allocFlags' in found_struct}}
@@ -20024,7 +20026,7 @@ cdef class CUmemPoolProps_st:
         return <void_ptr>self._pvt_ptr[0].win32SecurityAttributes
     @win32SecurityAttributes.setter
     def win32SecurityAttributes(self, win32SecurityAttributes):
-        _cwin32SecurityAttributes = utils.HelperInputVoidPtr(win32SecurityAttributes)
+        _cwin32SecurityAttributes = _HelperInputVoidPtr(win32SecurityAttributes)
         self._pvt_ptr[0].win32SecurityAttributes = <void*><void_ptr>_cwin32SecurityAttributes.cptr
     {{endif}}
     {{if 'CUmemPoolProps_st.maxSize' in found_struct}}
@@ -22246,7 +22248,7 @@ cdef class CUmemDecompressParams_st:
         return <void_ptr>self._pvt_ptr[0].src
     @src.setter
     def src(self, src):
-        _csrc = utils.HelperInputVoidPtr(src)
+        _csrc = _HelperInputVoidPtr(src)
         self._pvt_ptr[0].src = <void*><void_ptr>_csrc.cptr
     {{endif}}
     {{if 'CUmemDecompressParams_st.dst' in found_struct}}
@@ -22255,7 +22257,7 @@ cdef class CUmemDecompressParams_st:
         return <void_ptr>self._pvt_ptr[0].dst
     @dst.setter
     def dst(self, dst):
-        _cdst = utils.HelperInputVoidPtr(dst)
+        _cdst = _HelperInputVoidPtr(dst)
         self._pvt_ptr[0].dst = <void*><void_ptr>_cdst.cptr
     {{endif}}
     {{if 'CUmemDecompressParams_st.algo' in found_struct}}
@@ -23831,7 +23833,7 @@ def cuDeviceGetNvSciSyncAttributes(nvSciSyncAttrList, dev, int flags):
     else:
         pdev = int(CUdevice(dev))
     cydev = <cydriver.CUdevice>pdev
-    cynvSciSyncAttrList = utils.HelperInputVoidPtr(nvSciSyncAttrList)
+    cynvSciSyncAttrList = _HelperInputVoidPtr(nvSciSyncAttrList)
     cdef void* cynvSciSyncAttrList_ptr = <void*><void_ptr>cynvSciSyncAttrList.cptr
     with nogil:
         err = cydriver.cuDeviceGetNvSciSyncAttributes(cynvSciSyncAttrList_ptr, cydev, flags)
@@ -25937,7 +25939,7 @@ def cuModuleLoadData(image):
     :py:obj:`~.cuModuleGetFunction`, :py:obj:`~.cuModuleGetGlobal`, :py:obj:`~.cuModuleGetTexRef`, :py:obj:`~.cuModuleLoad`, :py:obj:`~.cuModuleLoadDataEx`, :py:obj:`~.cuModuleLoadFatBinary`, :py:obj:`~.cuModuleUnload`
     """
     cdef CUmodule module = CUmodule()
-    cyimage = utils.HelperInputVoidPtr(image)
+    cyimage = _HelperInputVoidPtr(image)
     cdef void* cyimage_ptr = <void*><void_ptr>cyimage.cptr
     with nogil:
         err = cydriver.cuModuleLoadData(<cydriver.CUmodule*>module._pvt_ptr, cyimage_ptr)
@@ -25984,13 +25986,13 @@ def cuModuleLoadDataEx(image, unsigned int numOptions, options : Optional[Tuple[
     if not all(isinstance(_x, (CUjit_option)) for _x in options):
         raise TypeError("Argument 'options' is not instance of type (expected Tuple[cydriver.CUjit_option] or List[cydriver.CUjit_option]")
     cdef CUmodule module = CUmodule()
-    cyimage = utils.HelperInputVoidPtr(image)
+    cyimage = _HelperInputVoidPtr(image)
     cdef void* cyimage_ptr = <void*><void_ptr>cyimage.cptr
     if numOptions > len(options): raise RuntimeError("List is too small: " + str(len(options)) + " < " + str(numOptions))
     if numOptions > len(optionValues): raise RuntimeError("List is too small: " + str(len(optionValues)) + " < " + str(numOptions))
     cdef vector[cydriver.CUjit_option] cyoptions = [pyoptions.value for pyoptions in (options)]
-    pylist = [utils.HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperoptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperoptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyoptionValues_ptr = <void**><void_ptr>voidStarHelperoptionValues.cptr
     with nogil:
         err = cydriver.cuModuleLoadDataEx(<cydriver.CUmodule*>module._pvt_ptr, cyimage_ptr, numOptions, cyoptions.data(), cyoptionValues_ptr)
@@ -26033,7 +26035,7 @@ def cuModuleLoadFatBinary(fatCubin):
     :py:obj:`~.cuModuleGetFunction`, :py:obj:`~.cuModuleGetGlobal`, :py:obj:`~.cuModuleGetTexRef`, :py:obj:`~.cuModuleLoad`, :py:obj:`~.cuModuleLoadData`, :py:obj:`~.cuModuleLoadDataEx`, :py:obj:`~.cuModuleUnload`
     """
     cdef CUmodule module = CUmodule()
-    cyfatCubin = utils.HelperInputVoidPtr(fatCubin)
+    cyfatCubin = _HelperInputVoidPtr(fatCubin)
     cdef void* cyfatCubin_ptr = <void*><void_ptr>cyfatCubin.cptr
     with nogil:
         err = cydriver.cuModuleLoadFatBinary(<cydriver.CUmodule*>module._pvt_ptr, cyfatCubin_ptr)
@@ -26357,8 +26359,8 @@ def cuLinkCreate(unsigned int numOptions, options : Optional[Tuple[CUjit_option]
     if numOptions > len(options): raise RuntimeError("List is too small: " + str(len(options)) + " < " + str(numOptions))
     if numOptions > len(optionValues): raise RuntimeError("List is too small: " + str(len(optionValues)) + " < " + str(numOptions))
     cdef vector[cydriver.CUjit_option] cyoptions = [pyoptions.value for pyoptions in (options)]
-    pylist = [utils.HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperoptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperoptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyoptionValues_ptr = <void**><void_ptr>voidStarHelperoptionValues.cptr
     cdef CUlinkState stateOut = CUlinkState()
     with nogil:
@@ -26432,13 +26434,13 @@ def cuLinkAddData(state, typename not None : CUjitInputType, data, size_t size,
         pstate = int(CUlinkState(state))
     cystate = <cydriver.CUlinkState><void_ptr>pstate
     cdef cydriver.CUjitInputType cytypename = typename.value
-    cydata = utils.HelperInputVoidPtr(data)
+    cydata = _HelperInputVoidPtr(data)
     cdef void* cydata_ptr = <void*><void_ptr>cydata.cptr
     if numOptions > len(options): raise RuntimeError("List is too small: " + str(len(options)) + " < " + str(numOptions))
     if numOptions > len(optionValues): raise RuntimeError("List is too small: " + str(len(optionValues)) + " < " + str(numOptions))
     cdef vector[cydriver.CUjit_option] cyoptions = [pyoptions.value for pyoptions in (options)]
-    pylist = [utils.HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperoptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperoptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyoptionValues_ptr = <void**><void_ptr>voidStarHelperoptionValues.cptr
     with nogil:
         err = cydriver.cuLinkAddData(cystate, cytypename, cydata_ptr, size, name, numOptions, cyoptions.data(), cyoptionValues_ptr)
@@ -26507,8 +26509,8 @@ def cuLinkAddFile(state, typename not None : CUjitInputType, char* path, unsigne
     if numOptions > len(options): raise RuntimeError("List is too small: " + str(len(options)) + " < " + str(numOptions))
     if numOptions > len(optionValues): raise RuntimeError("List is too small: " + str(len(optionValues)) + " < " + str(numOptions))
     cdef vector[cydriver.CUjit_option] cyoptions = [pyoptions.value for pyoptions in (options)]
-    pylist = [utils.HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperoptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(options, optionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperoptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyoptionValues_ptr = <void**><void_ptr>voidStarHelperoptionValues.cptr
     with nogil:
         err = cydriver.cuLinkAddFile(cystate, cytypename, path, numOptions, cyoptions.data(), cyoptionValues_ptr)
@@ -26768,17 +26770,17 @@ def cuLibraryLoadData(code, jitOptions : Optional[Tuple[CUjit_option] | List[CUj
     if not all(isinstance(_x, (CUjit_option)) for _x in jitOptions):
         raise TypeError("Argument 'jitOptions' is not instance of type (expected Tuple[cydriver.CUjit_option] or List[cydriver.CUjit_option]")
     cdef CUlibrary library = CUlibrary()
-    cycode = utils.HelperInputVoidPtr(code)
+    cycode = _HelperInputVoidPtr(code)
     cdef void* cycode_ptr = <void*><void_ptr>cycode.cptr
     cdef vector[cydriver.CUjit_option] cyjitOptions = [pyjitOptions.value for pyjitOptions in (jitOptions)]
-    pylist = [utils.HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyjitOptionsValues_ptr = <void**><void_ptr>voidStarHelperjitOptionsValues.cptr
     if numJitOptions > len(jitOptions): raise RuntimeError("List is too small: " + str(len(jitOptions)) + " < " + str(numJitOptions))
     if numJitOptions > len(jitOptionsValues): raise RuntimeError("List is too small: " + str(len(jitOptionsValues)) + " < " + str(numJitOptions))
     cdef vector[cydriver.CUlibraryOption] cylibraryOptions = [pylibraryOptions.value for pylibraryOptions in (libraryOptions)]
-    pylist = [utils.HelperCUlibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUlibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cylibraryOptionValues_ptr = <void**><void_ptr>voidStarHelperlibraryOptionValues.cptr
     if numLibraryOptions > len(libraryOptions): raise RuntimeError("List is too small: " + str(len(libraryOptions)) + " < " + str(numLibraryOptions))
     if numLibraryOptions > len(libraryOptionValues): raise RuntimeError("List is too small: " + str(len(libraryOptionValues)) + " < " + str(numLibraryOptions))
@@ -26868,14 +26870,14 @@ def cuLibraryLoadFromFile(char* fileName, jitOptions : Optional[Tuple[CUjit_opti
         raise TypeError("Argument 'jitOptions' is not instance of type (expected Tuple[cydriver.CUjit_option] or List[cydriver.CUjit_option]")
     cdef CUlibrary library = CUlibrary()
     cdef vector[cydriver.CUjit_option] cyjitOptions = [pyjitOptions.value for pyjitOptions in (jitOptions)]
-    pylist = [utils.HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUjit_option(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyjitOptionsValues_ptr = <void**><void_ptr>voidStarHelperjitOptionsValues.cptr
     if numJitOptions > len(jitOptions): raise RuntimeError("List is too small: " + str(len(jitOptions)) + " < " + str(numJitOptions))
     if numJitOptions > len(jitOptionsValues): raise RuntimeError("List is too small: " + str(len(jitOptionsValues)) + " < " + str(numJitOptions))
     cdef vector[cydriver.CUlibraryOption] cylibraryOptions = [pylibraryOptions.value for pylibraryOptions in (libraryOptions)]
-    pylist = [utils.HelperCUlibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUlibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cylibraryOptionValues_ptr = <void**><void_ptr>voidStarHelperlibraryOptionValues.cptr
     if numLibraryOptions > len(libraryOptions): raise RuntimeError("List is too small: " + str(len(libraryOptions)) + " < " + str(numLibraryOptions))
     if numLibraryOptions > len(libraryOptionValues): raise RuntimeError("List is too small: " + str(len(libraryOptionValues)) + " < " + str(numLibraryOptions))
@@ -28081,7 +28083,7 @@ def cuMemFreeHost(p):
     --------
     :py:obj:`~.cuArray3DCreate`, :py:obj:`~.cuArray3DGetDescriptor`, :py:obj:`~.cuArrayCreate`, :py:obj:`~.cuArrayDestroy`, :py:obj:`~.cuArrayGetDescriptor`, :py:obj:`~.cuMemAlloc`, :py:obj:`~.cuMemAllocHost`, :py:obj:`~.cuMemAllocPitch`, :py:obj:`~.cuMemcpy2D`, :py:obj:`~.cuMemcpy2DAsync`, :py:obj:`~.cuMemcpy2DUnaligned`, :py:obj:`~.cuMemcpy3D`, :py:obj:`~.cuMemcpy3DAsync`, :py:obj:`~.cuMemcpyAtoA`, :py:obj:`~.cuMemcpyAtoD`, :py:obj:`~.cuMemcpyAtoH`, :py:obj:`~.cuMemcpyAtoHAsync`, :py:obj:`~.cuMemcpyDtoA`, :py:obj:`~.cuMemcpyDtoD`, :py:obj:`~.cuMemcpyDtoDAsync`, :py:obj:`~.cuMemcpyDtoH`, :py:obj:`~.cuMemcpyDtoHAsync`, :py:obj:`~.cuMemcpyHtoA`, :py:obj:`~.cuMemcpyHtoAAsync`, :py:obj:`~.cuMemcpyHtoD`, :py:obj:`~.cuMemcpyHtoDAsync`, :py:obj:`~.cuMemFree`, :py:obj:`~.cuMemGetAddressRange`, :py:obj:`~.cuMemGetInfo`, :py:obj:`~.cuMemHostAlloc`, :py:obj:`~.cuMemHostGetDevicePointer`, :py:obj:`~.cuMemsetD2D8`, :py:obj:`~.cuMemsetD2D16`, :py:obj:`~.cuMemsetD2D32`, :py:obj:`~.cuMemsetD8`, :py:obj:`~.cuMemsetD16`, :py:obj:`~.cuMemsetD32`, :py:obj:`~.cudaFreeHost`
     """
-    cyp = utils.HelperInputVoidPtr(p)
+    cyp = _HelperInputVoidPtr(p)
     cdef void* cyp_ptr = <void*><void_ptr>cyp.cptr
     with nogil:
         err = cydriver.cuMemFreeHost(cyp_ptr)
@@ -28235,7 +28237,7 @@ def cuMemHostGetDevicePointer(p, unsigned int Flags):
     :py:obj:`~.cuArray3DCreate`, :py:obj:`~.cuArray3DGetDescriptor`, :py:obj:`~.cuArrayCreate`, :py:obj:`~.cuArrayDestroy`, :py:obj:`~.cuArrayGetDescriptor`, :py:obj:`~.cuMemAlloc`, :py:obj:`~.cuMemAllocHost`, :py:obj:`~.cuMemAllocPitch`, :py:obj:`~.cuMemcpy2D`, :py:obj:`~.cuMemcpy2DAsync`, :py:obj:`~.cuMemcpy2DUnaligned`, :py:obj:`~.cuMemcpy3D`, :py:obj:`~.cuMemcpy3DAsync`, :py:obj:`~.cuMemcpyAtoA`, :py:obj:`~.cuMemcpyAtoD`, :py:obj:`~.cuMemcpyAtoH`, :py:obj:`~.cuMemcpyAtoHAsync`, :py:obj:`~.cuMemcpyDtoA`, :py:obj:`~.cuMemcpyDtoD`, :py:obj:`~.cuMemcpyDtoDAsync`, :py:obj:`~.cuMemcpyDtoH`, :py:obj:`~.cuMemcpyDtoHAsync`, :py:obj:`~.cuMemcpyHtoA`, :py:obj:`~.cuMemcpyHtoAAsync`, :py:obj:`~.cuMemcpyHtoD`, :py:obj:`~.cuMemcpyHtoDAsync`, :py:obj:`~.cuMemFree`, :py:obj:`~.cuMemFreeHost`, :py:obj:`~.cuMemGetAddressRange`, :py:obj:`~.cuMemGetInfo`, :py:obj:`~.cuMemHostAlloc`, :py:obj:`~.cuMemsetD2D8`, :py:obj:`~.cuMemsetD2D16`, :py:obj:`~.cuMemsetD2D32`, :py:obj:`~.cuMemsetD8`, :py:obj:`~.cuMemsetD16`, :py:obj:`~.cuMemsetD32`, :py:obj:`~.cudaHostGetDevicePointer`
     """
     cdef CUdeviceptr pdptr = CUdeviceptr()
-    cyp = utils.HelperInputVoidPtr(p)
+    cyp = _HelperInputVoidPtr(p)
     cdef void* cyp_ptr = <void*><void_ptr>cyp.cptr
     with nogil:
         err = cydriver.cuMemHostGetDevicePointer(<cydriver.CUdeviceptr*>pdptr._pvt_ptr, cyp_ptr, Flags)
@@ -28274,7 +28276,7 @@ def cuMemHostGetFlags(p):
     :py:obj:`~.cuMemAllocHost`, :py:obj:`~.cuMemHostAlloc`, :py:obj:`~.cudaHostGetFlags`
     """
     cdef unsigned int pFlags = 0
-    cyp = utils.HelperInputVoidPtr(p)
+    cyp = _HelperInputVoidPtr(p)
     cdef void* cyp_ptr = <void*><void_ptr>cyp.cptr
     with nogil:
         err = cydriver.cuMemHostGetFlags(&pFlags, cyp_ptr)
@@ -28493,7 +28495,7 @@ def cuDeviceRegisterAsyncNotification(device, callbackFunc, userData):
     else:
         pdevice = int(CUdevice(device))
     cydevice = <cydriver.CUdevice>pdevice
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
 
     cdef cuAsyncCallbackData *cbData = NULL
@@ -29029,7 +29031,7 @@ def cuMemHostRegister(p, size_t bytesize, unsigned int Flags):
     --------
     :py:obj:`~.cuMemHostUnregister`, :py:obj:`~.cuMemHostGetFlags`, :py:obj:`~.cuMemHostGetDevicePointer`, :py:obj:`~.cudaHostRegister`
     """
-    cyp = utils.HelperInputVoidPtr(p)
+    cyp = _HelperInputVoidPtr(p)
     cdef void* cyp_ptr = <void*><void_ptr>cyp.cptr
     with nogil:
         err = cydriver.cuMemHostRegister(cyp_ptr, bytesize, Flags)
@@ -29062,7 +29064,7 @@ def cuMemHostUnregister(p):
     --------
     :py:obj:`~.cuMemHostRegister`, :py:obj:`~.cudaHostUnregister`
     """
-    cyp = utils.HelperInputVoidPtr(p)
+    cyp = _HelperInputVoidPtr(p)
     cdef void* cyp_ptr = <void*><void_ptr>cyp.cptr
     with nogil:
         err = cydriver.cuMemHostUnregister(cyp_ptr)
@@ -29228,7 +29230,7 @@ def cuMemcpyHtoD(dstDevice, srcHost, size_t ByteCount):
     else:
         pdstDevice = int(CUdeviceptr(dstDevice))
     cydstDevice = <cydriver.CUdeviceptr><void_ptr>pdstDevice
-    cysrcHost = utils.HelperInputVoidPtr(srcHost)
+    cysrcHost = _HelperInputVoidPtr(srcHost)
     cdef void* cysrcHost_ptr = <void*><void_ptr>cysrcHost.cptr
     with nogil:
         err = cydriver.cuMemcpyHtoD(cydstDevice, cysrcHost_ptr, ByteCount)
@@ -29271,7 +29273,7 @@ def cuMemcpyDtoH(dstHost, srcDevice, size_t ByteCount):
     else:
         psrcDevice = int(CUdeviceptr(srcDevice))
     cysrcDevice = <cydriver.CUdeviceptr><void_ptr>psrcDevice
-    cydstHost = utils.HelperInputVoidPtr(dstHost)
+    cydstHost = _HelperInputVoidPtr(dstHost)
     cdef void* cydstHost_ptr = <void*><void_ptr>cydstHost.cptr
     with nogil:
         err = cydriver.cuMemcpyDtoH(cydstHost_ptr, cysrcDevice, ByteCount)
@@ -29472,7 +29474,7 @@ def cuMemcpyHtoA(dstArray, size_t dstOffset, srcHost, size_t ByteCount):
     else:
         pdstArray = int(CUarray(dstArray))
     cydstArray = <cydriver.CUarray><void_ptr>pdstArray
-    cysrcHost = utils.HelperInputVoidPtr(srcHost)
+    cysrcHost = _HelperInputVoidPtr(srcHost)
     cdef void* cysrcHost_ptr = <void*><void_ptr>cysrcHost.cptr
     with nogil:
         err = cydriver.cuMemcpyHtoA(cydstArray, dstOffset, cysrcHost_ptr, ByteCount)
@@ -29518,7 +29520,7 @@ def cuMemcpyAtoH(dstHost, srcArray, size_t srcOffset, size_t ByteCount):
     else:
         psrcArray = int(CUarray(srcArray))
     cysrcArray = <cydriver.CUarray><void_ptr>psrcArray
-    cydstHost = utils.HelperInputVoidPtr(dstHost)
+    cydstHost = _HelperInputVoidPtr(dstHost)
     cdef void* cydstHost_ptr = <void*><void_ptr>cydstHost.cptr
     with nogil:
         err = cydriver.cuMemcpyAtoH(cydstHost_ptr, cysrcArray, srcOffset, ByteCount)
@@ -30185,7 +30187,7 @@ def cuMemcpyHtoDAsync(dstDevice, srcHost, size_t ByteCount, hStream):
     else:
         pdstDevice = int(CUdeviceptr(dstDevice))
     cydstDevice = <cydriver.CUdeviceptr><void_ptr>pdstDevice
-    cysrcHost = utils.HelperInputVoidPtr(srcHost)
+    cysrcHost = _HelperInputVoidPtr(srcHost)
     cdef void* cysrcHost_ptr = <void*><void_ptr>cysrcHost.cptr
     with nogil:
         err = cydriver.cuMemcpyHtoDAsync(cydstDevice, cysrcHost_ptr, ByteCount, cyhStream)
@@ -30238,7 +30240,7 @@ def cuMemcpyDtoHAsync(dstHost, srcDevice, size_t ByteCount, hStream):
     else:
         psrcDevice = int(CUdeviceptr(srcDevice))
     cysrcDevice = <cydriver.CUdeviceptr><void_ptr>psrcDevice
-    cydstHost = utils.HelperInputVoidPtr(dstHost)
+    cydstHost = _HelperInputVoidPtr(dstHost)
     cdef void* cydstHost_ptr = <void*><void_ptr>cydstHost.cptr
     with nogil:
         err = cydriver.cuMemcpyDtoHAsync(cydstHost_ptr, cysrcDevice, ByteCount, cyhStream)
@@ -30353,7 +30355,7 @@ def cuMemcpyHtoAAsync(dstArray, size_t dstOffset, srcHost, size_t ByteCount, hSt
     else:
         pdstArray = int(CUarray(dstArray))
     cydstArray = <cydriver.CUarray><void_ptr>pdstArray
-    cysrcHost = utils.HelperInputVoidPtr(srcHost)
+    cysrcHost = _HelperInputVoidPtr(srcHost)
     cdef void* cysrcHost_ptr = <void*><void_ptr>cysrcHost.cptr
     with nogil:
         err = cydriver.cuMemcpyHtoAAsync(cydstArray, dstOffset, cysrcHost_ptr, ByteCount, cyhStream)
@@ -30409,7 +30411,7 @@ def cuMemcpyAtoHAsync(dstHost, srcArray, size_t srcOffset, size_t ByteCount, hSt
     else:
         psrcArray = int(CUarray(srcArray))
     cysrcArray = <cydriver.CUarray><void_ptr>psrcArray
-    cydstHost = utils.HelperInputVoidPtr(dstHost)
+    cydstHost = _HelperInputVoidPtr(dstHost)
     cdef void* cydstHost_ptr = <void*><void_ptr>cydstHost.cptr
     with nogil:
         err = cydriver.cuMemcpyAtoHAsync(cydstHost_ptr, cysrcArray, srcOffset, ByteCount, cyhStream)
@@ -33314,7 +33316,7 @@ def cuMemExportToShareableHandle(handle, handleType not None : CUmemAllocationHa
     else:
         phandle = int(CUmemGenericAllocationHandle(handle))
     cyhandle = <cydriver.CUmemGenericAllocationHandle><void_ptr>phandle
-    cdef utils.HelperCUmemAllocationHandleType cyshareableHandle = utils.HelperCUmemAllocationHandleType(handleType)
+    cdef _HelperCUmemAllocationHandleType cyshareableHandle = _HelperCUmemAllocationHandleType(handleType)
     cdef void* cyshareableHandle_ptr = <void*><void_ptr>cyshareableHandle.cptr
     cdef cydriver.CUmemAllocationHandleType cyhandleType = handleType.value
     with nogil:
@@ -33364,7 +33366,7 @@ def cuMemImportFromShareableHandle(osHandle, shHandleType not None : CUmemAlloca
     Importing shareable handles exported from some graphics APIs(VUlkan, OpenGL, etc) created on devices under an SLI group may not be supported, and thus this API will return CUDA_ERROR_NOT_SUPPORTED. There is no guarantee that the contents of `handle` will be the same CUDA memory handle for the same given OS shareable handle, or the same underlying allocation.
     """
     cdef CUmemGenericAllocationHandle handle = CUmemGenericAllocationHandle()
-    cyosHandle = utils.HelperInputVoidPtr(osHandle)
+    cyosHandle = _HelperInputVoidPtr(osHandle)
     cdef void* cyosHandle_ptr = <void*><void_ptr>cyosHandle.cptr
     cdef cydriver.CUmemAllocationHandleType cyshHandleType = shHandleType.value
     with nogil:
@@ -33484,7 +33486,7 @@ def cuMemRetainAllocationHandle(addr):
     The address `addr`, can be any address in a range previously mapped by :py:obj:`~.cuMemMap`, and not necessarily the start address.
     """
     cdef CUmemGenericAllocationHandle handle = CUmemGenericAllocationHandle()
-    cyaddr = utils.HelperInputVoidPtr(addr)
+    cyaddr = _HelperInputVoidPtr(addr)
     cdef void* cyaddr_ptr = <void*><void_ptr>cyaddr.cptr
     with nogil:
         err = cydriver.cuMemRetainAllocationHandle(<cydriver.CUmemGenericAllocationHandle*>handle._pvt_ptr, cyaddr_ptr)
@@ -33713,7 +33715,7 @@ def cuMemPoolSetAttribute(pool, attr not None : CUmemPool_attribute, value):
         ppool = int(CUmemoryPool(pool))
     cypool = <cydriver.CUmemoryPool><void_ptr>ppool
     cdef cydriver.CUmemPool_attribute cyattr = attr.value
-    cdef utils.HelperCUmemPool_attribute cyvalue = utils.HelperCUmemPool_attribute(attr, value, is_getter=False)
+    cdef _HelperCUmemPool_attribute cyvalue = _HelperCUmemPool_attribute(attr, value, is_getter=False)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cydriver.cuMemPoolSetAttribute(cypool, cyattr, cyvalue_ptr)
@@ -33795,7 +33797,7 @@ def cuMemPoolGetAttribute(pool, attr not None : CUmemPool_attribute):
         ppool = int(CUmemoryPool(pool))
     cypool = <cydriver.CUmemoryPool><void_ptr>ppool
     cdef cydriver.CUmemPool_attribute cyattr = attr.value
-    cdef utils.HelperCUmemPool_attribute cyvalue = utils.HelperCUmemPool_attribute(attr, 0, is_getter=True)
+    cdef _HelperCUmemPool_attribute cyvalue = _HelperCUmemPool_attribute(attr, 0, is_getter=True)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cydriver.cuMemPoolGetAttribute(cypool, cyattr, cyvalue_ptr)
@@ -34301,7 +34303,7 @@ def cuMemPoolExportToShareableHandle(pool, handleType not None : CUmemAllocation
     else:
         ppool = int(CUmemoryPool(pool))
     cypool = <cydriver.CUmemoryPool><void_ptr>ppool
-    cdef utils.HelperCUmemAllocationHandleType cyhandle_out = utils.HelperCUmemAllocationHandleType(handleType)
+    cdef _HelperCUmemAllocationHandleType cyhandle_out = _HelperCUmemAllocationHandleType(handleType)
     cdef void* cyhandle_out_ptr = <void*><void_ptr>cyhandle_out.cptr
     cdef cydriver.CUmemAllocationHandleType cyhandleType = handleType.value
     with nogil:
@@ -34350,7 +34352,7 @@ def cuMemPoolImportFromShareableHandle(handle, handleType not None : CUmemAlloca
     Imported memory pools do not support creating new allocations. As such imported memory pools may not be used in cuDeviceSetMemPool or :py:obj:`~.cuMemAllocFromPoolAsync` calls.
     """
     cdef CUmemoryPool pool_out = CUmemoryPool()
-    cyhandle = utils.HelperInputVoidPtr(handle)
+    cyhandle = _HelperInputVoidPtr(handle)
     cdef void* cyhandle_ptr = <void*><void_ptr>cyhandle.cptr
     cdef cydriver.CUmemAllocationHandleType cyhandleType = handleType.value
     with nogil:
@@ -35030,7 +35032,7 @@ def cuPointerGetAttribute(attribute not None : CUpointer_attribute, ptr):
     else:
         pptr = int(CUdeviceptr(ptr))
     cyptr = <cydriver.CUdeviceptr><void_ptr>pptr
-    cdef utils.HelperCUpointer_attribute cydata = utils.HelperCUpointer_attribute(attribute, 0, is_getter=True)
+    cdef _HelperCUpointer_attribute cydata = _HelperCUpointer_attribute(attribute, 0, is_getter=True)
     cdef void* cydata_ptr = <void*><void_ptr>cydata.cptr
     cdef cydriver.CUpointer_attribute cyattribute = attribute.value
     with nogil:
@@ -35848,7 +35850,7 @@ def cuMemRangeGetAttribute(size_t dataSize, attribute not None : CUmem_range_att
     else:
         pdevPtr = int(CUdeviceptr(devPtr))
     cydevPtr = <cydriver.CUdeviceptr><void_ptr>pdevPtr
-    cdef utils.HelperCUmem_range_attribute cydata = utils.HelperCUmem_range_attribute(attribute, dataSize)
+    cdef _HelperCUmem_range_attribute cydata = _HelperCUmem_range_attribute(attribute, dataSize)
     cdef void* cydata_ptr = <void*><void_ptr>cydata.cptr
     cdef cydriver.CUmem_range_attribute cyattribute = attribute.value
     with nogil:
@@ -35931,8 +35933,8 @@ def cuMemRangeGetAttributes(dataSizes : Tuple[int] | List[int], attributes : Opt
         raise TypeError("Argument 'attributes' is not instance of type (expected Tuple[cydriver.CUmem_range_attribute] or List[cydriver.CUmem_range_attribute]")
     if not all(isinstance(_x, (int)) for _x in dataSizes):
         raise TypeError("Argument 'dataSizes' is not instance of type (expected Tuple[int] or List[int]")
-    pylist = [utils.HelperCUmem_range_attribute(pyattributes, pydataSizes) for (pyattributes, pydataSizes) in zip(attributes, dataSizes)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperdata = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUmem_range_attribute(pyattributes, pydataSizes) for (pyattributes, pydataSizes) in zip(attributes, dataSizes)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperdata = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyvoidStarHelper_ptr = <void**><void_ptr>voidStarHelperdata.cptr
     cdef vector[size_t] cydataSizes = dataSizes
     cdef vector[cydriver.CUmem_range_attribute] cyattributes = [pyattributes.value for pyattributes in (attributes)]
@@ -35993,7 +35995,7 @@ def cuPointerSetAttribute(value, attribute not None : CUpointer_attribute, ptr):
     else:
         pptr = int(CUdeviceptr(ptr))
     cyptr = <cydriver.CUdeviceptr><void_ptr>pptr
-    cdef utils.HelperCUpointer_attribute cyvalue = utils.HelperCUpointer_attribute(attribute, value, is_getter=False)
+    cdef _HelperCUpointer_attribute cyvalue = _HelperCUpointer_attribute(attribute, value, is_getter=False)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     cdef cydriver.CUpointer_attribute cyattribute = attribute.value
     with nogil:
@@ -36085,8 +36087,8 @@ def cuPointerGetAttributes(unsigned int numAttributes, attributes : Optional[Tup
         raise TypeError("Argument 'attributes' is not instance of type (expected Tuple[cydriver.CUpointer_attribute] or List[cydriver.CUpointer_attribute]")
     if numAttributes > len(attributes): raise RuntimeError("List is too small: " + str(len(attributes)) + " < " + str(numAttributes))
     cdef vector[cydriver.CUpointer_attribute] cyattributes = [pyattributes.value for pyattributes in (attributes)]
-    pylist = [utils.HelperCUpointer_attribute(pyattributes, 0, is_getter=True) for pyattributes in attributes]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperdata = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUpointer_attribute(pyattributes, 0, is_getter=True) for pyattributes in attributes]
+    cdef _InputVoidPtrPtrHelper voidStarHelperdata = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyvoidStarHelper_ptr = <void**><void_ptr>voidStarHelperdata.cptr
     with nogil:
         err = cydriver.cuPointerGetAttributes(numAttributes, cyattributes.data(), cyvoidStarHelper_ptr, cyptr)
@@ -36684,7 +36686,7 @@ def cuStreamAddCallback(hStream, callback, userData, unsigned int flags):
     else:
         phStream = int(CUstream(hStream))
     cyhStream = <cydriver.CUstream><void_ptr>phStream
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
 
     cdef cuStreamCallbackData *cbData = NULL
@@ -39793,7 +39795,7 @@ def cuLaunchKernel(f, unsigned int gridDimX, unsigned int gridDimY, unsigned int
     else:
         pf = int(CUfunction(f))
     cyf = <cydriver.CUfunction><void_ptr>pf
-    cykernelParams = utils.HelperKernelParams(kernelParams)
+    cykernelParams = _HelperKernelParams(kernelParams)
     cdef void** cykernelParams_ptr = <void**><void_ptr>cykernelParams.ckernelParams
     with nogil:
         err = cydriver.cuLaunchKernel(cyf, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ, sharedMemBytes, cyhStream, cykernelParams_ptr, <void**>extra)
@@ -40034,7 +40036,7 @@ def cuLaunchKernelEx(config : Optional[CUlaunchConfig], f, kernelParams, void_pt
         pf = int(CUfunction(f))
     cyf = <cydriver.CUfunction><void_ptr>pf
     cdef cydriver.CUlaunchConfig* cyconfig_ptr = config._pvt_ptr if config != None else NULL
-    cykernelParams = utils.HelperKernelParams(kernelParams)
+    cykernelParams = _HelperKernelParams(kernelParams)
     cdef void** cykernelParams_ptr = <void**><void_ptr>cykernelParams.ckernelParams
     with nogil:
         err = cydriver.cuLaunchKernelEx(cyconfig_ptr, cyf, cykernelParams_ptr, <void**>extra)
@@ -40147,7 +40149,7 @@ def cuLaunchCooperativeKernel(f, unsigned int gridDimX, unsigned int gridDimY, u
     else:
         pf = int(CUfunction(f))
     cyf = <cydriver.CUfunction><void_ptr>pf
-    cykernelParams = utils.HelperKernelParams(kernelParams)
+    cykernelParams = _HelperKernelParams(kernelParams)
     cdef void** cykernelParams_ptr = <void**><void_ptr>cykernelParams.ckernelParams
     with nogil:
         err = cydriver.cuLaunchCooperativeKernel(cyf, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ, sharedMemBytes, cyhStream, cykernelParams_ptr)
@@ -40420,7 +40422,7 @@ def cuLaunchHostFunc(hStream, fn, userData):
     else:
         phStream = int(CUstream(hStream))
     cyhStream = <cydriver.CUstream><void_ptr>phStream
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
 
     cdef cuHostCallbackData *cbData = NULL
@@ -40688,7 +40690,7 @@ def cuParamSetv(hfunc, int offset, ptr, unsigned int numbytes):
     else:
         phfunc = int(CUfunction(hfunc))
     cyhfunc = <cydriver.CUfunction><void_ptr>phfunc
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cydriver.cuParamSetv(cyhfunc, offset, cyptr_ptr, numbytes)
@@ -43163,7 +43165,7 @@ def cuDeviceGetGraphMemAttribute(device, attr not None : CUgraphMem_attribute):
         pdevice = int(CUdevice(device))
     cydevice = <cydriver.CUdevice>pdevice
     cdef cydriver.CUgraphMem_attribute cyattr = attr.value
-    cdef utils.HelperCUgraphMem_attribute cyvalue = utils.HelperCUgraphMem_attribute(attr, 0, is_getter=True)
+    cdef _HelperCUgraphMem_attribute cyvalue = _HelperCUgraphMem_attribute(attr, 0, is_getter=True)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cydriver.cuDeviceGetGraphMemAttribute(cydevice, cyattr, cyvalue_ptr)
@@ -43215,7 +43217,7 @@ def cuDeviceSetGraphMemAttribute(device, attr not None : CUgraphMem_attribute, v
         pdevice = int(CUdevice(device))
     cydevice = <cydriver.CUdevice>pdevice
     cdef cydriver.CUgraphMem_attribute cyattr = attr.value
-    cdef utils.HelperCUgraphMem_attribute cyvalue = utils.HelperCUgraphMem_attribute(attr, value, is_getter=False)
+    cdef _HelperCUgraphMem_attribute cyvalue = _HelperCUgraphMem_attribute(attr, value, is_getter=False)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cydriver.cuDeviceSetGraphMemAttribute(cydevice, cyattr, cyvalue_ptr)
@@ -45574,7 +45576,7 @@ def cuUserObjectCreate(ptr, destroy, unsigned int initialRefcount, unsigned int
         pdestroy = int(CUhostFn(destroy))
     cydestroy = <cydriver.CUhostFn><void_ptr>pdestroy
     cdef CUuserObject object_out = CUuserObject()
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cydriver.cuUserObjectCreate(<cydriver.CUuserObject*>object_out._pvt_ptr, cyptr_ptr, cydestroy, initialRefcount, flags)
@@ -48667,7 +48669,7 @@ def cuTensorMapEncodeTiled(tensorDataType not None : CUtensorMapDataType, tensor
     cytensorRank = <cydriver.cuuint32_t><void_ptr>ptensorRank
     cdef CUtensorMap tensorMap = CUtensorMap()
     cdef cydriver.CUtensorMapDataType cytensorDataType = tensorDataType.value
-    cyglobalAddress = utils.HelperInputVoidPtr(globalAddress)
+    cyglobalAddress = _HelperInputVoidPtr(globalAddress)
     cdef void* cyglobalAddress_ptr = <void*><void_ptr>cyglobalAddress.cptr
     cdef cydriver.cuuint64_t* cyglobalDim = NULL
     if len(globalDim) > 1:
@@ -49024,7 +49026,7 @@ def cuTensorMapEncodeIm2col(tensorDataType not None : CUtensorMapDataType, tenso
     cytensorRank = <cydriver.cuuint32_t><void_ptr>ptensorRank
     cdef CUtensorMap tensorMap = CUtensorMap()
     cdef cydriver.CUtensorMapDataType cytensorDataType = tensorDataType.value
-    cyglobalAddress = utils.HelperInputVoidPtr(globalAddress)
+    cyglobalAddress = _HelperInputVoidPtr(globalAddress)
     cdef void* cyglobalAddress_ptr = <void*><void_ptr>cyglobalAddress.cptr
     cdef cydriver.cuuint64_t* cyglobalDim = NULL
     if len(globalDim) > 1:
@@ -49352,7 +49354,7 @@ def cuTensorMapEncodeIm2colWide(tensorDataType not None : CUtensorMapDataType, t
     cytensorRank = <cydriver.cuuint32_t><void_ptr>ptensorRank
     cdef CUtensorMap tensorMap = CUtensorMap()
     cdef cydriver.CUtensorMapDataType cytensorDataType = tensorDataType.value
-    cyglobalAddress = utils.HelperInputVoidPtr(globalAddress)
+    cyglobalAddress = _HelperInputVoidPtr(globalAddress)
     cdef void* cyglobalAddress_ptr = <void*><void_ptr>cyglobalAddress.cptr
     cdef cydriver.cuuint64_t* cyglobalDim = NULL
     if len(globalDim) > 1:
@@ -49433,7 +49435,7 @@ def cuTensorMapReplaceAddress(tensorMap : Optional[CUtensorMap], globalAddress):
     :py:obj:`~.cuTensorMapEncodeTiled`, :py:obj:`~.cuTensorMapEncodeIm2col`, :py:obj:`~.cuTensorMapEncodeIm2colWide`
     """
     cdef cydriver.CUtensorMap* cytensorMap_ptr = tensorMap._pvt_ptr if tensorMap != None else NULL
-    cyglobalAddress = utils.HelperInputVoidPtr(globalAddress)
+    cyglobalAddress = _HelperInputVoidPtr(globalAddress)
     cdef void* cyglobalAddress_ptr = <void*><void_ptr>cyglobalAddress.cptr
     with nogil:
         err = cydriver.cuTensorMapReplaceAddress(cytensorMap_ptr, cyglobalAddress_ptr)
@@ -50358,7 +50360,7 @@ def cuCoredumpGetAttribute(attrib not None : CUcoredumpSettings):
     :py:obj:`~.cuCoredumpGetAttributeGlobal`, :py:obj:`~.cuCoredumpSetAttribute`, :py:obj:`~.cuCoredumpSetAttributeGlobal`
     """
     cdef cydriver.CUcoredumpSettings cyattrib = attrib.value
-    cdef utils.HelperCUcoredumpSettings cyvalue = utils.HelperCUcoredumpSettings(attrib, 0, is_getter=True)
+    cdef _HelperCUcoredumpSettings cyvalue = _HelperCUcoredumpSettings(attrib, 0, is_getter=True)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     cdef size_t size = cyvalue.size()
     with nogil:
@@ -50474,7 +50476,7 @@ def cuCoredumpGetAttributeGlobal(attrib not None : CUcoredumpSettings):
     :py:obj:`~.cuCoredumpGetAttribute`, :py:obj:`~.cuCoredumpSetAttribute`, :py:obj:`~.cuCoredumpSetAttributeGlobal`
     """
     cdef cydriver.CUcoredumpSettings cyattrib = attrib.value
-    cdef utils.HelperCUcoredumpSettings cyvalue = utils.HelperCUcoredumpSettings(attrib, 0, is_getter=True)
+    cdef _HelperCUcoredumpSettings cyvalue = _HelperCUcoredumpSettings(attrib, 0, is_getter=True)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     cdef size_t size = cyvalue.size()
     with nogil:
@@ -50597,7 +50599,7 @@ def cuCoredumpSetAttribute(attrib not None : CUcoredumpSettings, value):
     :py:obj:`~.cuCoredumpGetAttributeGlobal`, :py:obj:`~.cuCoredumpGetAttribute`, :py:obj:`~.cuCoredumpSetAttributeGlobal`
     """
     cdef cydriver.CUcoredumpSettings cyattrib = attrib.value
-    cdef utils.HelperCUcoredumpSettings cyvalue = utils.HelperCUcoredumpSettings(attrib, value, is_getter=False)
+    cdef _HelperCUcoredumpSettings cyvalue = _HelperCUcoredumpSettings(attrib, value, is_getter=False)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     cdef size_t size = cyvalue.size()
     with nogil:
@@ -50723,7 +50725,7 @@ def cuCoredumpSetAttributeGlobal(attrib not None : CUcoredumpSettings, value):
     :py:obj:`~.cuCoredumpGetAttribute`, :py:obj:`~.cuCoredumpGetAttributeGlobal`, :py:obj:`~.cuCoredumpSetAttribute`
     """
     cdef cydriver.CUcoredumpSettings cyattrib = attrib.value
-    cdef utils.HelperCUcoredumpSettings cyvalue = utils.HelperCUcoredumpSettings(attrib, value, is_getter=False)
+    cdef _HelperCUcoredumpSettings cyvalue = _HelperCUcoredumpSettings(attrib, value, is_getter=False)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     cdef size_t size = cyvalue.size()
     with nogil:
@@ -51585,7 +51587,7 @@ def cuLogsRegisterCallback(callbackFunc, userData):
     else:
         pcallbackFunc = int(CUlogsCallback(callbackFunc))
     cycallbackFunc = <cydriver.CUlogsCallback><void_ptr>pcallbackFunc
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
 
     cdef cuLogsCallbackData *cbData = NULL
diff --git a/cuda_bindings/cuda/bindings/nvrtc.pxd.in b/cuda_bindings/cuda/bindings/nvrtc.pxd.in
index c9f797520..e1f030921 100644
--- a/cuda_bindings/cuda/bindings/nvrtc.pxd.in
+++ b/cuda_bindings/cuda/bindings/nvrtc.pxd.in
@@ -3,7 +3,8 @@
 
 # This code was automatically generated with version 13.0.0. Do not modify it directly.
 cimport cuda.bindings.cynvrtc as cynvrtc
-cimport cuda.bindings._lib.utils as utils
+
+include "_lib/utils.pxd"
 
 {{if 'nvrtcProgram' in found_types}}
 
diff --git a/cuda_bindings/cuda/bindings/nvrtc.pyx.in b/cuda_bindings/cuda/bindings/nvrtc.pyx.in
index f852867a3..e2e2fb361 100644
--- a/cuda_bindings/cuda/bindings/nvrtc.pyx.in
+++ b/cuda_bindings/cuda/bindings/nvrtc.pyx.in
@@ -15,6 +15,8 @@ from libcpp.vector cimport vector
 from cpython.buffer cimport PyObject_CheckBuffer, PyObject_GetBuffer, PyBuffer_Release, PyBUF_SIMPLE, PyBUF_ANY_CONTIGUOUS
 from cpython.bytes cimport PyBytes_FromStringAndSize
 
+include "_lib/utils.pxi"
+
 ctypedef unsigned long long signed_char_ptr
 ctypedef unsigned long long unsigned_char_ptr
 ctypedef unsigned long long char_ptr
@@ -1026,9 +1028,9 @@ def nvrtcSetFlowCallback(prog, callback, payload):
     else:
         pprog = int(nvrtcProgram(prog))
     cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog
-    cycallback = utils.HelperInputVoidPtr(callback)
+    cycallback = _HelperInputVoidPtr(callback)
     cdef void* cycallback_ptr = <void*><void_ptr>cycallback.cptr
-    cypayload = utils.HelperInputVoidPtr(payload)
+    cypayload = _HelperInputVoidPtr(payload)
     cdef void* cypayload_ptr = <void*><void_ptr>cypayload.cptr
     with nogil:
         err = cynvrtc.nvrtcSetFlowCallback(cyprog, cycallback_ptr, cypayload_ptr)
diff --git a/cuda_bindings/cuda/bindings/runtime.pxd.in b/cuda_bindings/cuda/bindings/runtime.pxd.in
index 29687849b..05a7b8df5 100644
--- a/cuda_bindings/cuda/bindings/runtime.pxd.in
+++ b/cuda_bindings/cuda/bindings/runtime.pxd.in
@@ -3,7 +3,8 @@
 
 # This code was automatically generated with version 13.0.0. Do not modify it directly.
 cimport cuda.bindings.cyruntime as cyruntime
-cimport cuda.bindings._lib.utils as utils
+
+include "_lib/utils.pxd"
 cimport cuda.bindings.driver as driver
 
 {{if 'cudaArray_t' in found_types}}
@@ -2948,7 +2949,7 @@ cdef class cudaKernelNodeParams:
     cdef dim3 _blockDim
     {{endif}}
     {{if 'cudaKernelNodeParams.kernelParams' in found_struct}}
-    cdef utils.HelperKernelParams _cykernelParams
+    cdef _HelperKernelParams _cykernelParams
     {{endif}}
 {{endif}}
 {{if 'cudaKernelNodeParamsV2' in found_struct}}
@@ -2998,7 +2999,7 @@ cdef class cudaKernelNodeParamsV2:
     cdef dim3 _blockDim
     {{endif}}
     {{if 'cudaKernelNodeParamsV2.kernelParams' in found_struct}}
-    cdef utils.HelperKernelParams _cykernelParams
+    cdef _HelperKernelParams _cykernelParams
     {{endif}}
 {{endif}}
 {{if 'cudaExternalSemaphoreSignalNodeParams' in found_struct}}
diff --git a/cuda_bindings/cuda/bindings/runtime.pyx.in b/cuda_bindings/cuda/bindings/runtime.pyx.in
index ced5d39a6..f17436058 100644
--- a/cuda_bindings/cuda/bindings/runtime.pyx.in
+++ b/cuda_bindings/cuda/bindings/runtime.pyx.in
@@ -17,6 +17,8 @@ from cpython.bytes cimport PyBytes_FromStringAndSize
 import cuda.bindings.driver
 from libcpp.map cimport map
 
+include "_lib/utils.pxi"
+
 ctypedef unsigned long long signed_char_ptr
 ctypedef unsigned long long unsigned_char_ptr
 ctypedef unsigned long long char_ptr
@@ -6578,7 +6580,7 @@ cdef class cudaPitchedPtr:
         return <void_ptr>self._pvt_ptr[0].ptr
     @ptr.setter
     def ptr(self, ptr):
-        _cptr = utils.HelperInputVoidPtr(ptr)
+        _cptr = _HelperInputVoidPtr(ptr)
         self._pvt_ptr[0].ptr = <void*><void_ptr>_cptr.cptr
     {{endif}}
     {{if 'cudaPitchedPtr.pitch' in found_struct}}
@@ -7415,7 +7417,7 @@ cdef class cudaMemsetParams:
         return <void_ptr>self._pvt_ptr[0].dst
     @dst.setter
     def dst(self, dst):
-        _cdst = utils.HelperInputVoidPtr(dst)
+        _cdst = _HelperInputVoidPtr(dst)
         self._pvt_ptr[0].dst = <void*><void_ptr>_cdst.cptr
     {{endif}}
     {{if 'cudaMemsetParams.pitch' in found_struct}}
@@ -7556,7 +7558,7 @@ cdef class cudaMemsetParamsV2:
         return <void_ptr>self._pvt_ptr[0].dst
     @dst.setter
     def dst(self, dst):
-        _cdst = utils.HelperInputVoidPtr(dst)
+        _cdst = _HelperInputVoidPtr(dst)
         self._pvt_ptr[0].dst = <void*><void_ptr>_cdst.cptr
     {{endif}}
     {{if 'cudaMemsetParamsV2.pitch' in found_struct}}
@@ -7698,7 +7700,7 @@ cdef class cudaAccessPolicyWindow:
         return <void_ptr>self._pvt_ptr[0].base_ptr
     @base_ptr.setter
     def base_ptr(self, base_ptr):
-        _cbase_ptr = utils.HelperInputVoidPtr(base_ptr)
+        _cbase_ptr = _HelperInputVoidPtr(base_ptr)
         self._pvt_ptr[0].base_ptr = <void*><void_ptr>_cbase_ptr.cptr
     {{endif}}
     {{if 'cudaAccessPolicyWindow.num_bytes' in found_struct}}
@@ -7815,7 +7817,7 @@ cdef class cudaHostNodeParams:
         return <void_ptr>self._pvt_ptr[0].userData
     @userData.setter
     def userData(self, userData):
-        _cuserData = utils.HelperInputVoidPtr(userData)
+        _cuserData = _HelperInputVoidPtr(userData)
         self._pvt_ptr[0].userData = <void*><void_ptr>_cuserData.cptr
     {{endif}}
 {{endif}}
@@ -7896,7 +7898,7 @@ cdef class cudaHostNodeParamsV2:
         return <void_ptr>self._pvt_ptr[0].userData
     @userData.setter
     def userData(self, userData):
-        _cuserData = utils.HelperInputVoidPtr(userData)
+        _cuserData = _HelperInputVoidPtr(userData)
         self._pvt_ptr[0].userData = <void*><void_ptr>_cuserData.cptr
     {{endif}}
 {{endif}}
@@ -8082,7 +8084,7 @@ cdef class anon_struct3:
         return <void_ptr>self._pvt_ptr[0].res.linear.devPtr
     @devPtr.setter
     def devPtr(self, devPtr):
-        _cdevPtr = utils.HelperInputVoidPtr(devPtr)
+        _cdevPtr = _HelperInputVoidPtr(devPtr)
         self._pvt_ptr[0].res.linear.devPtr = <void*><void_ptr>_cdevPtr.cptr
     {{endif}}
     {{if 'cudaResourceDesc.res.linear.desc' in found_struct}}
@@ -8188,7 +8190,7 @@ cdef class anon_struct4:
         return <void_ptr>self._pvt_ptr[0].res.pitch2D.devPtr
     @devPtr.setter
     def devPtr(self, devPtr):
-        _cdevPtr = utils.HelperInputVoidPtr(devPtr)
+        _cdevPtr = _HelperInputVoidPtr(devPtr)
         self._pvt_ptr[0].res.pitch2D.devPtr = <void*><void_ptr>_cdevPtr.cptr
     {{endif}}
     {{if 'cudaResourceDesc.res.pitch2D.desc' in found_struct}}
@@ -8808,7 +8810,7 @@ cdef class cudaPointerAttributes:
         return <void_ptr>self._pvt_ptr[0].devicePointer
     @devicePointer.setter
     def devicePointer(self, devicePointer):
-        _cdevicePointer = utils.HelperInputVoidPtr(devicePointer)
+        _cdevicePointer = _HelperInputVoidPtr(devicePointer)
         self._pvt_ptr[0].devicePointer = <void*><void_ptr>_cdevicePointer.cptr
     {{endif}}
     {{if 'cudaPointerAttributes.hostPointer' in found_struct}}
@@ -8817,7 +8819,7 @@ cdef class cudaPointerAttributes:
         return <void_ptr>self._pvt_ptr[0].hostPointer
     @hostPointer.setter
     def hostPointer(self, hostPointer):
-        _chostPointer = utils.HelperInputVoidPtr(hostPointer)
+        _chostPointer = _HelperInputVoidPtr(hostPointer)
         self._pvt_ptr[0].hostPointer = <void*><void_ptr>_chostPointer.cptr
     {{endif}}
     {{if 'cudaPointerAttributes.reserved' in found_struct}}
@@ -9494,7 +9496,7 @@ cdef class cudaMemPoolProps:
         return <void_ptr>self._pvt_ptr[0].win32SecurityAttributes
     @win32SecurityAttributes.setter
     def win32SecurityAttributes(self, win32SecurityAttributes):
-        _cwin32SecurityAttributes = utils.HelperInputVoidPtr(win32SecurityAttributes)
+        _cwin32SecurityAttributes = _HelperInputVoidPtr(win32SecurityAttributes)
         self._pvt_ptr[0].win32SecurityAttributes = <void*><void_ptr>_cwin32SecurityAttributes.cptr
     {{endif}}
     {{if 'cudaMemPoolProps.maxSize' in found_struct}}
@@ -9723,7 +9725,7 @@ cdef class cudaMemAllocNodeParams:
         return <void_ptr>self._pvt_ptr[0].dptr
     @dptr.setter
     def dptr(self, dptr):
-        _cdptr = utils.HelperInputVoidPtr(dptr)
+        _cdptr = _HelperInputVoidPtr(dptr)
         self._pvt_ptr[0].dptr = <void*><void_ptr>_cdptr.cptr
     {{endif}}
 {{endif}}
@@ -9872,7 +9874,7 @@ cdef class cudaMemAllocNodeParamsV2:
         return <void_ptr>self._pvt_ptr[0].dptr
     @dptr.setter
     def dptr(self, dptr):
-        _cdptr = utils.HelperInputVoidPtr(dptr)
+        _cdptr = _HelperInputVoidPtr(dptr)
         self._pvt_ptr[0].dptr = <void*><void_ptr>_cdptr.cptr
     {{endif}}
 {{endif}}
@@ -9923,7 +9925,7 @@ cdef class cudaMemFreeNodeParams:
         return <void_ptr>self._pvt_ptr[0].dptr
     @dptr.setter
     def dptr(self, dptr):
-        _cdptr = utils.HelperInputVoidPtr(dptr)
+        _cdptr = _HelperInputVoidPtr(dptr)
         self._pvt_ptr[0].dptr = <void*><void_ptr>_cdptr.cptr
     {{endif}}
 {{endif}}
@@ -10206,7 +10208,7 @@ cdef class anon_struct6:
         return <void_ptr>self._pvt_ptr[0].op.ptr.ptr
     @ptr.setter
     def ptr(self, ptr):
-        _cptr = utils.HelperInputVoidPtr(ptr)
+        _cptr = _HelperInputVoidPtr(ptr)
         self._pvt_ptr[0].op.ptr.ptr = <void*><void_ptr>_cptr.cptr
     {{endif}}
     {{if 'cudaMemcpy3DOperand.op.ptr.rowLength' in found_struct}}
@@ -12613,7 +12615,7 @@ cdef class anon_struct8:
         return <void_ptr>self._pvt_ptr[0].handle.win32.handle
     @handle.setter
     def handle(self, handle):
-        _chandle = utils.HelperInputVoidPtr(handle)
+        _chandle = _HelperInputVoidPtr(handle)
         self._pvt_ptr[0].handle.win32.handle = <void*><void_ptr>_chandle.cptr
     {{endif}}
     {{if 'cudaExternalMemoryHandleDesc.handle.win32.name' in found_struct}}
@@ -12622,7 +12624,7 @@ cdef class anon_struct8:
         return <void_ptr>self._pvt_ptr[0].handle.win32.name
     @name.setter
     def name(self, name):
-        _cname = utils.HelperInputVoidPtr(name)
+        _cname = _HelperInputVoidPtr(name)
         self._pvt_ptr[0].handle.win32.name = <void*><void_ptr>_cname.cptr
     {{endif}}
 {{endif}}
@@ -12708,7 +12710,7 @@ cdef class anon_union2:
         return <void_ptr>self._pvt_ptr[0].handle.nvSciBufObject
     @nvSciBufObject.setter
     def nvSciBufObject(self, nvSciBufObject):
-        _cnvSciBufObject = utils.HelperInputVoidPtr(nvSciBufObject)
+        _cnvSciBufObject = _HelperInputVoidPtr(nvSciBufObject)
         self._pvt_ptr[0].handle.nvSciBufObject = <void*><void_ptr>_cnvSciBufObject.cptr
     {{endif}}
 {{endif}}
@@ -13146,7 +13148,7 @@ cdef class anon_struct9:
         return <void_ptr>self._pvt_ptr[0].handle.win32.handle
     @handle.setter
     def handle(self, handle):
-        _chandle = utils.HelperInputVoidPtr(handle)
+        _chandle = _HelperInputVoidPtr(handle)
         self._pvt_ptr[0].handle.win32.handle = <void*><void_ptr>_chandle.cptr
     {{endif}}
     {{if 'cudaExternalSemaphoreHandleDesc.handle.win32.name' in found_struct}}
@@ -13155,7 +13157,7 @@ cdef class anon_struct9:
         return <void_ptr>self._pvt_ptr[0].handle.win32.name
     @name.setter
     def name(self, name):
-        _cname = utils.HelperInputVoidPtr(name)
+        _cname = _HelperInputVoidPtr(name)
         self._pvt_ptr[0].handle.win32.name = <void*><void_ptr>_cname.cptr
     {{endif}}
 {{endif}}
@@ -13241,7 +13243,7 @@ cdef class anon_union3:
         return <void_ptr>self._pvt_ptr[0].handle.nvSciSyncObj
     @nvSciSyncObj.setter
     def nvSciSyncObj(self, nvSciSyncObj):
-        _cnvSciSyncObj = utils.HelperInputVoidPtr(nvSciSyncObj)
+        _cnvSciSyncObj = _HelperInputVoidPtr(nvSciSyncObj)
         self._pvt_ptr[0].handle.nvSciSyncObj = <void*><void_ptr>_cnvSciSyncObj.cptr
     {{endif}}
 {{endif}}
@@ -13455,7 +13457,7 @@ cdef class anon_union4:
         return <void_ptr>self._pvt_ptr[0].params.nvSciSync.fence
     @fence.setter
     def fence(self, fence):
-        _cfence = utils.HelperInputVoidPtr(fence)
+        _cfence = _HelperInputVoidPtr(fence)
         self._pvt_ptr[0].params.nvSciSync.fence = <void*><void_ptr>_cfence.cptr
     {{endif}}
     {{if 'cudaExternalSemaphoreSignalParams.params.nvSciSync.reserved' in found_struct}}
@@ -13817,7 +13819,7 @@ cdef class anon_union5:
         return <void_ptr>self._pvt_ptr[0].params.nvSciSync.fence
     @fence.setter
     def fence(self, fence):
-        _cfence = utils.HelperInputVoidPtr(fence)
+        _cfence = _HelperInputVoidPtr(fence)
         self._pvt_ptr[0].params.nvSciSync.fence = <void*><void_ptr>_cfence.cptr
     {{endif}}
     {{if 'cudaExternalSemaphoreWaitParams.params.nvSciSync.reserved' in found_struct}}
@@ -14173,7 +14175,7 @@ cdef class cudalibraryHostUniversalFunctionAndDataTable:
         return <void_ptr>self._pvt_ptr[0].functionTable
     @functionTable.setter
     def functionTable(self, functionTable):
-        _cfunctionTable = utils.HelperInputVoidPtr(functionTable)
+        _cfunctionTable = _HelperInputVoidPtr(functionTable)
         self._pvt_ptr[0].functionTable = <void*><void_ptr>_cfunctionTable.cptr
     {{endif}}
     {{if 'cudalibraryHostUniversalFunctionAndDataTable.functionWindowSize' in found_struct}}
@@ -14190,7 +14192,7 @@ cdef class cudalibraryHostUniversalFunctionAndDataTable:
         return <void_ptr>self._pvt_ptr[0].dataTable
     @dataTable.setter
     def dataTable(self, dataTable):
-        _cdataTable = utils.HelperInputVoidPtr(dataTable)
+        _cdataTable = _HelperInputVoidPtr(dataTable)
         self._pvt_ptr[0].dataTable = <void*><void_ptr>_cdataTable.cptr
     {{endif}}
     {{if 'cudalibraryHostUniversalFunctionAndDataTable.dataWindowSize' in found_struct}}
@@ -14305,7 +14307,7 @@ cdef class cudaKernelNodeParams:
         return <void_ptr>self._pvt_ptr[0].func
     @func.setter
     def func(self, func):
-        _cfunc = utils.HelperInputVoidPtr(func)
+        _cfunc = _HelperInputVoidPtr(func)
         self._pvt_ptr[0].func = <void*><void_ptr>_cfunc.cptr
     {{endif}}
     {{if 'cudaKernelNodeParams.gridDim' in found_struct}}
@@ -14338,7 +14340,7 @@ cdef class cudaKernelNodeParams:
         return <void_ptr>self._pvt_ptr[0].kernelParams
     @kernelParams.setter
     def kernelParams(self, kernelParams):
-        self._cykernelParams = utils.HelperKernelParams(kernelParams)
+        self._cykernelParams = _HelperKernelParams(kernelParams)
         self._pvt_ptr[0].kernelParams = <void**><void_ptr>self._cykernelParams.ckernelParams
     {{endif}}
     {{if 'cudaKernelNodeParams.extra' in found_struct}}
@@ -14453,7 +14455,7 @@ cdef class cudaKernelNodeParamsV2:
         return <void_ptr>self._pvt_ptr[0].func
     @func.setter
     def func(self, func):
-        _cfunc = utils.HelperInputVoidPtr(func)
+        _cfunc = _HelperInputVoidPtr(func)
         self._pvt_ptr[0].func = <void*><void_ptr>_cfunc.cptr
     {{endif}}
     {{if 'cudaKernelNodeParamsV2.gridDim' in found_struct}}
@@ -14486,7 +14488,7 @@ cdef class cudaKernelNodeParamsV2:
         return <void_ptr>self._pvt_ptr[0].kernelParams
     @kernelParams.setter
     def kernelParams(self, kernelParams):
-        self._cykernelParams = utils.HelperKernelParams(kernelParams)
+        self._cykernelParams = _HelperKernelParams(kernelParams)
         self._pvt_ptr[0].kernelParams = <void**><void_ptr>self._cykernelParams.ckernelParams
     {{endif}}
     {{if 'cudaKernelNodeParamsV2.extra' in found_struct}}
@@ -16137,7 +16139,7 @@ cdef class anon_struct16:
         return <void_ptr>self._pvt_ptr[0].updateData.param.pValue
     @pValue.setter
     def pValue(self, pValue):
-        _cpValue = utils.HelperInputVoidPtr(pValue)
+        _cpValue = _HelperInputVoidPtr(pValue)
         self._pvt_ptr[0].updateData.param.pValue = <void*><void_ptr>_cpValue.cptr
     {{endif}}
     {{if 'cudaGraphKernelNodeUpdate.updateData.param.offset' in found_struct}}
@@ -19104,7 +19106,7 @@ def cudaIpcGetMemHandle(devPtr):
     :py:obj:`~.cudaMalloc`, :py:obj:`~.cudaFree`, :py:obj:`~.cudaIpcGetEventHandle`, :py:obj:`~.cudaIpcOpenEventHandle`, :py:obj:`~.cudaIpcOpenMemHandle`, :py:obj:`~.cudaIpcCloseMemHandle`, :py:obj:`~.cuIpcGetMemHandle`
     """
     cdef cudaIpcMemHandle_t handle = cudaIpcMemHandle_t()
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaIpcGetMemHandle(<cyruntime.cudaIpcMemHandle_t*>handle._pvt_ptr, cydevPtr_ptr)
@@ -19222,7 +19224,7 @@ def cudaIpcCloseMemHandle(devPtr):
     --------
     :py:obj:`~.cudaMalloc`, :py:obj:`~.cudaFree`, :py:obj:`~.cudaIpcGetEventHandle`, :py:obj:`~.cudaIpcOpenEventHandle`, :py:obj:`~.cudaIpcGetMemHandle`, :py:obj:`~.cudaIpcOpenMemHandle`, :py:obj:`~.cuIpcCloseMemHandle`
     """
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaIpcCloseMemHandle(cydevPtr_ptr)
@@ -19337,7 +19339,7 @@ def cudaDeviceRegisterAsyncNotification(int device, callbackFunc, userData):
     else:
         pcallbackFunc = int(cudaAsyncCallback(callbackFunc))
     cycallbackFunc = <cyruntime.cudaAsyncCallback><void_ptr>pcallbackFunc
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
 
     cdef cudaAsyncCallbackData *cbData = NULL
@@ -19976,7 +19978,7 @@ def cudaDeviceGetNvSciSyncAttributes(nvSciSyncAttrList, int device, int flags):
     --------
     :py:obj:`~.cudaImportExternalSemaphore`, :py:obj:`~.cudaDestroyExternalSemaphore`, :py:obj:`~.cudaSignalExternalSemaphoresAsync`, :py:obj:`~.cudaWaitExternalSemaphoresAsync`
     """
-    cynvSciSyncAttrList = utils.HelperInputVoidPtr(nvSciSyncAttrList)
+    cynvSciSyncAttrList = _HelperInputVoidPtr(nvSciSyncAttrList)
     cdef void* cynvSciSyncAttrList_ptr = <void*><void_ptr>cynvSciSyncAttrList.cptr
     with nogil:
         err = cyruntime.cudaDeviceGetNvSciSyncAttributes(cynvSciSyncAttrList_ptr, device, flags)
@@ -21086,7 +21088,7 @@ def cudaStreamAddCallback(stream, callback, userData, unsigned int flags):
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
 
     cdef cudaStreamCallbackData *cbData = NULL
@@ -21285,7 +21287,7 @@ def cudaStreamAttachMemAsync(stream, devPtr, size_t length, unsigned int flags):
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaStreamAttachMemAsync(cystream, cydevPtr_ptr, length, flags)
@@ -23097,7 +23099,7 @@ def cudaFuncSetCacheConfig(func, cacheConfig not None : cudaFuncCache):
     -----
     This API does not accept a :py:obj:`~.cudaKernel_t` casted as void*. If cache config modification is required for a :py:obj:`~.cudaKernel_t` (or a global function), it can be replaced with a call to :py:obj:`~.cudaFuncSetAttributes` with the attribute :py:obj:`~.cudaFuncAttributePreferredSharedMemoryCarveout` to specify a more granular L1 cache and shared memory split configuration.
     """
-    cyfunc = utils.HelperInputVoidPtr(func)
+    cyfunc = _HelperInputVoidPtr(func)
     cdef void* cyfunc_ptr = <void*><void_ptr>cyfunc.cptr
     cdef cyruntime.cudaFuncCache cycacheConfig = cacheConfig.value
     with nogil:
@@ -23140,7 +23142,7 @@ def cudaFuncGetAttributes(func):
     :py:obj:`~.cudaFuncSetCacheConfig (C API)`, cudaFuncGetAttributes (C++ API), :py:obj:`~.cudaLaunchKernel (C API)`, :py:obj:`~.cuFuncGetAttribute`
     """
     cdef cudaFuncAttributes attr = cudaFuncAttributes()
-    cyfunc = utils.HelperInputVoidPtr(func)
+    cyfunc = _HelperInputVoidPtr(func)
     cdef void* cyfunc_ptr = <void*><void_ptr>cyfunc.cptr
     with nogil:
         err = cyruntime.cudaFuncGetAttributes(<cyruntime.cudaFuncAttributes*>attr._pvt_ptr, cyfunc_ptr)
@@ -23227,7 +23229,7 @@ def cudaFuncSetAttribute(func, attr not None : cudaFuncAttribute, int value):
     cudaError_t
         :py:obj:`~.cudaSuccess`, :py:obj:`~.cudaErrorInvalidDeviceFunction`, :py:obj:`~.cudaErrorInvalidValue`
     """
-    cyfunc = utils.HelperInputVoidPtr(func)
+    cyfunc = _HelperInputVoidPtr(func)
     cdef void* cyfunc_ptr = <void*><void_ptr>cyfunc.cptr
     cdef cyruntime.cudaFuncAttribute cyattr = attr.value
     with nogil:
@@ -23328,7 +23330,7 @@ def cudaLaunchHostFunc(stream, fn, userData):
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
 
     cdef cudaStreamHostCallbackData *cbData = NULL
@@ -23405,7 +23407,7 @@ def cudaFuncSetSharedMemConfig(func, config not None : cudaSharedMemConfig):
     --------
     :py:obj:`~.cudaDeviceSetSharedMemConfig`, :py:obj:`~.cudaDeviceGetSharedMemConfig`, :py:obj:`~.cudaDeviceSetCacheConfig`, :py:obj:`~.cudaDeviceGetCacheConfig`, :py:obj:`~.cudaFuncSetCacheConfig`, :py:obj:`~.cuFuncSetSharedMemConfig`
     """
-    cyfunc = utils.HelperInputVoidPtr(func)
+    cyfunc = _HelperInputVoidPtr(func)
     cdef void* cyfunc_ptr = <void*><void_ptr>cyfunc.cptr
     cdef cyruntime.cudaSharedMemConfig cyconfig = config.value
     with nogil:
@@ -23443,7 +23445,7 @@ def cudaOccupancyMaxActiveBlocksPerMultiprocessor(func, int blockSize, size_t dy
     :py:obj:`~.cudaOccupancyMaxActiveBlocksPerMultiprocessorWithFlags`, cudaOccupancyMaxPotentialBlockSize (C++ API), cudaOccupancyMaxPotentialBlockSizeWithFlags (C++ API), cudaOccupancyMaxPotentialBlockSizeVariableSMem (C++ API), cudaOccupancyMaxPotentialBlockSizeVariableSMemWithFlags (C++ API), cudaOccupancyAvailableDynamicSMemPerBlock (C++ API), :py:obj:`~.cuOccupancyMaxActiveBlocksPerMultiprocessor`
     """
     cdef int numBlocks = 0
-    cyfunc = utils.HelperInputVoidPtr(func)
+    cyfunc = _HelperInputVoidPtr(func)
     cdef void* cyfunc_ptr = <void*><void_ptr>cyfunc.cptr
     with nogil:
         err = cyruntime.cudaOccupancyMaxActiveBlocksPerMultiprocessor(&numBlocks, cyfunc_ptr, blockSize, dynamicSMemSize)
@@ -23482,7 +23484,7 @@ def cudaOccupancyAvailableDynamicSMemPerBlock(func, int numBlocks, int blockSize
     :py:obj:`~.cudaOccupancyMaxActiveBlocksPerMultiprocessorWithFlags`, cudaOccupancyMaxPotentialBlockSize (C++ API), cudaOccupancyMaxPotentialBlockSizeWithFlags (C++ API), cudaOccupancyMaxPotentialBlockSizeVariableSMem (C++ API), cudaOccupancyMaxPotentialBlockSizeVariableSMemWithFlags (C++ API), :py:obj:`~.cudaOccupancyAvailableDynamicSMemPerBlock`
     """
     cdef size_t dynamicSmemSize = 0
-    cyfunc = utils.HelperInputVoidPtr(func)
+    cyfunc = _HelperInputVoidPtr(func)
     cdef void* cyfunc_ptr = <void*><void_ptr>cyfunc.cptr
     with nogil:
         err = cyruntime.cudaOccupancyAvailableDynamicSMemPerBlock(&dynamicSmemSize, cyfunc_ptr, numBlocks, blockSize)
@@ -23538,7 +23540,7 @@ def cudaOccupancyMaxActiveBlocksPerMultiprocessorWithFlags(func, int blockSize,
     :py:obj:`~.cudaOccupancyMaxActiveBlocksPerMultiprocessor`, cudaOccupancyMaxPotentialBlockSize (C++ API), cudaOccupancyMaxPotentialBlockSizeWithFlags (C++ API), cudaOccupancyMaxPotentialBlockSizeVariableSMem (C++ API), cudaOccupancyMaxPotentialBlockSizeVariableSMemWithFlags (C++ API), cudaOccupancyAvailableDynamicSMemPerBlock (C++ API), :py:obj:`~.cuOccupancyMaxActiveBlocksPerMultiprocessorWithFlags`
     """
     cdef int numBlocks = 0
-    cyfunc = utils.HelperInputVoidPtr(func)
+    cyfunc = _HelperInputVoidPtr(func)
     cdef void* cyfunc_ptr = <void*><void_ptr>cyfunc.cptr
     with nogil:
         err = cyruntime.cudaOccupancyMaxActiveBlocksPerMultiprocessorWithFlags(&numBlocks, cyfunc_ptr, blockSize, dynamicSMemSize, flags)
@@ -23944,7 +23946,7 @@ def cudaFree(devPtr):
     --------
     :py:obj:`~.cudaMalloc`, :py:obj:`~.cudaMallocPitch`, :py:obj:`~.cudaMallocManaged`, :py:obj:`~.cudaMallocArray`, :py:obj:`~.cudaFreeArray`, :py:obj:`~.cudaMallocAsync`, :py:obj:`~.cudaMallocFromPoolAsync` :py:obj:`~.cudaMallocHost (C API)`, :py:obj:`~.cudaFreeHost`, :py:obj:`~.cudaMalloc3D`, :py:obj:`~.cudaMalloc3DArray`, :py:obj:`~.cudaFreeAsync` :py:obj:`~.cudaHostAlloc`, :py:obj:`~.cuMemFree`
     """
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaFree(cydevPtr_ptr)
@@ -23975,7 +23977,7 @@ def cudaFreeHost(ptr):
     --------
     :py:obj:`~.cudaMalloc`, :py:obj:`~.cudaMallocPitch`, :py:obj:`~.cudaFree`, :py:obj:`~.cudaMallocArray`, :py:obj:`~.cudaFreeArray`, :py:obj:`~.cudaMallocHost (C API)`, :py:obj:`~.cudaMalloc3D`, :py:obj:`~.cudaMalloc3DArray`, :py:obj:`~.cudaHostAlloc`, :py:obj:`~.cuMemFreeHost`
     """
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cyruntime.cudaFreeHost(cyptr_ptr)
@@ -24251,7 +24253,7 @@ def cudaHostRegister(ptr, size_t size, unsigned int flags):
     --------
     :py:obj:`~.cudaHostUnregister`, :py:obj:`~.cudaHostGetFlags`, :py:obj:`~.cudaHostGetDevicePointer`, :py:obj:`~.cuMemHostRegister`
     """
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cyruntime.cudaHostRegister(cyptr_ptr, size, flags)
@@ -24284,7 +24286,7 @@ def cudaHostUnregister(ptr):
     --------
     :py:obj:`~.cudaHostUnregister`, :py:obj:`~.cuMemHostUnregister`
     """
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cyruntime.cudaHostUnregister(cyptr_ptr)
@@ -24344,7 +24346,7 @@ def cudaHostGetDevicePointer(pHost, unsigned int flags):
     :py:obj:`~.cudaSetDeviceFlags`, :py:obj:`~.cudaHostAlloc`, :py:obj:`~.cuMemHostGetDevicePointer`
     """
     cdef void_ptr pDevice = 0
-    cypHost = utils.HelperInputVoidPtr(pHost)
+    cypHost = _HelperInputVoidPtr(pHost)
     cdef void* cypHost_ptr = <void*><void_ptr>cypHost.cptr
     with nogil:
         err = cyruntime.cudaHostGetDevicePointer(<void**>&pDevice, cypHost_ptr, flags)
@@ -24379,7 +24381,7 @@ def cudaHostGetFlags(pHost):
     :py:obj:`~.cudaHostAlloc`, :py:obj:`~.cuMemHostGetFlags`
     """
     cdef unsigned int pFlags = 0
-    cypHost = utils.HelperInputVoidPtr(pHost)
+    cypHost = _HelperInputVoidPtr(pHost)
     cdef void* cypHost_ptr = <void*><void_ptr>cypHost.cptr
     with nogil:
         err = cyruntime.cudaHostGetFlags(&pFlags, cypHost_ptr)
@@ -25403,9 +25405,9 @@ def cudaMemcpy(dst, src, size_t count, kind not None : cudaMemcpyKind):
     --------
     :py:obj:`~.cudaMemcpy2D`, :py:obj:`~.cudaMemcpy2DToArray`, :py:obj:`~.cudaMemcpy2DFromArray`, :py:obj:`~.cudaMemcpy2DArrayToArray`, :py:obj:`~.cudaMemcpyToSymbol`, :py:obj:`~.cudaMemcpyFromSymbol`, :py:obj:`~.cudaMemcpyAsync`, :py:obj:`~.cudaMemcpy2DAsync`, :py:obj:`~.cudaMemcpy2DToArrayAsync`, :py:obj:`~.cudaMemcpy2DFromArrayAsync`, :py:obj:`~.cudaMemcpyToSymbolAsync`, :py:obj:`~.cudaMemcpyFromSymbolAsync`, :py:obj:`~.cuMemcpyDtoH`, :py:obj:`~.cuMemcpyHtoD`, :py:obj:`~.cuMemcpyDtoD`, :py:obj:`~.cuMemcpy`
     """
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -25452,9 +25454,9 @@ def cudaMemcpyPeer(dst, int dstDevice, src, int srcDevice, size_t count):
     --------
     :py:obj:`~.cudaMemcpy`, :py:obj:`~.cudaMemcpyAsync`, :py:obj:`~.cudaMemcpyPeerAsync`, :py:obj:`~.cudaMemcpy3DPeerAsync`, :py:obj:`~.cuMemcpyPeer`
     """
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     with nogil:
         err = cyruntime.cudaMemcpyPeer(cydst_ptr, dstDevice, cysrc_ptr, srcDevice, count)
@@ -25511,9 +25513,9 @@ def cudaMemcpy2D(dst, size_t dpitch, src, size_t spitch, size_t width, size_t he
     --------
     :py:obj:`~.cudaMemcpy`, :py:obj:`~.cudaMemcpy2DToArray`, :py:obj:`~.cudaMemcpy2DFromArray`, :py:obj:`~.cudaMemcpy2DArrayToArray`, :py:obj:`~.cudaMemcpyToSymbol`, :py:obj:`~.cudaMemcpyFromSymbol`, :py:obj:`~.cudaMemcpyAsync`, :py:obj:`~.cudaMemcpy2DAsync`, :py:obj:`~.cudaMemcpy2DToArrayAsync`, :py:obj:`~.cudaMemcpy2DFromArrayAsync`, :py:obj:`~.cudaMemcpyToSymbolAsync`, :py:obj:`~.cudaMemcpyFromSymbolAsync`, :py:obj:`~.cuMemcpy2D`, :py:obj:`~.cuMemcpy2DUnaligned`
     """
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -25580,7 +25582,7 @@ def cudaMemcpy2DToArray(dst, size_t wOffset, size_t hOffset, src, size_t spitch,
     else:
         pdst = int(cudaArray_t(dst))
     cydst = <cyruntime.cudaArray_t><void_ptr>pdst
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -25647,7 +25649,7 @@ def cudaMemcpy2DFromArray(dst, size_t dpitch, src, size_t wOffset, size_t hOffse
     else:
         psrc = int(cudaArray_const_t(src))
     cysrc = <cyruntime.cudaArray_const_t><void_ptr>psrc
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -25788,9 +25790,9 @@ def cudaMemcpyAsync(dst, src, size_t count, kind not None : cudaMemcpyKind, stre
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -25845,9 +25847,9 @@ def cudaMemcpyPeerAsync(dst, int dstDevice, src, int srcDevice, size_t count, st
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     with nogil:
         err = cyruntime.cudaMemcpyPeerAsync(cydst_ptr, dstDevice, cysrc_ptr, srcDevice, count, cystream)
@@ -25975,11 +25977,11 @@ def cudaMemcpyBatchAsync(dsts : Optional[Tuple[Any] | List[Any]], srcs : Optiona
         raise TypeError("Argument 'sizes' is not instance of type (expected Tuple[int] or List[int]")
     srcs = [] if srcs is None else srcs
     dsts = [] if dsts is None else dsts
-    pylist = [utils.HelperInputVoidPtr(pydsts) for pydsts in dsts]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperdsts = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperInputVoidPtr(pydsts) for pydsts in dsts]
+    cdef _InputVoidPtrPtrHelper voidStarHelperdsts = _InputVoidPtrPtrHelper(pylist)
     cdef const void** cydsts_ptr = <const void**><void_ptr>voidStarHelperdsts.cptr
-    pylist = [utils.HelperInputVoidPtr(pysrcs) for pysrcs in srcs]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelpersrcs = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperInputVoidPtr(pysrcs) for pysrcs in srcs]
+    cdef _InputVoidPtrPtrHelper voidStarHelpersrcs = _InputVoidPtrPtrHelper(pylist)
     cdef const void** cysrcs_ptr = <const void**><void_ptr>voidStarHelpersrcs.cptr
     cdef vector[size_t] cysizes = sizes
     if count > <size_t>len(dsts): raise RuntimeError("List is too small: " + str(len(dsts)) + " < " + str(count))
@@ -26205,9 +26207,9 @@ def cudaMemcpy2DAsync(dst, size_t dpitch, src, size_t spitch, size_t width, size
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -26295,7 +26297,7 @@ def cudaMemcpy2DToArrayAsync(dst, size_t wOffset, size_t hOffset, src, size_t sp
     else:
         pdst = int(cudaArray_t(dst))
     cydst = <cyruntime.cudaArray_t><void_ptr>pdst
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -26382,7 +26384,7 @@ def cudaMemcpy2DFromArrayAsync(dst, size_t dpitch, src, size_t wOffset, size_t h
     else:
         psrc = int(cudaArray_const_t(src))
     cysrc = <cyruntime.cudaArray_const_t><void_ptr>psrc
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -26420,7 +26422,7 @@ def cudaMemset(devPtr, int value, size_t count):
     --------
     :py:obj:`~.cuMemsetD8`, :py:obj:`~.cuMemsetD16`, :py:obj:`~.cuMemsetD32`
     """
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaMemset(cydevPtr_ptr, value, count)
@@ -26464,7 +26466,7 @@ def cudaMemset2D(devPtr, size_t pitch, int value, size_t width, size_t height):
     --------
     :py:obj:`~.cudaMemset`, :py:obj:`~.cudaMemset3D`, :py:obj:`~.cudaMemsetAsync`, :py:obj:`~.cudaMemset2DAsync`, :py:obj:`~.cudaMemset3DAsync`, :py:obj:`~.cuMemsetD2D8`, :py:obj:`~.cuMemsetD2D16`, :py:obj:`~.cuMemsetD2D32`
     """
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaMemset2D(cydevPtr_ptr, pitch, value, width, height)
@@ -26571,7 +26573,7 @@ def cudaMemsetAsync(devPtr, int value, size_t count, stream):
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaMemsetAsync(cydevPtr_ptr, value, count, cystream)
@@ -26631,7 +26633,7 @@ def cudaMemset2DAsync(devPtr, size_t pitch, int value, size_t width, size_t heig
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaMemset2DAsync(cydevPtr_ptr, pitch, value, width, height, cystream)
@@ -26818,7 +26820,7 @@ def cudaMemPrefetchAsync(devPtr, size_t count, location not None : cudaMemLocati
     else:
         pstream = int(cudaStream_t(stream))
     cystream = <cyruntime.cudaStream_t><void_ptr>pstream
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaMemPrefetchAsync(cydevPtr_ptr, count, location._pvt_ptr[0], flags, cystream)
@@ -26909,8 +26911,8 @@ def cudaMemPrefetchBatchAsync(dptrs : Optional[Tuple[Any] | List[Any]], sizes :
     if not all(isinstance(_x, (int)) for _x in sizes):
         raise TypeError("Argument 'sizes' is not instance of type (expected Tuple[int] or List[int]")
     dptrs = [] if dptrs is None else dptrs
-    pylist = [utils.HelperInputVoidPtr(pydptrs) for pydptrs in dptrs]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperdptrs = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperInputVoidPtr(pydptrs) for pydptrs in dptrs]
+    cdef _InputVoidPtrPtrHelper voidStarHelperdptrs = _InputVoidPtrPtrHelper(pylist)
     cdef void** cydptrs_ptr = <void**><void_ptr>voidStarHelperdptrs.cptr
     cdef vector[size_t] cysizes = sizes
     if count > <size_t>len(dptrs): raise RuntimeError("List is too small: " + str(len(dptrs)) + " < " + str(count))
@@ -26997,8 +26999,8 @@ def cudaMemDiscardBatchAsync(dptrs : Optional[Tuple[Any] | List[Any]], sizes : T
     if not all(isinstance(_x, (int)) for _x in sizes):
         raise TypeError("Argument 'sizes' is not instance of type (expected Tuple[int] or List[int]")
     dptrs = [] if dptrs is None else dptrs
-    pylist = [utils.HelperInputVoidPtr(pydptrs) for pydptrs in dptrs]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperdptrs = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperInputVoidPtr(pydptrs) for pydptrs in dptrs]
+    cdef _InputVoidPtrPtrHelper voidStarHelperdptrs = _InputVoidPtrPtrHelper(pylist)
     cdef void** cydptrs_ptr = <void**><void_ptr>voidStarHelperdptrs.cptr
     cdef vector[size_t] cysizes = sizes
     if count > <size_t>len(dptrs): raise RuntimeError("List is too small: " + str(len(dptrs)) + " < " + str(count))
@@ -27100,8 +27102,8 @@ def cudaMemDiscardAndPrefetchBatchAsync(dptrs : Optional[Tuple[Any] | List[Any]]
     if not all(isinstance(_x, (int)) for _x in sizes):
         raise TypeError("Argument 'sizes' is not instance of type (expected Tuple[int] or List[int]")
     dptrs = [] if dptrs is None else dptrs
-    pylist = [utils.HelperInputVoidPtr(pydptrs) for pydptrs in dptrs]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperdptrs = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperInputVoidPtr(pydptrs) for pydptrs in dptrs]
+    cdef _InputVoidPtrPtrHelper voidStarHelperdptrs = _InputVoidPtrPtrHelper(pylist)
     cdef void** cydptrs_ptr = <void**><void_ptr>voidStarHelperdptrs.cptr
     cdef vector[size_t] cysizes = sizes
     if count > <size_t>len(dptrs): raise RuntimeError("List is too small: " + str(len(dptrs)) + " < " + str(count))
@@ -27314,7 +27316,7 @@ def cudaMemAdvise(devPtr, size_t count, advice not None : cudaMemoryAdvise, loca
     --------
     :py:obj:`~.cudaMemcpy`, :py:obj:`~.cudaMemcpyPeer`, :py:obj:`~.cudaMemcpyAsync`, :py:obj:`~.cudaMemcpy3DPeerAsync`, :py:obj:`~.cudaMemPrefetchAsync`, :py:obj:`~.cuMemAdvise`
     """
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     cdef cyruntime.cudaMemoryAdvise cyadvice = advice.value
     with nogil:
@@ -27460,10 +27462,10 @@ def cudaMemRangeGetAttribute(size_t dataSize, attribute not None : cudaMemRangeA
     --------
     :py:obj:`~.cudaMemRangeGetAttributes`, :py:obj:`~.cudaMemPrefetchAsync`, :py:obj:`~.cudaMemAdvise`, :py:obj:`~.cuMemRangeGetAttribute`
     """
-    cdef utils.HelperCUmem_range_attribute cydata = utils.HelperCUmem_range_attribute(attribute, dataSize)
+    cdef _HelperCUmem_range_attribute cydata = _HelperCUmem_range_attribute(attribute, dataSize)
     cdef void* cydata_ptr = <void*><void_ptr>cydata.cptr
     cdef cyruntime.cudaMemRangeAttribute cyattribute = attribute.value
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaMemRangeGetAttribute(cydata_ptr, dataSize, cyattribute, cydevPtr_ptr, count)
@@ -27537,14 +27539,14 @@ def cudaMemRangeGetAttributes(dataSizes : Tuple[int] | List[int], attributes : O
         raise TypeError("Argument 'attributes' is not instance of type (expected Tuple[cyruntime.cudaMemRangeAttribute] or List[cyruntime.cudaMemRangeAttribute]")
     if not all(isinstance(_x, (int)) for _x in dataSizes):
         raise TypeError("Argument 'dataSizes' is not instance of type (expected Tuple[int] or List[int]")
-    pylist = [utils.HelperCUmem_range_attribute(pyattributes, pydataSizes) for (pyattributes, pydataSizes) in zip(attributes, dataSizes)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperdata = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCUmem_range_attribute(pyattributes, pydataSizes) for (pyattributes, pydataSizes) in zip(attributes, dataSizes)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperdata = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyvoidStarHelper_ptr = <void**><void_ptr>voidStarHelperdata.cptr
     cdef vector[size_t] cydataSizes = dataSizes
     cdef vector[cyruntime.cudaMemRangeAttribute] cyattributes = [pyattributes.value for pyattributes in (attributes)]
     if numAttributes > <size_t>len(dataSizes): raise RuntimeError("List is too small: " + str(len(dataSizes)) + " < " + str(numAttributes))
     if numAttributes > <size_t>len(attributes): raise RuntimeError("List is too small: " + str(len(attributes)) + " < " + str(numAttributes))
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaMemRangeGetAttributes(cyvoidStarHelper_ptr, cydataSizes.data(), cyattributes.data(), numAttributes, cydevPtr_ptr, count)
@@ -27604,7 +27606,7 @@ def cudaMemcpyToArray(dst, size_t wOffset, size_t hOffset, src, size_t count, ki
     else:
         pdst = int(cudaArray_t(dst))
     cydst = <cyruntime.cudaArray_t><void_ptr>pdst
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -27663,7 +27665,7 @@ def cudaMemcpyFromArray(dst, src, size_t wOffset, size_t hOffset, size_t count,
     else:
         psrc = int(cudaArray_const_t(src))
     cysrc = <cyruntime.cudaArray_const_t><void_ptr>psrc
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -27809,7 +27811,7 @@ def cudaMemcpyToArrayAsync(dst, size_t wOffset, size_t hOffset, src, size_t coun
     else:
         pdst = int(cudaArray_t(dst))
     cydst = <cyruntime.cudaArray_t><void_ptr>pdst
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -27885,7 +27887,7 @@ def cudaMemcpyFromArrayAsync(dst, src, size_t wOffset, size_t hOffset, size_t co
     else:
         psrc = int(cudaArray_const_t(src))
     cysrc = <cyruntime.cudaArray_const_t><void_ptr>psrc
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -27987,7 +27989,7 @@ def cudaFreeAsync(devPtr, hStream):
     else:
         phStream = int(cudaStream_t(hStream))
     cyhStream = <cyruntime.cudaStream_t><void_ptr>phStream
-    cydevPtr = utils.HelperInputVoidPtr(devPtr)
+    cydevPtr = _HelperInputVoidPtr(devPtr)
     cdef void* cydevPtr_ptr = <void*><void_ptr>cydevPtr.cptr
     with nogil:
         err = cyruntime.cudaFreeAsync(cydevPtr_ptr, cyhStream)
@@ -28113,7 +28115,7 @@ def cudaMemPoolSetAttribute(memPool, attr not None : cudaMemPoolAttr, value):
         pmemPool = int(cudaMemPool_t(memPool))
     cymemPool = <cyruntime.cudaMemPool_t><void_ptr>pmemPool
     cdef cyruntime.cudaMemPoolAttr cyattr = attr.value
-    cdef utils.HelperCUmemPool_attribute cyvalue = utils.HelperCUmemPool_attribute(attr, value, is_getter=False)
+    cdef _HelperCUmemPool_attribute cyvalue = _HelperCUmemPool_attribute(attr, value, is_getter=False)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cyruntime.cudaMemPoolSetAttribute(cymemPool, cyattr, cyvalue_ptr)
@@ -28195,7 +28197,7 @@ def cudaMemPoolGetAttribute(memPool, attr not None : cudaMemPoolAttr):
         pmemPool = int(cudaMemPool_t(memPool))
     cymemPool = <cyruntime.cudaMemPool_t><void_ptr>pmemPool
     cdef cyruntime.cudaMemPoolAttr cyattr = attr.value
-    cdef utils.HelperCUmemPool_attribute cyvalue = utils.HelperCUmemPool_attribute(attr, 0, is_getter=True)
+    cdef _HelperCUmemPool_attribute cyvalue = _HelperCUmemPool_attribute(attr, 0, is_getter=True)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cyruntime.cudaMemPoolGetAttribute(cymemPool, cyattr, cyvalue_ptr)
@@ -28703,7 +28705,7 @@ def cudaMemPoolExportToShareableHandle(memPool, handleType not None : cudaMemAll
     else:
         pmemPool = int(cudaMemPool_t(memPool))
     cymemPool = <cyruntime.cudaMemPool_t><void_ptr>pmemPool
-    cdef utils.HelperCUmemAllocationHandleType cyshareableHandle = utils.HelperCUmemAllocationHandleType(handleType)
+    cdef _HelperCUmemAllocationHandleType cyshareableHandle = _HelperCUmemAllocationHandleType(handleType)
     cdef void* cyshareableHandle_ptr = <void*><void_ptr>cyshareableHandle.cptr
     cdef cyruntime.cudaMemAllocationHandleType cyhandleType = handleType.value
     with nogil:
@@ -28747,7 +28749,7 @@ def cudaMemPoolImportFromShareableHandle(shareableHandle, handleType not None :
     Imported memory pools do not support creating new allocations. As such imported memory pools may not be used in :py:obj:`~.cudaDeviceSetMemPool` or :py:obj:`~.cudaMallocFromPoolAsync` calls.
     """
     cdef cudaMemPool_t memPool = cudaMemPool_t()
-    cyshareableHandle = utils.HelperInputVoidPtr(shareableHandle)
+    cyshareableHandle = _HelperInputVoidPtr(shareableHandle)
     cdef void* cyshareableHandle_ptr = <void*><void_ptr>cyshareableHandle.cptr
     cdef cyruntime.cudaMemAllocationHandleType cyhandleType = handleType.value
     with nogil:
@@ -28785,7 +28787,7 @@ def cudaMemPoolExportPointer(ptr):
     :py:obj:`~.cuMemPoolExportPointer`, :py:obj:`~.cudaMemPoolExportToShareableHandle`, :py:obj:`~.cudaMemPoolImportFromShareableHandle`, :py:obj:`~.cudaMemPoolImportPointer`
     """
     cdef cudaMemPoolPtrExportData exportData = cudaMemPoolPtrExportData()
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cyruntime.cudaMemPoolExportPointer(<cyruntime.cudaMemPoolPtrExportData*>exportData._pvt_ptr, cyptr_ptr)
@@ -28910,7 +28912,7 @@ def cudaPointerGetAttributes(ptr):
     In CUDA 11.0 forward passing host pointer will return :py:obj:`~.cudaMemoryTypeUnregistered` in :py:obj:`~.cudaPointerAttributes.type` and call will return :py:obj:`~.cudaSuccess`.
     """
     cdef cudaPointerAttributes attributes = cudaPointerAttributes()
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cyruntime.cudaPointerGetAttributes(<cyruntime.cudaPointerAttributes*>attributes._pvt_ptr, cyptr_ptr)
@@ -30113,7 +30115,7 @@ def cudaLogsRegisterCallback(callbackFunc, userData):
     else:
         pcallbackFunc = int(cudaLogsCallback_t(callbackFunc))
     cycallbackFunc = <cyruntime.cudaLogsCallback_t><void_ptr>pcallbackFunc
-    cyuserData = utils.HelperInputVoidPtr(userData)
+    cyuserData = _HelperInputVoidPtr(userData)
     cdef void* cyuserData_ptr = <void*><void_ptr>cyuserData.cptr
     cdef cudaLogsCallbackHandle callback_out = cudaLogsCallbackHandle()
     with nogil:
@@ -30811,9 +30813,9 @@ def cudaGraphAddMemcpyNode1D(graph, pDependencies : Optional[Tuple[cudaGraphNode
                 cypDependencies[idx] = <cyruntime.cudaGraphNode_t>(<cudaGraphNode_t>pDependencies[idx])._pvt_ptr[0]
     elif len(pDependencies) == 1:
         cypDependencies = <cyruntime.cudaGraphNode_t*>(<cudaGraphNode_t>pDependencies[0])._pvt_ptr
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -30955,9 +30957,9 @@ def cudaGraphMemcpyNodeSetParams1D(node, dst, src, size_t count, kind not None :
     else:
         pnode = int(cudaGraphNode_t(node))
     cynode = <cyruntime.cudaGraphNode_t><void_ptr>pnode
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -32339,7 +32341,7 @@ def cudaGraphAddMemFreeNode(graph, pDependencies : Optional[Tuple[cudaGraphNode_
     elif len(pDependencies) == 1:
         cypDependencies = <cyruntime.cudaGraphNode_t*>(<cudaGraphNode_t>pDependencies[0])._pvt_ptr
     if numDependencies > <size_t>len(pDependencies): raise RuntimeError("List is too small: " + str(len(pDependencies)) + " < " + str(numDependencies))
-    cydptr = utils.HelperInputVoidPtr(dptr)
+    cydptr = _HelperInputVoidPtr(dptr)
     cdef void* cydptr_ptr = <void*><void_ptr>cydptr.cptr
     with nogil:
         err = cyruntime.cudaGraphAddMemFreeNode(<cyruntime.cudaGraphNode_t*>pGraphNode._pvt_ptr, cygraph, cypDependencies, numDependencies, cydptr_ptr)
@@ -32462,7 +32464,7 @@ def cudaDeviceGetGraphMemAttribute(int device, attr not None : cudaGraphMemAttri
     :py:obj:`~.cudaDeviceSetGraphMemAttribute`, :py:obj:`~.cudaGraphAddMemAllocNode`, :py:obj:`~.cudaGraphAddMemFreeNode`, :py:obj:`~.cudaDeviceGraphMemTrim`, :py:obj:`~.cudaMallocAsync`, :py:obj:`~.cudaFreeAsync`
     """
     cdef cyruntime.cudaGraphMemAttributeType cyattr = attr.value
-    cdef utils.HelperCUgraphMem_attribute cyvalue = utils.HelperCUgraphMem_attribute(attr, 0, is_getter=True)
+    cdef _HelperCUgraphMem_attribute cyvalue = _HelperCUgraphMem_attribute(attr, 0, is_getter=True)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cyruntime.cudaDeviceGetGraphMemAttribute(device, cyattr, cyvalue_ptr)
@@ -32506,7 +32508,7 @@ def cudaDeviceSetGraphMemAttribute(int device, attr not None : cudaGraphMemAttri
     :py:obj:`~.cudaDeviceGetGraphMemAttribute`, :py:obj:`~.cudaGraphAddMemAllocNode`, :py:obj:`~.cudaGraphAddMemFreeNode`, :py:obj:`~.cudaDeviceGraphMemTrim`, :py:obj:`~.cudaMallocAsync`, :py:obj:`~.cudaFreeAsync`
     """
     cdef cyruntime.cudaGraphMemAttributeType cyattr = attr.value
-    cdef utils.HelperCUgraphMem_attribute cyvalue = utils.HelperCUgraphMem_attribute(attr, value, is_getter=False)
+    cdef _HelperCUgraphMem_attribute cyvalue = _HelperCUgraphMem_attribute(attr, value, is_getter=False)
     cdef void* cyvalue_ptr = <void*><void_ptr>cyvalue.cptr
     with nogil:
         err = cyruntime.cudaDeviceSetGraphMemAttribute(device, cyattr, cyvalue_ptr)
@@ -33860,9 +33862,9 @@ def cudaGraphExecMemcpyNodeSetParams1D(hGraphExec, node, dst, src, size_t count,
     else:
         phGraphExec = int(cudaGraphExec_t(hGraphExec))
     cyhGraphExec = <cyruntime.cudaGraphExec_t><void_ptr>phGraphExec
-    cydst = utils.HelperInputVoidPtr(dst)
+    cydst = _HelperInputVoidPtr(dst)
     cdef void* cydst_ptr = <void*><void_ptr>cydst.cptr
-    cysrc = utils.HelperInputVoidPtr(src)
+    cysrc = _HelperInputVoidPtr(src)
     cdef void* cysrc_ptr = <void*><void_ptr>cysrc.cptr
     cdef cyruntime.cudaMemcpyKind cykind = kind.value
     with nogil:
@@ -34875,7 +34877,7 @@ def cudaUserObjectCreate(ptr, destroy, unsigned int initialRefcount, unsigned in
         pdestroy = int(cudaHostFn_t(destroy))
     cydestroy = <cyruntime.cudaHostFn_t><void_ptr>pdestroy
     cdef cudaUserObject_t object_out = cudaUserObject_t()
-    cyptr = utils.HelperInputVoidPtr(ptr)
+    cyptr = _HelperInputVoidPtr(ptr)
     cdef void* cyptr_ptr = <void*><void_ptr>cyptr.cptr
     with nogil:
         err = cyruntime.cudaUserObjectCreate(<cyruntime.cudaUserObject_t*>object_out._pvt_ptr, cyptr_ptr, cydestroy, initialRefcount, flags)
@@ -35620,17 +35622,17 @@ def cudaLibraryLoadData(code, jitOptions : Optional[Tuple[cudaJitOption] | List[
     if not all(isinstance(_x, (cudaJitOption)) for _x in jitOptions):
         raise TypeError("Argument 'jitOptions' is not instance of type (expected Tuple[cyruntime.cudaJitOption] or List[cyruntime.cudaJitOption]")
     cdef cudaLibrary_t library = cudaLibrary_t()
-    cycode = utils.HelperInputVoidPtr(code)
+    cycode = _HelperInputVoidPtr(code)
     cdef void* cycode_ptr = <void*><void_ptr>cycode.cptr
     cdef vector[cyruntime.cudaJitOption] cyjitOptions = [pyjitOptions.value for pyjitOptions in (jitOptions)]
-    pylist = [utils.HelperCudaJitOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCudaJitOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyjitOptionsValues_ptr = <void**><void_ptr>voidStarHelperjitOptionsValues.cptr
     if numJitOptions > len(jitOptions): raise RuntimeError("List is too small: " + str(len(jitOptions)) + " < " + str(numJitOptions))
     if numJitOptions > len(jitOptionsValues): raise RuntimeError("List is too small: " + str(len(jitOptionsValues)) + " < " + str(numJitOptions))
     cdef vector[cyruntime.cudaLibraryOption] cylibraryOptions = [pylibraryOptions.value for pylibraryOptions in (libraryOptions)]
-    pylist = [utils.HelperCudaLibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCudaLibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cylibraryOptionValues_ptr = <void**><void_ptr>voidStarHelperlibraryOptionValues.cptr
     if numLibraryOptions > len(libraryOptions): raise RuntimeError("List is too small: " + str(len(libraryOptions)) + " < " + str(numLibraryOptions))
     if numLibraryOptions > len(libraryOptionValues): raise RuntimeError("List is too small: " + str(len(libraryOptionValues)) + " < " + str(numLibraryOptions))
@@ -35720,14 +35722,14 @@ def cudaLibraryLoadFromFile(char* fileName, jitOptions : Optional[Tuple[cudaJitO
         raise TypeError("Argument 'jitOptions' is not instance of type (expected Tuple[cyruntime.cudaJitOption] or List[cyruntime.cudaJitOption]")
     cdef cudaLibrary_t library = cudaLibrary_t()
     cdef vector[cyruntime.cudaJitOption] cyjitOptions = [pyjitOptions.value for pyjitOptions in (jitOptions)]
-    pylist = [utils.HelperCudaJitOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCudaJitOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(jitOptions, jitOptionsValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperjitOptionsValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cyjitOptionsValues_ptr = <void**><void_ptr>voidStarHelperjitOptionsValues.cptr
     if numJitOptions > len(jitOptions): raise RuntimeError("List is too small: " + str(len(jitOptions)) + " < " + str(numJitOptions))
     if numJitOptions > len(jitOptionsValues): raise RuntimeError("List is too small: " + str(len(jitOptionsValues)) + " < " + str(numJitOptions))
     cdef vector[cyruntime.cudaLibraryOption] cylibraryOptions = [pylibraryOptions.value for pylibraryOptions in (libraryOptions)]
-    pylist = [utils.HelperCudaLibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
-    cdef utils.InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = utils.InputVoidPtrPtrHelper(pylist)
+    pylist = [_HelperCudaLibraryOption(pyoptions, pyoptionValues) for pyoptions, pyoptionValues in zip(libraryOptions, libraryOptionValues)]
+    cdef _InputVoidPtrPtrHelper voidStarHelperlibraryOptionValues = _InputVoidPtrPtrHelper(pylist)
     cdef void** cylibraryOptionValues_ptr = <void**><void_ptr>voidStarHelperlibraryOptionValues.cptr
     if numLibraryOptions > len(libraryOptions): raise RuntimeError("List is too small: " + str(len(libraryOptions)) + " < " + str(numLibraryOptions))
     if numLibraryOptions > len(libraryOptionValues): raise RuntimeError("List is too small: " + str(len(libraryOptionValues)) + " < " + str(numLibraryOptions))
@@ -36217,7 +36219,7 @@ def cudaGetKernel(entryFuncAddr):
     cudaGetKernel (C++ API)
     """
     cdef cudaKernel_t kernelPtr = cudaKernel_t()
-    cyentryFuncAddr = utils.HelperInputVoidPtr(entryFuncAddr)
+    cyentryFuncAddr = _HelperInputVoidPtr(entryFuncAddr)
     cdef void* cyentryFuncAddr_ptr = <void*><void_ptr>cyentryFuncAddr.cptr
     with nogil:
         err = cyruntime.cudaGetKernel(<cyruntime.cudaKernel_t*>kernelPtr._pvt_ptr, cyentryFuncAddr_ptr)
@@ -36257,7 +36259,7 @@ def make_cudaPitchedPtr(d, size_t p, size_t xsz, size_t ysz):
     --------
     make_cudaExtent, make_cudaPos
     """
-    cyd = utils.HelperInputVoidPtr(d)
+    cyd = _HelperInputVoidPtr(d)
     cdef void* cyd_ptr = <void*><void_ptr>cyd.cptr
     with nogil:
         err = cyruntime.make_cudaPitchedPtr(cyd_ptr, p, xsz, ysz)
diff --git a/cuda_bindings/setup.py b/cuda_bindings/setup.py
index 8ebd68a75..486452a24 100644
--- a/cuda_bindings/setup.py
+++ b/cuda_bindings/setup.py
@@ -343,7 +343,6 @@ def do_cythonize(extensions):
     (["cuda/bindings/_bindings/cyruntime.pyx"], static_runtime_libraries),
     (["cuda/bindings/_bindings/cyruntime_ptds.pyx"], static_runtime_libraries),
     # utils
-    (["cuda/bindings/_lib/utils.pyx", "cuda/bindings/_lib/param_packer.cpp"], None),
     (["cuda/bindings/_lib/cyruntime/cyruntime.pyx"], None),
     (["cuda/bindings/_lib/cyruntime/utils.pyx"], None),
     (["cuda/bindings/utils/*.pyx"], None),

From 57dafb4fed515d6c83f35eaaeefa3e11c83ddf32 Mon Sep 17 00:00:00 2001
From: Michael Droettboom <mdboom@gmail.com>
Date: Mon, 25 Aug 2025 12:39:12 -0400
Subject: [PATCH 3/4] Add a test for cyclical imports

---
 cuda_bindings/tests/test_utils.py             | 28 +++++++++++++
 .../tests/utils/check_cyclical_import.py      | 40 +++++++++++++++++++
 2 files changed, 68 insertions(+)
 create mode 100644 cuda_bindings/tests/utils/check_cyclical_import.py

diff --git a/cuda_bindings/tests/test_utils.py b/cuda_bindings/tests/test_utils.py
index 3da7272ed..7ed4fd753 100644
--- a/cuda_bindings/tests/test_utils.py
+++ b/cuda_bindings/tests/test_utils.py
@@ -2,6 +2,9 @@
 # SPDX-License-Identifier: LicenseRef-NVIDIA-SOFTWARE-LICENSE
 
 import random
+import subprocess  # nosec B404
+import sys
+from pathlib import Path
 
 import pytest
 
@@ -87,3 +90,28 @@ def test_get_handle(target):
 def test_get_handle_error(target):
     with pytest.raises(TypeError) as e:
         handle = get_cuda_native_handle(target)
+
+
+@pytest.mark.parametrize(
+    "module",
+    [
+        # Top-level modules for external Python use
+        # TODO: Import cycle detected: (('numeric',), ''), stack: [((),
+        # 'cuda.bindings.cufile'), ((), 'cuda.bindings.cycufile'),
+        # (('show_config',), 'numpy.__config__'), (('__cpu_features__',
+        # '__cpu_baseline__', '__cpu_dispatch__'),
+        # 'numpy._core._multiarray_umath'), (('numeric',), ''),
+        # (('shape_base',), '')]
+        # "cufile",
+        "driver",
+        "nvjitlink",
+        "nvrtc",
+        "nvvm",
+        # TODO: cuda.bindings.cyruntime -> cuda.bindings._lib.cyruntime.cyruntime cycle
+        # "runtime",
+    ],
+)
+def test_cyclical_imports(module):
+    subprocess.check_call(  # nosec B603
+        [sys.executable, Path(__file__).parent / "utils" / "check_cyclical_import.py", f"cuda.bindings.{module}"],
+    )
diff --git a/cuda_bindings/tests/utils/check_cyclical_import.py b/cuda_bindings/tests/utils/check_cyclical_import.py
new file mode 100644
index 000000000..19e72d59d
--- /dev/null
+++ b/cuda_bindings/tests/utils/check_cyclical_import.py
@@ -0,0 +1,40 @@
+# SPDX-FileCopyrightText: Copyright (c) 2025 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+# SPDX-License-Identifier: LicenseRef-NVIDIA-SOFTWARE-LICENSE
+
+"""
+Tests whether importing a specific module leads to cyclical imports.
+
+See https://github.com/NVIDIA/cuda-python/issues/789 for more info.
+"""
+
+import argparse
+
+orig_import = __builtins__.__import__
+
+import_stack = []
+
+
+def import_hook(name, globals=None, locals=None, fromlist=(), *args, **kwargs):
+    """Approximate a custom import system that does not allow import cycles."""
+
+    stack_entry = (tuple(fromlist) if fromlist is not None else None, name)
+    if stack_entry in import_stack:
+        raise ImportError(f"Import cycle detected: {stack_entry}, stack: {import_stack}")
+    import_stack.append(stack_entry)
+    res = orig_import(name, globals, locals, fromlist, *args, **kwargs)
+    import_stack.pop()
+    return res
+
+
+__builtins__.__import__ = import_hook
+
+
+if __name__ == "__main__":
+    parser = argparse.ArgumentParser()
+    parser.add_argument(
+        "module",
+        type=str,
+    )
+    args = parser.parse_args()
+
+    __import__(args.module)

From ab28510ca383f3e347b437326c666caf54b33592 Mon Sep 17 00:00:00 2001
From: Michael Droettboom <mdboom@gmail.com>
Date: Mon, 25 Aug 2025 13:15:20 -0400
Subject: [PATCH 4/4] Only raise on cuda.bindings modules

---
 cuda_bindings/tests/utils/check_cyclical_import.py | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/cuda_bindings/tests/utils/check_cyclical_import.py b/cuda_bindings/tests/utils/check_cyclical_import.py
index 19e72d59d..4466a5c76 100644
--- a/cuda_bindings/tests/utils/check_cyclical_import.py
+++ b/cuda_bindings/tests/utils/check_cyclical_import.py
@@ -18,7 +18,7 @@ def import_hook(name, globals=None, locals=None, fromlist=(), *args, **kwargs):
     """Approximate a custom import system that does not allow import cycles."""
 
     stack_entry = (tuple(fromlist) if fromlist is not None else None, name)
-    if stack_entry in import_stack:
+    if stack_entry in import_stack and name.startswith("cuda.bindings."):
         raise ImportError(f"Import cycle detected: {stack_entry}, stack: {import_stack}")
     import_stack.append(stack_entry)
     res = orig_import(name, globals, locals, fromlist, *args, **kwargs)