diff --git a/cuda_core/cuda/core/_utils/cuda_utils.pxd b/cuda_core/cuda/core/_utils/cuda_utils.pxd index 11e464e6381..9b485597912 100644 --- a/cuda_core/cuda/core/_utils/cuda_utils.pxd +++ b/cuda_core/cuda/core/_utils/cuda_utils.pxd @@ -18,11 +18,35 @@ ctypedef fused integer_t: cdef const cydriver.CUcontext CU_CONTEXT_INVALID = (-2) -cdef int HANDLE_RETURN(cydriver.CUresult err) except?-1 nogil -cdef int HANDLE_RETURN_NVRTC(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) except?-1 nogil -cdef int HANDLE_RETURN_NVVM(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) except?-1 nogil -cdef int HANDLE_RETURN_NVJITLINK( - cynvjitlink.nvJitLinkHandle handle, cynvjitlink.nvJitLinkResult err) except?-1 nogil +cdef inline int HANDLE_RETURN(cydriver.CUresult err) except?-1 nogil: + if err != cydriver.CUresult.CUDA_SUCCESS: + return _check_driver_error(err) + return 0 + + +cdef inline int HANDLE_RETURN_NVRTC(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) except?-1 nogil: + """Handle NVRTC result codes, raising NVRTCError with program log on failure.""" + if err == cynvrtc.nvrtcResult.NVRTC_SUCCESS: + return 0 + with gil: + _raise_nvrtc_error(prog, err) + + +cdef inline int HANDLE_RETURN_NVVM(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) except?-1 nogil: + """Handle NVVM result codes, raising nvvmError with program log on failure.""" + if err == cynvvm.nvvmResult.NVVM_SUCCESS: + return 0 + with gil: + _raise_nvvm_error(prog, err) + + +cdef inline int HANDLE_RETURN_NVJITLINK( + cynvjitlink.nvJitLinkHandle handle, cynvjitlink.nvJitLinkResult err) except?-1 nogil: + """Handle nvJitLink result codes, raising nvJitLinkError with error log on failure.""" + if err == cynvjitlink.nvJitLinkResult.NVJITLINK_SUCCESS: + return 0 + with gil: + _raise_nvjitlink_error(handle, err) # Helper for retrieving the current CUDA device. Raises if no active context @@ -34,7 +58,9 @@ cdef int _get_current_device_id() except? -1 cpdef int _check_driver_error(cydriver.CUresult error) except?-1 nogil cpdef int _check_runtime_error(error) except?-1 cpdef int _check_nvrtc_error(error) except?-1 - +cdef int _raise_nvrtc_error(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) except -1 +cdef int _raise_nvvm_error(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) except -1 +cdef int _raise_nvjitlink_error(cynvjitlink.nvJitLinkHandle handle, cynvjitlink.nvJitLinkResult err) except -1 cpdef check_or_create_options(type cls, options, str options_description=*, bint keep_none=*) diff --git a/cuda_core/cuda/core/_utils/cuda_utils.pyx b/cuda_core/cuda/core/_utils/cuda_utils.pyx index 318d4466bee..cf3415b3458 100644 --- a/cuda_core/cuda/core/_utils/cuda_utils.pyx +++ b/cuda_core/cuda/core/_utils/cuda_utils.pyx @@ -63,12 +63,6 @@ def cast_to_3_tuple(label: str, cfg: int | tuple[int, ...]) -> tuple[int, int, i return cfg + (1,) * (3 - len(cfg)) -cdef int HANDLE_RETURN(cydriver.CUresult err) except?-1 nogil: - if err != cydriver.CUresult.CUDA_SUCCESS: - return _check_driver_error(err) - return 0 - - cdef int _get_current_device_id() except? -1: """Return the current thread's bound CUdevice ordinal.""" cdef cydriver.CUdevice dev @@ -77,14 +71,6 @@ cdef int _get_current_device_id() except? -1: return dev -cdef int HANDLE_RETURN_NVRTC(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) except?-1 nogil: - """Handle NVRTC result codes, raising NVRTCError with program log on failure.""" - if err == cynvrtc.nvrtcResult.NVRTC_SUCCESS: - return 0 - with gil: - _raise_nvrtc_error(prog, err) - - cdef int _raise_nvrtc_error(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) except -1: """Build error message with program log and raise NVRTCError.""" cdef const char* err_str = cynvrtc.nvrtcGetErrorString(err) @@ -103,14 +89,6 @@ cdef int _raise_nvrtc_error(cynvrtc.nvrtcProgram prog, cynvrtc.nvrtcResult err) raise NVRTCError(err_msg) -cdef int HANDLE_RETURN_NVVM(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) except?-1 nogil: - """Handle NVVM result codes, raising nvvmError with program log on failure.""" - if err == cynvvm.nvvmResult.NVVM_SUCCESS: - return 0 - with gil: - _raise_nvvm_error(prog, err) - - cdef int _raise_nvvm_error(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) except -1: """Raise nvvmError annotated with the program log.""" cdef size_t logsize = 0 @@ -128,15 +106,6 @@ cdef int _raise_nvvm_error(cynvvm.nvvmProgram prog, cynvvm.nvvmResult err) excep raise exc -cdef int HANDLE_RETURN_NVJITLINK( - cynvjitlink.nvJitLinkHandle handle, cynvjitlink.nvJitLinkResult err) except?-1 nogil: - """Handle nvJitLink result codes, raising nvJitLinkError with error log on failure.""" - if err == cynvjitlink.nvJitLinkResult.NVJITLINK_SUCCESS: - return 0 - with gil: - _raise_nvjitlink_error(handle, err) - - cdef int _raise_nvjitlink_error( cynvjitlink.nvJitLinkHandle handle, cynvjitlink.nvJitLinkResult err) except -1: """Raise nvJitLinkError annotated with the error log."""