fix : hack around missing datetime64 in dlpack

Author: cpcloud

numba_cuda/numba/cuda/cudadrv/devicearray.py

@@ -1009,9 +1009,16 @@ def _to_strided_memory_view(
 ) -> tuple[StridedMemoryView, bool]:
     if _driver.is_device_memory(obj):
         return obj._strided_memory_view_shim, False
-    elif not isinstance(
-        obj, (np.ndarray, _UNSUPPORTED_DLPACK_TYPES)
-    ) and hasattr(obj, "__dlpack__"):
+    elif (
+        not isinstance(obj, (np.ndarray, _UNSUPPORTED_DLPACK_TYPES))
+        and hasattr(obj, "__dlpack__")
+        and (
+            (dtype := getattr(obj, "dtype", None)) is None
+            or not issubclass(
+                getattr(dtype, "type", None), _UNSUPPORTED_DLPACK_TYPES
+            )
+        )
+    ):
         # numpy arrays need to be copied to the device
         # so we can't view them as SMVs until then
         #

numba_cuda/numba/cuda/tests/cudapy/test_datetime.py

@@ -8,6 +8,8 @@
 from numba.cuda.testing import CUDATestCase, skip_on_cudasim
 import unittest
 
+import pytest
+
 
 class TestCudaDateTime(CUDATestCase):
     def test_basic_datetime_kernel(self):
@@ -58,6 +60,26 @@ def timediff(start, end):
 
         self.assertPreciseEqual(delta, arr2 - arr1)
 
+    @skip_on_cudasim("API unsupported in the simulator")
+    def test_datetime_cupy_inputs(self):
+        cp = pytest.importorskip("cupy")
+        datetime_t = from_dtype(cp.dtype("datetime64[D]"))
+
+        @cuda.jit
+        def assign(out, arr):
+            for i in range(arr.shape[0]):
+                out[i] = arr[i]
+
+        # TODO: cupy doesn't allow passing the datetime64[D] array directly
+        arr = cp.array(
+            np.arange("2005-02", "2006-02", dtype="datetime64[D]").view("int64")
+        ).view("datetime64[D]")
+
+        out = cp.empty_like(arr)
+        assign[1, 1](out, arr)
+
+        self.assertPreciseEqual(arr.get(), out.get())
+
     @skip_on_cudasim("ufunc API unsupported in the simulator")
     def test_gufunc(self):
         datetime_t = from_dtype(np.dtype("datetime64[D]"))