Skip to content

Commit 885f91a

Browse files
dpnp_take_c uses SYCL kernel, no need to use no_sycl parameter in adapter
The reason this crashed with CPU device and gave incorrect results on Windows was deeper. 1. Adapter call allocated USM-shared buffer and copies data into it 2. Kernel is submitted to work on USM-shared pointer 3. dpnp_take_c returns kernel submission even 4. Adapter class goes out of scope and frees USM allocation without making sure that the kernel that works on it has completed its execution 5. If kernel execution was in progress we got a crash on CPU, or incorrect result on GPU If kernel execution was complete it worked as expected. This change fixes the problem because it removes creation of unprotected USM-shared temporary.
1 parent 51369c3 commit 885f91a

File tree

1 file changed

+1
-1
lines changed

1 file changed

+1
-1
lines changed

dpnp/backend/kernels/dpnp_krnl_indexing.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -901,7 +901,7 @@ DPCTLSyclEventRef dpnp_take_c(DPCTLSyclQueueRef q_ref,
901901
DPCTLSyclEventRef event_ref = nullptr;
902902
sycl::queue q = *(reinterpret_cast<sycl::queue*>(q_ref));
903903

904-
DPNPC_ptr_adapter<_DataType> input1_ptr(q_ref, array1_in, array1_size, true);
904+
DPNPC_ptr_adapter<_DataType> input1_ptr(q_ref, array1_in, array1_size);
905905
DPNPC_ptr_adapter<_IndecesType> input2_ptr(q_ref, indices1, size);
906906
_DataType* array_1 = input1_ptr.get_ptr();
907907
_IndecesType* indices = input2_ptr.get_ptr();

0 commit comments

Comments
 (0)