Skip to content

Commit 2b12021

Browse files
densmirnoleksandr-pavlyk
densmirn
authored and
oleksandr-pavlyk
committed
Add tests for compute follows data for math funcs
1 parent ca114f2 commit 2b12021

9 files changed

+369
-113
lines changed

‎dpnp/backend/include/dpnp_gen_1arg_1type_tbl.hpp

Lines changed: 7 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -90,20 +90,20 @@
9090

9191
#endif
9292

93-
MACRO_1ARG_1TYPE_OP(dpnp_conjugate_c, std::conj(input_elem), DPNP_QUEUE.submit(kernel_func))
94-
MACRO_1ARG_1TYPE_OP(dpnp_copy_c, input_elem, DPNP_QUEUE.submit(kernel_func))
93+
MACRO_1ARG_1TYPE_OP(dpnp_conjugate_c, std::conj(input_elem), q.submit(kernel_func))
94+
MACRO_1ARG_1TYPE_OP(dpnp_copy_c, input_elem, q.submit(kernel_func))
9595
MACRO_1ARG_1TYPE_OP(dpnp_erf_c,
9696
sycl::erf((double)input_elem),
97-
oneapi::mkl::vm::erf(DPNP_QUEUE, input1_size, input1_data, result)) // no sycl::erf for int and long
98-
MACRO_1ARG_1TYPE_OP(dpnp_negative_c, -input_elem, DPNP_QUEUE.submit(kernel_func))
97+
oneapi::mkl::vm::erf(q, input1_size, input1_data, result)) // no sycl::erf for int and long
98+
MACRO_1ARG_1TYPE_OP(dpnp_negative_c, -input_elem, q.submit(kernel_func))
9999
MACRO_1ARG_1TYPE_OP(dpnp_recip_c,
100100
_DataType(1) / input_elem,
101-
DPNP_QUEUE.submit(kernel_func)) // error: no member named 'recip' in namespace 'sycl'
101+
q.submit(kernel_func)) // error: no member named 'recip' in namespace 'sycl'
102102
MACRO_1ARG_1TYPE_OP(dpnp_sign_c,
103103
sycl::sign((double)input_elem),
104-
DPNP_QUEUE.submit(kernel_func)) // no sycl::sign for int and long
104+
q.submit(kernel_func)) // no sycl::sign for int and long
105105
MACRO_1ARG_1TYPE_OP(dpnp_square_c,
106106
input_elem* input_elem,
107-
oneapi::mkl::vm::sqr(DPNP_QUEUE, input1_size, input1_data, result))
107+
oneapi::mkl::vm::sqr(q, input1_size, input1_data, result))
108108

109109
#undef MACRO_1ARG_1TYPE_OP

‎dpnp/backend/include/dpnp_gen_1arg_2type_tbl.hpp

Lines changed: 28 additions & 28 deletions
Original file line numberDiff line numberDiff line change
@@ -92,81 +92,81 @@
9292

9393
MACRO_1ARG_2TYPES_OP(dpnp_acos_c,
9494
sycl::acos(input_elem),
95-
oneapi::mkl::vm::acos(DPNP_QUEUE, input1_size, input1_data, result))
95+
oneapi::mkl::vm::acos(q, input1_size, input1_data, result))
9696
MACRO_1ARG_2TYPES_OP(dpnp_acosh_c,
9797
sycl::acosh(input_elem),
98-
oneapi::mkl::vm::acosh(DPNP_QUEUE, input1_size, input1_data, result))
98+
oneapi::mkl::vm::acosh(q, input1_size, input1_data, result))
9999
MACRO_1ARG_2TYPES_OP(dpnp_asin_c,
100100
sycl::asin(input_elem),
101-
oneapi::mkl::vm::asin(DPNP_QUEUE, input1_size, input1_data, result))
101+
oneapi::mkl::vm::asin(q, input1_size, input1_data, result))
102102
MACRO_1ARG_2TYPES_OP(dpnp_asinh_c,
103103
sycl::asinh(input_elem),
104-
oneapi::mkl::vm::asinh(DPNP_QUEUE, input1_size, input1_data, result))
104+
oneapi::mkl::vm::asinh(q, input1_size, input1_data, result))
105105
MACRO_1ARG_2TYPES_OP(dpnp_atan_c,
106106
sycl::atan(input_elem),
107-
oneapi::mkl::vm::atan(DPNP_QUEUE, input1_size, input1_data, result))
107+
oneapi::mkl::vm::atan(q, input1_size, input1_data, result))
108108
MACRO_1ARG_2TYPES_OP(dpnp_atanh_c,
109109
sycl::atanh(input_elem),
110-
oneapi::mkl::vm::atanh(DPNP_QUEUE, input1_size, input1_data, result))
110+
oneapi::mkl::vm::atanh(q, input1_size, input1_data, result))
111111
MACRO_1ARG_2TYPES_OP(dpnp_cbrt_c,
112112
sycl::cbrt(input_elem),
113-
oneapi::mkl::vm::cbrt(DPNP_QUEUE, input1_size, input1_data, result))
113+
oneapi::mkl::vm::cbrt(q, input1_size, input1_data, result))
114114
MACRO_1ARG_2TYPES_OP(dpnp_ceil_c,
115115
sycl::ceil(input_elem),
116-
oneapi::mkl::vm::ceil(DPNP_QUEUE, input1_size, input1_data, result))
117-
MACRO_1ARG_2TYPES_OP(dpnp_copyto_c, input_elem, DPNP_QUEUE.submit(kernel_func))
116+
oneapi::mkl::vm::ceil(q, input1_size, input1_data, result))
117+
MACRO_1ARG_2TYPES_OP(dpnp_copyto_c, input_elem, q.submit(kernel_func))
118118
MACRO_1ARG_2TYPES_OP(dpnp_cos_c,
119119
sycl::cos(input_elem),
120-
oneapi::mkl::vm::cos(DPNP_QUEUE, input1_size, input1_data, result))
120+
oneapi::mkl::vm::cos(q, input1_size, input1_data, result))
121121
MACRO_1ARG_2TYPES_OP(dpnp_cosh_c,
122122
sycl::cosh(input_elem),
123-
oneapi::mkl::vm::cosh(DPNP_QUEUE, input1_size, input1_data, result))
124-
MACRO_1ARG_2TYPES_OP(dpnp_degrees_c, sycl::degrees(input_elem), DPNP_QUEUE.submit(kernel_func))
123+
oneapi::mkl::vm::cosh(q, input1_size, input1_data, result))
124+
MACRO_1ARG_2TYPES_OP(dpnp_degrees_c, sycl::degrees(input_elem), q.submit(kernel_func))
125125
MACRO_1ARG_2TYPES_OP(dpnp_exp2_c,
126126
sycl::exp2(input_elem),
127-
oneapi::mkl::vm::exp2(DPNP_QUEUE, input1_size, input1_data, result))
127+
oneapi::mkl::vm::exp2(q, input1_size, input1_data, result))
128128
MACRO_1ARG_2TYPES_OP(dpnp_exp_c,
129129
sycl::exp(input_elem),
130-
oneapi::mkl::vm::exp(DPNP_QUEUE, input1_size, input1_data, result))
130+
oneapi::mkl::vm::exp(q, input1_size, input1_data, result))
131131
MACRO_1ARG_2TYPES_OP(dpnp_expm1_c,
132132
sycl::expm1(input_elem),
133-
oneapi::mkl::vm::expm1(DPNP_QUEUE, input1_size, input1_data, result))
133+
oneapi::mkl::vm::expm1(q, input1_size, input1_data, result))
134134
MACRO_1ARG_2TYPES_OP(dpnp_fabs_c,
135135
sycl::fabs(input_elem),
136-
oneapi::mkl::vm::abs(DPNP_QUEUE, input1_size, input1_data, result))
136+
oneapi::mkl::vm::abs(q, input1_size, input1_data, result))
137137
MACRO_1ARG_2TYPES_OP(dpnp_floor_c,
138138
sycl::floor(input_elem),
139-
oneapi::mkl::vm::floor(DPNP_QUEUE, input1_size, input1_data, result))
139+
oneapi::mkl::vm::floor(q, input1_size, input1_data, result))
140140
MACRO_1ARG_2TYPES_OP(dpnp_log10_c,
141141
sycl::log10(input_elem),
142-
oneapi::mkl::vm::log10(DPNP_QUEUE, input1_size, input1_data, result))
142+
oneapi::mkl::vm::log10(q, input1_size, input1_data, result))
143143
MACRO_1ARG_2TYPES_OP(dpnp_log1p_c,
144144
sycl::log1p(input_elem),
145-
oneapi::mkl::vm::log1p(DPNP_QUEUE, input1_size, input1_data, result))
145+
oneapi::mkl::vm::log1p(q, input1_size, input1_data, result))
146146
MACRO_1ARG_2TYPES_OP(dpnp_log2_c,
147147
sycl::log2(input_elem),
148-
oneapi::mkl::vm::log2(DPNP_QUEUE, input1_size, input1_data, result))
148+
oneapi::mkl::vm::log2(q, input1_size, input1_data, result))
149149
MACRO_1ARG_2TYPES_OP(dpnp_log_c,
150150
sycl::log(input_elem),
151-
oneapi::mkl::vm::ln(DPNP_QUEUE, input1_size, input1_data, result))
152-
MACRO_1ARG_2TYPES_OP(dpnp_radians_c, sycl::radians(input_elem), DPNP_QUEUE.submit(kernel_func))
151+
oneapi::mkl::vm::ln(q, input1_size, input1_data, result))
152+
MACRO_1ARG_2TYPES_OP(dpnp_radians_c, sycl::radians(input_elem), q.submit(kernel_func))
153153
MACRO_1ARG_2TYPES_OP(dpnp_sin_c,
154154
sycl::sin(input_elem),
155-
oneapi::mkl::vm::sin(DPNP_QUEUE, input1_size, input1_data, result))
155+
oneapi::mkl::vm::sin(q, input1_size, input1_data, result))
156156
MACRO_1ARG_2TYPES_OP(dpnp_sinh_c,
157157
sycl::sinh(input_elem),
158-
oneapi::mkl::vm::sinh(DPNP_QUEUE, input1_size, input1_data, result))
158+
oneapi::mkl::vm::sinh(q, input1_size, input1_data, result))
159159
MACRO_1ARG_2TYPES_OP(dpnp_sqrt_c,
160160
sycl::sqrt(input_elem),
161-
oneapi::mkl::vm::sqrt(DPNP_QUEUE, input1_size, input1_data, result))
161+
oneapi::mkl::vm::sqrt(q, input1_size, input1_data, result))
162162
MACRO_1ARG_2TYPES_OP(dpnp_tan_c,
163163
sycl::tan(input_elem),
164-
oneapi::mkl::vm::tan(DPNP_QUEUE, input1_size, input1_data, result))
164+
oneapi::mkl::vm::tan(q, input1_size, input1_data, result))
165165
MACRO_1ARG_2TYPES_OP(dpnp_tanh_c,
166166
sycl::tanh(input_elem),
167-
oneapi::mkl::vm::tanh(DPNP_QUEUE, input1_size, input1_data, result))
167+
oneapi::mkl::vm::tanh(q, input1_size, input1_data, result))
168168
MACRO_1ARG_2TYPES_OP(dpnp_trunc_c,
169169
sycl::trunc(input_elem),
170-
oneapi::mkl::vm::trunc(DPNP_QUEUE, input1_size, input1_data, result))
170+
oneapi::mkl::vm::trunc(q, input1_size, input1_data, result))
171171

172172
#undef MACRO_1ARG_2TYPES_OP

‎dpnp/dpnp_algo/dpnp_algo.pyx

Lines changed: 12 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -379,6 +379,8 @@ cdef utils.dpnp_descriptor call_fptr_1in_1out(DPNPFuncName fptr_name,
379379

380380
result = out
381381

382+
utils.get_common_usm_allocation(x1, result) # check USM allocation is common
383+
382384
result_sycl_queue = result.get_array().sycl_queue
383385

384386
cdef c_dpctl.SyclQueue q = <c_dpctl.SyclQueue> result_sycl_queue
@@ -432,6 +434,8 @@ cdef utils.dpnp_descriptor call_fptr_1in_1out_strides(DPNPFuncName fptr_name,
432434

433435
result = out
434436

437+
utils.get_common_usm_allocation(x1, result) # check USM allocation is common
438+
435439
result_sycl_queue = result.get_array().sycl_queue
436440

437441
cdef c_dpctl.SyclQueue q = <c_dpctl.SyclQueue> result_sycl_queue
@@ -484,9 +488,10 @@ cdef utils.dpnp_descriptor call_fptr_2in_1out(DPNPFuncName fptr_name,
484488
cdef shape_type_c result_shape = utils.get_common_shape(x1_shape, x2_shape)
485489
cdef utils.dpnp_descriptor result
486490

491+
result_sycl_device, result_usm_type, result_sycl_queue = utils.get_common_usm_allocation(x1_obj, x2_obj)
492+
487493
if out is None:
488494
""" Create result array with type given by FPTR data """
489-
result_sycl_device, result_usm_type, result_sycl_queue = utils.get_common_usm_allocation(x1_obj, x2_obj)
490495
result = utils.create_output_descriptor(result_shape,
491496
kernel_data.return_type,
492497
None,
@@ -501,6 +506,8 @@ cdef utils.dpnp_descriptor call_fptr_2in_1out(DPNPFuncName fptr_name,
501506

502507
result = out
503508

509+
utils.get_common_usm_allocation(x1_obj, result) # check USM allocation is common
510+
504511
cdef c_dpctl.SyclQueue q = <c_dpctl.SyclQueue> result_sycl_queue
505512
cdef c_dpctl.DPCTLSyclQueueRef q_ref = q.get_queue_ref()
506513

@@ -551,9 +558,10 @@ cdef utils.dpnp_descriptor call_fptr_2in_1out_strides(DPNPFuncName fptr_name,
551558
cdef shape_type_c result_shape = utils.get_common_shape(x1_shape, x2_shape)
552559
cdef utils.dpnp_descriptor result
553560

561+
result_sycl_device, result_usm_type, result_sycl_queue = utils.get_common_usm_allocation(x1_obj, x2_obj)
562+
554563
if out is None:
555564
""" Create result array with type given by FPTR data """
556-
result_sycl_device, result_usm_type, result_sycl_queue = utils.get_common_usm_allocation(x1_obj, x2_obj)
557565
result = utils.create_output_descriptor(result_shape,
558566
kernel_data.return_type,
559567
None,
@@ -568,6 +576,8 @@ cdef utils.dpnp_descriptor call_fptr_2in_1out_strides(DPNPFuncName fptr_name,
568576

569577
result = out
570578

579+
utils.get_common_usm_allocation(x1_obj, result) # check USM allocation is common
580+
571581
cdef shape_type_c result_strides = utils.strides_to_vector(result.strides, result_shape)
572582

573583
result_obj = result.get_array()

‎dpnp/dpnp_algo/dpnp_algo_mathematical.pyx

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -451,7 +451,7 @@ cpdef utils.dpnp_descriptor dpnp_nancumprod(utils.dpnp_descriptor x1):
451451
if dpnp.isnan(cur_x1_flatiter[i]):
452452
cur_x1_flatiter[i] = 1
453453

454-
x1_desc = dpnp.get_dpnp_descriptor(cur_x1)
454+
x1_desc = dpnp.get_dpnp_descriptor(cur_x1, copy_when_nondefault_queue=False)
455455
return dpnp_cumprod(x1_desc)
456456

457457

@@ -464,7 +464,7 @@ cpdef utils.dpnp_descriptor dpnp_nancumsum(utils.dpnp_descriptor x1):
464464
if dpnp.isnan(cur_x1_flatiter[i]):
465465
cur_x1_flatiter[i] = 0
466466

467-
x1_desc = dpnp.get_dpnp_descriptor(cur_x1)
467+
x1_desc = dpnp.get_dpnp_descriptor(cur_x1, copy_when_nondefault_queue=False)
468468
return dpnp_cumsum(x1_desc)
469469

470470

‎dpnp/dpnp_iface_arraycreation.py

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -390,7 +390,7 @@ def copy(x1, order='K', subok=False):
390390
391391
"""
392392

393-
x1_desc = dpnp.get_dpnp_descriptor(x1, copy_when_strides=False)
393+
x1_desc = dpnp.get_dpnp_descriptor(x1, copy_when_strides=False, copy_when_nondefault_queue=False)
394394
if x1_desc:
395395
if order != 'K':
396396
pass

0 commit comments

Comments
 (0)