Skip to content

Commit 67a325d

Browse files
oleksandr-pavlykoleksandr-pavlyk
committed
Add DPCTLQueue_MemcpyWithEvents
This is the copy operation where one can specify list of events the copy operation requires before start of its execution. DPCTLQueue_MemcpyWithEvents( __dpctl_keep DPCTLSyclQueueRef QRef, void *dst, const void *src, size_t nbytes, const DPCTLSyclEventRef *depEvents, size_t nDE ) Uses this function in tests.
1 parent 306ff96 commit 67a325d

File tree

3 files changed

+146
-49
lines changed

3 files changed

+146
-49
lines changed

‎libsyclinterface/include/dpctl_sycl_queue_interface.h‎

Lines changed: 23 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -294,6 +294,29 @@ DPCTLQueue_Memcpy(__dpctl_keep const DPCTLSyclQueueRef QRef,
294294
const void *Src,
295295
size_t Count);
296296

297+
/*!
298+
* @brief C-API wrapper for ``sycl::queue::memcpy``.
299+
*
300+
* @param QRef An opaque pointer to the ``sycl::queue``.
301+
* @param Dest An USM pointer to the destination memory.
302+
* @param Src An USM pointer to the source memory.
303+
* @param Count A number of bytes to copy.
304+
* @param DepEvents A pointer to array of DPCTLSyclEventRef opaque
305+
* pointers to dependent events.
306+
* @param DepEventsCount A number of dependent events.
307+
* @return An opaque pointer to the ``sycl::event`` returned by the
308+
* ``sycl::queue::memcpy`` function.
309+
* @ingroup QueueInterface
310+
*/
311+
DPCTL_API
312+
__dpctl_give DPCTLSyclEventRef
313+
DPCTLQueue_MemcpyWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef,
314+
void *Dest,
315+
const void *Src,
316+
size_t Count,
317+
__dpctl_keep const DPCTLSyclEventRef *DepEvents,
318+
size_t DepEventsCount);
319+
297320
/*!
298321
* @brief C-API wrapper for ``sycl::queue::prefetch``.
299322
*

‎libsyclinterface/source/dpctl_sycl_queue_interface.cpp‎

Lines changed: 42 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -410,9 +410,12 @@ DPCTLQueue_SubmitNDRange(__dpctl_keep const DPCTLSyclKernelRef KRef,
410410
try {
411411
e = Queue->submit([&](handler &cgh) {
412412
// Depend on any event that was specified by the caller.
413-
if (NDepEvents)
414-
for (auto i = 0ul; i < NDepEvents; ++i)
415-
cgh.depends_on(*unwrap<event>(DepEvents[i]));
413+
if (DepEvents)
414+
for (auto i = 0ul; i < NDepEvents; ++i) {
415+
auto ei = unwrap<event>(DepEvents[i]);
416+
if (ei)
417+
cgh.depends_on(*ei);
418+
}
416419

417420
for (auto i = 0ul; i < NArgs; ++i) {
418421
// \todo add support for Sycl buffers
@@ -485,6 +488,42 @@ DPCTLQueue_Memcpy(__dpctl_keep const DPCTLSyclQueueRef QRef,
485488
}
486489
}
487490

491+
__dpctl_give DPCTLSyclEventRef
492+
DPCTLQueue_MemcpyWithEvents(__dpctl_keep const DPCTLSyclQueueRef QRef,
493+
void *Dest,
494+
const void *Src,
495+
size_t Count,
496+
const DPCTLSyclEventRef *DepEvents,
497+
size_t DepEventsCount)
498+
{
499+
event ev;
500+
auto Q = unwrap<queue>(QRef);
501+
if (Q) {
502+
try {
503+
ev = Q->submit([&](handler &cgh) {
504+
if (DepEvents)
505+
for (size_t i = 0; i < DepEventsCount; ++i) {
506+
event *ei = unwrap<event>(DepEvents[i]);
507+
if (ei)
508+
cgh.depends_on(*ei);
509+
}
510+
511+
cgh.memcpy(Dest, Src, Count);
512+
});
513+
} catch (const std::exception &ex) {
514+
error_handler(ex, __FILE__, __func__, __LINE__);
515+
return nullptr;
516+
}
517+
}
518+
else {
519+
error_handler("QRef passed to memcpy was NULL.", __FILE__, __func__,
520+
__LINE__);
521+
return nullptr;
522+
}
523+
524+
return wrap<event>(new event(ev));
525+
}
526+
488527
__dpctl_give DPCTLSyclEventRef
489528
DPCTLQueue_Prefetch(__dpctl_keep DPCTLSyclQueueRef QRef,
490529
const void *Ptr,

‎libsyclinterface/tests/test_sycl_queue_interface.cpp‎

Lines changed: 81 additions & 46 deletions
Original file line numberDiff line numberDiff line change
@@ -340,6 +340,10 @@ TEST(TestDPCTLSyclQueueInterface, CheckMemOpsZeroQRef)
340340
ASSERT_NO_FATAL_FAILURE(ERef = DPCTLQueue_Memcpy(QRef, p1, p2, n_bytes));
341341
ASSERT_FALSE(bool(ERef));
342342

343+
ASSERT_NO_FATAL_FAILURE(
344+
ERef = DPCTLQueue_MemcpyWithEvents(QRef, p1, p2, n_bytes, NULL, 0));
345+
ASSERT_FALSE(bool(ERef));
346+
343347
ASSERT_NO_FATAL_FAILURE(ERef = DPCTLQueue_Prefetch(QRef, p1, n_bytes));
344348
ASSERT_FALSE(bool(ERef));
345349

@@ -391,6 +395,10 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckMemOpsNullPtr)
391395
ASSERT_NO_FATAL_FAILURE(ERef = DPCTLQueue_Memcpy(QRef, p1, p2, n_bytes));
392396
ASSERT_FALSE(bool(ERef));
393397

398+
ASSERT_NO_FATAL_FAILURE(
399+
ERef = DPCTLQueue_MemcpyWithEvents(QRef, p1, p2, n_bytes, NULL, 0));
400+
ASSERT_FALSE(bool(ERef));
401+
394402
ASSERT_NO_FATAL_FAILURE(ERef = DPCTLQueue_Prefetch(QRef, p1, n_bytes));
395403
if (ERef) {
396404
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
@@ -450,6 +458,38 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckMemset)
450458
delete[] host_arr;
451459
}
452460

461+
TEST_P(TestDPCTLQueueMemberFunctions, CheckMemset2)
462+
{
463+
DPCTLSyclUSMRef p = nullptr;
464+
DPCTLSyclEventRef Memset_ERef = nullptr;
465+
DPCTLSyclEventRef Memcpy_ERef = nullptr;
466+
uint8_t val = 42;
467+
size_t nbytes = 256;
468+
uint8_t *host_arr = new uint8_t[nbytes];
469+
470+
ASSERT_FALSE(host_arr == nullptr);
471+
472+
ASSERT_NO_FATAL_FAILURE(p = DPCTLmalloc_device(nbytes, QRef));
473+
ASSERT_FALSE(p == nullptr);
474+
475+
ASSERT_NO_FATAL_FAILURE(
476+
Memset_ERef = DPCTLQueue_Memset(QRef, (void *)p, val, nbytes));
477+
478+
ASSERT_NO_FATAL_FAILURE(Memcpy_ERef = DPCTLQueue_MemcpyWithEvents(
479+
QRef, host_arr, p, nbytes, &Memset_ERef, 1));
480+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(Memcpy_ERef));
481+
482+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memset_ERef));
483+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memcpy_ERef));
484+
485+
ASSERT_NO_FATAL_FAILURE(DPCTLfree_with_queue(p, QRef));
486+
487+
for (size_t i = 0; i < nbytes; ++i) {
488+
ASSERT_TRUE(host_arr[i] == val);
489+
}
490+
delete[] host_arr;
491+
}
492+
453493
TEST(TestDPCTLSyclQueueInterface, CheckFillNullQRef)
454494
{
455495
DPCTLSyclQueueRef QRef = nullptr;
@@ -481,7 +521,8 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill8)
481521
{
482522
using T = uint8_t;
483523
DPCTLSyclUSMRef p = nullptr;
484-
DPCTLSyclEventRef ERef = nullptr;
524+
DPCTLSyclEventRef Fill8_ERef = nullptr;
525+
DPCTLSyclEventRef Memcpy_ERef = nullptr;
485526
T val = static_cast<T>(0xB);
486527
size_t nelems = 256;
487528
T *host_arr = new T[nelems];
@@ -492,17 +533,15 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill8)
492533
ASSERT_NO_FATAL_FAILURE(p = DPCTLmalloc_device(nbytes, QRef));
493534
ASSERT_FALSE(p == nullptr);
494535

495-
ASSERT_NO_FATAL_FAILURE(ERef =
536+
ASSERT_NO_FATAL_FAILURE(Fill8_ERef =
496537
DPCTLQueue_Fill8(QRef, (void *)p, val, nelems));
497-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
498-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
499538

500-
ERef = nullptr;
539+
ASSERT_NO_FATAL_FAILURE(Memcpy_ERef = DPCTLQueue_MemcpyWithEvents(
540+
QRef, host_arr, p, nbytes, &Fill8_ERef, 1));
541+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(Memcpy_ERef));
501542

502-
ASSERT_NO_FATAL_FAILURE(ERef =
503-
DPCTLQueue_Memcpy(QRef, host_arr, p, nbytes));
504-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
505-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
543+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Fill8_ERef));
544+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memcpy_ERef));
506545

507546
ASSERT_NO_FATAL_FAILURE(DPCTLfree_with_queue(p, QRef));
508547

@@ -517,7 +556,8 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill16)
517556
using T = uint16_t;
518557

519558
DPCTLSyclUSMRef p = nullptr;
520-
DPCTLSyclEventRef ERef = nullptr;
559+
DPCTLSyclEventRef Fill16_ERef = nullptr;
560+
DPCTLSyclEventRef Memcpy_ERef = nullptr;
521561
T val = static_cast<T>(0xAB);
522562
size_t nelems = 256;
523563
T *host_arr = new T[nelems];
@@ -529,16 +569,14 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill16)
529569
ASSERT_FALSE(p == nullptr);
530570

531571
ASSERT_NO_FATAL_FAILURE(
532-
ERef = DPCTLQueue_Fill16(QRef, (void *)p, val, nelems));
533-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
534-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
572+
Fill16_ERef = DPCTLQueue_Fill16(QRef, (void *)p, val, nelems));
535573

536-
ERef = nullptr;
574+
ASSERT_NO_FATAL_FAILURE(Memcpy_ERef = DPCTLQueue_MemcpyWithEvents(
575+
QRef, host_arr, p, nbytes, &Fill16_ERef, 1));
576+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(Memcpy_ERef));
537577

538-
ASSERT_NO_FATAL_FAILURE(ERef =
539-
DPCTLQueue_Memcpy(QRef, host_arr, p, nbytes));
540-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
541-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
578+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Fill16_ERef));
579+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memcpy_ERef));
542580

543581
ASSERT_NO_FATAL_FAILURE(DPCTLfree_with_queue(p, QRef));
544582

@@ -553,7 +591,8 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill32)
553591
using T = uint32_t;
554592

555593
DPCTLSyclUSMRef p = nullptr;
556-
DPCTLSyclEventRef ERef = nullptr;
594+
DPCTLSyclEventRef Fill32_ERef = nullptr;
595+
DPCTLSyclEventRef Memcpy_ERef = nullptr;
557596
T val = static_cast<T>(0xABCD);
558597
size_t nelems = 256;
559598
T *host_arr = new T[nelems];
@@ -565,16 +604,14 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill32)
565604
ASSERT_FALSE(p == nullptr);
566605

567606
ASSERT_NO_FATAL_FAILURE(
568-
ERef = DPCTLQueue_Fill32(QRef, (void *)p, val, nelems));
569-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
570-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
607+
Fill32_ERef = DPCTLQueue_Fill32(QRef, (void *)p, val, nelems));
571608

572-
ERef = nullptr;
609+
ASSERT_NO_FATAL_FAILURE(Memcpy_ERef = DPCTLQueue_MemcpyWithEvents(
610+
QRef, host_arr, p, nbytes, &Fill32_ERef, 1));
611+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(Memcpy_ERef));
573612

574-
ASSERT_NO_FATAL_FAILURE(ERef =
575-
DPCTLQueue_Memcpy(QRef, host_arr, p, nbytes));
576-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
577-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
613+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Fill32_ERef));
614+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memcpy_ERef));
578615

579616
ASSERT_NO_FATAL_FAILURE(DPCTLfree_with_queue(p, QRef));
580617

@@ -589,7 +626,8 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill64)
589626
using T = uint64_t;
590627

591628
DPCTLSyclUSMRef p = nullptr;
592-
DPCTLSyclEventRef ERef = nullptr;
629+
DPCTLSyclEventRef Fill64_ERef = nullptr;
630+
DPCTLSyclEventRef Memcpy_ERef = nullptr;
593631
T val = static_cast<T>(0xABCDEF73);
594632
size_t nelems = 256;
595633
T *host_arr = new T[nelems];
@@ -601,16 +639,14 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill64)
601639
ASSERT_FALSE(p == nullptr);
602640

603641
ASSERT_NO_FATAL_FAILURE(
604-
ERef = DPCTLQueue_Fill64(QRef, (void *)p, val, nelems));
605-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
606-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
642+
Fill64_ERef = DPCTLQueue_Fill64(QRef, (void *)p, val, nelems));
607643

608-
ERef = nullptr;
644+
ASSERT_NO_FATAL_FAILURE(Memcpy_ERef = DPCTLQueue_MemcpyWithEvents(
645+
QRef, host_arr, p, nbytes, &Fill64_ERef, 1));
646+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(Memcpy_ERef));
609647

610-
ASSERT_NO_FATAL_FAILURE(ERef =
611-
DPCTLQueue_Memcpy(QRef, host_arr, p, nbytes));
612-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
613-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
648+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Fill64_ERef));
649+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memcpy_ERef));
614650

615651
ASSERT_NO_FATAL_FAILURE(DPCTLfree_with_queue(p, QRef));
616652

@@ -639,7 +675,8 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill128)
639675
using T = value128_t;
640676

641677
DPCTLSyclUSMRef p = nullptr;
642-
DPCTLSyclEventRef ERef = nullptr;
678+
DPCTLSyclEventRef Fill128_ERef = nullptr;
679+
DPCTLSyclEventRef Memcpy_ERef = nullptr;
643680
T val{static_cast<uint64_t>(0xABCDEF73), static_cast<uint64_t>(0x3746AF05)};
644681
size_t nelems = 256;
645682
T *host_arr = new T[nelems];
@@ -651,17 +688,15 @@ TEST_P(TestDPCTLQueueMemberFunctions, CheckFill128)
651688
ASSERT_FALSE(p == nullptr);
652689

653690
ASSERT_NO_FATAL_FAILURE(
654-
ERef = DPCTLQueue_Fill128(QRef, (void *)p,
655-
reinterpret_cast<uint64_t *>(&val), nelems));
656-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
657-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
691+
Fill128_ERef = DPCTLQueue_Fill128(
692+
QRef, (void *)p, reinterpret_cast<uint64_t *>(&val), nelems));
658693

659-
ERef = nullptr;
694+
ASSERT_NO_FATAL_FAILURE(Memcpy_ERef = DPCTLQueue_MemcpyWithEvents(
695+
QRef, host_arr, p, nbytes, &Fill128_ERef, 1));
696+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(Memcpy_ERef));
660697

661-
ASSERT_NO_FATAL_FAILURE(ERef =
662-
DPCTLQueue_Memcpy(QRef, host_arr, p, nbytes));
663-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Wait(ERef));
664-
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(ERef));
698+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Fill128_ERef));
699+
ASSERT_NO_FATAL_FAILURE(DPCTLEvent_Delete(Memcpy_ERef));
665700

666701
ASSERT_NO_FATAL_FAILURE(DPCTLfree_with_queue(p, QRef));
667702

0 commit comments

Comments
 (0)