Merge remote-tracking branch 'origin/master' into custom-is-finite

Author: oleksandr-pavlyk

cmake/FindDpctl.cmake

@@ -17,15 +17,8 @@
 #
 
 if(NOT Dpctl_FOUND)
-  set(_find_extra_args)
-  if(Dpctl_FIND_REQUIRED)
-    list(APPEND _find_extra_args REQUIRED)
-  endif()
-  if(Dpctl_FIND_QUIET)
-    list(APPEND _find_extra_args QUIET)
-  endif()
-  find_package(PythonInterp ${_find_extra_args})
-  find_package(PythonLibs ${_find_extra_args})
+  find_package(Python 3.9 REQUIRED
+    COMPONENTS Interpreter Development.Module)
 
   if(PYTHON_EXECUTABLE)
     execute_process(COMMAND "${PYTHON_EXECUTABLE}"

conda-recipe/run_test.bat

@@ -3,8 +3,8 @@
 "%PYTHON%" -c "import dpctl; print(dpctl.__version__)"
 if errorlevel 1 exit 1
 
-"%PYTHON%" -c "import dpctl; dpctl.lsplatform()"
+"%PYTHON%" -m dpctl -f
 if errorlevel 1 exit 1
 
-python -m pytest -q -p no:faulthandler -ra --disable-warnings --pyargs dpctl -vv
+python -m pytest -q -ra --disable-warnings --pyargs dpctl -vv
 if errorlevel 1 exit 1

conda-recipe/run_test.sh

@@ -3,5 +3,5 @@
 set -e
 
 ${PYTHON} -c "import dpctl; print(dpctl.__version__)"
-${PYTHON} -c "import dpctl; dpctl.lsplatform(verbosity=2)"
-${PYTHON} -m pytest -q -ra --disable-warnings -p no:faulthandler --cov dpctl --cov-report term-missing --pyargs dpctl -vv
+${PYTHON} -m dpctl -f
+${PYTHON} -m pytest -q -ra --disable-warnings --cov dpctl --cov-report term-missing --pyargs dpctl -vv

dpctl/_sycl_context.pyx

@@ -114,7 +114,7 @@ cdef class SyclContext(_SyclContext):
                 ctx = dpctl.SyclContext()
                 print(ctx.get_devices())
 
-        - Invoking the constuctor with a specific filter string that creates a
+        - Invoking the constructor with a specific filter string that creates a
           context for the device corresponding to the filter string.
 
         :Example:
@@ -127,7 +127,7 @@ cdef class SyclContext(_SyclContext):
                 d = ctx.get_devices()[0]
                 assert(d.is_gpu)
 
-        - Invoking the constuctor with a :class:`dpctl.SyclDevice` object
+        - Invoking the constructor with a :class:`dpctl.SyclDevice` object
           creates a context for that device.
 
         :Example:
@@ -141,7 +141,7 @@ cdef class SyclContext(_SyclContext):
                 d = ctx.get_devices()[0]
                 assert(d.is_gpu)
 
-        - Invoking the constuctor with a list of :class:`dpctl.SyclDevice`
+        - Invoking the constructor with a list of :class:`dpctl.SyclDevice`
           objects creates a common context for all the devices. This
           constructor call is especially useful when creation a context for
           multiple sub-devices.
@@ -159,7 +159,7 @@ cdef class SyclContext(_SyclContext):
                 ctx = dpctl.SyclContext(sub_devices)
                 assert(len(ctx.get_devices) == len(sub_devices))
 
-        - Invoking the constuctor with a named ``PyCapsule`` with name
+        - Invoking the constructor with a named ``PyCapsule`` with name
           **"SyclContextRef"** that carries a pointer to a ``sycl::context``
           object. The capsule will be renamed upon successful consumption
           to ensure one-time use. A new named capsule can be constructed by
@@ -430,7 +430,7 @@ cdef class SyclContext(_SyclContext):
             return num_devs
         else:
             raise ValueError(
-                "An error was encountered quering the number of devices "
+                "An error was encountered querying the number of devices "
                 "associated with this context"
             )
 

dpctl/_sycl_device.pyx

@@ -1163,7 +1163,7 @@ cdef class SyclDevice(_SyclDevice):
     def create_sub_devices(self, **kwargs):
         """create_sub_devices(partition=parition_spec)
         Creates a list of sub-devices by partitioning a root device based on the
-        provided partion specifier.
+        provided partition specifier.
 
         A partition specifier must be provided using a "partition"
         keyword argument. Possible values for the specifier are: an int, a

dpctl/_sycl_event.pyx

@@ -119,7 +119,7 @@ cdef class SyclEvent(_SyclEvent):
                 # Create a default SyclEvent
                 e = dpctl.SyclEvent()
 
-        - Invoking the constuctor with a named ``PyCapsule`` with name
+        - Invoking the constructor with a named ``PyCapsule`` with name
           **"SyclEventRef"** that carries a pointer to a ``sycl::event``
           object. The capsule will be renamed upon successful consumption
           to ensure one-time use. A new named capsule can be constructed by

dpctl/_sycl_queue.pyx

@@ -246,7 +246,7 @@ cdef class SyclQueue(_SyclQueue):
                 # create a queue for each sub-device using the common context
                 queues = [dpctl.SyclQueue(ctx, sub_d) for sub_d in sub_devices]
 
-        - Invoking the constuctor with a named ``PyCapsule`` with the name
+        - Invoking the constructor with a named ``PyCapsule`` with the name
           **"SyclQueueRef"** that carries a pointer to a ``sycl::queue``
           object. The capsule will be renamed upon successful consumption
           to ensure one-time use. A new named capsule can be constructed by

dpctl/_sycl_timer.py

@@ -37,14 +37,14 @@ class SyclTimer:
             q = dpctl.SyclQueue(property='enable_profiling')
 
             # create the timer
-            miliseconds_sc = 1e-3
-            timer = dpctl.SyclTimer(time_scale = miliseconds_sc)
+            milliseconds_sc = 1e-3
+            timer = dpctl.SyclTimer(time_scale = milliseconds_sc)
 
             # use the timer
             with timer(queue=q):
                 code_block
 
-            # retrieve elapsed times in miliseconds
+            # retrieve elapsed times in milliseconds
             sycl_dt, wall_dt = timer.dt
 
     Remark:

dpctl/apis/include/dpctl4pybind11.hpp

@@ -742,12 +742,12 @@ class usm_memory : public py::object
             return nullptr;
         }
 
-        auto convertor =
+        auto converter =
             ::dpctl::detail::dpctl_capi::get().as_usm_memory_pyobj();
 
         py::object res;
         try {
-            res = convertor(py::handle(o));
+            res = converter(py::handle(o));
         } catch (const py::error_already_set &e) {
             return nullptr;
         }

dpctl/memory/_memory.pyx

@@ -92,7 +92,7 @@ cdef void copy_via_host(void *dest_ptr, SyclQueue dest_queue,
                         void *src_ptr, SyclQueue src_queue, size_t nbytes):
     """
     Copies `nbytes` bytes from `src_ptr` USM memory to
-    `dest_ptr` USM memory using host as the intemediary.
+    `dest_ptr` USM memory using host as the intermediary.
 
     This is useful when `src_ptr` and `dest_ptr` are bound to incompatible
     SYCL contexts.

dpctl/tensor/_copy_utils.py

@@ -56,7 +56,7 @@ def _copy_to_numpy(ary):
 
 def _copy_from_numpy(np_ary, usm_type="device", sycl_queue=None):
     "Copies numpy array `np_ary` into a new usm_ndarray"
-    # This may peform a copy to meet stated requirements
+    # This may perform a copy to meet stated requirements
     Xnp = np.require(np_ary, requirements=["A", "E"])
     alloc_q = normalize_queue_device(sycl_queue=sycl_queue, device=None)
     dt = Xnp.dtype

dpctl/tensor/_device.py

@@ -96,7 +96,7 @@ def sycl_context(self):
     @property
     def sycl_device(self):
         """
-        :class:`dpctl.SyclDevice` targed by this :class:`.Device`.
+        :class:`dpctl.SyclDevice` targeted by this :class:`.Device`.
         """
         return self.sycl_queue_.sycl_device
 

dpctl/tensor/_elementwise_funcs.py

@@ -1628,7 +1628,7 @@
 _subtract_docstring_ = """
 subtract(x1, x2, out=None, order='K')
 
-Calculates the difference bewteen each element `x1_i` of the input
+Calculates the difference between each element `x1_i` of the input
 array `x1` and the respective element `x2_i` of the input array `x2`.
 
 Args:

dpctl/tensor/_manipulation_functions.py

@@ -315,7 +315,7 @@ def broadcast_to(X, shape):
         raise TypeError(f"Expected usm_ndarray type, got {type(X)}.")
 
     # Use numpy.broadcast_to to check the validity of the input
-    # parametr 'shape'. Raise ValueError if 'X' is not compatible
+    # parameter 'shape'. Raise ValueError if 'X' is not compatible
     # with 'shape' according to NumPy's broadcasting rules.
     new_array = np.broadcast_to(
         np.broadcast_to(np.empty(tuple(), dtype="u1"), X.shape), shape
@@ -757,7 +757,7 @@ def iinfo(dtype):
 
     Returns:
         iinfo_object:
-            An object with the followign attributes
+            An object with the following attributes
             * bits: int
                 number of bits occupied by the data type
             * max: int

dpctl/tensor/_print.py

@@ -148,7 +148,7 @@ def set_print_options(
         suppress (bool, optional): If `True,` numbers equal to zero
             in the current precision will print as zero.
             Default: `False`.
-        nanstr (str, optional): String used to repesent nan.
+        nanstr (str, optional): String used to represent nan.
             Raises `TypeError` if nanstr is not a string.
             Default: `"nan"`.
         infstr (str, optional): String used to represent infinity.

dpctl/tensor/_slicing.pxi

@@ -104,7 +104,7 @@ def _basic_slice_meta(ind, shape : tuple, strides : tuple, offset : int):
     Give basic slicing index `ind` and array layout information produce
     a 5-tuple (resulting_shape, resulting_strides, resulting_offset,
        advanced_ind, resulting_advanced_ind_pos)
-    used to contruct a view into underlying array over which advanced
+    used to construct a view into underlying array over which advanced
     indexing, if any, is to be performed.
 
     Raises IndexError for invalid index `ind`.
@@ -201,7 +201,7 @@ def _basic_slice_meta(ind, shape : tuple, strides : tuple, offset : int):
                 raise TypeError
         if ellipses_count > 1:
             raise IndexError(
-                "an index can only have a sinlge ellipsis ('...')")
+                "an index can only have a single ellipsis ('...')")
         if axes_referenced > len(shape):
             raise IndexError(
                 "too many indices for an array, array is "

dpctl/tensor/_stride_utils.pxi

@@ -55,7 +55,7 @@ cdef int _from_input_shape_strides(
         nelems - Number of elements in array
         min_disp = min( dot(strides, index), index for shape)
         max_disp = max( dor(strides, index), index for shape)
-        contig = enumation for array contiguity
+        contig = enumeration for array contiguity
     Returns: 0 on success, error code otherwise.
         On success pointers point to allocated arrays,
         Otherwise they are set to NULL

dpctl/tensor/_usmarray.pyx

@@ -425,7 +425,7 @@ cdef class usm_ndarray:
         cdef char *ary_ptr = NULL
         if (not isinstance(self.base_, dpmem._memory._Memory)):
             raise InternalUSMArrayError(
-                "Invalid instance of usm_ndarray ecountered. "
+                "Invalid instance of usm_ndarray encountered. "
                 "Private field base_ has an unexpected type {}.".format(
                     type(self.base_)
                 )
@@ -566,7 +566,7 @@ cdef class usm_ndarray:
             elif (self.flags_ & USM_ARRAY_F_CONTIGUOUS):
                 return _f_contig_strides(self.nd_, self.shape_)
             else:
-                raise ValueError("Inconsitent usm_ndarray data")
+                raise ValueError("Inconsistent usm_ndarray data")
 
     @property
     def flags(self):
@@ -653,7 +653,7 @@ cdef class usm_ndarray:
 
     @property
     def T(self):
-        """ Returns tranposed array for 2D array, raises `ValueError`
+        """ Returns transposed array for 2D array, raises `ValueError`
         otherwise.
         """
         if self.nd_ == 2:
@@ -1376,8 +1376,8 @@ cdef api object UsmNDArray_MakeSimpleFromMemory(
         QRef: DPCTLSyclQueueRef associated with the allocation
         offset: distance between element with zero multi-index and the
                 start of allocation
-        oder: Memory layout of the array. Use 'C' for C-contiguous or
-              row-major layout; 'F' for F-contiguous or column-major layout
+        order: Memory layout of the array. Use 'C' for C-contiguous or
+               row-major layout; 'F' for F-contiguous or column-major layout
     Returns:
         Created usm_ndarray instance
     """

dpctl/tensor/include/dlpack/README.md

@@ -4,4 +4,4 @@ The header `dlpack.h` downloaded from `https://github.com/dmlc/dlpack.git` remot
 
 The file can also be viewed using github web interface at https://github.com/dmlc/dlpack/blob/e2bdd3bee8cb6501558042633fa59144cc8b7f5f/include/dlpack/dlpack.h
 
-License file was retrived from https://github.com/dmlc/dlpack/blob/main/LICENSE
+License file was retrieved from https://github.com/dmlc/dlpack/blob/main/LICENSE

dpctl/tensor/include/dlpack/dlpack.h

@@ -168,7 +168,7 @@ typedef struct {
    * `byte_offset` field should be used to point to the beginning of the data.
    *
    * Note that as of Nov 2021, multiply libraries (CuPy, PyTorch, TensorFlow,
-   * TVM, perhaps others) do not adhere to this 256 byte aligment requirement
+   * TVM, perhaps others) do not adhere to this 256 byte alignment requirement
    * on CPU/CUDA/ROCm, and always use `byte_offset=0`.  This must be fixed
    * (after which this note will be updated); at the moment it is recommended
    * to not rely on the data pointer being correctly aligned.

dpctl/tensor/libtensor/include/kernels/copy_and_cast.hpp

@@ -118,7 +118,7 @@ typedef sycl::event (*copy_and_cast_generic_fn_ptr_t)(
  * @brief Generic function to copy `nelems` elements from `src` usm_ndarray to
  `dst` usm_ndarray while casting from `srcTy` to `dstTy`.
 
-   Both arrays have array dimensionality specied via argument `nd`. The
+   Both arrays have array dimensionality specified via argument `nd`. The
  `shape_and_strides` is kernel accessible USM array of length `3*nd`, where the
  first `nd` elements encode common shape, second `nd` elements contain strides
  of `src` array, and the trailing `nd` elements contain strides of `dst` array.
@@ -696,7 +696,7 @@ typedef sycl::event (*copy_for_reshape_fn_ptr_t)(
  * @param  src_nd Array dimension of the source array
  * @param  dst_nd Array dimension of the destination array
  * @param  packed_shapes_and_strides Kernel accessible USM array of size
- * `2*src_nd + 2*dst_nd` with contant `[src_shape, src_strides, dst_shape,
+ * `2*src_nd + 2*dst_nd` with content `[src_shape, src_strides, dst_shape,
  * dst_strides]`.
  * @param  src_p  Typeless USM pointer to the buffer of the source array
  * @param  dst_p  Typeless USM pointer to the buffer of the destination array

dpctl/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp

@@ -116,8 +116,12 @@ template <typename argT, typename resT> struct Expm1Functor
             }
 
             // x, y finite numbers
-            const realT sinY_val = std::sin(y);
-            const realT cosY_val = std::cos(y);
+            realT cosY_val;
+            auto cosY_val_multi_ptr = sycl::address_space_cast<
+                sycl::access::address_space::private_space,
+                sycl::access::decorated::yes>(&cosY_val);
+            const realT sinY_val = sycl::sincos(y, cosY_val_multi_ptr);
+
             const realT sinhalfY_val = std::sin(y / 2);
 
             const realT res_re =