Merge pull request #497 from rwgk/fix_segfault_char_ptr_to_bytes

cuda.bindings: Fix segfault when converting `char*` `NULL` to `bytes`

Author: leofang

cuda_bindings/cuda/bindings/driver.pyx.in

@@ -22246,7 +22246,7 @@ def cuGetErrorString(error not None : CUresult):
     cdef cydriver.CUresult cyerror = error.value
     cdef const char* pStr = NULL
     err = cydriver.cuGetErrorString(cyerror, &pStr)
-    return (CUresult(err), <bytes>pStr)
+    return (CUresult(err), <bytes>pStr if pStr != NULL else None)
 {{endif}}
 
 {{if 'cuGetErrorName' in found_functions}}
@@ -22279,7 +22279,7 @@ def cuGetErrorName(error not None : CUresult):
     cdef cydriver.CUresult cyerror = error.value
     cdef const char* pStr = NULL
     err = cydriver.cuGetErrorName(cyerror, &pStr)
-    return (CUresult(err), <bytes>pStr)
+    return (CUresult(err), <bytes>pStr if pStr != NULL else None)
 {{endif}}
 
 {{if 'cuInit' in found_functions}}
@@ -27132,7 +27132,7 @@ def cuKernelGetName(hfunc):
         cyhfunc = <cydriver.CUkernel><void_ptr>phfunc
     cdef const char* name = NULL
     err = cydriver.cuKernelGetName(&name, cyhfunc)
-    return (CUresult(err), <bytes>name)
+    return (CUresult(err), <bytes>name if name != NULL else None)
 {{endif}}
 
 {{if 'cuKernelGetParamInfo' in found_functions}}
@@ -38744,7 +38744,7 @@ def cuFuncGetName(hfunc):
         cyhfunc = <cydriver.CUfunction><void_ptr>phfunc
     cdef const char* name = NULL
     err = cydriver.cuFuncGetName(&name, cyhfunc)
-    return (CUresult(err), <bytes>name)
+    return (CUresult(err), <bytes>name if name != NULL else None)
 {{endif}}
 
 {{if 'cuFuncGetParamInfo' in found_functions}}

cuda_bindings/cuda/bindings/nvrtc.pyx.in

@@ -870,7 +870,7 @@ def nvrtcGetLoweredName(prog, char* name_expression):
         cyprog = <cynvrtc.nvrtcProgram><void_ptr>pprog
     cdef const char* lowered_name = NULL
     err = cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name)
-    return (nvrtcResult(err), <bytes>lowered_name)
+    return (nvrtcResult(err), <bytes>lowered_name if lowered_name != NULL else None)
 {{endif}}
 
 {{if 'nvrtcGetPCHHeapSize' in found_functions}}

cuda_bindings/tests/test_cuda.py

@@ -948,3 +948,47 @@ def test_conditional():
 def test_CUmemDecompressParams_st():
     desc = cuda.CUmemDecompressParams_st()
     assert int(desc.dstActBytes) == 0
+
+
+def test_all_CUresult_codes():
+    max_code = int(max(cuda.CUresult))
+    # Smoke test. CUDA_ERROR_UNKNOWN = 999, but intentionally using literal value.
+    assert max_code >= 999
+    num_good = 0
+    for code in range(max_code + 2):  # One past max_code
+        try:
+            error = cuda.CUresult(code)
+        except ValueError:
+            pass  # cython-generated enum does not exist for this code
+        else:
+            err_name, name = cuda.cuGetErrorName(error)
+            if err_name == cuda.CUresult.CUDA_SUCCESS:
+                assert name
+                err_desc, desc = cuda.cuGetErrorString(error)
+                assert err_desc == cuda.CUresult.CUDA_SUCCESS
+                assert desc
+                num_good += 1
+            else:
+                # cython-generated enum exists but is not known to an older driver
+                # (example: cuda-bindings built with CTK 12.8, driver from CTK 12.0)
+                assert name is None
+                assert err_name == cuda.CUresult.CUDA_ERROR_INVALID_VALUE
+                err_desc, desc = cuda.cuGetErrorString(error)
+                assert err_desc == cuda.CUresult.CUDA_ERROR_INVALID_VALUE
+                assert desc is None
+    # Smoke test: Do we have at least some "good" codes?
+    # The number will increase over time as new enums are added and support for
+    # old CTKs is dropped, but it is not critical that this number is updated.
+    assert num_good >= 76  # CTK 11.0.3_450.51.06
+
+
+def test_cuKernelGetName_failure():
+    err, name = cuda.cuKernelGetName(0)
+    assert err == cuda.CUresult.CUDA_ERROR_INVALID_VALUE
+    assert name is None
+
+
+def test_cuFuncGetName_failure():
+    err, name = cuda.cuFuncGetName(0)
+    assert err == cuda.CUresult.CUDA_ERROR_INVALID_VALUE
+    assert name is None