events

Author: brandon-b-miller

numba_cuda/numba/cuda/cudadrv/driver.py

@@ -2221,7 +2221,7 @@ def record(self, stream=0):
         queued in the stream at the time of the call to ``record()`` has been
         completed.
         """
-        hstream = stream.handle.value if stream else binding.CUstream(0)
+        hstream = _stream_handle(stream)
         handle = self.handle.value
         driver.cuEventRecord(handle, hstream)
 
@@ -2236,7 +2236,7 @@ def wait(self, stream=0):
         """
         All future works submitted to stream will wait util the event completes.
         """
-        hstream = stream.handle.value if stream else binding.CUstream(0)
+        hstream = _stream_handle(stream)
         handle = self.handle.value
         flags = 0
         driver.cuStreamWaitEvent(hstream, handle, flags)

numba_cuda/numba/cuda/tests/cudadrv/test_events.py

@@ -4,6 +4,7 @@
 import numpy as np
 from numba import cuda
 from numba.cuda.testing import unittest, CUDATestCase
+from cuda.core.experimental import Device
 
 
 class TestCudaEvent(CUDATestCase):
@@ -22,8 +23,17 @@ def test_event_elapsed(self):
         evtstart.elapsed_time(evtend)
 
     def test_event_elapsed_stream(self):
-        N = 32
         stream = cuda.stream()
+        self.event_elapsed_inner(stream)
+
+    def test_event_elapsed_cuda_core_stream(self):
+        dev = Device()
+        dev.set_current()
+        stream = dev.create_stream()
+        self.event_elapsed_inner(stream)
+
+    def event_elapsed_inner(self, stream):
+        N = 32
         dary = cuda.device_array(N, dtype=np.double)
         evtstart = cuda.event()
         evtend = cuda.event()