Fix race condition in test_cudart_cudaMemcpy3DPeerAsync

The test was experiencing flaky failures, particularly on Windows when
per-thread default stream (PTDS) mode is enabled via
CUDA_PYTHON_CUDA_PER_THREAD_DEFAULT_STREAM=1.

Root cause:
The test performs a synchronous cudaMemcpy from pageable host memory to
device memory, followed immediately by cudaMemcpy3DPeerAsync in an
explicit stream. When using pageable host memory, cudaMemcpy may return
before the DMA transfer to device memory has completed. The subsequent
async copy operation in a different stream can then read from device
memory before the host-to-device transfer finishes, resulting in
incorrect (zero) data being copied.

Why it appeared flaky:
- In legacy default stream mode, blocking streams created with
  cudaStreamCreate() implicitly synchronize with the default stream,
  masking the race condition.
- In PTDS mode, the per-thread default stream does not synchronize with
  other streams, exposing the race and causing intermittent failures.

Fix:
Add cudaStreamSynchronize(0) after cudaMemcpy to explicitly ensure the
DMA transfer to device memory completes before launching the async copy
operation. This establishes proper ordering between the default stream
and the explicit stream, making the test reliable under both legacy
and PTDS stream semantics.

This fix aligns with documented CUDA behavior: synchronous cudaMemcpy
from pageable host memory does not guarantee completion before return,
and explicit synchronization is required when coordinating with work in
other streams.

Author: rwgk

cuda_bindings/tests/test_cudart.py

@@ -1158,6 +1158,10 @@ def test_cudart_cudaMemcpy3DPeerAsync():
     (err,) = cudart.cudaMemcpy(dptr, h1, size, cudart.cudaMemcpyKind.cudaMemcpyHostToDevice)
     assertSuccess(err)
 
+    # ensure the DMA to device memory has completed
+    (err,) = cudart.cudaStreamSynchronize(0)
+    assertSuccess(err)
+
     # D to arr
     (err,) = cudart.cudaMemcpy3DPeerAsync(params, stream)
     assertSuccess(err)