Sycl cython2 check sasha

.github/pull_request_template.md

@@ -0,0 +1,5 @@
+- [ ] Have you provided a meaningful PR description?
+- [ ] Have you added a test, reproducer or referred to issue with a reproducer?
+- [ ] Have you tested your changes locally for CPU and GPU devices?
+- [ ] Have you made sure that new changes do not introduce compiler warnings?
+- [ ] If this PR is a work in progress, are you filing the PR as a draft?

.gitignore

@@ -0,0 +1,7 @@
+build
+build_cython
+__pycache__/
+*dpnp_backend*
+dpnp/**/*.cpython*.so
+dpnp/**/*.pyd
+*~

dpnp/backend/include/dpnp_gen_1arg_1type_tbl.hpp

@@ -90,20 +90,20 @@
 
 #endif
 
-MACRO_1ARG_1TYPE_OP(dpnp_conjugate_c, std::conj(input_elem), DPNP_QUEUE.submit(kernel_func))
-MACRO_1ARG_1TYPE_OP(dpnp_copy_c, input_elem, DPNP_QUEUE.submit(kernel_func))
+MACRO_1ARG_1TYPE_OP(dpnp_conjugate_c, std::conj(input_elem), q.submit(kernel_func))
+MACRO_1ARG_1TYPE_OP(dpnp_copy_c, input_elem, q.submit(kernel_func))
 MACRO_1ARG_1TYPE_OP(dpnp_erf_c,
                     sycl::erf((double)input_elem),
-                    oneapi::mkl::vm::erf(DPNP_QUEUE, input1_size, input1_data, result)) // no sycl::erf for int and long
-MACRO_1ARG_1TYPE_OP(dpnp_negative_c, -input_elem, DPNP_QUEUE.submit(kernel_func))
+                    oneapi::mkl::vm::erf(q, input1_size, input1_data, result)) // no sycl::erf for int and long
+MACRO_1ARG_1TYPE_OP(dpnp_negative_c, -input_elem, q.submit(kernel_func))
 MACRO_1ARG_1TYPE_OP(dpnp_recip_c,
                     _DataType(1) / input_elem,
-                    DPNP_QUEUE.submit(kernel_func)) // error: no member named 'recip' in namespace 'sycl'
+                    q.submit(kernel_func)) // error: no member named 'recip' in namespace 'sycl'
 MACRO_1ARG_1TYPE_OP(dpnp_sign_c,
                     sycl::sign((double)input_elem),
-                    DPNP_QUEUE.submit(kernel_func)) // no sycl::sign for int and long
+                    q.submit(kernel_func)) // no sycl::sign for int and long
 MACRO_1ARG_1TYPE_OP(dpnp_square_c,
                     input_elem* input_elem,
-                    oneapi::mkl::vm::sqr(DPNP_QUEUE, input1_size, input1_data, result))
+                    oneapi::mkl::vm::sqr(q, input1_size, input1_data, result))
 
 #undef MACRO_1ARG_1TYPE_OP

dpnp/backend/include/dpnp_gen_1arg_2type_tbl.hpp

@@ -92,81 +92,81 @@
 
 MACRO_1ARG_2TYPES_OP(dpnp_acos_c,
                      sycl::acos(input_elem),
-                     oneapi::mkl::vm::acos(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::acos(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_acosh_c,
                      sycl::acosh(input_elem),
-                     oneapi::mkl::vm::acosh(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::acosh(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_asin_c,
                      sycl::asin(input_elem),
-                     oneapi::mkl::vm::asin(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::asin(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_asinh_c,
                      sycl::asinh(input_elem),
-                     oneapi::mkl::vm::asinh(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::asinh(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_atan_c,
                      sycl::atan(input_elem),
-                     oneapi::mkl::vm::atan(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::atan(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_atanh_c,
                      sycl::atanh(input_elem),
-                     oneapi::mkl::vm::atanh(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::atanh(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_cbrt_c,
                      sycl::cbrt(input_elem),
-                     oneapi::mkl::vm::cbrt(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::cbrt(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_ceil_c,
                      sycl::ceil(input_elem),
-                     oneapi::mkl::vm::ceil(DPNP_QUEUE, input1_size, input1_data, result))
-MACRO_1ARG_2TYPES_OP(dpnp_copyto_c, input_elem, DPNP_QUEUE.submit(kernel_func))
+                     oneapi::mkl::vm::ceil(q, input1_size, input1_data, result))
+MACRO_1ARG_2TYPES_OP(dpnp_copyto_c, input_elem, q.submit(kernel_func))
 MACRO_1ARG_2TYPES_OP(dpnp_cos_c,
                      sycl::cos(input_elem),
-                     oneapi::mkl::vm::cos(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::cos(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_cosh_c,
                      sycl::cosh(input_elem),
-                     oneapi::mkl::vm::cosh(DPNP_QUEUE, input1_size, input1_data, result))
-MACRO_1ARG_2TYPES_OP(dpnp_degrees_c, sycl::degrees(input_elem), DPNP_QUEUE.submit(kernel_func))
+                     oneapi::mkl::vm::cosh(q, input1_size, input1_data, result))
+MACRO_1ARG_2TYPES_OP(dpnp_degrees_c, sycl::degrees(input_elem), q.submit(kernel_func))
 MACRO_1ARG_2TYPES_OP(dpnp_exp2_c,
                      sycl::exp2(input_elem),
-                     oneapi::mkl::vm::exp2(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::exp2(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_exp_c,
                      sycl::exp(input_elem),
-                     oneapi::mkl::vm::exp(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::exp(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_expm1_c,
                      sycl::expm1(input_elem),
-                     oneapi::mkl::vm::expm1(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::expm1(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_fabs_c,
                      sycl::fabs(input_elem),
-                     oneapi::mkl::vm::abs(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::abs(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_floor_c,
                      sycl::floor(input_elem),
-                     oneapi::mkl::vm::floor(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::floor(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_log10_c,
                      sycl::log10(input_elem),
-                     oneapi::mkl::vm::log10(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::log10(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_log1p_c,
                      sycl::log1p(input_elem),
-                     oneapi::mkl::vm::log1p(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::log1p(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_log2_c,
                      sycl::log2(input_elem),
-                     oneapi::mkl::vm::log2(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::log2(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_log_c,
                      sycl::log(input_elem),
-                     oneapi::mkl::vm::ln(DPNP_QUEUE, input1_size, input1_data, result))
-MACRO_1ARG_2TYPES_OP(dpnp_radians_c, sycl::radians(input_elem), DPNP_QUEUE.submit(kernel_func))
+                     oneapi::mkl::vm::ln(q, input1_size, input1_data, result))
+MACRO_1ARG_2TYPES_OP(dpnp_radians_c, sycl::radians(input_elem), q.submit(kernel_func))
 MACRO_1ARG_2TYPES_OP(dpnp_sin_c,
                      sycl::sin(input_elem),
-                     oneapi::mkl::vm::sin(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::sin(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_sinh_c,
                      sycl::sinh(input_elem),
-                     oneapi::mkl::vm::sinh(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::sinh(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_sqrt_c,
                      sycl::sqrt(input_elem),
-                     oneapi::mkl::vm::sqrt(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::sqrt(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_tan_c,
                      sycl::tan(input_elem),
-                     oneapi::mkl::vm::tan(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::tan(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_tanh_c,
                      sycl::tanh(input_elem),
-                     oneapi::mkl::vm::tanh(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::tanh(q, input1_size, input1_data, result))
 MACRO_1ARG_2TYPES_OP(dpnp_trunc_c,
                      sycl::trunc(input_elem),
-                     oneapi::mkl::vm::trunc(DPNP_QUEUE, input1_size, input1_data, result))
+                     oneapi::mkl::vm::trunc(q, input1_size, input1_data, result))
 
 #undef MACRO_1ARG_2TYPES_OP

dpnp/backend/include/dpnp_iface.hpp

@@ -533,7 +533,7 @@ INP_DLLEXPORT void dpnp_cumprod_c(void* array1_in, void* result1, size_t size);
  * @param [in]  q_ref               Reference to SYCL queue.
  * @param [in]  array1_in           Input array.
  * @param [out] result1             Output array.
- * @param [in]  size                 Number of elements in input arrays.
+ * @param [in]  size                Number of elements in input arrays.
  * @param [in]  dep_event_vec_ref   Reference to vector of SYCL events.
  *
  */
@@ -551,7 +551,7 @@ INP_DLLEXPORT void dpnp_cumsum_c(void* array1_in, void* result1, size_t size);
  * @ingroup BACKEND_API
  * @brief The differences between consecutive elements of an array.
  *
- * @param [in]  q_ref             Reference to SYCL queue.
+ * @param [in]  q_ref               Reference to SYCL queue.
  * @param [out] result_out          Output array.
  * @param [in]  result_size         Size of output array.
  * @param [in]  result_ndim         Number of output array dimensions.

dpnp/backend/include/dpnp_iface_fptr.hpp

@@ -381,7 +381,6 @@ enum class DPNPFuncName : size_t
     DPNP_FN_ZEROS_LIKE,                   /**< Used in numpy.zeros_like() impl */
     DPNP_FN_ZEROS_LIKE_EXT,               /**< Used in numpy.zeros_like() impl, requires extra parameters */
     DPNP_FN_LAST,                         /**< The latest element of the enumeration */
-    DPNP_FN_LAST_EXT                      /**< The latest element of the enumeration, requires extra parameters */
 };
 
 /**

dpnp/backend/include/dpnp_iface_random.hpp

@@ -335,23 +335,23 @@ INP_DLLEXPORT void dpnp_rng_lognormal_c(void* result, const _DataType mean, cons
  * @param [in]  q_ref               Reference to SYCL queue.
  * @param [out] result              Output array.
  * @param [in]  ntrial              Number of independent trials.
- * @param [in]  p_vector            Probability vector of possible outcomes (k length).
- * @param [in]  p_vector_size       Length of `p_vector`.
+ * @param [in]  p_in                Probability of possible outcomes (k length).
+ * @param [in]  p_size              Length of `p_in`.
  * @param [in]  size                Number of elements in `result` arrays.
  * @param [in]  dep_event_vec_ref   Reference to vector of SYCL events.
  */
 template <typename _DataType>
 INP_DLLEXPORT DPCTLSyclEventRef dpnp_rng_multinomial_c(DPCTLSyclQueueRef q_ref,
                                                        void* result,
                                                        const int ntrial,
-                                                       const double* p_vector,
-                                                       const size_t p_vector_size,
+                                                       const double* p_in,
+                                                       const size_t p_size,
                                                        const size_t size,
                                                        const DPCTLEventVectorRef dep_event_vec_ref);
 
 template <typename _DataType>
 INP_DLLEXPORT void dpnp_rng_multinomial_c(
-    void* result, const int ntrial, const double* p_vector, const size_t p_vector_size, const size_t size);
+    void* result, const int ntrial, const double* p_in, const size_t p_size, const size_t size);
 
 /**
  * @ingroup BACKEND_RANDOM_API
@@ -360,31 +360,31 @@ INP_DLLEXPORT void dpnp_rng_multinomial_c(
  * @param [in]  q_ref               Reference to SYCL queue.
  * @param [out] result              Output array.
  * @param [in]  dimen               Dimension of output random vectors.
- * @param [in]  mean_vector         Mean vector a of dimension.
- * @param [in]  mean_vector_size    Length of `mean_vector`.
- * @param [in]  cov_vector          Variance-covariance matrix.
- * @param [in]  cov_vector_size     Length of `cov_vector`.
+ * @param [in]  mean_in             Mean arry of dimension.
+ * @param [in]  mean_size           Length of `mean_in`.
+ * @param [in]  cov                 Variance-covariance matrix.
+ * @param [in]  cov_size            Length of `cov_in`.
  * @param [in]  size                Number of elements in `result` arrays.
  * @param [in]  dep_event_vec_ref   Reference to vector of SYCL events.
  */
 template <typename _DataType>
 INP_DLLEXPORT DPCTLSyclEventRef dpnp_rng_multivariate_normal_c(DPCTLSyclQueueRef q_ref,
                                                                void* result,
                                                                const int dimen,
-                                                               const double* mean_vector,
-                                                               const size_t mean_vector_size,
-                                                               const double* cov_vector,
-                                                               const size_t cov_vector_size,
+                                                               const double* mean_in,
+                                                               const size_t mean_size,
+                                                               const double* cov_in,
+                                                               const size_t cov_size,
                                                                const size_t size,
                                                                const DPCTLEventVectorRef dep_event_vec_ref);
 
 template <typename _DataType>
 INP_DLLEXPORT void dpnp_rng_multivariate_normal_c(void* result,
                                                   const int dimen,
-                                                  const double* mean_vector,
-                                                  const size_t mean_vector_size,
-                                                  const double* cov_vector,
-                                                  const size_t cov_vector_size,
+                                                  const double* mean_in,
+                                                  const size_t mean_size,
+                                                  const double* cov_in,
+                                                  const size_t cov_size,
                                                   const size_t size);
 
 /**

dpnp/backend/kernels/dpnp_krnl_arraycreation.cpp

@@ -73,9 +73,9 @@ DPCTLSyclEventRef dpnp_arange_c(DPCTLSyclQueueRef q_ref,
 
     event = q.submit(kernel_func);
 
-    event.wait();
+    event_ref = reinterpret_cast<DPCTLSyclEventRef>(&event);
 
-    return event_ref;
+    return DPCTLEvent_Copy(event_ref);
 }
 
 template <typename _DataType>
@@ -133,15 +133,15 @@ DPCTLSyclEventRef dpnp_diag_c(DPCTLSyclQueueRef q_ref,
 
     if (ndim == 1)
     {
-        for (size_t i = 0; i < shape[0]; ++i)
+        for (size_t i = 0; i < static_cast<size_t>(shape[0]); ++i)
         {
             size_t ind = (init0 + i) * res_shape[1] + init1 + i;
             result[ind] = v[i];
         }
     }
     else
     {
-        for (size_t i = 0; i < res_shape[0]; ++i)
+        for (size_t i = 0; i < static_cast<size_t>(res_shape[0]); ++i)
         {
             size_t ind = (init0 + i) * shape[1] + init1 + i;
             result[i] = v[ind];
@@ -617,9 +617,6 @@ DPCTLSyclEventRef dpnp_vander_c(DPCTLSyclQueueRef q_ref,
                                 const int increasing,
                                 const DPCTLEventVectorRef dep_event_vec_ref)
 {
-    // avoid warning unused variable
-    (void)dep_event_vec_ref;
-
     DPCTLSyclEventRef event_ref = nullptr;
 
     if ((array1_in == nullptr) || (result1 == nullptr))
@@ -637,8 +634,7 @@ DPCTLSyclEventRef dpnp_vander_c(DPCTLSyclQueueRef q_ref,
 
     if (N == 1)
     {
-        dpnp_ones_c<_DataType_output>(result, size_in);
-        return event_ref;
+        return dpnp_ones_c<_DataType_output>(q_ref, result, size_in, dep_event_vec_ref);
     }
 
     if (increasing)
@@ -762,9 +758,10 @@ DPCTLSyclEventRef dpnp_trace_c(DPCTLSyclQueueRef q_ref,
     };
 
     auto event = q.submit(kernel_func);
-    event.wait();
 
-    return event_ref;
+    event_ref = reinterpret_cast<DPCTLSyclEventRef>(&event);
+
+    return DPCTLEvent_Copy(event_ref);
 }
 
 template <typename _DataType, typename _ResultType>
@@ -849,9 +846,9 @@ DPCTLSyclEventRef dpnp_tri_c(DPCTLSyclQueueRef q_ref,
 
     event = q.submit(kernel_func);
 
-    event.wait();
+    event_ref = reinterpret_cast<DPCTLSyclEventRef>(&event);
 
-    return event_ref;
+    return DPCTLEvent_Copy(event_ref);
 }
 
 template <typename _DataType>