From d6b53a3f1a783193687d2423f812b25a47973ed1 Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Thu, 6 Mar 2025 22:03:16 -0800 Subject: [PATCH 1/8] PR #497 squash-merged --- cuda_bindings/cuda/bindings/driver.pyx.in | 8 ++--- cuda_bindings/cuda/bindings/nvrtc.pyx.in | 2 +- cuda_bindings/tests/test_cuda.py | 41 +++++++++++++++++++++++ 3 files changed, 46 insertions(+), 5 deletions(-) diff --git a/cuda_bindings/cuda/bindings/driver.pyx.in b/cuda_bindings/cuda/bindings/driver.pyx.in index 6be529571..7eec936ba 100644 --- a/cuda_bindings/cuda/bindings/driver.pyx.in +++ b/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), pStr) + return (CUresult(err), 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), pStr) + return (CUresult(err), pStr if pStr != NULL else None) {{endif}} {{if 'cuInit' in found_functions}} @@ -27132,7 +27132,7 @@ def cuKernelGetName(hfunc): cyhfunc = phfunc cdef const char* name = NULL err = cydriver.cuKernelGetName(&name, cyhfunc) - return (CUresult(err), name) + return (CUresult(err), name if name != NULL else None) {{endif}} {{if 'cuKernelGetParamInfo' in found_functions}} @@ -38744,7 +38744,7 @@ def cuFuncGetName(hfunc): cyhfunc = phfunc cdef const char* name = NULL err = cydriver.cuFuncGetName(&name, cyhfunc) - return (CUresult(err), name) + return (CUresult(err), name if name != NULL else None) {{endif}} {{if 'cuFuncGetParamInfo' in found_functions}} diff --git a/cuda_bindings/cuda/bindings/nvrtc.pyx.in b/cuda_bindings/cuda/bindings/nvrtc.pyx.in index 67808422d..6c2704c59 100644 --- a/cuda_bindings/cuda/bindings/nvrtc.pyx.in +++ b/cuda_bindings/cuda/bindings/nvrtc.pyx.in @@ -870,7 +870,7 @@ def nvrtcGetLoweredName(prog, char* name_expression): cyprog = pprog cdef const char* lowered_name = NULL err = cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name) - return (nvrtcResult(err), lowered_name) + return (nvrtcResult(err), lowered_name if lowered_name != NULL else None) {{endif}} {{if 'nvrtcGetPCHHeapSize' in found_functions}} diff --git a/cuda_bindings/tests/test_cuda.py b/cuda_bindings/tests/test_cuda.py index 7be1b0b95..c58952d14 100644 --- a/cuda_bindings/tests/test_cuda.py +++ b/cuda_bindings/tests/test_cuda.py @@ -948,3 +948,44 @@ def test_conditional(): def test_CUmemDecompressParams_st(): desc = cuda.CUmemDecompressParams_st() assert int(desc.dstActBytes) == 0 + + +def test_all_CUresult_codes(max_code=1000): + num_good = 0 + for code in range(max_code + 1): + 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 + # Super-simple smoke check: 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 From ff5deecb63c2200c53a44df5051e035df156bee0 Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Thu, 6 Mar 2025 22:04:11 -0800 Subject: [PATCH 2/8] Bring back test_nvrtcGetLoweredName_failure() (it was originally under PR #497). --- cuda_bindings/tests/test_nvrtc.py | 6 ++++++ 1 file changed, 6 insertions(+) diff --git a/cuda_bindings/tests/test_nvrtc.py b/cuda_bindings/tests/test_nvrtc.py index 077484010..27dec963a 100644 --- a/cuda_bindings/tests/test_nvrtc.py +++ b/cuda_bindings/tests/test_nvrtc.py @@ -29,3 +29,9 @@ def test_nvrtcGetSupportedArchs(): err, supportedArchs = nvrtc.nvrtcGetSupportedArchs() ASSERT_DRV(err) assert len(supportedArchs) != 0 + + +def test_nvrtcGetLoweredName_failure(): + err, name = nvrtc.nvrtcGetLoweredName(0, b"I'm an elevated name!") + assert err == nvrtc.nvrtcResult.NVRTC_ERROR_INVALID_PROGRAM + assert name is None From 9dd48fd8752a6ce53a686776e70ad254ce98d287 Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Thu, 6 Mar 2025 22:30:24 -0800 Subject: [PATCH 3/8] Add code for debugging --- cuda_bindings/cuda/bindings/nvrtc.pyx.in | 8 ++++++++ cuda_bindings/tests/test_nvrtc.py | 5 ++++- 2 files changed, 12 insertions(+), 1 deletion(-) diff --git a/cuda_bindings/cuda/bindings/nvrtc.pyx.in b/cuda_bindings/cuda/bindings/nvrtc.pyx.in index 6c2704c59..b38ca3e1e 100644 --- a/cuda_bindings/cuda/bindings/nvrtc.pyx.in +++ b/cuda_bindings/cuda/bindings/nvrtc.pyx.in @@ -859,17 +859,25 @@ def nvrtcGetLoweredName(prog, char* name_expression): -------- nvrtcAddNameExpression """ + print("\nLOOOK nvrtcGetLoweredName ENTRY", flush=True) cdef cynvrtc.nvrtcProgram cyprog if prog is None: + print("LOOOK nvrtcGetLoweredName prog is None", flush=True) cyprog = 0 elif isinstance(prog, (nvrtcProgram,)): + print("LOOOK nvrtcGetLoweredName isinstance(prog, (nvrtcProgram,))", flush=True) pprog = int(prog) + print("LOOOK nvrtcGetLoweredName AFTER int(prog)", flush=True) cyprog = pprog else: + print("LOOOK nvrtcGetLoweredName else", flush=True) pprog = int(nvrtcProgram(prog)) + print("LOOOK nvrtcGetLoweredName AFTER int(nvrtcProgram(prog))", flush=True) cyprog = pprog cdef const char* lowered_name = NULL + print("LOOOK nvrtcGetLoweredName BEFORE cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name)", flush=True) err = cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name) + print("LOOOK nvrtcGetLoweredName RETURN\n", flush=True) return (nvrtcResult(err), lowered_name if lowered_name != NULL else None) {{endif}} diff --git a/cuda_bindings/tests/test_nvrtc.py b/cuda_bindings/tests/test_nvrtc.py index 27dec963a..1e3e628af 100644 --- a/cuda_bindings/tests/test_nvrtc.py +++ b/cuda_bindings/tests/test_nvrtc.py @@ -32,6 +32,9 @@ def test_nvrtcGetSupportedArchs(): def test_nvrtcGetLoweredName_failure(): - err, name = nvrtc.nvrtcGetLoweredName(0, b"I'm an elevated name!") + 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 From 8fd065f6c2476d8cfed84de01ce4ec3380da41ad Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Fri, 7 Mar 2025 08:24:29 -0800 Subject: [PATCH 4/8] test_all_CUresult_codes(): max_code = int(max(cuda.CUresult)) as suggested by at-leofang --- cuda_bindings/tests/test_cuda.py | 9 ++++++--- 1 file changed, 6 insertions(+), 3 deletions(-) diff --git a/cuda_bindings/tests/test_cuda.py b/cuda_bindings/tests/test_cuda.py index c58952d14..d89420cd9 100644 --- a/cuda_bindings/tests/test_cuda.py +++ b/cuda_bindings/tests/test_cuda.py @@ -950,9 +950,12 @@ def test_CUmemDecompressParams_st(): assert int(desc.dstActBytes) == 0 -def test_all_CUresult_codes(max_code=1000): +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 + 1): + for code in range(max_code + 2): # One past max_code try: error = cuda.CUresult(code) except ValueError: @@ -973,7 +976,7 @@ def test_all_CUresult_codes(max_code=1000): err_desc, desc = cuda.cuGetErrorString(error) assert err_desc == cuda.CUresult.CUDA_ERROR_INVALID_VALUE assert desc is None - # Super-simple smoke check: Do we have at least some "good" codes? + # 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 From b0464e7cee7d7d9595a2e00cef876652602efc1a Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Fri, 7 Mar 2025 09:24:05 -0800 Subject: [PATCH 5/8] Change pytest options, mostly to disable output capturing (of both stdout and stderr) --- .github/workflows/test-wheel-linux.yml | 8 ++++---- .github/workflows/test-wheel-windows.yml | 4 ++-- 2 files changed, 6 insertions(+), 6 deletions(-) diff --git a/.github/workflows/test-wheel-linux.yml b/.github/workflows/test-wheel-linux.yml index 322f859e3..19c78c8cc 100644 --- a/.github/workflows/test-wheel-linux.yml +++ b/.github/workflows/test-wheel-linux.yml @@ -194,7 +194,7 @@ jobs: pushd ./cuda_bindings pip install -r requirements.txt - pytest -rxXs -v tests/ + pytest -ra -s -v tests/ # It is a bit convoluted to run the Cython tests against CTK wheels, # so let's just skip them. @@ -205,7 +205,7 @@ jobs: # TODO: enable this once win-64 runners are up exit 1 fi - pytest -rxXs -v tests/cython + pytest -ra -s -v tests/cython fi popd @@ -229,7 +229,7 @@ jobs: pushd ./cuda_core pip install -r "tests/requirements-cu${TEST_CUDA_MAJOR}.txt" - pytest -rxXs -v tests/ + pytest -ra -s -v tests/ # It is a bit convoluted to run the Cython tests against CTK wheels, # so let's just skip them. Also, currently our CI always installs the @@ -243,7 +243,7 @@ jobs: # TODO: enable this once win-64 runners are up exit 1 fi - pytest -rxXs -v tests/cython + pytest -ra -s -v tests/cython fi popd diff --git a/.github/workflows/test-wheel-windows.yml b/.github/workflows/test-wheel-windows.yml index 4e48590a3..233f56e4f 100644 --- a/.github/workflows/test-wheel-windows.yml +++ b/.github/workflows/test-wheel-windows.yml @@ -186,7 +186,7 @@ jobs: Push-Location ./cuda_bindings pip install -r requirements.txt - pytest -rxXs -v tests/ + pytest -ra -s -v tests/ # skip Cython tests for now (NVIDIA/cuda-python#466) Pop-Location @@ -210,7 +210,7 @@ jobs: Push-Location ./cuda_core pip install -r "tests/requirements-cu${TEST_CUDA_MAJOR}.txt" - pytest -rxXs -v tests/ + pytest -ra -s -v tests/ Pop-Location - name: Ensure cuda-python installable From 550f18b165a0f89e5aa85f0886cc81312357e887 Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Fri, 7 Mar 2025 12:10:43 -0800 Subject: [PATCH 6/8] Undo debugging changes in nvrtc.pyx.in --- cuda_bindings/cuda/bindings/nvrtc.pyx.in | 8 -------- 1 file changed, 8 deletions(-) diff --git a/cuda_bindings/cuda/bindings/nvrtc.pyx.in b/cuda_bindings/cuda/bindings/nvrtc.pyx.in index b38ca3e1e..6c2704c59 100644 --- a/cuda_bindings/cuda/bindings/nvrtc.pyx.in +++ b/cuda_bindings/cuda/bindings/nvrtc.pyx.in @@ -859,25 +859,17 @@ def nvrtcGetLoweredName(prog, char* name_expression): -------- nvrtcAddNameExpression """ - print("\nLOOOK nvrtcGetLoweredName ENTRY", flush=True) cdef cynvrtc.nvrtcProgram cyprog if prog is None: - print("LOOOK nvrtcGetLoweredName prog is None", flush=True) cyprog = 0 elif isinstance(prog, (nvrtcProgram,)): - print("LOOOK nvrtcGetLoweredName isinstance(prog, (nvrtcProgram,))", flush=True) pprog = int(prog) - print("LOOOK nvrtcGetLoweredName AFTER int(prog)", flush=True) cyprog = pprog else: - print("LOOOK nvrtcGetLoweredName else", flush=True) pprog = int(nvrtcProgram(prog)) - print("LOOOK nvrtcGetLoweredName AFTER int(nvrtcProgram(prog))", flush=True) cyprog = pprog cdef const char* lowered_name = NULL - print("LOOOK nvrtcGetLoweredName BEFORE cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name)", flush=True) err = cynvrtc.nvrtcGetLoweredName(cyprog, name_expression, &lowered_name) - print("LOOOK nvrtcGetLoweredName RETURN\n", flush=True) return (nvrtcResult(err), lowered_name if lowered_name != NULL else None) {{endif}} From 4fbcb9b88bdb91b55b4629d3b4aad49b20472098 Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Fri, 7 Mar 2025 12:11:57 -0800 Subject: [PATCH 7/8] Revert "Change pytest options, mostly to disable output capturing (of both stdout and stderr)" This reverts commit b0464e7cee7d7d9595a2e00cef876652602efc1a. --- .github/workflows/test-wheel-linux.yml | 8 ++++---- .github/workflows/test-wheel-windows.yml | 4 ++-- 2 files changed, 6 insertions(+), 6 deletions(-) diff --git a/.github/workflows/test-wheel-linux.yml b/.github/workflows/test-wheel-linux.yml index 19c78c8cc..322f859e3 100644 --- a/.github/workflows/test-wheel-linux.yml +++ b/.github/workflows/test-wheel-linux.yml @@ -194,7 +194,7 @@ jobs: pushd ./cuda_bindings pip install -r requirements.txt - pytest -ra -s -v tests/ + pytest -rxXs -v tests/ # It is a bit convoluted to run the Cython tests against CTK wheels, # so let's just skip them. @@ -205,7 +205,7 @@ jobs: # TODO: enable this once win-64 runners are up exit 1 fi - pytest -ra -s -v tests/cython + pytest -rxXs -v tests/cython fi popd @@ -229,7 +229,7 @@ jobs: pushd ./cuda_core pip install -r "tests/requirements-cu${TEST_CUDA_MAJOR}.txt" - pytest -ra -s -v tests/ + pytest -rxXs -v tests/ # It is a bit convoluted to run the Cython tests against CTK wheels, # so let's just skip them. Also, currently our CI always installs the @@ -243,7 +243,7 @@ jobs: # TODO: enable this once win-64 runners are up exit 1 fi - pytest -ra -s -v tests/cython + pytest -rxXs -v tests/cython fi popd diff --git a/.github/workflows/test-wheel-windows.yml b/.github/workflows/test-wheel-windows.yml index 233f56e4f..4e48590a3 100644 --- a/.github/workflows/test-wheel-windows.yml +++ b/.github/workflows/test-wheel-windows.yml @@ -186,7 +186,7 @@ jobs: Push-Location ./cuda_bindings pip install -r requirements.txt - pytest -ra -s -v tests/ + pytest -rxXs -v tests/ # skip Cython tests for now (NVIDIA/cuda-python#466) Pop-Location @@ -210,7 +210,7 @@ jobs: Push-Location ./cuda_core pip install -r "tests/requirements-cu${TEST_CUDA_MAJOR}.txt" - pytest -ra -s -v tests/ + pytest -rxXs -v tests/ Pop-Location - name: Ensure cuda-python installable From 73d8f4af31823da3272d8c5f6cc88327c45c7c9b Mon Sep 17 00:00:00 2001 From: "Ralf W. Grosse-Kunstleve" Date: Fri, 7 Mar 2025 12:21:39 -0800 Subject: [PATCH 8/8] Skip new test if nvrtc version < 12.1 --- cuda_bindings/tests/test_nvrtc.py | 1 + 1 file changed, 1 insertion(+) diff --git a/cuda_bindings/tests/test_nvrtc.py b/cuda_bindings/tests/test_nvrtc.py index 1e3e628af..013d75f72 100644 --- a/cuda_bindings/tests/test_nvrtc.py +++ b/cuda_bindings/tests/test_nvrtc.py @@ -31,6 +31,7 @@ def test_nvrtcGetSupportedArchs(): 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