From 30834c0c898e0395f3def1d131cb497e11f3c563 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 06:58:02 -0700 Subject: [PATCH 01/20] Add SyclQueue.memset method --- dpctl/_sycl_queue.pxd | 1 + dpctl/_sycl_queue.pyx | 43 +++++++++++++++++++++++++++++++++++++++++++ 2 files changed, 44 insertions(+) diff --git a/dpctl/_sycl_queue.pxd b/dpctl/_sycl_queue.pxd index c2102a52e9..66d80178f5 100644 --- a/dpctl/_sycl_queue.pxd +++ b/dpctl/_sycl_queue.pxd @@ -107,6 +107,7 @@ cdef public api class SyclQueue (_SyclQueue) [ cpdef SyclEvent copy_async( self, dest, src, size_t count, list dEvents=*, str dtype=* ) + cpdef memset(self, mem, int val, size_t count=*) cpdef prefetch(self, ptr, size_t count=*) cpdef mem_advise(self, ptr, size_t count, int mem) cpdef SyclEvent submit_barrier(self, dependent_events=*) diff --git a/dpctl/_sycl_queue.pyx b/dpctl/_sycl_queue.pyx index a22586b71e..2c609e788a 100644 --- a/dpctl/_sycl_queue.pyx +++ b/dpctl/_sycl_queue.pyx @@ -49,6 +49,7 @@ from ._backend cimport ( # noqa: E211 DPCTLQueue_MemAdvise, DPCTLQueue_Memcpy, DPCTLQueue_MemcpyWithEvents, + DPCTLQueue_Memset, DPCTLQueue_Prefetch, DPCTLQueue_SubmitBarrierForEvents, DPCTLQueue_SubmitNDRange, @@ -1594,6 +1595,48 @@ cdef class SyclQueue(_SyclQueue): return SyclEvent._create(ERef) + cpdef memset(self, mem, int val, size_t count=0): + """Fill USM allocation ``mem`` with the byte value ``val`` and wait. + + Internally, this dispatches ``sycl::queue::memset``. The operation is + byte-wise: ``count`` bytes are set, each to the same value ``val``. + + Args: + mem: + Destination USM allocation, an instance of + :class:`dpctl.memory._Memory`. + val (int): + Value to fill ``mem`` with. Following ``sycl::queue::memset``, + it is interpreted as an ``unsigned char``, i.e. only the least + significant byte is used. + count (int, optional): + Number of bytes to fill. If ``0`` or greater than the size of + ``mem``, the whole allocation is filled. Default: ``0``. + + Raises: + TypeError: + If ``mem`` is not an instance of :class:`dpctl.memory._Memory`. + """ + cdef void *ptr + cdef DPCTLSyclEventRef ERef = NULL + + if isinstance(mem, _Memory): + ptr = (<_Memory>mem).get_data_ptr() + else: + raise TypeError("Parameter `mem` should have type _Memory") + + if (count <= 0 or count > mem.nbytes): + count = mem.nbytes + + ERef = DPCTLQueue_Memset(self._queue_ref, ptr, val, count) + if (ERef is NULL): + raise RuntimeError( + "SyclQueue.memset operation encountered an error" + ) + with nogil: + DPCTLEvent_Wait(ERef) + DPCTLEvent_Delete(ERef) + cpdef prefetch(self, mem, size_t count=0): cdef void *ptr cdef DPCTLSyclEventRef ERef = NULL From 4f08b2f2416837d4f89247ae3e7efe07af62ea7b Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 06:59:13 -0700 Subject: [PATCH 02/20] Add tests for SyclQueue.memset --- dpctl/tests/test_sycl_queue_memset.py | 106 ++++++++++++++++++++++++++ 1 file changed, 106 insertions(+) create mode 100644 dpctl/tests/test_sycl_queue_memset.py diff --git a/dpctl/tests/test_sycl_queue_memset.py b/dpctl/tests/test_sycl_queue_memset.py new file mode 100644 index 0000000000..c97413a6b9 --- /dev/null +++ b/dpctl/tests/test_sycl_queue_memset.py @@ -0,0 +1,106 @@ +# Data Parallel Control (dpctl) +# +# Copyright 2026 Intel Corporation +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +"""Defines unit test cases for the SyclQueue.memset.""" + +import pytest + +import dpctl +import dpctl.memory + + +def _create_memory(q, nbytes=1024): + return dpctl.memory.MemoryUSMShared(nbytes, queue=q) + + +def test_memset_fills_whole_allocation(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 256 + mobj = _create_memory(q, nbytes) + + q.memset(mobj, 0xAB) + + assert bytes(memoryview(mobj)) == b"\xab" * nbytes + + +def test_memset_zero_count_fills_whole_allocation(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 64 + mobj = _create_memory(q, nbytes) + + q.memset(mobj, 0x01, 0) + + assert bytes(memoryview(mobj)) == b"\x01" * nbytes + + +def test_memset_partial_count(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 16 + mobj = _create_memory(q, nbytes) + + # zero-out first, then fill only the leading 4 bytes + q.memset(mobj, 0x00) + q.memset(mobj, 0x7F, 4) + + assert bytes(memoryview(mobj)) == b"\x7f" * 4 + b"\x00" * (nbytes - 4) + + +def test_memset_count_clamped_to_allocation(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 8 + mobj = _create_memory(q, nbytes) + + # requesting more bytes than allocated must not overrun; it is clamped + q.memset(mobj, 0x02, 4 * nbytes) + + assert bytes(memoryview(mobj)) == b"\x02" * nbytes + + +def test_memset_zero_value(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 32 + mobj = _create_memory(q, nbytes) + + q.memset(mobj, 0xFF) + q.memset(mobj, 0) + + assert bytes(memoryview(mobj)) == b"\x00" * nbytes + + +def test_memset_type_error(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + + with pytest.raises(TypeError) as cm: + q.memset(None, 1) + assert "_Memory" in str(cm.value) From a591bcf31aaf2a3d90c45ac1f2e9bfa535afcbcf Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 07:04:10 -0700 Subject: [PATCH 03/20] Add gh-2361 to changelog --- CHANGELOG.md | 1 + 1 file changed, 1 insertion(+) diff --git a/CHANGELOG.md b/CHANGELOG.md index 6eff2d59ac..a23ce6c315 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -11,6 +11,7 @@ and this project adheres to [Semantic Versioning](https://semver.org/spec/v2.0.0 * Added a number of `sycl::device` info queries to `dpctl.SyclDevice` [gh-2324](https://github.com/IntelPython/dpctl/pull/2324) * Added `sycl::info::context` queries `sycl_platform`, `atomic_memory_order_capabilities`, `atomic_fence_order_capabilities`, `atomic_memory_scope_capabilities`, and `atomic_fence_scope_capabilities` to `dpctl.SyclContext` [gh-2354](https://github.com/IntelPython/dpctl/pull/2354) * Added `create_kernel_bundle_from_sycl_source`, `is_sycl_source_compilation_available`, and `dpctl.SyclDevice.can_compile` for supporting the creation of `dpctl.SyclKernelBundle`s from SYCL source strings via DPC++ extension, as well as corresponding C-API functions to support it [gh-2206](https://github.com/IntelPython/dpctl/pull/2206) +* Added `dpctl.SyclQueue.memset` method [gh-2361](https://github.com/IntelPython/dpctl/pull/2361) ### Changed * Bump minimum NumPy version to 1.26 [gh-2192](https://github.com/IntelPython/dpctl/pull/2192) From af2da99124a405c6082065a5293fe8d7a6529b4d Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 08:17:08 -0700 Subject: [PATCH 04/20] Add DPCTLQueue_MemsetWithEvents C-API function --- .../dpctl_sycl_queue_interface.h | 23 ++++++++++++ .../source/dpctl_sycl_queue_interface.cpp | 36 +++++++++++++++++++ 2 files changed, 59 insertions(+) diff --git a/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h b/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h index afd2b3240d..6d35984abf 100644 --- a/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h +++ b/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h @@ -481,6 +481,29 @@ DPCTLQueue_Memset(__dpctl_keep const DPCTLSyclQueueRef QRef, uint8_t Value, size_t Count); +/*! + * @brief C-API wrapper for ``sycl::queue::memset``. + * + * @param QRef An opaque pointer to the ``sycl::queue``. + * @param USMRef An USM pointer to the memory to fill. + * @param Value A value to fill. + * @param Count A number of uint8_t elements to fill. + * @param DepEvents A pointer to array of DPCTLSyclEventRef opaque + * pointers to dependent events. + * @param DepEventsCount A number of dependent events. + * @return An opaque pointer to the ``sycl::event`` returned by the + * ``sycl::queue::memset`` function. + * @ingroup QueueInterface + */ +DPCTL_API +__dpctl_give DPCTLSyclEventRef +DPCTLQueue_MemsetWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef, + void *USMRef, + uint8_t Value, + size_t Count, + __dpctl_keep const DPCTLSyclEventRef *DepEvents, + size_t DepEventsCount); + /*! * @brief C-API wrapper for ``sycl::queue::fill``. * diff --git a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp index 575a3c13fa..c2f6701999 100644 --- a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp @@ -905,6 +905,42 @@ DPCTLQueue_Memset(__dpctl_keep const DPCTLSyclQueueRef QRef, } }; +__dpctl_give DPCTLSyclEventRef +DPCTLQueue_MemsetWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef, + void *USMRef, + uint8_t Value, + size_t Count, + const DPCTLSyclEventRef *DepEvents, + size_t DepEventsCount) +{ + event ev; + auto Q = unwrap(QRef); + if (Q && USMRef) { + try { + ev = Q->submit([&](handler &cgh) { + if (DepEvents) + for (size_t i = 0; i < DepEventsCount; ++i) { + event *ei = unwrap(DepEvents[i]); + if (ei) + cgh.depends_on(*ei); + } + + cgh.memset(USMRef, static_cast(Value), Count); + }); + } catch (const std::exception &ex) { + error_handler(ex, __FILE__, __func__, __LINE__); + return nullptr; + } + } + else { + error_handler("QRef or USMRef passed to memset were NULL.", __FILE__, + __func__, __LINE__); + return nullptr; + } + + return wrap(new event(ev)); +}; + __dpctl_give DPCTLSyclEventRef DPCTLQueue_Fill8(__dpctl_keep const DPCTLSyclQueueRef QRef, void *USMRef, From 4b197d5847640d66305f0cc936dd66aaaaf665d9 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 08:18:51 -0700 Subject: [PATCH 05/20] Add tests for DPCTLQueue_MemsetWithEvents --- .../tests/test_sycl_queue_interface.cpp | 44 +++++++++++++++++++ 1 file changed, 44 insertions(+) diff --git a/libsyclinterface/tests/test_sycl_queue_interface.cpp b/libsyclinterface/tests/test_sycl_queue_interface.cpp index b95aae3e76..e140008ff3 100644 --- a/libsyclinterface/tests/test_sycl_queue_interface.cpp +++ b/libsyclinterface/tests/test_sycl_queue_interface.cpp @@ -467,6 +467,10 @@ TEST(TestDPCTLSyclQueueInterface, CheckMemsetNullQRef) ASSERT_NO_FATAL_FAILURE(ERef = DPCTLQueue_Memset(QRef, p, val8, 1)); ASSERT_FALSE(bool(ERef)); + + ASSERT_NO_FATAL_FAILURE( + ERef = DPCTLQueue_MemsetWithEvents(QRef, p, val8, 1, NULL, 0)); + ASSERT_FALSE(bool(ERef)); } TEST_P(TestDPCTLQueueMemberFunctions, CheckMemset) @@ -534,6 +538,46 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckMemset2) delete[] host_arr; } +TEST_P(TestDPCTLQueueMemberFunctions, CheckMemsetWithEvents) +{ + DPCTLSyclUSMRef p = nullptr; + DPCTLSyclEventRef Memset_ERef = nullptr; + DPCTLSyclEventRef MemsetWithEvents_ERef = nullptr; + DPCTLSyclEventRef Memcpy_ERef = nullptr; + uint8_t val1 = 42; + uint8_t val2 = 73; + size_t nbytes = 256; + uint8_t *host_arr = new uint8_t[nbytes]; + + ASSERT_FALSE(host_arr == nullptr); + + ASSERT_NO_FATAL_FAILURE(p = DPCTLmalloc_device(nbytes, QRef)); + ASSERT_FALSE(p == nullptr); + + ASSERT_NO_FATAL_FAILURE( + Memset_ERef = DPCTLQueue_Memset(QRef, (void *)p, val1, nbytes)); + + ASSERT_NO_FATAL_FAILURE( + MemsetWithEvents_ERef = DPCTLQueue_MemsetWithEvents( + QRef, (void *)p, val2, nbytes, &Memset_ERef, 1)); + + ASSERT_NO_FATAL_FAILURE( + Memcpy_ERef = DPCTLQueue_MemcpyWithEvents(QRef, host_arr, p, nbytes, + &MemsetWithEvents_ERef, 1)); + ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(Memcpy_ERef)); + + ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memset_ERef)); + ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(MemsetWithEvents_ERef)); + ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memcpy_ERef)); + + ASSERT_NO_FATAL_FAILURE(DPCTLfree_with_queue(p, QRef)); + + for (size_t i = 0; i < nbytes; ++i) { + ASSERT_TRUE(host_arr[i] == val2); + } + delete[] host_arr; +} + TEST(TestDPCTLSyclQueueInterface, CheckFillNullQRef) { DPCTLSyclQueueRef QRef = nullptr; From 9d63c23e822d3da0d33f22d5abe3ec77e655ccf7 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 08:21:30 -0700 Subject: [PATCH 06/20] Add DPCTLQueue_MemsetWithEvents declaration to _backend.pxd --- dpctl/_backend.pxd | 7 +++++++ 1 file changed, 7 insertions(+) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 21a301d94d..29da03ece7 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -679,6 +679,13 @@ cdef extern from "syclinterface/dpctl_sycl_queue_interface.h": void *Dest, int Val, size_t Count) + cdef DPCTLSyclEventRef DPCTLQueue_MemsetWithEvents( + const DPCTLSyclQueueRef Q, + void *Dest, + int Val, + size_t Count, + const DPCTLSyclEventRef *depEvents, + size_t depEventsCount) cdef DPCTLSyclEventRef DPCTLQueue_Prefetch( const DPCTLSyclQueueRef Q, const void *Src, From fd541c008a91c65e29d2445ee5ba24b265c090a0 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 08:22:56 -0700 Subject: [PATCH 07/20] Add SyclQueue.memset_async method --- dpctl/_sycl_queue.pxd | 3 ++ dpctl/_sycl_queue.pyx | 105 ++++++++++++++++++++++++++++++++++++++---- 2 files changed, 98 insertions(+), 10 deletions(-) diff --git a/dpctl/_sycl_queue.pxd b/dpctl/_sycl_queue.pxd index 66d80178f5..ff5e7b8471 100644 --- a/dpctl/_sycl_queue.pxd +++ b/dpctl/_sycl_queue.pxd @@ -108,6 +108,9 @@ cdef public api class SyclQueue (_SyclQueue) [ self, dest, src, size_t count, list dEvents=*, str dtype=* ) cpdef memset(self, mem, int val, size_t count=*) + cpdef SyclEvent memset_async( + self, mem, int val, size_t count=*, list dEvents=* + ) cpdef prefetch(self, ptr, size_t count=*) cpdef mem_advise(self, ptr, size_t count, int mem) cpdef SyclEvent submit_barrier(self, dependent_events=*) diff --git a/dpctl/_sycl_queue.pyx b/dpctl/_sycl_queue.pyx index 2c609e788a..4289d25d1c 100644 --- a/dpctl/_sycl_queue.pyx +++ b/dpctl/_sycl_queue.pyx @@ -50,6 +50,7 @@ from ._backend cimport ( # noqa: E211 DPCTLQueue_Memcpy, DPCTLQueue_MemcpyWithEvents, DPCTLQueue_Memset, + DPCTLQueue_MemsetWithEvents, DPCTLQueue_Prefetch, DPCTLQueue_SubmitBarrierForEvents, DPCTLQueue_SubmitNDRange, @@ -603,6 +604,35 @@ cdef DPCTLSyclEventRef _copy_impl( ) +cdef DPCTLSyclEventRef _memset_impl( + SyclQueue q, + object mem, + int val, + size_t count, + DPCTLSyclEventRef *dep_events, + size_t dep_events_count, +) except *: + cdef void *ptr = NULL + cdef DPCTLSyclEventRef ERef = NULL + + if isinstance(mem, _Memory): + ptr = (<_Memory>mem).get_data_ptr() + else: + raise TypeError("Parameter `mem` should have type _Memory") + + if count <= 0 or count > mem.nbytes: + count = mem.nbytes + + if dep_events_count == 0 or dep_events is NULL: + ERef = DPCTLQueue_Memset(q._queue_ref, ptr, val, count) + else: + ERef = DPCTLQueue_MemsetWithEvents( + q._queue_ref, ptr, val, count, dep_events, dep_events_count + ) + + return ERef + + cdef class _SyclQueue: """ Barebone data owner class used by SyclQueue. """ @@ -1617,18 +1647,9 @@ cdef class SyclQueue(_SyclQueue): TypeError: If ``mem`` is not an instance of :class:`dpctl.memory._Memory`. """ - cdef void *ptr cdef DPCTLSyclEventRef ERef = NULL - if isinstance(mem, _Memory): - ptr = (<_Memory>mem).get_data_ptr() - else: - raise TypeError("Parameter `mem` should have type _Memory") - - if (count <= 0 or count > mem.nbytes): - count = mem.nbytes - - ERef = DPCTLQueue_Memset(self._queue_ref, ptr, val, count) + ERef = _memset_impl(self, mem, val, count, NULL, 0) if (ERef is NULL): raise RuntimeError( "SyclQueue.memset operation encountered an error" @@ -1637,6 +1658,70 @@ cdef class SyclQueue(_SyclQueue): DPCTLEvent_Wait(ERef) DPCTLEvent_Delete(ERef) + cpdef SyclEvent memset_async( + self, mem, int val, size_t count=0, list dEvents=None + ): + """Fill USM allocation ``mem`` with the byte value ``val`` + asynchronously. + + Internally, this dispatches ``sycl::queue::memset``. The operation is + byte-wise: ``count`` bytes are set, each to the same value ``val``. + + Args: + mem: + Destination USM allocation, an instance of + :class:`dpctl.memory._Memory`. + val (int): + Value to fill ``mem`` with. Following ``sycl::queue::memset``, + it is interpreted as an ``unsigned char``, i.e. only the least + significant byte is used. + count (int, optional): + Number of bytes to fill. If ``0`` or greater than the size of + ``mem``, the whole allocation is filled. Default: ``0``. + dEvents (List[dpctl.SyclEvent], optional): + Events that this operation depends on. + + Returns: + dpctl.SyclEvent: + Event associated with the memset operation. + + Raises: + TypeError: + If ``mem`` is not an instance of :class:`dpctl.memory._Memory`, + or ``dEvents`` is not a sequence of :class:`dpctl.SyclEvent`. + """ + cdef DPCTLSyclEventRef ERef = NULL + cdef DPCTLSyclEventRef *depEvents = NULL + cdef size_t nDE = 0 + + if dEvents is None: + ERef = _memset_impl(self, mem, val, count, NULL, 0) + else: + nDE = len(dEvents) + depEvents = ( + malloc(nDE*sizeof(DPCTLSyclEventRef)) + ) + if depEvents is NULL: + raise MemoryError() + try: + for idx, de in enumerate(dEvents): + if isinstance(de, SyclEvent): + depEvents[idx] = (de).get_event_ref() + else: + raise TypeError( + "A sequence of dpctl.SyclEvent is expected" + ) + ERef = _memset_impl(self, mem, val, count, depEvents, nDE) + finally: + free(depEvents) + + if (ERef is NULL): + raise RuntimeError( + "SyclQueue.memset operation encountered an error" + ) + + return SyclEvent._create(ERef) + cpdef prefetch(self, mem, size_t count=0): cdef void *ptr cdef DPCTLSyclEventRef ERef = NULL From 9a3fbe17c200de42f378881b0ad01b04314837fb Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 08:33:14 -0700 Subject: [PATCH 08/20] Fix typos and improve DPCTLQueue_Memset doc --- .../include/syclinterface/dpctl_sycl_queue_interface.h | 6 +++--- libsyclinterface/source/dpctl_sycl_queue_interface.cpp | 2 +- 2 files changed, 4 insertions(+), 4 deletions(-) diff --git a/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h b/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h index 6d35984abf..ee0b37888c 100644 --- a/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h +++ b/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h @@ -468,10 +468,10 @@ __dpctl_give DPCTLSyclEventRef DPCTLQueue_SubmitBarrierForEvents( * * @param QRef An opaque pointer to the ``sycl::queue``. * @param USMRef An USM pointer to the memory to fill. - * @param Value A value to fill. + * @param Value A value to fill, interpreted as an unsigned char. * @param Count A number of uint8_t elements to fill. * @return An opaque pointer to the ``sycl::event`` returned by the - * ``sycl::queue::fill`` function. + * ``sycl::queue::memset`` function. * @ingroup QueueInterface */ DPCTL_API @@ -486,7 +486,7 @@ DPCTLQueue_Memset(__dpctl_keep const DPCTLSyclQueueRef QRef, * * @param QRef An opaque pointer to the ``sycl::queue``. * @param USMRef An USM pointer to the memory to fill. - * @param Value A value to fill. + * @param Value A value to fill, interpreted as an unsigned char. * @param Count A number of uint8_t elements to fill. * @param DepEvents A pointer to array of DPCTLSyclEventRef opaque * pointers to dependent events. diff --git a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp index c2f6701999..f9207b6cec 100644 --- a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp @@ -899,7 +899,7 @@ DPCTLQueue_Memset(__dpctl_keep const DPCTLSyclQueueRef QRef, return wrap(new event(std::move(ev))); } else { - error_handler("QRef or USMRef passed to fill8 were NULL.", __FILE__, + error_handler("QRef or USMRef passed to memset were NULL.", __FILE__, __func__, __LINE__); return nullptr; } From c905349fc3fc4b94029796848e39788c8286a93e Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 08:33:59 -0700 Subject: [PATCH 09/20] Add tests for SyclQueue.memset_async --- dpctl/tests/test_sycl_queue_memset.py | 60 +++++++++++++++++++++++++++ 1 file changed, 60 insertions(+) diff --git a/dpctl/tests/test_sycl_queue_memset.py b/dpctl/tests/test_sycl_queue_memset.py index c97413a6b9..f6545fbd47 100644 --- a/dpctl/tests/test_sycl_queue_memset.py +++ b/dpctl/tests/test_sycl_queue_memset.py @@ -104,3 +104,63 @@ def test_memset_type_error(): with pytest.raises(TypeError) as cm: q.memset(None, 1) assert "_Memory" in str(cm.value) + + +def test_memset_async(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 64 + mobj = _create_memory(q, nbytes) + + e = q.memset_async(mobj, 0xAB) + assert isinstance(e, dpctl.SyclEvent) + e.wait() + + assert bytes(memoryview(mobj)) == b"\xab" * nbytes + + +def test_memset_async_with_dependent_events(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 64 + mobj = _create_memory(q, nbytes) + + e1 = q.memset_async(mobj, 0x01) + e2 = q.memset_async(mobj, 0x02, nbytes, [e1]) + e2.wait() + + assert bytes(memoryview(mobj)) == b"\x02" * nbytes + + +def test_memset_async_partial_count(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 16 + mobj = _create_memory(q, nbytes) + + q.memset(mobj, 0x00) + e = q.memset_async(mobj, 0x7F, 4) + e.wait() + + assert bytes(memoryview(mobj)) == b"\x7f" * 4 + b"\x00" * (nbytes - 4) + + +def test_memset_async_type_error(): + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + mobj = _create_memory(q) + + with pytest.raises(TypeError) as cm: + q.memset_async(None, 1) + assert "_Memory" in str(cm.value) + + with pytest.raises(TypeError): + q.memset_async(mobj, 1, 0, [None]) From 8e1cfde7e3731e357e0bfa3de28d48f31035dcaf Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 13 Aug 2026 08:37:10 -0700 Subject: [PATCH 10/20] Update gh-2361 changelog entry --- CHANGELOG.md | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/CHANGELOG.md b/CHANGELOG.md index a23ce6c315..f51c7785ed 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -11,7 +11,8 @@ and this project adheres to [Semantic Versioning](https://semver.org/spec/v2.0.0 * Added a number of `sycl::device` info queries to `dpctl.SyclDevice` [gh-2324](https://github.com/IntelPython/dpctl/pull/2324) * Added `sycl::info::context` queries `sycl_platform`, `atomic_memory_order_capabilities`, `atomic_fence_order_capabilities`, `atomic_memory_scope_capabilities`, and `atomic_fence_scope_capabilities` to `dpctl.SyclContext` [gh-2354](https://github.com/IntelPython/dpctl/pull/2354) * Added `create_kernel_bundle_from_sycl_source`, `is_sycl_source_compilation_available`, and `dpctl.SyclDevice.can_compile` for supporting the creation of `dpctl.SyclKernelBundle`s from SYCL source strings via DPC++ extension, as well as corresponding C-API functions to support it [gh-2206](https://github.com/IntelPython/dpctl/pull/2206) -* Added `dpctl.SyclQueue.memset` method [gh-2361](https://github.com/IntelPython/dpctl/pull/2361) +* Added `dpctl.SyclQueue.memset` and `dpctl.SyclQueue.memset_async` methods [gh-2361](https://github.com/IntelPython/dpctl/pull/2361) +* Added `DPCTLQueue_MemsetWithEvents` C-API function to support `dpctl.SyclQueue.memset_async` [gh-2361](https://github.com/IntelPython/dpctl/pull/2361) ### Changed * Bump minimum NumPy version to 1.26 [gh-2192](https://github.com/IntelPython/dpctl/pull/2192) From 24d58fa9a64d3de5dc9f67c27b3ee24da41f3d70 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 02:35:22 -0700 Subject: [PATCH 11/20] Fix memset_async error message and improve memset docstrings --- dpctl/_sycl_queue.pyx | 9 ++++++++- 1 file changed, 8 insertions(+), 1 deletion(-) diff --git a/dpctl/_sycl_queue.pyx b/dpctl/_sycl_queue.pyx index 4289d25d1c..df54850456 100644 --- a/dpctl/_sycl_queue.pyx +++ b/dpctl/_sycl_queue.pyx @@ -1631,6 +1631,9 @@ cdef class SyclQueue(_SyclQueue): Internally, this dispatches ``sycl::queue::memset``. The operation is byte-wise: ``count`` bytes are set, each to the same value ``val``. + This is a synchronizing variant corresponding to + :meth:`dpctl.SyclQueue.memset_async`. + Args: mem: Destination USM allocation, an instance of @@ -1646,6 +1649,8 @@ cdef class SyclQueue(_SyclQueue): Raises: TypeError: If ``mem`` is not an instance of :class:`dpctl.memory._Memory`. + RuntimeError: + If the memset operation encountered an error. """ cdef DPCTLSyclEventRef ERef = NULL @@ -1689,6 +1694,8 @@ cdef class SyclQueue(_SyclQueue): TypeError: If ``mem`` is not an instance of :class:`dpctl.memory._Memory`, or ``dEvents`` is not a sequence of :class:`dpctl.SyclEvent`. + RuntimeError: + If the memset operation encountered an error. """ cdef DPCTLSyclEventRef ERef = NULL cdef DPCTLSyclEventRef *depEvents = NULL @@ -1717,7 +1724,7 @@ cdef class SyclQueue(_SyclQueue): if (ERef is NULL): raise RuntimeError( - "SyclQueue.memset operation encountered an error" + "SyclQueue.memset_async operation encountered an error" ) return SyclEvent._create(ERef) From 5661ab9936e1ce2b1ee4c8ccc8fed2816aca980c Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 02:43:30 -0700 Subject: [PATCH 12/20] Fix memset value type to uint8_t --- dpctl/_backend.pxd | 6 +++--- dpctl/_sycl_queue.pyx | 11 +++++++---- 2 files changed, 10 insertions(+), 7 deletions(-) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 29da03ece7..aa7769aaa0 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -21,7 +21,7 @@ types defined by dpctl's C API. """ -from libc.stdint cimport int64_t, uint32_t, uint64_t +from libc.stdint cimport int64_t, uint8_t, uint32_t, uint64_t from libcpp cimport bool @@ -677,12 +677,12 @@ cdef extern from "syclinterface/dpctl_sycl_queue_interface.h": cdef DPCTLSyclEventRef DPCTLQueue_Memset( const DPCTLSyclQueueRef Q, void *Dest, - int Val, + uint8_t Val, size_t Count) cdef DPCTLSyclEventRef DPCTLQueue_MemsetWithEvents( const DPCTLSyclQueueRef Q, void *Dest, - int Val, + uint8_t Val, size_t Count, const DPCTLSyclEventRef *depEvents, size_t depEventsCount) diff --git a/dpctl/_sycl_queue.pyx b/dpctl/_sycl_queue.pyx index df54850456..0c135a078d 100644 --- a/dpctl/_sycl_queue.pyx +++ b/dpctl/_sycl_queue.pyx @@ -88,6 +88,7 @@ from cpython.buffer cimport ( PyObject_GetBuffer, ) from cpython.ref cimport Py_INCREF, PyObject +from libc.stdint cimport uint8_t from libc.stdlib cimport free, malloc import collections.abc @@ -607,7 +608,7 @@ cdef DPCTLSyclEventRef _copy_impl( cdef DPCTLSyclEventRef _memset_impl( SyclQueue q, object mem, - int val, + uint8_t val, size_t count, DPCTLSyclEventRef *dep_events, size_t dep_events_count, @@ -1653,8 +1654,9 @@ cdef class SyclQueue(_SyclQueue): If the memset operation encountered an error. """ cdef DPCTLSyclEventRef ERef = NULL + cdef uint8_t byte_val = val - ERef = _memset_impl(self, mem, val, count, NULL, 0) + ERef = _memset_impl(self, mem, byte_val, count, NULL, 0) if (ERef is NULL): raise RuntimeError( "SyclQueue.memset operation encountered an error" @@ -1700,9 +1702,10 @@ cdef class SyclQueue(_SyclQueue): cdef DPCTLSyclEventRef ERef = NULL cdef DPCTLSyclEventRef *depEvents = NULL cdef size_t nDE = 0 + cdef uint8_t byte_val = val if dEvents is None: - ERef = _memset_impl(self, mem, val, count, NULL, 0) + ERef = _memset_impl(self, mem, byte_val, count, NULL, 0) else: nDE = len(dEvents) depEvents = ( @@ -1718,7 +1721,7 @@ cdef class SyclQueue(_SyclQueue): raise TypeError( "A sequence of dpctl.SyclEvent is expected" ) - ERef = _memset_impl(self, mem, val, count, depEvents, nDE) + ERef = _memset_impl(self, mem, byte_val, count, depEvents, nDE) finally: free(depEvents) From 51c36df0092737edfd5b07f9a8ad6cca1107b7f1 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 02:54:42 -0700 Subject: [PATCH 13/20] Cover host/device memory and value truncation in memset tests --- dpctl/tests/test_sycl_queue_memset.py | 53 ++++++++++++++++++++++++--- 1 file changed, 47 insertions(+), 6 deletions(-) diff --git a/dpctl/tests/test_sycl_queue_memset.py b/dpctl/tests/test_sycl_queue_memset.py index f6545fbd47..a956cf1853 100644 --- a/dpctl/tests/test_sycl_queue_memset.py +++ b/dpctl/tests/test_sycl_queue_memset.py @@ -26,17 +26,32 @@ def _create_memory(q, nbytes=1024): return dpctl.memory.MemoryUSMShared(nbytes, queue=q) -def test_memset_fills_whole_allocation(): +_MEMORY_CLASSES = [ + dpctl.memory.MemoryUSMShared, + dpctl.memory.MemoryUSMHost, + dpctl.memory.MemoryUSMDevice, +] + + +def _read_back(q, mobj, nbytes): + """Copy USM memory to host and return it as bytes (works for device).""" + host = bytearray(nbytes) + q.memcpy(host, mobj, nbytes) + return bytes(host) + + +@pytest.mark.parametrize("mem_cls", _MEMORY_CLASSES) +def test_memset_fills_whole_allocation(mem_cls): try: q = dpctl.SyclQueue() except dpctl.SyclQueueCreationError: pytest.skip("Default constructor for SyclQueue failed") nbytes = 256 - mobj = _create_memory(q, nbytes) + mobj = mem_cls(nbytes, queue=q) q.memset(mobj, 0xAB) - assert bytes(memoryview(mobj)) == b"\xab" * nbytes + assert _read_back(q, mobj, nbytes) == b"\xab" * nbytes def test_memset_zero_count_fills_whole_allocation(): @@ -106,19 +121,20 @@ def test_memset_type_error(): assert "_Memory" in str(cm.value) -def test_memset_async(): +@pytest.mark.parametrize("mem_cls", _MEMORY_CLASSES) +def test_memset_async(mem_cls): try: q = dpctl.SyclQueue() except dpctl.SyclQueueCreationError: pytest.skip("Default constructor for SyclQueue failed") nbytes = 64 - mobj = _create_memory(q, nbytes) + mobj = mem_cls(nbytes, queue=q) e = q.memset_async(mobj, 0xAB) assert isinstance(e, dpctl.SyclEvent) e.wait() - assert bytes(memoryview(mobj)) == b"\xab" * nbytes + assert _read_back(q, mobj, nbytes) == b"\xab" * nbytes def test_memset_async_with_dependent_events(): @@ -164,3 +180,28 @@ def test_memset_async_type_error(): with pytest.raises(TypeError): q.memset_async(mobj, 1, 0, [None]) + + +@pytest.mark.parametrize( + "val, expected", + [ + (0xAB, 0xAB), + (0, 0x00), + (255, 0xFF), + (256, 0x00), + (-1, 0xFF), + (300, 0x2C), + ], +) +def test_memset_value_truncated_to_byte(val, expected): + # ``val`` is used as a single byte, so values wrap modulo 256 + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nbytes = 8 + mobj = _create_memory(q, nbytes) + + q.memset(mobj, val) + + assert bytes(memoryview(mobj)) == bytes([expected]) * nbytes From 0f08fe20bb994c02906b7bfa2d67204410dd2f63 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 05:32:45 -0700 Subject: [PATCH 14/20] Fix stray semicolons and Count docstring in memset C-API --- .../include/syclinterface/dpctl_sycl_queue_interface.h | 4 ++-- libsyclinterface/source/dpctl_sycl_queue_interface.cpp | 4 ++-- 2 files changed, 4 insertions(+), 4 deletions(-) diff --git a/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h b/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h index ee0b37888c..47185c3fa5 100644 --- a/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h +++ b/libsyclinterface/include/syclinterface/dpctl_sycl_queue_interface.h @@ -469,7 +469,7 @@ __dpctl_give DPCTLSyclEventRef DPCTLQueue_SubmitBarrierForEvents( * @param QRef An opaque pointer to the ``sycl::queue``. * @param USMRef An USM pointer to the memory to fill. * @param Value A value to fill, interpreted as an unsigned char. - * @param Count A number of uint8_t elements to fill. + * @param Count A number of bytes to fill. * @return An opaque pointer to the ``sycl::event`` returned by the * ``sycl::queue::memset`` function. * @ingroup QueueInterface @@ -487,7 +487,7 @@ DPCTLQueue_Memset(__dpctl_keep const DPCTLSyclQueueRef QRef, * @param QRef An opaque pointer to the ``sycl::queue``. * @param USMRef An USM pointer to the memory to fill. * @param Value A value to fill, interpreted as an unsigned char. - * @param Count A number of uint8_t elements to fill. + * @param Count A number of bytes to fill. * @param DepEvents A pointer to array of DPCTLSyclEventRef opaque * pointers to dependent events. * @param DepEventsCount A number of dependent events. diff --git a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp index f9207b6cec..7486581356 100644 --- a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp @@ -903,7 +903,7 @@ DPCTLQueue_Memset(__dpctl_keep const DPCTLSyclQueueRef QRef, __func__, __LINE__); return nullptr; } -}; +} __dpctl_give DPCTLSyclEventRef DPCTLQueue_MemsetWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef, @@ -939,7 +939,7 @@ DPCTLQueue_MemsetWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef, } return wrap(new event(ev)); -}; +} __dpctl_give DPCTLSyclEventRef DPCTLQueue_Fill8(__dpctl_keep const DPCTLSyclQueueRef QRef, From 0689902eb45aff5ace63c69cc9677787499c49c3 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 05:38:07 -0700 Subject: [PATCH 15/20] Add a note about keeping mem alive in memset_async docs --- dpctl/_sycl_queue.pyx | 5 +++++ 1 file changed, 5 insertions(+) diff --git a/dpctl/_sycl_queue.pyx b/dpctl/_sycl_queue.pyx index 0c135a078d..4ebd48ffa8 100644 --- a/dpctl/_sycl_queue.pyx +++ b/dpctl/_sycl_queue.pyx @@ -1674,6 +1674,11 @@ cdef class SyclQueue(_SyclQueue): Internally, this dispatches ``sycl::queue::memset``. The operation is byte-wise: ``count`` bytes are set, each to the same value ``val``. + Note: + The returned event does not keep ``mem`` alive. Keep ``mem`` + alive until the event completes, otherwise its USM allocation + may be freed mid-operation, causing a use-after-free. + Args: mem: Destination USM allocation, an instance of From 7240bad7269c5ed0d1324fc6e469c1c7d8ecf78b Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 05:40:13 -0700 Subject: [PATCH 16/20] Cast self to SyclQueue in memset_async --- dpctl/_sycl_queue.pyx | 4 +++- 1 file changed, 3 insertions(+), 1 deletion(-) diff --git a/dpctl/_sycl_queue.pyx b/dpctl/_sycl_queue.pyx index 4ebd48ffa8..46d2cc8b27 100644 --- a/dpctl/_sycl_queue.pyx +++ b/dpctl/_sycl_queue.pyx @@ -1726,7 +1726,9 @@ cdef class SyclQueue(_SyclQueue): raise TypeError( "A sequence of dpctl.SyclEvent is expected" ) - ERef = _memset_impl(self, mem, byte_val, count, depEvents, nDE) + ERef = _memset_impl( + self, mem, byte_val, count, depEvents, nDE + ) finally: free(depEvents) From d584ca628cd9ba74b6f68e4f9ad9b41e51638baa Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 05:44:59 -0700 Subject: [PATCH 17/20] Add test_memset_fills_bytewise --- dpctl/tests/test_sycl_queue_memset.py | 16 ++++++++++++++++ 1 file changed, 16 insertions(+) diff --git a/dpctl/tests/test_sycl_queue_memset.py b/dpctl/tests/test_sycl_queue_memset.py index a956cf1853..d02a15829d 100644 --- a/dpctl/tests/test_sycl_queue_memset.py +++ b/dpctl/tests/test_sycl_queue_memset.py @@ -16,6 +16,7 @@ """Defines unit test cases for the SyclQueue.memset.""" +import numpy as np import pytest import dpctl @@ -205,3 +206,18 @@ def test_memset_value_truncated_to_byte(val, expected): q.memset(mobj, val) assert bytes(memoryview(mobj)) == bytes([expected]) * nbytes + + +def test_memset_fills_bytewise(): + # memset fills byte-by-byte, so 0xAB reads back as 0xABABABAB per uint32 + try: + q = dpctl.SyclQueue() + except dpctl.SyclQueueCreationError: + pytest.skip("Default constructor for SyclQueue failed") + nelems = 16 + mobj = _create_memory(q, nelems * 4) + + q.memset(mobj, 0xAB) + + view = np.frombuffer(memoryview(mobj), dtype=np.uint32) + assert np.all(view == 0xABABABAB) From 5e0630b05df8cb944df194d9396ccb25b885984c Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 05:54:53 -0700 Subject: [PATCH 18/20] Update an error message in DPCTLQueue_MemsetWithEvents --- libsyclinterface/source/dpctl_sycl_queue_interface.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp index 7486581356..2e27e25f87 100644 --- a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp @@ -933,8 +933,8 @@ DPCTLQueue_MemsetWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef, } } else { - error_handler("QRef or USMRef passed to memset were NULL.", __FILE__, - __func__, __LINE__); + error_handler("QRef or USMRef passed to memset_async were NULL.", + __FILE__, __func__, __LINE__); return nullptr; } From 7b4bebca8ef7a6094e6bf5e1b002b05b6077ec40 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 08:24:49 -0700 Subject: [PATCH 19/20] Apply remarks for _sycl_queue.pyx --- dpctl/_sycl_queue.pyx | 6 ++++++ libsyclinterface/source/dpctl_sycl_queue_interface.cpp | 2 +- 2 files changed, 7 insertions(+), 1 deletion(-) diff --git a/dpctl/_sycl_queue.pyx b/dpctl/_sycl_queue.pyx index 46d2cc8b27..b27f8601c1 100644 --- a/dpctl/_sycl_queue.pyx +++ b/dpctl/_sycl_queue.pyx @@ -1650,6 +1650,9 @@ cdef class SyclQueue(_SyclQueue): Raises: TypeError: If ``mem`` is not an instance of :class:`dpctl.memory._Memory`. + OverflowError: + If ``val`` does not fit in a C ``int`` or ``count`` is + negative. RuntimeError: If the memset operation encountered an error. """ @@ -1701,6 +1704,9 @@ cdef class SyclQueue(_SyclQueue): TypeError: If ``mem`` is not an instance of :class:`dpctl.memory._Memory`, or ``dEvents`` is not a sequence of :class:`dpctl.SyclEvent`. + OverflowError: + If ``val`` does not fit in a C ``int`` or ``count`` is + negative. RuntimeError: If the memset operation encountered an error. """ diff --git a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp index 2e27e25f87..cc9c6e5909 100644 --- a/libsyclinterface/source/dpctl_sycl_queue_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_queue_interface.cpp @@ -938,7 +938,7 @@ DPCTLQueue_MemsetWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef, return nullptr; } - return wrap(new event(ev)); + return wrap(new event(std::move(ev))); } __dpctl_give DPCTLSyclEventRef From 2359b0cbc71155b7e224ce8ab32a887b4bba66bd Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 19 Aug 2026 08:31:42 -0700 Subject: [PATCH 20/20] Update test_sycl_queue_memset.py --- dpctl/tests/test_sycl_queue_memset.py | 19 +++++++++++++++++-- 1 file changed, 17 insertions(+), 2 deletions(-) diff --git a/dpctl/tests/test_sycl_queue_memset.py b/dpctl/tests/test_sycl_queue_memset.py index d02a15829d..fefdbe2537 100644 --- a/dpctl/tests/test_sycl_queue_memset.py +++ b/dpctl/tests/test_sycl_queue_memset.py @@ -41,12 +41,24 @@ def _read_back(q, mobj, nbytes): return bytes(host) +def _skip_if_usm_unsupported(q, mem_cls): + dev = q.sycl_device + supported = { + dpctl.memory.MemoryUSMShared: dev.has_aspect_usm_shared_allocations, + dpctl.memory.MemoryUSMHost: dev.has_aspect_usm_host_allocations, + dpctl.memory.MemoryUSMDevice: dev.has_aspect_usm_device_allocations, + } + if not supported[mem_cls]: + pytest.skip(f"{mem_cls.__name__} is not supported on this device") + + @pytest.mark.parametrize("mem_cls", _MEMORY_CLASSES) def test_memset_fills_whole_allocation(mem_cls): try: q = dpctl.SyclQueue() except dpctl.SyclQueueCreationError: pytest.skip("Default constructor for SyclQueue failed") + _skip_if_usm_unsupported(q, mem_cls) nbytes = 256 mobj = mem_cls(nbytes, queue=q) @@ -128,6 +140,7 @@ def test_memset_async(mem_cls): q = dpctl.SyclQueue() except dpctl.SyclQueueCreationError: pytest.skip("Default constructor for SyclQueue failed") + _skip_if_usm_unsupported(q, mem_cls) nbytes = 64 mobj = mem_cls(nbytes, queue=q) @@ -144,13 +157,15 @@ def test_memset_async_with_dependent_events(): except dpctl.SyclQueueCreationError: pytest.skip("Default constructor for SyclQueue failed") nbytes = 64 + half = nbytes // 2 mobj = _create_memory(q, nbytes) e1 = q.memset_async(mobj, 0x01) - e2 = q.memset_async(mobj, 0x02, nbytes, [e1]) + e2 = q.memset_async(mobj, 0x02, half, [e1]) e2.wait() - assert bytes(memoryview(mobj)) == b"\x02" * nbytes + expected = b"\x02" * half + b"\x01" * (nbytes - half) + assert bytes(memoryview(mobj)) == expected def test_memset_async_partial_count():