fix(examples): standardize explicit resource cleanup paths

Ensure cuda.core and cuda.bindings examples deterministically release streams, buffers, and loaded modules so examples remain safe when copied into long-running programs.

Made-with: Cursor

Author: cpcloud

cuda_bindings/examples/0_Introduction/clock_nvrtc_test.py

@@ -102,6 +102,7 @@ def main():
     checkCudaErrors(cuda.cuMemFree(dinput))
     checkCudaErrors(cuda.cuMemFree(doutput))
     checkCudaErrors(cuda.cuMemFree(dtimer))
+    kernelHelper.close()
 
     avgElapsedClocks = 0.0
 

cuda_bindings/examples/0_Introduction/simpleCubemapTexture_test.py

@@ -205,6 +205,7 @@ def main():
     checkCudaErrors(cudart.cudaMemcpy(h_odata, d_data, size, cudart.cudaMemcpyKind.cudaMemcpyDeviceToHost))
 
     checkCudaErrors(cudart.cudaDestroyTextureObject(tex))
+    kernelHelper.close()
     checkCudaErrors(cudart.cudaFree(d_data))
     checkCudaErrors(cudart.cudaFreeArray(cu_3darray))
 

cuda_bindings/examples/0_Introduction/simpleP2P_test.py

@@ -227,6 +227,9 @@ def main():
 
     # Cleanup and shutdown
     print("Shutting down...")
+    for helper in kernelHelper:
+        if helper is not None:
+            helper.close()
     checkCudaErrors(cudart.cudaEventDestroy(start_event))
     checkCudaErrors(cudart.cudaEventDestroy(stop_event))
     checkCudaErrors(cudart.cudaSetDevice(gpuid[0]))

cuda_bindings/examples/0_Introduction/simpleZeroCopy_test.py

@@ -163,6 +163,8 @@ def main():
     errorNorm = math.sqrt(errorNorm)
     refNorm = math.sqrt(refNorm)
 
+    kernelHelper.close()
+
     # Memory clean up
 
     print("Releasing CPU memory...")

cuda_bindings/examples/0_Introduction/systemWideAtomics_test.py

@@ -234,6 +234,7 @@ def main():
 
     # Compute & verify reference solution
     testResult = verify(atom_arr_h, numThreads * numBlocks)
+    kernelHelper.close()
 
     if device_prop.pageableMemoryAccess:
         pass

cuda_bindings/examples/0_Introduction/vectorAddDrv_test.py

@@ -104,6 +104,7 @@ def main():
     checkCudaErrors(cuda.cuMemFree(d_A))
     checkCudaErrors(cuda.cuMemFree(d_B))
     checkCudaErrors(cuda.cuMemFree(d_C))
+    kernelHelper.close()
 
     checkCudaErrors(cuda.cuCtxDestroy(cuContext))
     print("{}".format("Result = PASS" if i + 1 == N else "Result = FAIL"))

cuda_bindings/examples/0_Introduction/vectorAddMMAP_test.py

@@ -295,6 +295,7 @@ def main():
     checkCudaErrors(simpleFreeMultiDeviceMmap(d_A, allocationSize))
     checkCudaErrors(simpleFreeMultiDeviceMmap(d_B, allocationSize))
     checkCudaErrors(simpleFreeMultiDeviceMmap(d_C, allocationSize))
+    kernelHelper.close()
 
     checkCudaErrors(cuda.cuCtxDestroy(cuContext))
 

cuda_bindings/examples/2_Concepts_and_Techniques/streamOrderedAllocation_test.py

@@ -239,6 +239,7 @@ def main():
 
     ret1 = basicStreamOrderedAllocation(dev, nelem, a, b, c)
     ret2 = streamOrderedAllocationPostSync(dev, nelem, a, b, c)
+    kernelHelper.close()
 
     if not ret1 or not ret2:
         sys.exit(-1)

cuda_bindings/examples/3_CUDA_Features/globalToShmemAsyncCopy_test.py

@@ -1211,6 +1211,7 @@ def main():
     _MatrixMulNaiveLargeChunk = kernelHelper.getFunction(b"MatrixMulNaiveLargeChunk")
 
     matrix_result = MatrixMultiply(dimsA, dimsB, selected_kernel)
+    kernelHelper.close()
 
     if matrix_result != 0:
         sys.exit(-1)

cuda_bindings/examples/3_CUDA_Features/simpleCudaGraphs_test.py

@@ -356,6 +356,10 @@ def cudaGraphsUsingStreamCapture(inputVec_h, inputVec_d, outputVec_d, result_d,
     checkCudaErrors(cudart.cudaGraphExecDestroy(clonedGraphExec))
     checkCudaErrors(cudart.cudaGraphDestroy(graph))
     checkCudaErrors(cudart.cudaGraphDestroy(clonedGraph))
+    checkCudaErrors(cudart.cudaEventDestroy(memsetEvent2))
+    checkCudaErrors(cudart.cudaEventDestroy(memsetEvent1))
+    checkCudaErrors(cudart.cudaEventDestroy(forkStreamEvent))
+    checkCudaErrors(cudart.cudaStreamDestroy(stream3))
     checkCudaErrors(cudart.cudaStreamDestroy(stream1))
     checkCudaErrors(cudart.cudaStreamDestroy(stream2))
     checkCudaErrors(cudart.cudaStreamDestroy(streamForGraph))
@@ -387,6 +391,7 @@ def main():
 
     cudaGraphsManual(inputVec_h, inputVec_d, outputVec_d, result_d, size, maxBlocks)
     cudaGraphsUsingStreamCapture(inputVec_h, inputVec_d, outputVec_d, result_d, size, maxBlocks)
+    kernelHelper.close()
 
     checkCudaErrors(cudart.cudaFree(inputVec_d))
     checkCudaErrors(cudart.cudaFree(outputVec_d))