Backport #497 and #499

cuda_bindings/cuda/bindings/driver.pyx.in

@@ -13208,7 +13208,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}}
@@ -13241,7 +13241,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}}

cuda_bindings/cuda/bindings/nvrtc.pyx.in

@@ -715,7 +715,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}}
 
 @cython.embedsignature(True)

cuda_bindings/tests/test_cuda.py

@@ -644,3 +644,35 @@ def test_char_range():
     for x in range(0, 256):
         val.reserved = [x] * 64
         assert(val.reserved[0] == x)
+
+
+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

cuda_bindings/tests/test_nvrtc.py

@@ -25,3 +25,13 @@ def test_nvrtcGetSupportedArchs():
     err, supportedArchs = nvrtc.nvrtcGetSupportedArchs()
     ASSERT_DRV(err)
     assert len(supportedArchs) != 0
+
+
+@pytest.mark.skipif(nvrtcVersionLessThan(12, 1), reason="Preempt Segmentation Fault (see #499)")
+def test_nvrtcGetLoweredName_failure():
+    err, name = nvrtc.nvrtcGetLoweredName(None, b"I'm an elevated name!")
+    assert err == nvrtc.nvrtcResult.NVRTC_ERROR_INVALID_PROGRAM
+    assert name is None
+    err, name = nvrtc.nvrtcGetLoweredName(0, b"I'm another elevated name!")
+    assert err == nvrtc.nvrtcResult.NVRTC_ERROR_INVALID_PROGRAM
+    assert name is None