diff --git a/docs/sources/heterogeneous_computing.rst b/docs/sources/heterogeneous_computing.rst
index e1499c7..afe5890 100644
--- a/docs/sources/heterogeneous_computing.rst
+++ b/docs/sources/heterogeneous_computing.rst
@@ -1,15 +1,15 @@
.. _heterogeneous_computing:
.. include:: ./ext_links.txt
-Heterogeneous computing
+Heterogeneous Computing
=======================
Device Offload
**************
-Python is an interpreted language, which implies that most of Python codes will run on CPU,
+Python is an interpreted language, which implies that most of the Python script will run on CPU,
and only a few data parallel regions will execute on data parallel devices.
-That is why the concept of host and offload devices is useful when it comes to conceptualizing
+That is why the concept of the host and offload devices is helpful when it comes to conceptualizing
a heterogeneous programming model in Python.
.. image:: ./_images/hetero-devices.png
@@ -19,76 +19,77 @@ a heterogeneous programming model in Python.
The above diagram illustrates the *host* (the CPU which runs Python interpreter) and three *devices*
(two GPU devices and one attached accelerator device). **Data Parallel Extensions for Python**
-offer a programming model where a script executed by Python interpreter on host can *offload* data
-parallel kernels to user-specified device. A *kernel* is the *data parallel region* of a program submitted
-for execution on the device. There can be multiple data parallel regions, and hence multiple *offload kernels*.
+offer a programming model where a script executed by Python interpreter on the host can *offload* data-parallel
+kernels to a user-specified device. A *kernel* is the *data-parallel region* of a program submitted
+for execution on the device. There can be multiple data-parallel regions, and hence multiple *offload kernels*.
-Kernels can be pre-compiled into a library, such as ``dpnp``, or, alternatively, directly coded
+Kernels can be pre-compiled into a library, such as ``dpnp``, or directly coded
in a programming language for heterogeneous computing, such as `OpenCl*`_ or `DPC++`_ .
**Data Parallel Extensions for Python** offer the way of writing kernels directly in Python
using `Numba*`_ compiler along with ``numba-dpex``, the `Data Parallel Extension for Numba*`_.
One or more kernels are submitted for execution into a *queue* targeting an *offload device*.
-For each device one or more queues can be created. In most cases you won’t need to work
+For each device, you can create one or more queues. In most cases, you do not need to work
with device queues directly. Data Parallel Extensions for Python will do necessary underlying
work with queues for you through the :ref:`Compute-Follows-Data`.
Unified Shared Memory
*********************
-Each device has its own memory, not necessarily accessible from another device.
+Each device has its memory, not necessarily accessible from another device.
.. image:: ./_images/hetero-devices.png
:width: 600px
:align: center
:alt: SIMD
-For example, **Device 1** memory may not be directly accessible from the host, but only accessible
+For example, **Device 1** memory may not be directly accessible from the host but accessible
via expensive copying by a driver software. Similarly, depending on the architecture, direct data
-exchange between **Device 2** and **Device 1** may be impossible, and only possible via expensive
+exchange between **Device 2** and **Device 1** may be only impossible possible via expensive
copying through the host memory. These aspects must be taken into consideration when programming
data parallel devices.
-In the above illustration the **Device 2** logically consists of two sub-devices, **Sub-Device 1**
+On the illustration above the **Device 2** logically consists of two sub-devices: **Sub-Device 1**
and **Sub-Device 2**. The programming model allows accessing **Device 2** as a single logical device, or
by working with each individual sub-devices. For the former case a programmer needs to create
a queue for **Device 2**. For the latter case a programmer needs to create 2 queues, one for each sub-device.
`SYCL*`_ standard introduces a concept of the *Unified Shared Memory* (USM). USM requires hardware support
for unified virtual address space, which allows coherency between the host and the device
-pointers. All memory is allocated by the host, but it offers three distinct allocation types:
+pointers. The host allocates all memory, but offers three distinct allocation types:
* **Host: located on the host, accessible by the host or device.** This type of memory is useful in a situation
- when you need to stream a read-only data from the host to the device once.
+ when you need to stream read-only data from the host to the device once.
-* **Device: located on the device, accessibly only by the device.** This type of memory is the fastest one.
- Useful in a situation when most of data crunching happens on the device.
+* **Device: located on the device, accessible by device only.** The fastest type of memory.
+ Useful in a situation when most of the data crunching happens on the device.
-* **Shared: location is both host and device (copies are synchronized by underlying software), accessible by
- the host or device.** Shared allocations are useful when data are accessed by both host and devices,
- since a user does not need to explicitly manage data migration. However, it is much slower than USM Device memory type.
+* **Shared: location is both host and device, accessible by the host and device**.
+ Shared allocations are useful when both host and device access data,
+ since a user does not need to manage data migration explicitly.
+ However, it is much slower than the USM Device memory type.
Compute-Follows-Data
********************
Since data copying between devices is typically very expensive, for performance reasons it is essential
to process data close to where it is allocated. This is the premise of the *Compute-Follows-Data* programming model,
-which states that the compute will happen where the data resides. Tensors implemented in ``dpctl`` and ``dpnp``
+which states that the compute happens where the data resides. Tensors implemented in ``dpctl`` and ``dpnp``
carry information about allocation queues, and hence, about the device on which an array is allocated.
-Based on tensor input arguments of the offload kernel, it deduces the queue on which the execution takes place.
+Based on tensor input arguments of the offload kernel, it deduces the queue on which the execution happens.
.. image:: ./_images/kernel-queue-device.png
:width: 600px
:align: center
:alt: SIMD
-The above picture illustrates the *Compute-Follows-Data* concept. Arrays ``A`` and ``B`` are inputs to the
+The picture above illustrates the *Compute-Follows-Data* concept. Arrays ``A`` and ``B`` are inputs to the
**Offload Kernel**. These arrays carry information about their *allocation queue* (**Device Queue**) and the
*device* (**Device 1**) where they were created. According to the Compute-Follows-Data paradigm
the **Offload Kernel** will be submitted to this **Device Queue**, and the resulting array ``C`` will
be created on the **Device Queue** associated with the **Device 1**.
-**Data Parallel Extensions for Python** require all input tensor arguments to have the **same** allocation queue,
-otherwise an exception will be thrown. For example, the following usages will result in the exception.
+**Data Parallel Extensions for Python** require all input tensor arguments to have the **same** allocation queue.
+Otherwise, an exception is thrown. For example, the following usages will result in an exception.
.. figure:: ./_images/queue-exception1.png
:width: 600px
@@ -102,7 +103,7 @@ otherwise an exception will be thrown. For example, the following usages will re
:align: center
:alt: SIMD
- Input tensors are on the same device but queues are different. Exception is thrown.
+ Input tensors are on the same device, but queues are different. Exception is thrown.
.. figure:: ./_images/queue-exception3.png
:width: 600px
@@ -111,19 +112,19 @@ otherwise an exception will be thrown. For example, the following usages will re
Data belongs to the same device, but queues are different and associated with different sub-devices.
-Copying data between devices and queues
+Copying Data Between Devices and Queues
***************************************
-**Data Parallel Extensions for Python** create **one** *canonical queue* per device so that in
-normal circumstances you do not need to directly manage queues. Having one canonical queue per device
-allows you to copy data between devices using to_device() method:
+**Data Parallel Extensions for Python** create **one** *canonical queue* per device. Normally,
+you do not need to directly manage queues. Having one canonical queue per device
+allows you to copy data between devices using the ``to_device()`` method:
.. code-block:: python
a_new = a.to_device(b.device)
-Array ``a`` will be copied to the device associated with array ``b`` into the new array ``a_new``.
-The same queue will be associated with ``b`` and ``a_new``.
+Array ``a`` is copied to the device associated with array ``b`` into the new array ``a_new``.
+The same queue is associated with ``b`` and ``a_new``.
Alternatively, you can do this as follows:
@@ -137,13 +138,11 @@ Alternatively, you can do this as follows:
a_new = dpctl.tensor.asarray(a, device=b.device)
-Creating additional queues
+Creating Additional Queues
**************************
-As previously indicated **Data Parallel Extensions for Python** automatically create one canonical queue per device,
+As said before, **Data Parallel Extensions for Python** automatically creates one canonical queue per device,
and you normally work with this queue implicitly. However, you can always create as many additional queues per device
-as needed, and work with them explicitly.
+as needed and work explicitly with them, for example, for profiling purposes.
-A typical situation when you will want to create the queue explicitly is for profiling purposes.
Read `Data Parallel Control`_ documentation for more details about queues.
-
diff --git a/docs/sources/index.rst b/docs/sources/index.rst
index 1e5c20e..e0d0a2a 100644
--- a/docs/sources/index.rst
+++ b/docs/sources/index.rst
@@ -10,13 +10,13 @@ Data Parallel Extensions for Python
===================================
Data Parallel Extensions for Python* extend numerical Python capabilities beyond CPU and allow even higher performance
-gains on data parallel devices such as GPUs. It consists of three foundational packages:
+gains on data parallel devices, such as GPUs. It consists of three foundational packages:
* **dpnp** - Data Parallel Extensions for `Numpy*`_ - a library that implements a subset of
Numpy that can be executed on any data parallel device. The subset is a drop-in replacement
of core Numpy functions and numerical data types.
-* **numba_dpex** - Data Parallel Extensions for `Numba*`_ - extension for Numba compiler
- that enables programming data parallel devices the same way you program CPU with Numba.
+* **numba_dpex** - Data Parallel Extensions for `Numba*`_ - an extension for Numba compiler
+ that lets you program data-parallel devices as you program CPU with Numba.
* **dpctl - Data Parallel Control library** that provides utilities for device selection,
allocation of data on devices, tensor data structure along with `Python* Array API Standard`_ implementation, and support for creation of user-defined data-parallel extensions.
diff --git a/docs/sources/parallelism.rst b/docs/sources/parallelism.rst
index d238286..835da8f 100644
--- a/docs/sources/parallelism.rst
+++ b/docs/sources/parallelism.rst
@@ -1,17 +1,17 @@
.. _parallelism:
.. include:: ./ext_links.txt
-Parallelism in modern data parallel architectures
+Parallelism in Modern Data-Parallel Architectures
=================================================
Python is loved for its productivity and interactivity. But when it comes to dealing with
-computationally heavy codes Python performance cannot be compromised. Intel and Python numerical
-computing communities, such as `NumFOCUS `_, dedicated attention to
+computationally heavy codes, Python performance cannot be compromised. Intel and Python numerical
+computing communities, such as `NumFOCUS `_, dedicate attention to
optimizing core numerical and data science packages for leveraging parallelism available in modern CPUs:
-* **Multiple computational cores:** Several computational cores allow processing data concurrently.
- Compared to a single core CPU, *N* cores can process either *N* times bigger data in a fixed time, or
- reduce a computation time *N* times for a fixed amount of data.
+* **Multiple computational cores:** Several computational cores allow to process the data concurrently.
+ Compared to a single-core CPU, *N* cores can process either *N* times bigger data in a fixed time, or
+ reduce a computation time *N* times for a set amount of data.
.. image:: ./_images/dpep-cores.png
:width: 600px
@@ -19,12 +19,12 @@ optimizing core numerical and data science packages for leveraging parallelism a
:alt: Multiple CPU Cores
* **SIMD parallelism:** SIMD (Single Instruction Multiple Data) is a special type of instructions
- that perform operations on vectors of data elements at the same time. The size of vectors is called SIMD width.
- If SIMD width is *K* then a SIMD instruction can process *K* data elements in parallel.
+ that perform operations on vectors of data elements at the same time. The size of vectors is called the SIMD width.
+ If a SIMD width is *K* then a SIMD instruction can process *K* data elements in parallel.
- In the following diagram the SIMD width is 2, which means that a single instruction processes two elements simultaneously.
+ In the following diagram, the SIMD width is 2, which means that a single instruction processes two elements simultaneously.
Compared to regular instructions that process one element at a time, 2-wide SIMD instruction performs
- 2 times more data in fixed time, or, respectively, process a fixed amount of data 2 times faster.
+ two times more data in fixed time, or, respectively, process a fixed amount of data two times faster.
.. image:: ./_images/dpep-simd.png
:width: 150px
@@ -32,9 +32,9 @@ optimizing core numerical and data science packages for leveraging parallelism a
:alt: SIMD
* **Instruction-Level Parallelism:** Modern CISC architectures, such as x86, allow performing data independent
- instructions in parallel. In the following example, we compute :math:`a * b + (c - d)`.
+ instructions in parallel. In the following example, see how to compute :math:`a * b + (c - d)`.
Operations :math:`*` and :math:`-` can be executed in parallel, the last instruction
- :math:`+` depends on availability of :math:`a * b` and :math:`c - d` and hence cannot be executed in parallel
+ :math:`+` depends on availability of :math:`a * b` and :math:`c - d` and cannot be executed in parallel
with :math:`*` and :math:`-`.
.. image:: ./_images/dpep-ilp.png
diff --git a/docs/sources/prerequisites_and_installation.rst b/docs/sources/prerequisites_and_installation.rst
index 800a558..37d64ea 100644
--- a/docs/sources/prerequisites_and_installation.rst
+++ b/docs/sources/prerequisites_and_installation.rst
@@ -5,15 +5,15 @@
.. |trade| unicode:: U+2122
-Prerequisites and installation
+Prerequisites and Installation
==============================
-1. Device drivers
+1. Device Drivers
******************
-Since you are about to start programming data parallel devices beyond CPU, you will need an appropriate hardware.
-For example, Data Parallel Extensions for Python work fine on Intel laptops with integrated graphics.
-In majority of cases your laptop already has all necessary device drivers installed. But if you want the most
+To start programming data parallel devices beyond CPU, you will need an appropriate hardware.
+For example, Data Parallel Extensions for Python work fine on Intel |copy| laptops with integrated graphics.
+In majority of cases, your Windows*-based laptop already has all necessary device drivers installed. But if you want the most
up-to-date driver, you can always
`update it to the latest one `_.
Follow device driver installation instructions
@@ -22,29 +22,29 @@ to complete this step.
All other necessary components for programming data parallel devices will be installed with
Data Parallel Extensions for Python.
-2. Python interpreter
+2. Python Interpreter
**********************
You will need Python 3.8, 3.9, or 3.10 installed on your system. If you do not have one yet the easiest way to do
that is to install `Intel Distribution for Python*`_.
-It will install all essential Python numerical and machine
-learning packages optimized for Intel hardware, including Data Parallel Extensions for Python*.
+It installs all essential Python numerical and machine
+learning packages optimized for the Intel hardware, including Data Parallel Extensions for Python*.
If you have Python installation from another vendor, it is fine too. All you need is to install Data Parallel
-Extensions for Python manually.
+Extensions for Python manually as shown in the next section.
3. Data Parallel Extensions for Python
***************************************
-You can skip this step if you already installed Intel |copy| Distribution for Python or Intel |copy| AI Analytics Toolkit.
+Skip this step if you already installed Intel |copy| Distribution for Python.
The easiest way to install Data Parallel Extensions for Python is to install numba-dpex:
-Conda: ``conda install numba-dpex``
+* Conda: ``conda install numba-dpex``
-Pip: ``pip install numba-dpex``
+* Pip: ``pip install numba-dpex``
-The above commands will install ``numba-dpex`` along with its dependencies, including ``dpnp``, ``dpctl``,
-and required compiler runtimes and drivers.
+These commands install ``numba-dpex`` along with its dependencies, including ``dpnp``, ``dpctl``,
+and required compiler runtimes.
.. WARNING::
- Before installing with conda or pip it is strongly advised to update ``conda`` and ``pip`` to latest versions
\ No newline at end of file
+ Before installing with conda or pip it is strongly advised to update ``conda`` and ``pip`` to latest versions
diff --git a/docs/sources/programming_dpep.rst b/docs/sources/programming_dpep.rst
index db2af12..92c9b3a 100644
--- a/docs/sources/programming_dpep.rst
+++ b/docs/sources/programming_dpep.rst
@@ -1,6 +1,11 @@
.. _programming_dpep:
.. include:: ./ext_links.txt
+
+.. |copy| unicode:: U+000A9
+.. |trade| unicode:: U+2122
+
+
Programming with Data Parallel Extensions for Python
====================================================
@@ -9,20 +14,20 @@ Programming with Data Parallel Extensions for Python
:align: center
:alt: Data Parallel Extensions for Python
-As we briefly outlined, **Data Parallel Extensions for Python** consist of three foundational packages:
+**Data Parallel Extensions for Python** consist of three foundational packages:
-* the `Numpy*`_-like library, ``dpnp``;
-* the compiler extension for `Numba*`_, ``numba-dpex``
-* the library for managing devices, queues, and heterogeneous data, ``dpctl``.
+* The `Numpy*`_-like library, ``dpnp``;
+* The compiler extension for `Numba*`_, ``numba-dpex``
+* The library for managing devices, queues, and heterogeneous data, ``dpctl``.
Their underlying implementation is based on `SYCL*`_ standard, which is a cross-platform abstraction layer
-for heterogeneous computing on data parallel devices, such as CPU, GPU, or domain specific accelerators.
+for heterogeneous computing on data parallel devices, such as CPU, GPU, or domain-specific accelerators.
-Scalars vs. 0-dimensional arrays
+Scalars vs. 0-Dimensional Arrays
********************************
-Primitive types such as Python’s and Numpy’s ``float``, ``int``, or ``complex``, used to represent scalars,
-have the host storage. In contrast, ``dpctl.tensor.usm_ndarray`` and ``dpnp.ndarray`` have USM storage
+Primitive types, such as Python’s and Numpy’s ``float``, ``int``, or ``complex``, used to represent scalars,
+have the host storage. In contrast, ``dpctl.tensor.usm_ndarray`` and ``dpnp.ndarray`` have the USM storage
and carry associated allocation queue. For the :ref:`Compute-Follows-Data` consistent behavior
all ``dpnp`` operations that produce scalars will instead produce respective 0-dimensional arrays.
@@ -32,9 +37,9 @@ That implies, that some code changes may be needed to replace scalar math operat
Data Parallel Extension for Numpy - dpnp
****************************************
-The ``dpnp`` library is a bare minimum to start programming numerical codes for data parallel devices.
-You may already have a Python script written in `Numpy*`_. Being a drop-in replacement of (a subset of) `Numpy*`_,
-to execute your `Numpy*`_ script on GPU usually requires changing just a few lines of the code:
+The ``dpnp`` library is a bare minimum to start programming numerical codes for data-parallel devices.
+If you already have a Python script written in `Numpy*`_, then running it on a GPU will typically require
+changing just a few lines of the code:
.. literalinclude:: ../../examples/01-hello_dpnp.py
:language: python
@@ -43,16 +48,16 @@ to execute your `Numpy*`_ script on GPU usually requires changing just a few lin
:name: ex_01_hello_dpnp
-In this example ``np.asarray()`` creates an array on the default `SYCL*`_ device, which is ``"gpu"`` on systems
-with integrated or discrete GPU (it is ``"host"`` on systems that do not have GPU).
-The queue associated with this array is now carried with ``x``, and ``np.sum(x)`` will derive it from ``x``,
-and respective pre-compiled kernel implementing ``np.sum()`` will be submitted to that queue.
-The result ``y`` will be allocated on the device 0-dimensional array associated with that queue too.
+In this example, ``np.asarray()`` creates an array on the default `SYCL*`_ device, which is a ``"gpu"`` on systems
+with integrated or discrete GPU (it is the ``"host"`` on systems that do not have GPU).
+The queue associated with this array is now carried with ``x``, and ``np.sum(x)`` derives it from ``x``,
+and respective pre-compiled kernel implementing ``np.sum()`` is submitted to that queue.
+The result ``y`` is allocated on the device 0-dimensional array associated with that queue too.
-All ``dpnp`` array creation routines as well as random number generators have additional optional keyword arguments
+All ``dpnp`` array creation routines and random number generators have additional optional keyword arguments:
``device``, ``queue``, and ``usm_type``, using which you can explicitly specify on which device or queue you want
-the tensor data to be created along with USM memory type to be used (``"host"``, ``"device"``, or ``"shared"``).
-In the following example we create the array ``x`` on the GPU device, and perform a reduction sum on it:
+the tensor data to be created along with the USM memory type to be used (``"host"``, ``"device"``, or ``"shared"``).
+In the following example, the array ``x`` is created on the GPU device, and reduction sum is done on it:
.. literalinclude:: ../../examples/02-dpnp_device.py
:language: python
@@ -65,10 +70,10 @@ Data Parallel Extension for Numba - numba-dpex
**********************************************
`Numba*`_ is a powerful Just-In-Time (JIT) compiler that works best on `Numpy*`_ arrays, `Numpy*`_ functions, and loops.
-Data parallel loops is where the data parallelism resides. It allows leveraging all available CPU cores,
+Data-parallel loops is where the data parallelism resides. It allows leveraging all available CPU cores,
SIMD instructions, and schedules those in a way that exploits maximum instruction-level parallelism.
-The ``numba-dpex`` extension allows to compile and offload data parallel regions to any data parallel device.
-It takes just a few lines to modify your CPU `Numba*`_ script to run on GPU.
+The ``numba-dpex`` extension allows to compile and offload data-parallel regions to any data parallel device.
+It takes just a few lines to modify your CPU `Numba*`_ script to run on GPU:
.. literalinclude:: ../../examples/03-dpnp2numba-dpex.py
:language: python
@@ -76,32 +81,27 @@ It takes just a few lines to modify your CPU `Numba*`_ script to run on GPU.
:caption: **EXAMPLE 03:** Compile dpnp code with numba-dpex
:name: ex_03_dpnp2numba_dpex
-In this example we implement a custom function ``sum_it()`` that takes an array input. We compile it with
-`Data Parallel Extension for Numba*`_. Being just-in-time compiler, Numba derives the queue from input argument ``x``,
+In this example, you can see a custom function ``sum_it()`` that takes an array input. By compiling it with
+`Data Parallel Extension for Numba*`_, the queue information is derived from input argument ``x``,
which is associated with the default device (``"gpu"`` on systems with integrated or discrete GPU) and
-dynamically compiles the kernel submitted to that queue. The result will reside as a 0-dimensional array on the device
-associated with the queue, and on exit from the offload kernel it will be assigned to the tensor y.
-
-The ``parallel=True`` setting in ``@njit` is essential to enable generation of data parallel kernels.
-Please also note that we use ``fastmath=True`` in ``@njit`` decorator. This is an important setting
-to instruct the compiler that you’re okay NOT preserving the order of floating-point operations.
-This will enable generation of instructions (such as SIMD) for greater performance.
+dynamically compiles the kernel submitted to that queue. The result resides as a 0-dimensional array on the device
+associated with the queue, and on exit from the offload kernel it will be assigned to the tensor ``y``.
Data Parallel Control - dpctl
*****************************
Both ``dpnp`` and ``numba-dpex`` provide enough API versatility for programming data parallel devices but
-there are some situations when you will need to use dpctl advanced capabilities:
+there are some situations when you need to use ``dpctl`` advanced capabilities:
1. **Advanced device management.** Both ``dpnp`` and ``numba-dpex`` support Numpy array creation routines
with additional parameters that specify the device on which the data is allocated and the type of memory to be used
(``"device"``, ``"host"``, or ``"shared"``). However, if you need some more advanced device and data management
- capabilities you will also need to import ``dpctl`` in addition to ``dpnp`` and/or ``numba-dpex``.
+ capabilities, you need to import ``dpctl`` in addition to ``dpnp`` and/or ``numba-dpex``.
One of frequent usages of ``dpctl`` is to query the list devices present on the system, available driver backend
(such as ``"opencl"``, ``"level_zero"``, ``"cuda"``, etc.)
- Another frequent usage is the creation additional queues for the purpose of profiling or choosing an out-of-order
+ Another frequent usage is the creation of additional queues for the purpose of profiling or choosing an out-of-order
execution of offload kernels.
.. literalinclude:: ../../examples/04-dpctl_device_query.py
@@ -110,11 +110,11 @@ there are some situations when you will need to use dpctl advanced capabilities:
:caption: **EXAMPLE 04:** Get information about devices
:name: ex_04_dpctl_device_query
-2. **Cross-platform development using Python Array API standard.** If you’re a Python developer
- programming Numpy-like codes and targeting different hardware vendors and different tensor implementations,
+2. **Cross-platform development using Python Array API standard.** If you
+ program Numpy-like codes and target different hardware vendors and different tensor implementations,
then going with `Python* Array API Standard`_ is a good choice for writing a portable Numpy-like code.
The ``dpctl.tensor`` implements `Python* Array API Standard`_ for `SYCL*`_ devices. Accompanied with
- respective SYCL device drivers from different vendors ``dpctl.tensor`` becomes a portable solution
+ respective SYCL device, driver from different vendors, ``dpctl.tensor`` becomes a portable solution
for writing numerical codes for any SYCL device.
For example, some Python communities, such as
@@ -122,51 +122,50 @@ there are some situations when you will need to use dpctl advanced capabilities:
a path for having algorithms (re-)implemented using `Python* Array API Standard`_ .
This is a reliable path for extending their capabilities beyond CPU only, or beyond certain GPU vendor only.
-3. **Zero-copy data exchange between tensor implementations.** Certain Python projects may have own tensor
- implementations not relying on ``dpctl.tensor`` or ``dpnp.ndarray`` tensors. Can users still exchange data
- between these tensors not copying it back and forth through the host?
+3. **Zero-copy data exchange between tensor implementations.** Certain Python projects may have their tensor
+ implementations not relying on ``dpctl.tensor`` or ``dpnp.ndarray`` tensors. However, you can still exchange data
+ between these tensors not copying it back and forth through the host.
`Python* Array API Standard`_ specifies the data exchange protocol for zero-copy exchange
between tensors through ``dlpack``. Being the `Python* Array API Standard`_ implementation
``dpctl`` provides ``dpctl.tensor.from_dlpack()`` function used for zero-copy view of another tensor input.
-Debugging and profiling Data Parallel Extensions for Python
+Debugging and Profiling Data Parallel Extensions for Python
***********************************************************
`Intel oneAPI Base Toolkit`_ provides two tools to assist programmers to analyze performance issues in programs
that use **Data Parallel Extensions for Python**. They are `Intel VTune Profiler`_ and
`Intel Advisor`_.
-Intel VTune Profiler examines various performance aspects of a program like, the most time-consuming parts,
+Intel |copy| VTune |trade| Profiler examines various performance aspects of a program like, the most time-consuming parts,
efficiency of offloaded code, impact of memory sub-systems, etc.
-Intel Advisor provides insights on the performance of offloaded code relative to the peak performance and
+Intel |copy| Advisor |trade| provides insights on the performance of offloaded code with respect to the peak performance and
memory bandwidth.
-Next, we will detail the steps involved in using Intel VTune Profiler and Intel Advisor with
+Below you can find how to use Intel VTune Profiler and Intel Advisor with
heterogenous programs that use **Data Parallel Extensions for Python**.
Profiling with Intel VTune Profiler
-----------------------------------
-.. |copy| unicode:: U+000A9
-
-.. |trade| unicode:: U+2122
-
-Intel |copy| VTune |trade| Profiler provides two mechanisms, called *GPU offload* and *GPU hotspots*, to profile heterogeneous programs
+Intel |copy| VTune |trade| Profiler provides two mechanisms to profile heterogeneous programs
targeted to GPUs.
+* GPU offload
+* GPU hotspots
+
The *GPU offload* analysis profiles the entire application (both GPU and host code) and helps to identify
-if the application is CPU or GPU bound. It provides information on the proportion of the execution time spent
+if the application is CPU- or GPU-bound. It provides information on the proportion of the execution time spent
in GPU execution. It also provides information about various hotspots in the program. The key goal of the *GPU offload*
analysis is to identify the parts of the program that can benefit from offloading to GPUs.
The *GPU hotspots* analysis focuses on providing insights into the performance of GPU-offloaded code.
-It provides insights about the parallelism in the GPU kernel, the efficiency of the kernel, SIMD utilization
+It provides insights about the parallelism in the GPU kernel, the efficiency of the kernel, SIMD utilization,
and memory latency. It also provides performance data regarding synchronization operations like GPU barriers and
atomic operations.
-The following instructions are used execute the two Intel VTune Profiler analyses on programs written
-using **Data Parallel Extensions for Python**.
+The following instructions are used to execute the two Intel VTune Profiler analyses on programs written
+using **Data Parallel Extensions for Python**:
.. code-block:: console
:caption: **GPU Offload**
@@ -184,17 +183,17 @@ After collecting the data using the above commands, the Intel VTune Profiler GUI
performance characteristics. In addition to the GUI, it provides mechanisms to generate reports through
the command line and setup a web server for post processing the data.
-Further details on viewing Intel VTune Profiler output along with other use-cases can be found in the
-`Intel VTune Profiler User Guide `_.
+See `Intel VTune Profiler User Guide `_
+for more details.
Profiling with Intel Advisor
----------------------------
-The primary goal of Intel |copy| Advisor is to help programmers make targeted optimizations by identifying
-appropriate kernels and characterizing the performance limiting factors. Intel Advisor provides mechanisms
-to analyze the performance of GPU kernels against the hardware roof-line performance. It provides information
+The primary goal of Intel |copy| Advisor is to help you make targeted optimizations by identifying
+appropriate kernels and characterizing the performance limiting factors. It provides mechanisms
+to analyze the performance of GPU kernels against the hardware roof-line performance. Intel Advisor gives you information
about the maximum achievable performance with the given hardware conditions and helps identify the best
-kernels for optimization. Further, it helps the programmer characterize if a GPU kernel is bound by
+kernels for optimization. Further, it helps you to characterize if a GPU kernel is bound by
compute capacity or by memory bandwidth.
The following instructions are used to generate GPU roof-line performance graphs using Intel Advisor.
@@ -207,7 +206,7 @@ The following instructions are used to generate GPU roof-line performance graphs
This command collects the GPU roof-line data from executing the application written using
**Data Parallel Extensions for Python**.
-The next command generates the roof-line graph as a html file in the output directory.
+The next command generates the roof-line graph as a HTML file in the output directory:
.. code-block:: console
:caption: **Generate Roofline HTML-File**
@@ -219,7 +218,7 @@ The next command generates the roof-line graph as a html file in the output dire
:align: center
:alt: Advisor roofline analysis example on Gen9 integrated GPU
-The above figure shows an example roof-line graph generated using Intel Advisor.
+The figure above shows the example of roof-line graph generated using Intel Advisor.
The X-axis in the graph represents arithmetic intensity and the Y-axis represents performance in GFLOPS.
The horizontal lines parallel to the X-axis represent the roof-line compute capacity for the given hardware.
The cross-diagonal lines represent the peak memory bandwidth of different layers of the memory hierarchy.
@@ -227,17 +226,17 @@ The red colored dot corresponds to the executed GPU kernel. The graph shows the
to the peak compute capacity and memory bandwidth. It also shows whether the GPU kernel is memory or compute
bound depending on the roof-line that is limiting the GPU kernel.
-For further details on Intel Advisor and its extended capabilities refer to the
+For further details on Intel Advisor and its extended capabilities, refer to the
`Intel |copy| Advisor User Guide `_.
.. todo::
Document debugging section
-Writing robust numerical codes for heterogeneous computing
+Writing Robust Numerical Codes for Heterogeneous Computing
**********************************************************
-Default primitive type (``dtype``) in `Numpy*`_ is double precision (``float64``), which is supported by
+Default primitive type (``dtype``) in `Numpy*`_ is the double precision (``float64``), which is supported by
majority of modern CPUs. When it comes to program GPUs and especially specialized accelerators,
the set of supported primitive data types may be limited. For example, certain GPUs may not support
double precision or half-precision. **Data Parallel Extensions for Python** select default ``dtype`` depending on
@@ -247,22 +246,22 @@ careful management of hardware peculiarities to keep the Python script portable
There are several hints how to make the numerical code portable and robust.
-Sensitivity to floating-point errors
+Sensitivity to Floating-Point Errors
------------------------------------
Floating-point arithmetic has a finite precision, which implies that only a tiny fraction of real numbers can be
represented in floating-point arithmetic. It is almost certain that every floating-point operation
-will induce a rounding error because the result cannot be accurately represented as a floating-point number.
+induces a rounding error because the result cannot be accurately represented as a floating-point number.
The `IEEE 754-2019 Standard for Floating-Point Arithmetic`_ sets the upper bound for rounding errors in each
arithmetic operation to 0.5 *ulp*, meaning that each arithmetic operation must be accurate to the last bit of
floating-point mantissa, which is an order of :math:`10^-16` in double precision and :math:`10^-7`
in single precision.
-In robust numerical codes these errors tend to accumulate slowly so that single precision is enough to
-calculate the result accurate to 3-5 decimal digits.
+In robust numerical codes, these errors tend to accumulate slowly so that single precision is enough to
+calculate the result accurate to three-five decimal digits.
However, there is a situation known as a *catastrophic cancellation*, when small accumulated errors
-result in a significant (or even a complete) loss of accuracy. The catastrophic cancellation happens
+result in a significant, or even complete, loss of accuracy. The catastrophic cancellation happens
when two close floating-point numbers with small rounding errors are subtracted. As a result the original
rounding errors amplify by the number of identical leading digits:
@@ -271,7 +270,7 @@ rounding errors amplify by the number of identical leading digits:
:align: center
:alt: Floating-Point Cancellation
-In the above example, green digits are accurate digits, a few trailing digits in red are inaccurate due to
+In the example above, green digits are accurate digits, a few trailing digits in red are inaccurate due to
induced errors. As a result of subtraction, only one accurate digit remains.
Situations with catastrophic cancellations must be carefully handled. An example where catastrophic
@@ -282,14 +281,14 @@ to approximate the derivative:
df/dx \approx \frac{f(x+\delta) - f(x-\delta)}{2\delta}
-Smaller you take :math:`\delta` is greater the catastrophic cancellation. At the same time bigger :math:`\delta`
+Smaller you take :math:`\delta` is greater the catastrophic cancellation. At the same time, bigger :math:`\delta`
results in bigger approximation error. Books on numerical computing and floating-point arithmetic discuss
variety of technics to make catastrophic cancellations controllable. For more details about floating-point
-arithmetic please refer to `IEEE 754-2019 Standard for Floating-Point Arithmetic`_ and the article by
+arithmetic, refer to `IEEE 754-2019 Standard for Floating-Point Arithmetic`_ and the article by
`David Goldberg, What every computer scientist should know about floating-point arithmetic`_.
-Switching between single and double precision
+Switching Between Single and Double Precision
---------------------------------------------
1. Implement your code to switch easily between single and double precision in a controlled fashion.
@@ -298,10 +297,10 @@ Switching between single and double precision
2. Run your code on a representative set of inputs in single and double precisions.
Observe sensitivity of computed results to the switching between single and double precisions.
- If results remain identical to 3-5 digits for different inputs, it is a good sign that your code
+ If results remain identical to three-five digits for different inputs, it is a good sign that your code
is not sensitive to floating-point errors.
3. Write your code with catastrophic cancellations in mind. These blocks of code will require special
care such as the use of extended precision or other techniques to control cancellations.
- It is likely that this part of the code will require a hardware specific implementation.
+ It is likely that this part of the code requires a hardware specific implementation.
diff --git a/docs/sources/useful_links.rst b/docs/sources/useful_links.rst
index 9bc7244..5ebb865 100644
--- a/docs/sources/useful_links.rst
+++ b/docs/sources/useful_links.rst
@@ -11,30 +11,30 @@ Useful links
* - Document
- Description
* - `Data Parallel Extension for Numpy*`_
- - Documentation for programming NumPy-like codes on data parallel devices
+ - Documentation for programming NumPy-like codes on data-parallel devices
* - `Data Parallel Extension for Numba*`_
- - Documentation for programming Numba codes on data parallel devices the same way as you program Numba on CPU
+ - Documentation for programming Numba codes on data-parallel devices as you program Numba on CPU
* - `Data Parallel Control`_
- Documentation how to manage data and devices, how to interchange data between different tensor implementations,
and how to write data parallel extensions
* - `Intel VTune Profiler`_
- - Performance profiler supporting analysis of bottlenecks from function leve down to low level instructions.
+ - Performance profiler supporting analysis of bottlenecks from function leve down to low level instructions.
Supports Python and Numba
* - `Intel Advisor`_
- - Analyzes native and Python codes and provides an advice for better composition of heterogeneous algorithms
+ - Analyzes native and Python codes and provides the advice for better composition of heterogeneous algorithms
* - `Python* Array API Standard`_
- Standard for writing portable Numpy-like codes targeting different hardware vendors and frameworks
operating with tensor data
* - `SYCL*`_
- Standard for writing C++-like codes for heterogeneous computing
* - `DPC++`_
- - Free e-book how to program data parallel devices using Data Parallel C++
+ - Free e-book on how to program data-parallel devices using Data Parallel C++
* - `OpenCl*`_
- OpenCl* Standard for heterogeneous programming
* - `IEEE 754-2019 Standard for Floating-Point Arithmetic`_
- Standard for floating-point arithmetic, essential for writing robust numerical codes
* - `David Goldberg, What every computer scientist should know about floating-point arithmetic`_
- - Scientific paper important for understanding how to write robust numerical code
+ - Scientific paper. Important for understanding how to write robust numerical code
* - `Numpy*`_
- Documentation for Numpy - foundational CPU library for array programming. Used in conjunction with
`Data Parallel Extension for Numpy*`_.
diff --git a/examples/01-get_started.ipynb b/examples/01-get_started.ipynb
deleted file mode 100644
index 7b4ec35..0000000
--- a/examples/01-get_started.ipynb
+++ /dev/null
@@ -1,581 +0,0 @@
-{
- "cells": [
- {
- "cell_type": "markdown",
- "id": "a002ea61",
- "metadata": {},
- "source": [
- "# Getting Started"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "905c742a",
- "metadata": {},
- "source": [
- "# Few Line Changes to Run on GPU"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "41bf2099",
- "metadata": {},
- "source": [
- "Let's see the example how you can easily in a few lines of code switch computations from CPU to GPU device.\n",
- "\n",
- "Please look on the original example.\n",
- "We allocate 2 matrices on the Host (CPU) device usnig NumPy array function, all future calculations will be performed as well on the allocated Host(CPU) device."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 3,
- "id": "d9e8711d",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "res = [[2 2]\n",
- " [2 2]]\n"
- ]
- }
- ],
- "source": [
- "# Original CPU script\n",
- "\n",
- "# Call numpy library\n",
- "import numpy as np\n",
- "\n",
- "# Data alocated on the CPU device\n",
- "x = np.array([[1, 1], [1, 1]])\n",
- "y = np.array([[1, 1], [1, 1]])\n",
- "\n",
- "# Compute performed on the CPU device, where data is allocated\n",
- "res = np.matmul(x, y)\n",
- "\n",
- "print (\"res = \", res)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "402c3d61",
- "metadata": {},
- "source": [
- "Now let's try to modify our code in a way when all calculations occur on the GPU device.\n",
- "To do it, you need just to switch to the dpnp library and see on the result."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 1,
- "id": "c10b7d83",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Array x is located on the device: Device(level_zero:gpu:0)\n",
- "Array y is located on the device: Device(level_zero:gpu:0)\n",
- "res is located on the device: Device(level_zero:gpu:0)\n",
- "res = [[2 2]\n",
- " [2 2]]\n"
- ]
- }
- ],
- "source": [
- "# Modified XPU script\n",
- "\n",
- "# Drop-in replacement via single line change\n",
- "import dpnp as np\n",
- "\n",
- "# Data alocated on default SYCL device\n",
- "x = np.array([[1, 1], [1, 1]])\n",
- "y = np.array([[1, 1], [1, 1]])\n",
- "\n",
- "# Compute performed on the device, where data is allocated\n",
- "res = np.matmul(x, y)\n",
- "\n",
- "\n",
- "print (\"Array x is located on the device:\", x.device)\n",
- "print (\"Array y is located on the device:\", y.device)\n",
- "print (\"res is located on the device:\", res.device)\n",
- "print (\"res = \", res)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "d6a58f17",
- "metadata": {},
- "source": [
- "As you may see changing only one line of code help us to perform all calculations on the GPU device.\n",
- "In this example np.array() creates an array on the default SYCL* device, which is \"gpu\" on systems with integrated or discrete GPU (it is \"host\" on systems that do not have GPU). The queue associated with this array is now carried with x and y, and np.matmul(x, y) will do matrix product of two arrays x and y, and respective pre-compiled kernel implementing np.matmul() will be submitted to that queue. The result res will be allocated on the device array associated with that queue too.\n",
- "\n",
- "Now let's make a few improvements in our code and see how we can control and specify exact device on which we want to perform our calculations and which USM memory type to use."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "be340585",
- "metadata": {},
- "source": [
- "# dpnp simple examples with popular functions"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "26d8c3c6",
- "metadata": {},
- "source": [
- "1. Example to return an array with evenly spaced values within a given interval."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 35,
- "id": "0435d7f0",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "282 µs ± 27.6 µs per loop (mean ± std. dev. of 7 runs, 10,000 loops each)\n",
- "Result a is located on the device: Device(level_zero:gpu:0)\n",
- "a = [ 3 9 15 21 27]\n"
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "# Create an array of values from 3 till 30 with step 6\n",
- "a = np.arange(3, 30, step = 6)\n",
- "\n",
- "print (\"Result a is located on the device:\", a.device)\n",
- "print (\"a = \", a)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "a6765095",
- "metadata": {},
- "source": [
- "In this example np.arange() creates an array on the default SYCL* device, which is \"gpu\" on systems with integrated or discrete GPU (it is \"host\" on systems that do not have GPU)."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "35081461",
- "metadata": {},
- "source": [
- "2. Example which calculates on the GPU the sum of the array elements"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 3,
- "id": "e613398f",
- "metadata": {
- "scrolled": true
- },
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Result x is located on the device: Device(level_zero:gpu:0)\n",
- "Result y is located on the device: Device(level_zero:gpu:0)\n",
- "The sum of the array elements is: 6\n"
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "x = np.empty(3)\n",
- "\n",
- "try:\n",
- " # Using filter selector strings to specify root devices for a new array\n",
- " x = np.asarray ([1, 2, 3], device=\"gpu\")\n",
- " print (\"Result x is located on the device:\", x.device)\n",
- "except:\n",
- " print (\"GPU device is not available\")\n",
- "\n",
- "# Return the sum of the array elements\n",
- "y = np.sum (x) # Expect 6\n",
- "\n",
- "print (\"Result y is located on the device:\", y.device)\n",
- "print (\"The sum of the array elements is: \", y )"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "88b4915e",
- "metadata": {},
- "source": [
- "In this example np.asarray() creates an array on the default GPU device. The queue associated with this array is now carried with x, and np.sum(x) will derive it from x, and respective pre-compiled kernel implementing np.sum() will be submitted to that queue. The result y will be allocated on the device 0-dimensional array associated with that queue too."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "b49f4c62",
- "metadata": {},
- "source": [
- "3. Example of inversion of an array"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 1,
- "id": "4b53afed",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Array a is located on the device: Device(level_zero:gpu:0)\n",
- "Result x is located on the device: Device(level_zero:gpu:0)\n",
- "Array x is: [[-2 -2]\n",
- " [-3 -2]\n",
- " [-2 -1]\n",
- " [ 0 -1]]\n"
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "try:\n",
- " \n",
- " # Using filter selector strings to specify root devices for an array\n",
- " a = np.array([[1, 1], [2, 1], [1, 0], [-1, 0]], device = \"gpu\")\n",
- " print (\"Array a is located on the device:\", a.device) \n",
- "\n",
- " # Do inversion of an array \"a\"\n",
- " x = np.invert(a)\n",
- "\n",
- " print (\"Result x is located on the device:\", x.device)\n",
- " print (\"Array x is:\", x) \n",
- "\n",
- "except:\n",
- " print (\"GPU device is not available\")\n"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "e10f226a",
- "metadata": {},
- "source": [
- "In this example np.array() creates an array on the default GPU device. The queue associated with this array is now carried with a, and np.invert(a) will derive it from a, and respective pre-compiled kernel implementing np.invert() will be submitted to that queue. The result x will be allocated on the device array associated with that queue too."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "94cb3b9b",
- "metadata": {},
- "source": [
- "# dpctl simple examples"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "9fb4b1b5",
- "metadata": {},
- "source": [
- "Here you may find a list of simple examples which explain how to understand how many devices you have in the systen and how to operate with them\n",
- "Let's print the list of all available SYCL devices."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 5,
- "id": "b890ae71",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Intel(R) OpenCL HD Graphics OpenCL 3.0 \n",
- "Intel(R) FPGA Emulation Platform for OpenCL(TM) OpenCL 1.2 Intel(R) FPGA SDK for OpenCL(TM), Version 20.3\n",
- "Intel(R) OpenCL OpenCL 3.0 WINDOWS\n",
- "Intel(R) Level-Zero 1.3\n"
- ]
- }
- ],
- "source": [
- "# See the list of available SYCL platforms and extra metadata about each platform.\n",
- "import dpctl\n",
- "\n",
- "dpctl.lsplatform() # Print platform information"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "db5e8db4",
- "metadata": {},
- "source": [
- "Let's look on the output.\n",
- "On my platform is available OpenCL GPU driver, Intel(R) FPGA Emulation Device, OpenCL CPU driver and Level Zero GPU driver.\n",
- "If i play with verbocity parameter, i can get more information about the devices i have."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 6,
- "id": "ffcf0cfb",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Platform 0 ::\n",
- " Name Intel(R) OpenCL HD Graphics\n",
- " Version OpenCL 3.0 \n",
- " Vendor Intel(R) Corporation\n",
- " Backend opencl\n",
- " Num Devices 1\n",
- " # 0\n",
- " Name Intel(R) Iris(R) Xe Graphics\n",
- " Version 31.0.101.3430\n",
- " Filter string opencl:gpu:0\n",
- "Platform 1 ::\n",
- " Name Intel(R) FPGA Emulation Platform for OpenCL(TM)\n",
- " Version OpenCL 1.2 Intel(R) FPGA SDK for OpenCL(TM), Version 20.3\n",
- " Vendor Intel(R) Corporation\n",
- " Backend opencl\n",
- " Num Devices 1\n",
- " # 0\n",
- " Name Intel(R) FPGA Emulation Device\n",
- " Version 2022.15.11.0.18_160000\n",
- " Filter string opencl:accelerator:0\n",
- "Platform 2 ::\n",
- " Name Intel(R) OpenCL\n",
- " Version OpenCL 3.0 WINDOWS\n",
- " Vendor Intel(R) Corporation\n",
- " Backend opencl\n",
- " Num Devices 1\n",
- " # 0\n",
- " Name 11th Gen Intel(R) Core(TM) i7-1185G7 @ 3.00GHz\n",
- " Version 2022.15.11.0.18_160000\n",
- " Filter string opencl:cpu:0\n",
- "Platform 3 ::\n",
- " Name Intel(R) Level-Zero\n",
- " Version 1.3\n",
- " Vendor Intel(R) Corporation\n",
- " Backend ext_oneapi_level_zero\n",
- " Num Devices 1\n",
- " # 0\n",
- " Name Intel(R) Iris(R) Xe Graphics\n",
- " Version 1.3.23904\n",
- " Filter string level_zero:gpu:0\n"
- ]
- }
- ],
- "source": [
- "# See the list of available SYCL platforms and extra metadata about each platform.\n",
- "import dpctl\n",
- "\n",
- "dpctl.lsplatform(2) # Print platform information with verbocitz level 2 (highest level)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "f63525d3",
- "metadata": {},
- "source": [
- "Having information about available SYCL platforms you can specify which type of devices you want to work with"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 3,
- "id": "84ff47e9",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "[, ]\n"
- ]
- }
- ],
- "source": [
- "# See the list of available gpu devices and their extra metadata.\n",
- "import dpctl\n",
- "\n",
- "if dpctl.has_gpu_devices():\n",
- " print (dpctl.get_devices(device_type='gpu'))\n",
- "else:\n",
- " print(\"GPU device is not available\")"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 7,
- "id": "a93e7cf8",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "[]\n"
- ]
- }
- ],
- "source": [
- "# See the list of available gpu devices and their extra metadata.\n",
- "import dpctl\n",
- "\n",
- "if dpctl.has_cpu_devices():\n",
- " print (dpctl.get_devices(device_type='cpu'))\n",
- "else:\n",
- " print(\"CPU device is not available\")"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "efcc3f45",
- "metadata": {},
- "source": [
- "And you can make selection of the specific device in your system using the default selctor"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 9,
- "id": "1c068447",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- " Name Intel(R) Iris(R) Xe Graphics\n",
- " Driver version 1.3.23904\n",
- " Vendor Intel(R) Corporation\n",
- " Filter string level_zero:gpu:0\n",
- "\n"
- ]
- }
- ],
- "source": [
- "import dpctl\n",
- "\n",
- "try:\n",
- " # Create a SyclDevice of type GPU based on whatever is returned\n",
- " # by the SYCL `gpu_selector` device selector class.\n",
- " gpu = dpctl.select_gpu_device()\n",
- " gpu.print_device_info() # print GPU device information\n",
- "\n",
- "except:\n",
- " print (\"GPU device is not available\")"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "c378b79d",
- "metadata": {},
- "source": [
- "Or by using the infromation in filter string of the device create abd explicit SyclDevice "
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 7,
- "id": "ad83abb5",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- " Name Intel(R) Iris(R) Xe Graphics\n",
- " Driver version 1.3.23904\n",
- " Vendor Intel(R) Corporation\n",
- " Profile FULL_PROFILE\n",
- " Filter string level_zero:gpu:0\n",
- "\n"
- ]
- }
- ],
- "source": [
- "import dpctl\n",
- "\n",
- "# Create a SyclDevice with an explicit filter string,\n",
- "# in this case the first level_zero gpu device.\n",
- "try:\n",
- " level_zero_gpu = dpctl.SyclDevice(\"level_zero:gpu:0\")\n",
- " level_zero_gpu.print_device_info()\n",
- "except:\n",
- " print(\"The first level_zero GPU device is not available\") "
- ]
- },
- {
- "cell_type": "markdown",
- "id": "eadefe0b",
- "metadata": {},
- "source": [
- "Let's check if your gpu device support double precision. To do this we need to selcet gpu device and check the parameter has_aspect_fp64:"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 16,
- "id": "a94756d7",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- " Name Intel(R) Iris(R) Xe Graphics\n",
- " Driver version 1.3.23904\n",
- " Vendor Intel(R) Corporation\n",
- " Filter string level_zero:gpu:0\n",
- "\n",
- "Double precision support is False\n"
- ]
- }
- ],
- "source": [
- "import dpctl\n",
- "# Select GPU device and check double precision support\n",
- "try:\n",
- " gpu = dpctl.select_gpu_device()\n",
- " gpu.print_device_info()\n",
- " print(\"Double precision support is\", gpu.has_aspect_fp64)\n",
- "except:\n",
- " print(\"The GPU device is not available\") "
- ]
- }
- ],
- "metadata": {
- "kernelspec": {
- "display_name": "Python 3 (ipykernel)",
- "language": "python",
- "name": "python3"
- },
- "language_info": {
- "codemirror_mode": {
- "name": "ipython",
- "version": 3
- },
- "file_extension": ".py",
- "mimetype": "text/x-python",
- "name": "python",
- "nbconvert_exporter": "python",
- "pygments_lexer": "ipython3",
- "version": "3.8.15"
- }
- },
- "nbformat": 4,
- "nbformat_minor": 5
-}
diff --git a/examples/02-dpnp_numpy_fallback.ipynb b/examples/02-dpnp_numpy_fallback.ipynb
deleted file mode 100644
index ca6bcf5..0000000
--- a/examples/02-dpnp_numpy_fallback.ipynb
+++ /dev/null
@@ -1,140 +0,0 @@
-{
- "cells": [
- {
- "cell_type": "markdown",
- "id": "be340585",
- "metadata": {},
- "source": [
- "# Usage of numpy functions in dpnp library"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "9008dbe5",
- "metadata": {},
- "source": [
- "1. Example of the usage of the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` environment variable and finding specific values in array based on condition using dpnp library"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "5d4300fe",
- "metadata": {},
- "source": [
- "Not all functions were yet implemented in the dpnp library, some of the functions require enabling of the direct fallback to the NumPy library. \n",
- "One of the example can be the function \"dpnp.full ()\" and parameter like in non default state. \n",
- "Let's look on the example where we want to create an two dimencial array with singular element and array like option. "
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 1,
- "id": "c78511cd",
- "metadata": {},
- "outputs": [
- {
- "ename": "NotImplementedError",
- "evalue": "Requested funtion=full with args=((2, 2), 3, None, 'C') and kwargs={'like': } isn't currently supported and would fall back on NumPy implementation. Define enviroment variable `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` to `0` if the fall back is required to be supported without rasing an exception.",
- "output_type": "error",
- "traceback": [
- "\u001b[1;31m---------------------------------------------------------------------------\u001b[0m",
- "\u001b[1;31mNotImplementedError\u001b[0m Traceback (most recent call last)",
- "Cell \u001b[1;32mIn[1], line 4\u001b[0m\n\u001b[0;32m 1\u001b[0m \u001b[38;5;28;01mimport\u001b[39;00m \u001b[38;5;21;01mdpnp\u001b[39;00m \u001b[38;5;28;01mas\u001b[39;00m \u001b[38;5;21;01mnp\u001b[39;00m\n\u001b[0;32m 3\u001b[0m \u001b[38;5;66;03m# Create an two dimencial array with singular element and array like option\u001b[39;00m\n\u001b[1;32m----> 4\u001b[0m a \u001b[38;5;241m=\u001b[39m \u001b[43mnp\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mfull\u001b[49m\u001b[43m(\u001b[49m\u001b[43m(\u001b[49m\u001b[38;5;241;43m2\u001b[39;49m\u001b[43m,\u001b[49m\u001b[38;5;241;43m2\u001b[39;49m\u001b[43m)\u001b[49m\u001b[43m,\u001b[49m\u001b[38;5;241;43m3\u001b[39;49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mlike\u001b[49m\u001b[43m \u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43m \u001b[49m\u001b[43mnp\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mzeros\u001b[49m\u001b[43m(\u001b[49m\u001b[43m(\u001b[49m\u001b[38;5;241;43m2\u001b[39;49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[38;5;241;43m0\u001b[39;49m\u001b[43m)\u001b[49m\u001b[43m)\u001b[49m\u001b[43m)\u001b[49m\n\u001b[0;32m 5\u001b[0m \u001b[38;5;28mprint\u001b[39m (\u001b[38;5;124m\"\u001b[39m\u001b[38;5;124mArray a is located on the device:\u001b[39m\u001b[38;5;124m\"\u001b[39m, a\u001b[38;5;241m.\u001b[39mdevice)\n\u001b[0;32m 6\u001b[0m \u001b[38;5;28mprint\u001b[39m (\u001b[38;5;124m\"\u001b[39m\u001b[38;5;124mArray a is\u001b[39m\u001b[38;5;124m\"\u001b[39m, a)\n",
- "File \u001b[1;32m~\\Anaconda3\\envs\\my_env\\lib\\site-packages\\dpnp\\dpnp_iface_arraycreation.py:732\u001b[0m, in \u001b[0;36mfull\u001b[1;34m(shape, fill_value, dtype, order, like, device, usm_type, sycl_queue)\u001b[0m\n\u001b[0;32m 723\u001b[0m \u001b[38;5;28;01melse\u001b[39;00m:\n\u001b[0;32m 724\u001b[0m \u001b[38;5;28;01mreturn\u001b[39;00m dpnp_container\u001b[38;5;241m.\u001b[39mfull(shape,\n\u001b[0;32m 725\u001b[0m fill_value,\n\u001b[0;32m 726\u001b[0m dtype\u001b[38;5;241m=\u001b[39mdtype,\n\u001b[1;32m (...)\u001b[0m\n\u001b[0;32m 729\u001b[0m usm_type\u001b[38;5;241m=\u001b[39musm_type,\n\u001b[0;32m 730\u001b[0m sycl_queue\u001b[38;5;241m=\u001b[39msycl_queue)\n\u001b[1;32m--> 732\u001b[0m \u001b[38;5;28;01mreturn\u001b[39;00m \u001b[43mcall_origin\u001b[49m\u001b[43m(\u001b[49m\u001b[43mnumpy\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mfull\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mshape\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mfill_value\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mdtype\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43morder\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mlike\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43mlike\u001b[49m\u001b[43m)\u001b[49m\n",
- "File \u001b[1;32mdpnp\\dpnp_utils\\dpnp_algo_utils.pyx:132\u001b[0m, in \u001b[0;36mdpnp.dpnp_utils.dpnp_algo_utils.call_origin\u001b[1;34m()\u001b[0m\n",
- "\u001b[1;31mNotImplementedError\u001b[0m: Requested funtion=full with args=((2, 2), 3, None, 'C') and kwargs={'like': } isn't currently supported and would fall back on NumPy implementation. Define enviroment variable `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` to `0` if the fall back is required to be supported without rasing an exception."
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "# Create an two dimencial array with singular element and array like option\n",
- "a = np.full((2,2),3, like = np.zeros((2, 0)))\n",
- "print (\"Array a is located on the device:\", a.device)\n",
- "print (\"Array a is\", a)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "ae026021",
- "metadata": {},
- "source": [
- "As you can see the function \"dpnp.full ()\" and parameter like in non default state is not implemented in the dpnp library.\n",
- "We got the following error message: \"Requested funtion=full with args=((2, 2), 3, None, 'C') and kwargs={'like': } isn't currently supported and would fall back on NumPy implementation. Define environment variable `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` to `0` if the fall back is required to be supported without raising an exception.\" \n",
- "By default the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` environment variable is not null and it allows to use only dpnp library. \n",
- "If we want to call NumPy library for functions not supported under dpnp, we need to change this environment variable to `0`."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "210440c3",
- "metadata": {},
- "source": [
- "Let's overwrite the same example using the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK = 0` environment variable."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "bca7490d",
- "metadata": {},
- "source": [
- "`Note:` Please pay attention that if you are working in the Jupyter Notebook, before running the example with setting the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK = 0` you need to restart the kernel in the Jupyter Notebook. "
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 1,
- "id": "05ad1565",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK = 0\n",
- "Array a is located on the device: Device(level_zero:gpu:0)\n",
- "Array a is [[3 3]\n",
- " [3 3]]\n"
- ]
- }
- ],
- "source": [
- "import os\n",
- "# call numpy if not null than we will use dpnp, by default not null\n",
- "os.environ[\"DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK\"] = \"0\" \n",
- "\n",
- "import dpnp as np\n",
- "\n",
- "# Expect result 0\n",
- "print (\"DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK =\", np.config.__DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK__) \n",
- "\n",
- "# Create an two dimencial array with singular element and array like option\n",
- "a = np.full((2,2),3, like = np.zeros((2, 0)))\n",
- "print (\"Array a is located on the device:\", a.device)\n",
- "print (\"Array a is\", a)"
- ]
- }
- ],
- "metadata": {
- "kernelspec": {
- "display_name": "Python 3 (ipykernel)",
- "language": "python",
- "name": "python3"
- },
- "language_info": {
- "codemirror_mode": {
- "name": "ipython",
- "version": 3
- },
- "file_extension": ".py",
- "mimetype": "text/x-python",
- "name": "python",
- "nbconvert_exporter": "python",
- "pygments_lexer": "ipython3",
- "version": "3.8.15"
- }
- },
- "nbformat": 4,
- "nbformat_minor": 5
-}
diff --git a/examples/03-dpnp2numba-dpex.py b/examples/03-dpnp2numba-dpex.py
index c4fb147..d3161f5 100644
--- a/examples/03-dpnp2numba-dpex.py
+++ b/examples/03-dpnp2numba-dpex.py
@@ -25,10 +25,10 @@
# *****************************************************************************
import dpnp as np
-from numba_dpex import njit
+from numba_dpex import dpjit as njit
-@njit(fastmath=True)
+@njit()
def sum_it(x): # Device queue is inferred from x. The kernel is submitted to that queue
return np.sum(x)
diff --git a/examples/05-dpctl_dpnp_test.py b/examples/05-dpctl_dpnp_test.py
deleted file mode 100644
index 1424a60..0000000
--- a/examples/05-dpctl_dpnp_test.py
+++ /dev/null
@@ -1,46 +0,0 @@
-# *****************************************************************************
-# Copyright (c) 2022, Intel Corporation All rights reserved.
-#
-# Redistribution and use in source and binary forms, with or without
-# modification, are permitted provided that the following conditions are met:
-#
-# Redistributions of source code must retain the above copyright notice,
-# this list of conditions and the following disclaimer.
-#
-# Redistributions in binary form must reproduce the above copyright notice,
-# this list of conditions and the following disclaimer in the documentation
-# and/or other materials provided with the distribution.
-#
-# THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS"
-# AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO,
-# THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
-# PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR
-# CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
-# EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
-# PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS;
-# OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY,
-# WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR
-# OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE,
-# EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
-# *****************************************************************************
-
-# Test installation of dpctl and dpnp
-import dpctl
-import dpnp as np
-
-try:
- gpu = dpctl.select_gpu_device() # Select gpu device
- print (gpu.print_device_info()) # Print gpu information
-
- x = np.array([1, 2, 3], device = gpu) # Create an array on gpu device
- print ("Array x is located on the device:", x.device)
- print ("Array x is:", x)
-
- y = np.sum(x) # Calculate sum of array elements
-
- print ("Result y is located on the device:", y.device)
- print ("y = ", y) # Expect 6
- print ("Test passed")
-except:
- print ("GPU device is not available")
- print ("Test failed")
\ No newline at end of file
diff --git a/examples/06-dpctl_dpnp_test.py b/examples/06-dpctl_dpnp_test.py
deleted file mode 100644
index c41105d..0000000
--- a/examples/06-dpctl_dpnp_test.py
+++ /dev/null
@@ -1,41 +0,0 @@
-# *****************************************************************************
-# Copyright (c) 2022, Intel Corporation All rights reserved.
-#
-# Redistribution and use in source and binary forms, with or without
-# modification, are permitted provided that the following conditions are met:
-#
-# Redistributions of source code must retain the above copyright notice,
-# this list of conditions and the following disclaimer.
-#
-# Redistributions in binary form must reproduce the above copyright notice,
-# this list of conditions and the following disclaimer in the documentation
-# and/or other materials provided with the distribution.
-#
-# THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS"
-# AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO,
-# THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
-# PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR
-# CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
-# EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
-# PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS;
-# OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY,
-# WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR
-# OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE,
-# EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
-# *****************************************************************************
-
-# Test installation of dpctl and dpnp
-import dpctl
-import dpnp as np
-
-gpu = dpctl.select_gpu_device() # Select gpu device
-print (gpu.print_device_info()) # Print gpu information
-
-x = np.array([1, 2, 3], device = gpu) # Create an array on gpu device
-print ("Array x is located on the device:", x.device)
-print ("Array x is:", x)
-
-y = np.sum(x) # Calculate sum of array elements
-
-print ("Result y is located on the device:", y.device)
-print ("y = ", y) # Expect 6
diff --git a/examples/06-numba-dpex_test.py b/examples/06-numba-dpex_test.py
deleted file mode 100644
index c33c395..0000000
--- a/examples/06-numba-dpex_test.py
+++ /dev/null
@@ -1,45 +0,0 @@
-# *****************************************************************************
-# Copyright (c) 2022, Intel Corporation All rights reserved.
-#
-# Redistribution and use in source and binary forms, with or without
-# modification, are permitted provided that the following conditions are met:
-#
-# Redistributions of source code must retain the above copyright notice,
-# this list of conditions and the following disclaimer.
-#
-# Redistributions in binary form must reproduce the above copyright notice,
-# this list of conditions and the following disclaimer in the documentation
-# and/or other materials provided with the distribution.
-#
-# THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS"
-# AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO,
-# THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
-# PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR
-# CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
-# EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
-# PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS;
-# OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY,
-# WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR
-# OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE,
-# EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
-# *****************************************************************************
-
-# Test installation of numba_dpex
-from numba import njit
-import numba_dpex
-import numpy as np
-
-@njit
-def test(a, b, c):
- return a*b+c
-
-
-
-a = np.array(np.random.random(10), dtype=np.float32)
-b = np.array(np.random.random(10), dtype=np.float32)
-c = np.array(np.random.random(10), dtype=np.float32)
-print ("Array a is", a)
-print ("Array b is", b)
-print ("Array c is", c)
-res = test(a, b, c)
-print ("Rasult is", res)
\ No newline at end of file
diff --git a/examples/jupyter_examples.ipynb b/examples/jupyter_examples.ipynb
deleted file mode 100644
index cb5c937..0000000
--- a/examples/jupyter_examples.ipynb
+++ /dev/null
@@ -1,1645 +0,0 @@
-{
- "cells": [
- {
- "cell_type": "markdown",
- "id": "905c742a",
- "metadata": {},
- "source": [
- "# Few Line Changes to Run on GPU"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "41bf2099",
- "metadata": {},
- "source": [
- "Let see the example how you can easily in a few lines of code switch calculations from CPU to GPU device.\n",
- "\n",
- "Please look on the original example.\n",
- "We call the NumPy library and allocate 2 arrays on the Host (CPU) device, all future calculations will be performed as well on the allocated Host(CPU) device."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 4,
- "id": "d9e8711d",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "1 µs ± 8.64 ns per loop (mean ± std. dev. of 7 runs, 1,000,000 loops each)\n",
- "res = [[2 2]\n",
- " [2 2]]\n"
- ]
- }
- ],
- "source": [
- "# Original CPU script\n",
- "\n",
- "# Call numpy library\n",
- "import numpy as np\n",
- "\n",
- "# Data alocated on the CPU device\n",
- "x = np.array ([[1, 1], [1, 1]])\n",
- "y = np.array ([[1, 1], [1, 1]])\n",
- "\n",
- "# Compute performed on the CPU device, where data is allocated\n",
- "%timeit res = np.matmul(x, y)\n",
- "\n",
- "print (\"res = \", res)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "402c3d61",
- "metadata": {},
- "source": [
- "Now let's try to modify our code in a way when all calculations occur on the GPU device.\n",
- "To do it, you need just to switch to the dpnp library and see on the result."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 15,
- "id": "c10b7d83",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "201 µs ± 13.7 µs per loop (mean ± std. dev. of 7 runs, 1,000 loops each)\n",
- "Array x is located on the device: Device(level_zero:gpu:0)\n",
- "Array y is located on the device: Device(level_zero:gpu:0)\n",
- "res is located on the device: Device(level_zero:gpu:0)\n",
- "res = [[2 2]\n",
- " [2 2]]\n"
- ]
- }
- ],
- "source": [
- "# Modified XPU script\n",
- "\n",
- "# Drop-in replacement via single line change\n",
- "import dpnp as np\n",
- "\n",
- "# Data alocated on default SYCL device\n",
- "x = np.array([[1, 1], [1, 1]])\n",
- "y = np.array([[1, 1], [1, 1]])\n",
- "\n",
- "# Compute performed on the device, where data is allocated\n",
- "%timeit res = np.matmul(x, y)\n",
- "\n",
- "\n",
- "print (\"Array x is located on the device:\", x.device)\n",
- "print (\"Array y is located on the device:\", y.device)\n",
- "print (\"res is located on the device:\", res.device)\n",
- "print (\"res = \", res)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "d6a58f17",
- "metadata": {},
- "source": [
- "As you may see changing only one line of code help us to perform all calculations on the GPU device.\n",
- "In this example np.array() creates an array on the default SYCL* device, which is \"gpu\" on systems with integrated or discrete GPU (it is \"host\" on systems that do not have GPU). The queue associated with this array is now carried with x and y, and np.matmul(x, y) will do matrix product of two arrays x and y, and respective pre-compiled kernel implementing np.matmul() will be submitted to that queue. The result res will be allocated on the device array associated with that queue too.\n",
- "\n",
- "Now let's make a few improvements in our code and see how we can control and specify exact device on which we want to perform our calculations and which USM memory type to use."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 6,
- "id": "65bb686b",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Array x is located on the device: Device(level_zero:gpu:0)\n",
- "Array y is located on the device: Device(level_zero:gpu:0)\n",
- "res is located on the device: Device(level_zero:gpu:0) (2, 2)\n",
- "res = [[2 2]\n",
- " [2 2]]\n"
- ]
- }
- ],
- "source": [
- "# Modified XPU script\n",
- "import dpnp as np\n",
- "\n",
- "# Using filter selector strings one can finely specify root devices. USM memory type is default(device)\n",
- "x = np.array([[1, 1], [1, 1]], device = \"level_zero:gpu:0\")\n",
- "# Default device (Level0, GPU0), USM memory type explicitely set to shared\n",
- "y = np.array([[1, 1],[1, 1]], usm_type = \"shared\")\n",
- "\n",
- "%timeit res = np.matmul(x, y)\n",
- "\n",
- "print (\"Array x is located on the device:\", x.device)\n",
- "print (\"Array y is located on the device:\", y.device)\n",
- "print (\"res is located on the device:\", res.device)\n",
- "print (\"res = \", res)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "f3c9afeb",
- "metadata": {},
- "source": [
- "All dpnp array creation routines as well as random number generators have additional optional keyword arguments device, queue, and usm_type, using which you can explicitly specify on which device or queue you want the tensor data to be created along with USM memory type to be used (\"host\", \"device\", or \"shared\"). \n",
- "In the above example we create the array x on the GPU device and for the array y we use shared USM memory type (It means that location is both host and device (copies are synchronized by underlying software), accessible by the host or device.) and perform our calculations on the gpu device."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "be340585",
- "metadata": {},
- "source": [
- "# dpnp simple examples with popular functions"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "9008dbe5",
- "metadata": {},
- "source": [
- "1. Example of the usage of the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` environment variable and finding specific values in array based on condition using dpnp library"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "5d4300fe",
- "metadata": {},
- "source": [
- "Not all functions were yet implemented in the dpnp library, some of the functions require enabling of the direct fallback to the NumPy library. \n",
- "One of the example can be the function \"dpnp.where ()\". \n",
- "Let's look on the example where we want to select certain values from an available array and find max, mean, and calculate the absolute value of the certain values."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 11,
- "id": "c78511cd",
- "metadata": {},
- "outputs": [
- {
- "ename": "NotImplementedError",
- "evalue": "Requested funtion=less_equal with args=(, 3.0) and kwargs={} isn't currently supported and would fall back on NumPy implementation. Define enviroment variable `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` to `0` if the fall back is required to be supported without rasing an exception.",
- "output_type": "error",
- "traceback": [
- "\u001b[1;31m---------------------------------------------------------------------------\u001b[0m",
- "\u001b[1;31mNotImplementedError\u001b[0m Traceback (most recent call last)",
- "Input \u001b[1;32mIn [11]\u001b[0m, in \u001b[0;36m\u001b[1;34m()\u001b[0m\n\u001b[0;32m 4\u001b[0m a \u001b[38;5;241m=\u001b[39m np\u001b[38;5;241m.\u001b[39marray([\u001b[38;5;241m1.0\u001b[39m, \u001b[38;5;241m0.5\u001b[39m, \u001b[38;5;241m3.0\u001b[39m, \u001b[38;5;241m2.0\u001b[39m, \u001b[38;5;241m5.0\u001b[39m, \u001b[38;5;241m6.0\u001b[39m])\n\u001b[0;32m 6\u001b[0m \u001b[38;5;66;03m# Select all variants which are lower or equal 3.0\u001b[39;00m\n\u001b[1;32m----> 7\u001b[0m b \u001b[38;5;241m=\u001b[39m np\u001b[38;5;241m.\u001b[39mwhere(\u001b[43ma\u001b[49m\u001b[43m \u001b[49m\u001b[38;5;241;43m<\u001b[39;49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43m \u001b[49m\u001b[38;5;241;43m3.0\u001b[39;49m)\n\u001b[0;32m 9\u001b[0m \u001b[38;5;28mprint\u001b[39m (\u001b[38;5;124m\"\u001b[39m\u001b[38;5;124mArray a is located on the device:\u001b[39m\u001b[38;5;124m\"\u001b[39m, a\u001b[38;5;241m.\u001b[39mdevice)\n\u001b[0;32m 10\u001b[0m \u001b[38;5;28mprint\u001b[39m (\u001b[38;5;124m\"\u001b[39m\u001b[38;5;124mResult b is located on the device:\u001b[39m\u001b[38;5;124m\"\u001b[39m, b\u001b[38;5;241m.\u001b[39mdevice)\n",
- "File \u001b[1;32m~\\Anaconda3\\envs\\my_env\\lib\\site-packages\\dpnp\\dpnp_array.py:207\u001b[0m, in \u001b[0;36mdpnp_array.__le__\u001b[1;34m(self, other)\u001b[0m\n\u001b[0;32m 206\u001b[0m \u001b[38;5;28;01mdef\u001b[39;00m \u001b[38;5;21m__le__\u001b[39m(\u001b[38;5;28mself\u001b[39m, other):\n\u001b[1;32m--> 207\u001b[0m \u001b[38;5;28;01mreturn\u001b[39;00m \u001b[43mdpnp\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mless_equal\u001b[49m\u001b[43m(\u001b[49m\u001b[38;5;28;43mself\u001b[39;49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mother\u001b[49m\u001b[43m)\u001b[49m\n",
- "File \u001b[1;32m~\\Anaconda3\\envs\\my_env\\lib\\site-packages\\dpnp\\dpnp_iface_logic.py:620\u001b[0m, in \u001b[0;36mless_equal\u001b[1;34m(x1, x2)\u001b[0m\n\u001b[0;32m 580\u001b[0m \u001b[38;5;124;03m\"\"\"\u001b[39;00m\n\u001b[0;32m 581\u001b[0m \u001b[38;5;124;03mReturn (x1 <= x2) element-wise.\u001b[39;00m\n\u001b[0;32m 582\u001b[0m \n\u001b[1;32m (...)\u001b[0m\n\u001b[0;32m 607\u001b[0m \n\u001b[0;32m 608\u001b[0m \u001b[38;5;124;03m\"\"\"\u001b[39;00m\n\u001b[0;32m 610\u001b[0m \u001b[38;5;66;03m# x1_desc = dpnp.get_dpnp_descriptor(x1)\u001b[39;00m\n\u001b[0;32m 611\u001b[0m \u001b[38;5;66;03m# x2_desc = dpnp.get_dpnp_descriptor(x2)\u001b[39;00m\n\u001b[0;32m 612\u001b[0m \u001b[38;5;66;03m# if x1_desc and x2_desc:\u001b[39;00m\n\u001b[1;32m (...)\u001b[0m\n\u001b[0;32m 617\u001b[0m \u001b[38;5;66;03m# else:\u001b[39;00m\n\u001b[0;32m 618\u001b[0m \u001b[38;5;66;03m# return dpnp_less_equal(x1_desc, x2_desc).get_pyobj()\u001b[39;00m\n\u001b[1;32m--> 620\u001b[0m \u001b[38;5;28;01mreturn\u001b[39;00m \u001b[43mcall_origin\u001b[49m\u001b[43m(\u001b[49m\u001b[43mnumpy\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mless_equal\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mx1\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mx2\u001b[49m\u001b[43m)\u001b[49m\n",
- "File \u001b[1;32mdpnp\\dpnp_utils\\dpnp_algo_utils.pyx:120\u001b[0m, in \u001b[0;36mdpnp.dpnp_utils.dpnp_algo_utils.call_origin\u001b[1;34m()\u001b[0m\n",
- "\u001b[1;31mNotImplementedError\u001b[0m: Requested funtion=less_equal with args=(, 3.0) and kwargs={} isn't currently supported and would fall back on NumPy implementation. Define enviroment variable `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` to `0` if the fall back is required to be supported without rasing an exception."
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "# Create an array\n",
- "a = np.array([1.0, 0.5, 3.0, 2.0, 5.0, 6.0])\n",
- "\n",
- "# Select all variants which are lower or equal 3.0\n",
- "b = np.where(a <= 3.0)\n",
- "\n",
- "print (\"Array a is located on the device:\", a.device)\n",
- "print (\"Result b is located on the device:\", b.device)\n",
- "print (\"np.max(b) = \", np.max(b), \"np.mean(b) = \", np.mean(b), \"np.abs(b) = \", np.abs(b))"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "ae026021",
- "metadata": {},
- "source": [
- "As you can see the function \"dpnp.where ()\" is not yet implemented in the dpnp library.\n",
- "We got the following error message: \"Function isn't currently supported and would fall back on NumPy implementation. Define environment variable `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` to `0` if the fall back is required to be supported without raising an exception.\" \n",
- "By default the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK` environment variable is not null and it allows to use only dpnp library. \n",
- "If we want to call NumPy library for functions not supported under dpnp, we need to change this environment variable to `0`."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "210440c3",
- "metadata": {},
- "source": [
- "Let's overwrite the same example using the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK = 0` environment variable."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "c5a6f449",
- "metadata": {},
- "source": [
- "`Note:` Please pay attention that if you are working in the Jupyter Notebook, before running the example with setting the `DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK = 0` you need to restart the kernel in the Jupyter Notebook. "
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 2,
- "id": "05ad1565",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK = 0\n"
- ]
- },
- {
- "ename": "TypeError",
- "evalue": "can only concatenate str (not \"numpy.dtype[object_]\") to str",
- "output_type": "error",
- "traceback": [
- "\u001b[1;31m---------------------------------------------------------------------------\u001b[0m",
- "\u001b[1;31mTypeError\u001b[0m Traceback (most recent call last)",
- "Input \u001b[1;32mIn [2]\u001b[0m, in \u001b[0;36m\u001b[1;34m()\u001b[0m\n\u001b[0;32m 11\u001b[0m a \u001b[38;5;241m=\u001b[39m np\u001b[38;5;241m.\u001b[39marray([\u001b[38;5;241m1.0\u001b[39m, \u001b[38;5;241m0.5\u001b[39m, \u001b[38;5;241m3.0\u001b[39m, \u001b[38;5;241m2.0\u001b[39m, \u001b[38;5;241m5.0\u001b[39m, \u001b[38;5;241m6.0\u001b[39m])\n\u001b[0;32m 13\u001b[0m \u001b[38;5;66;03m# Select all variants which are lower or equal 3.0\u001b[39;00m\n\u001b[1;32m---> 14\u001b[0m b \u001b[38;5;241m=\u001b[39m \u001b[43mnp\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mwhere\u001b[49m\u001b[43m(\u001b[49m\u001b[43ma\u001b[49m\u001b[43m \u001b[49m\u001b[38;5;241;43m<\u001b[39;49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43m \u001b[49m\u001b[38;5;241;43m3.0\u001b[39;49m\u001b[43m)\u001b[49m\n\u001b[0;32m 16\u001b[0m \u001b[38;5;28mprint\u001b[39m (\u001b[38;5;124m\"\u001b[39m\u001b[38;5;124mArray a is located on the device:\u001b[39m\u001b[38;5;124m\"\u001b[39m, a\u001b[38;5;241m.\u001b[39mdevice)\n\u001b[0;32m 17\u001b[0m \u001b[38;5;28mprint\u001b[39m (\u001b[38;5;124m\"\u001b[39m\u001b[38;5;124mResult b is located on the device:\u001b[39m\u001b[38;5;124m\"\u001b[39m, b\u001b[38;5;241m.\u001b[39mdevice)\n",
- "File \u001b[1;32m~\\Anaconda3\\envs\\my_env\\lib\\site-packages\\dpnp\\dpnp_iface_searching.py:187\u001b[0m, in \u001b[0;36mwhere\u001b[1;34m(condition, x, y)\u001b[0m\n\u001b[0;32m 179\u001b[0m \u001b[38;5;28;01mdef\u001b[39;00m \u001b[38;5;21mwhere\u001b[39m(condition, x\u001b[38;5;241m=\u001b[39m\u001b[38;5;28;01mNone\u001b[39;00m, y\u001b[38;5;241m=\u001b[39m\u001b[38;5;28;01mNone\u001b[39;00m):\n\u001b[0;32m 180\u001b[0m \u001b[38;5;124;03m\"\"\"\u001b[39;00m\n\u001b[0;32m 181\u001b[0m \u001b[38;5;124;03m Find indices where elements should be inserted to maintain order.\u001b[39;00m\n\u001b[0;32m 182\u001b[0m \n\u001b[0;32m 183\u001b[0m \u001b[38;5;124;03m For full documentation refer to :obj:`numpy.searchsorted`.\u001b[39;00m\n\u001b[0;32m 184\u001b[0m \n\u001b[0;32m 185\u001b[0m \u001b[38;5;124;03m \"\"\"\u001b[39;00m\n\u001b[1;32m--> 187\u001b[0m \u001b[38;5;28;01mreturn\u001b[39;00m \u001b[43mcall_origin\u001b[49m\u001b[43m(\u001b[49m\u001b[43mnumpy\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mwhere\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mcondition\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mx\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43my\u001b[49m\u001b[43m)\u001b[49m\n",
- "File \u001b[1;32mdpnp\\dpnp_utils\\dpnp_algo_utils.pyx:179\u001b[0m, in \u001b[0;36mdpnp.dpnp_utils.dpnp_algo_utils.call_origin\u001b[1;34m()\u001b[0m\n",
- "File \u001b[1;32m~\\Anaconda3\\envs\\my_env\\lib\\site-packages\\dpnp\\dpnp_container.py:109\u001b[0m, in \u001b[0;36mempty\u001b[1;34m(shape, dtype, order, device, usm_type, sycl_queue)\u001b[0m\n\u001b[0;32m 106\u001b[0m sycl_queue_normalized \u001b[38;5;241m=\u001b[39m dpnp\u001b[38;5;241m.\u001b[39mget_normalized_queue_device(device\u001b[38;5;241m=\u001b[39mdevice, sycl_queue\u001b[38;5;241m=\u001b[39msycl_queue)\n\u001b[0;32m 108\u001b[0m \u001b[38;5;124;03m\"\"\"Creates `dpnp_array` from uninitialized USM allocation.\"\"\"\u001b[39;00m\n\u001b[1;32m--> 109\u001b[0m array_obj \u001b[38;5;241m=\u001b[39m \u001b[43mdpt\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43mempty\u001b[49m\u001b[43m(\u001b[49m\u001b[43mshape\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 110\u001b[0m \u001b[43m \u001b[49m\u001b[43mdtype\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43mdtype\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 111\u001b[0m \u001b[43m \u001b[49m\u001b[43morder\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43morder\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 112\u001b[0m \u001b[43m \u001b[49m\u001b[43musm_type\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43musm_type\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 113\u001b[0m \u001b[43m \u001b[49m\u001b[43msycl_queue\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43msycl_queue_normalized\u001b[49m\u001b[43m)\u001b[49m\n\u001b[0;32m 114\u001b[0m \u001b[38;5;28;01mreturn\u001b[39;00m dpnp_array(array_obj\u001b[38;5;241m.\u001b[39mshape, buffer\u001b[38;5;241m=\u001b[39marray_obj, order\u001b[38;5;241m=\u001b[39morder)\n",
- "File \u001b[1;32m~\\Anaconda3\\envs\\my_env\\lib\\site-packages\\dpctl\\tensor\\_ctors.py:446\u001b[0m, in \u001b[0;36mempty\u001b[1;34m(sh, dtype, order, device, usm_type, sycl_queue)\u001b[0m\n\u001b[0;32m 444\u001b[0m sycl_queue \u001b[38;5;241m=\u001b[39m normalize_queue_device(sycl_queue\u001b[38;5;241m=\u001b[39msycl_queue, device\u001b[38;5;241m=\u001b[39mdevice)\n\u001b[0;32m 445\u001b[0m dtype \u001b[38;5;241m=\u001b[39m _get_dtype(dtype, sycl_queue)\n\u001b[1;32m--> 446\u001b[0m res \u001b[38;5;241m=\u001b[39m \u001b[43mdpt\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43musm_ndarray\u001b[49m\u001b[43m(\u001b[49m\n\u001b[0;32m 447\u001b[0m \u001b[43m \u001b[49m\u001b[43msh\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 448\u001b[0m \u001b[43m \u001b[49m\u001b[43mdtype\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43mdtype\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 449\u001b[0m \u001b[43m \u001b[49m\u001b[43mbuffer\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43musm_type\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 450\u001b[0m \u001b[43m \u001b[49m\u001b[43morder\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43morder\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 451\u001b[0m \u001b[43m \u001b[49m\u001b[43mbuffer_ctor_kwargs\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43m{\u001b[49m\u001b[38;5;124;43m\"\u001b[39;49m\u001b[38;5;124;43mqueue\u001b[39;49m\u001b[38;5;124;43m\"\u001b[39;49m\u001b[43m:\u001b[49m\u001b[43m \u001b[49m\u001b[43msycl_queue\u001b[49m\u001b[43m}\u001b[49m\u001b[43m,\u001b[49m\n\u001b[0;32m 452\u001b[0m \u001b[43m\u001b[49m\u001b[43m)\u001b[49m\n\u001b[0;32m 453\u001b[0m \u001b[38;5;28;01mreturn\u001b[39;00m res\n",
- "File \u001b[1;32mdpctl\\tensor\\_usmarray.pyx:195\u001b[0m, in \u001b[0;36mdpctl.tensor._usmarray.usm_ndarray.__cinit__\u001b[1;34m()\u001b[0m\n",
- "\u001b[1;31mTypeError\u001b[0m: can only concatenate str (not \"numpy.dtype[object_]\") to str"
- ]
- }
- ],
- "source": [
- "import os\n",
- "# call numpy if not null than we will use dpnp, by default not null\n",
- "os.environ[\"DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK\"] = \"0\" \n",
- "\n",
- "import dpnp as np\n",
- "\n",
- "# Expect result 0\n",
- "print (\"DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK =\", np.config.__DPNP_RAISE_EXCEPION_ON_NUMPY_FALLBACK__) \n",
- "\n",
- "# Create an array\n",
- "a = np.array([1.0, 0.5, 3.0, 2.0, 5.0, 6.0])\n",
- "\n",
- "# Select all variants which are lower or equal 3.0\n",
- "b = np.where(a <= 3.0)\n",
- "\n",
- "print (\"Array a is located on the device:\", a.device)\n",
- "print (\"Result b is located on the device:\", b.device)\n",
- "print (\"np.max(b) = \", np.max(b), \"np.mean(b) = \", np.mean(b), \"np.abs(b) = \", np.abs(b))"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "26d8c3c6",
- "metadata": {},
- "source": [
- "2. Example to return an array with evenly spaced values within a given interval."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 35,
- "id": "0435d7f0",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "282 µs ± 27.6 µs per loop (mean ± std. dev. of 7 runs, 10,000 loops each)\n",
- "Result a is located on the device: Device(level_zero:gpu:0)\n",
- "a = [ 3 9 15 21 27]\n"
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "# Create an array of values from 3 till 30 with step 6\n",
- "%timeit a = np.arange(3, 30, step = 6)\n",
- "\n",
- "print (\"Result a is located on the device:\", a.device)\n",
- "print (\"a = \", a)"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "f8a6eb78",
- "metadata": {},
- "source": [
- "In this example np.arange() creates an array on the default SYCL* device, which is \"gpu\" on systems with integrated or discrete GPU (it is \"host\" on systems that do not have GPU)."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "35081461",
- "metadata": {},
- "source": [
- "3. Example which calculates on the GPU the sum of the array elements"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 3,
- "id": "e613398f",
- "metadata": {
- "scrolled": true
- },
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Result x is located on the device: Device(level_zero:gpu:0)\n",
- "Result y is located on the device: Device(level_zero:gpu:0)\n",
- "The sum of the array elements is: 6\n"
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "x = np.empty(3)\n",
- "\n",
- "try:\n",
- " # Using filter selector strings to specify root devices for a new array\n",
- " x = np.asarray ([1, 2, 3], device=\"gpu\")\n",
- " print (\"Result x is located on the device:\", x.device)\n",
- "except:\n",
- " print (\"GPU device is not available\")\n",
- "\n",
- "# Return the sum of the array elements\n",
- "y = np.sum (x) # Expect 6\n",
- "\n",
- "print (\"Result y is located on the device:\", y.device)\n",
- "print (\"The sum of the array elements is: \", y )"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "ece0d3f6",
- "metadata": {},
- "source": [
- "In this example np.asarray() creates an array on the selected device - \"gpu\". The queue associated with this array is now carried with x, and np.sum(x) will derive it from x, and respective pre-compiled kernel implementing np.sum() will be submitted to that queue. The result y will be allocated on the device 0-dimensional array associated with that queue too."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "b49f4c62",
- "metadata": {},
- "source": [
- "4. Example of inversion of an array"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 1,
- "id": "4b53afed",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Array a is located on the device: Device(level_zero:gpu:0)\n",
- "Result x is located on the device: Device(level_zero:gpu:0)\n",
- "Array x is: [[-2 -2]\n",
- " [-3 -2]\n",
- " [-2 -1]\n",
- " [ 0 -1]]\n"
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "try:\n",
- " \n",
- " # Using filter selector strings to specify root devices for an array\n",
- " a = np.array([[1, 1], [2, 1], [1, 0], [-1, 0]], device = \"gpu\")\n",
- " print (\"Array a is located on the device:\", a.device) \n",
- "\n",
- " # Do inversion of an array \"a\"\n",
- " x = np.invert(a)\n",
- "\n",
- " print (\"Result x is located on the device:\", x.device)\n",
- " print (\"Array x is:\", x) \n",
- "\n",
- "except:\n",
- " print (\"GPU device is not available\")\n"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "07029d53",
- "metadata": {},
- "source": [
- "In this example np.array() creates an array on the selected device - \"gpu\". The queue associated with this array is now carried with a, and np.invert(a) will derive it from a, and respective pre-compiled kernel implementing np.invert() will be submitted to that queue. The result x will be allocated on the device array associated with that queue too."
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 12,
- "id": "d55bf7e8",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "Result a is located on the device: Device(level_zero:gpu:0)\n",
- "Result x is located on the device: Device(level_zero:gpu:0)\n",
- "The array ´a` with axes transposed is equal to: [[1 1 0 1]\n",
- " [2 1 2 1]]\n"
- ]
- }
- ],
- "source": [
- "import dpnp as np\n",
- "\n",
- "try:\n",
- " # Using filter selector strings to specify root devices for an array\n",
- " a = np.array ([[1, 2],[1, 1],[0, 2],[1, 1]], device = \"gpu\")\n",
- " print (\"Result a is located on the device:\", a.device)\n",
- "\n",
- " # Returns a view of the array \"a\" with axes transposed.\n",
- " x = np.transpose(a)\n",
- "\n",
- " print (\"Result x is located on the device:\", x.device)\n",
- " print (\"The array ´a` with axes transposed is equal to:\", x) \n",
- " \n",
- "except:\n",
- " print(\"GPU device is not available\")"
- ]
- },
- {
- "cell_type": "markdown",
- "id": "4399f5c7",
- "metadata": {},
- "source": [
- "In this example np.array() creates an array on the selected device - \"gpu\". The queue associated with this array is now carried with a, and np.transpose(a) will derive it from a, and respective pre-compiled kernel implementing np.transpose() will be submitted to that queue. The result x will be allocated on the device array associated with that queue too."
- ]
- },
- {
- "cell_type": "markdown",
- "id": "c7c0429e",
- "metadata": {},
- "source": [
- "5. Example which creates a function f(x) = sin(x) in NumPy and dpnp "
- ]
- },
- {
- "cell_type": "markdown",
- "id": "ee1d3dd7",
- "metadata": {},
- "source": [
- "`Prerequisites:` Please install matplotlib library to execute next examples"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 13,
- "id": "9d5ba0e0",
- "metadata": {},
- "outputs": [
- {
- "name": "stdout",
- "output_type": "stream",
- "text": [
- "16.1 µs ± 1.03 µs per loop (mean ± std. dev. of 7 runs, 100,000 loops each)\n"
- ]
- },
- {
- "data": {
- "application/javascript": [
- "/* Put everything inside the global mpl namespace */\n",
- "window.mpl = {};\n",
- "\n",
- "\n",
- "mpl.get_websocket_type = function() {\n",
- " if (typeof(WebSocket) !== 'undefined') {\n",
- " return WebSocket;\n",
- " } else if (typeof(MozWebSocket) !== 'undefined') {\n",
- " return MozWebSocket;\n",
- " } else {\n",
- " alert('Your browser does not have WebSocket support. ' +\n",
- " 'Please try Chrome, Safari or Firefox ≥ 6. ' +\n",
- " 'Firefox 4 and 5 are also supported but you ' +\n",
- " 'have to enable WebSockets in about:config.');\n",
- " };\n",
- "}\n",
- "\n",
- "mpl.figure = function(figure_id, websocket, ondownload, parent_element) {\n",
- " this.id = figure_id;\n",
- "\n",
- " this.ws = websocket;\n",
- "\n",
- " this.supports_binary = (this.ws.binaryType != undefined);\n",
- "\n",
- " if (!this.supports_binary) {\n",
- " var warnings = document.getElementById(\"mpl-warnings\");\n",
- " if (warnings) {\n",
- " warnings.style.display = 'block';\n",
- " warnings.textContent = (\n",
- " \"This browser does not support binary websocket messages. \" +\n",
- " \"Performance may be slow.\");\n",
- " }\n",
- " }\n",
- "\n",
- " this.imageObj = new Image();\n",
- "\n",
- " this.context = undefined;\n",
- " this.message = undefined;\n",
- " this.canvas = undefined;\n",
- " this.rubberband_canvas = undefined;\n",
- " this.rubberband_context = undefined;\n",
- " this.format_dropdown = undefined;\n",
- "\n",
- " this.image_mode = 'full';\n",
- "\n",
- " this.root = $('');\n",
- " this._root_extra_style(this.root)\n",
- " this.root.attr('style', 'display: inline-block');\n",
- "\n",
- " $(parent_element).append(this.root);\n",
- "\n",
- " this._init_header(this);\n",
- " this._init_canvas(this);\n",
- " this._init_toolbar(this);\n",
- "\n",
- " var fig = this;\n",
- "\n",
- " this.waiting = false;\n",
- "\n",
- " this.ws.onopen = function () {\n",
- " fig.send_message(\"supports_binary\", {value: fig.supports_binary});\n",
- " fig.send_message(\"send_image_mode\", {});\n",
- " if (mpl.ratio != 1) {\n",
- " fig.send_message(\"set_dpi_ratio\", {'dpi_ratio': mpl.ratio});\n",
- " }\n",
- " fig.send_message(\"refresh\", {});\n",
- " }\n",
- "\n",
- " this.imageObj.onload = function() {\n",
- " if (fig.image_mode == 'full') {\n",
- " // Full images could contain transparency (where diff images\n",
- " // almost always do), so we need to clear the canvas so that\n",
- " // there is no ghosting.\n",
- " fig.context.clearRect(0, 0, fig.canvas.width, fig.canvas.height);\n",
- " }\n",
- " fig.context.drawImage(fig.imageObj, 0, 0);\n",
- " };\n",
- "\n",
- " this.imageObj.onunload = function() {\n",
- " fig.ws.close();\n",
- " }\n",
- "\n",
- " this.ws.onmessage = this._make_on_message_function(this);\n",
- "\n",
- " this.ondownload = ondownload;\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._init_header = function() {\n",
- " var titlebar = $(\n",
- " '');\n",
- " var titletext = $(\n",
- " '');\n",
- " titlebar.append(titletext)\n",
- " this.root.append(titlebar);\n",
- " this.header = titletext[0];\n",
- "}\n",
- "\n",
- "\n",
- "\n",
- "mpl.figure.prototype._canvas_extra_style = function(canvas_div) {\n",
- "\n",
- "}\n",
- "\n",
- "\n",
- "mpl.figure.prototype._root_extra_style = function(canvas_div) {\n",
- "\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._init_canvas = function() {\n",
- " var fig = this;\n",
- "\n",
- " var canvas_div = $('');\n",
- "\n",
- " canvas_div.attr('style', 'position: relative; clear: both; outline: 0');\n",
- "\n",
- " function canvas_keyboard_event(event) {\n",
- " return fig.key_event(event, event['data']);\n",
- " }\n",
- "\n",
- " canvas_div.keydown('key_press', canvas_keyboard_event);\n",
- " canvas_div.keyup('key_release', canvas_keyboard_event);\n",
- " this.canvas_div = canvas_div\n",
- " this._canvas_extra_style(canvas_div)\n",
- " this.root.append(canvas_div);\n",
- "\n",
- " var canvas = $('');\n",
- " canvas.addClass('mpl-canvas');\n",
- " canvas.attr('style', \"left: 0; top: 0; z-index: 0; outline: 0\")\n",
- "\n",
- " this.canvas = canvas[0];\n",
- " this.context = canvas[0].getContext(\"2d\");\n",
- "\n",
- " var backingStore = this.context.backingStorePixelRatio ||\n",
- "\tthis.context.webkitBackingStorePixelRatio ||\n",
- "\tthis.context.mozBackingStorePixelRatio ||\n",
- "\tthis.context.msBackingStorePixelRatio ||\n",
- "\tthis.context.oBackingStorePixelRatio ||\n",
- "\tthis.context.backingStorePixelRatio || 1;\n",
- "\n",
- " mpl.ratio = (window.devicePixelRatio || 1) / backingStore;\n",
- "\n",
- " var rubberband = $('');\n",
- " rubberband.attr('style', \"position: absolute; left: 0; top: 0; z-index: 1;\")\n",
- "\n",
- " var pass_mouse_events = true;\n",
- "\n",
- " canvas_div.resizable({\n",
- " start: function(event, ui) {\n",
- " pass_mouse_events = false;\n",
- " },\n",
- " resize: function(event, ui) {\n",
- " fig.request_resize(ui.size.width, ui.size.height);\n",
- " },\n",
- " stop: function(event, ui) {\n",
- " pass_mouse_events = true;\n",
- " fig.request_resize(ui.size.width, ui.size.height);\n",
- " },\n",
- " });\n",
- "\n",
- " function mouse_event_fn(event) {\n",
- " if (pass_mouse_events)\n",
- " return fig.mouse_event(event, event['data']);\n",
- " }\n",
- "\n",
- " rubberband.mousedown('button_press', mouse_event_fn);\n",
- " rubberband.mouseup('button_release', mouse_event_fn);\n",
- " // Throttle sequential mouse events to 1 every 20ms.\n",
- " rubberband.mousemove('motion_notify', mouse_event_fn);\n",
- "\n",
- " rubberband.mouseenter('figure_enter', mouse_event_fn);\n",
- " rubberband.mouseleave('figure_leave', mouse_event_fn);\n",
- "\n",
- " canvas_div.on(\"wheel\", function (event) {\n",
- " event = event.originalEvent;\n",
- " event['data'] = 'scroll'\n",
- " if (event.deltaY < 0) {\n",
- " event.step = 1;\n",
- " } else {\n",
- " event.step = -1;\n",
- " }\n",
- " mouse_event_fn(event);\n",
- " });\n",
- "\n",
- " canvas_div.append(canvas);\n",
- " canvas_div.append(rubberband);\n",
- "\n",
- " this.rubberband = rubberband;\n",
- " this.rubberband_canvas = rubberband[0];\n",
- " this.rubberband_context = rubberband[0].getContext(\"2d\");\n",
- " this.rubberband_context.strokeStyle = \"#000000\";\n",
- "\n",
- " this._resize_canvas = function(width, height) {\n",
- " // Keep the size of the canvas, canvas container, and rubber band\n",
- " // canvas in synch.\n",
- " canvas_div.css('width', width)\n",
- " canvas_div.css('height', height)\n",
- "\n",
- " canvas.attr('width', width * mpl.ratio);\n",
- " canvas.attr('height', height * mpl.ratio);\n",
- " canvas.attr('style', 'width: ' + width + 'px; height: ' + height + 'px;');\n",
- "\n",
- " rubberband.attr('width', width);\n",
- " rubberband.attr('height', height);\n",
- " }\n",
- "\n",
- " // Set the figure to an initial 600x600px, this will subsequently be updated\n",
- " // upon first draw.\n",
- " this._resize_canvas(600, 600);\n",
- "\n",
- " // Disable right mouse context menu.\n",
- " $(this.rubberband_canvas).bind(\"contextmenu\",function(e){\n",
- " return false;\n",
- " });\n",
- "\n",
- " function set_focus () {\n",
- " canvas.focus();\n",
- " canvas_div.focus();\n",
- " }\n",
- "\n",
- " window.setTimeout(set_focus, 100);\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._init_toolbar = function() {\n",
- " var fig = this;\n",
- "\n",
- " var nav_element = $('');\n",
- " nav_element.attr('style', 'width: 100%');\n",
- " this.root.append(nav_element);\n",
- "\n",
- " // Define a callback function for later on.\n",
- " function toolbar_event(event) {\n",
- " return fig.toolbar_button_onclick(event['data']);\n",
- " }\n",
- " function toolbar_mouse_event(event) {\n",
- " return fig.toolbar_button_onmouseover(event['data']);\n",
- " }\n",
- "\n",
- " for(var toolbar_ind in mpl.toolbar_items) {\n",
- " var name = mpl.toolbar_items[toolbar_ind][0];\n",
- " var tooltip = mpl.toolbar_items[toolbar_ind][1];\n",
- " var image = mpl.toolbar_items[toolbar_ind][2];\n",
- " var method_name = mpl.toolbar_items[toolbar_ind][3];\n",
- "\n",
- " if (!name) {\n",
- " // put a spacer in here.\n",
- " continue;\n",
- " }\n",
- " var button = $('');\n",
- " button.addClass('ui-button ui-widget ui-state-default ui-corner-all ' +\n",
- " 'ui-button-icon-only');\n",
- " button.attr('role', 'button');\n",
- " button.attr('aria-disabled', 'false');\n",
- " button.click(method_name, toolbar_event);\n",
- " button.mouseover(tooltip, toolbar_mouse_event);\n",
- "\n",
- " var icon_img = $('');\n",
- " icon_img.addClass('ui-button-icon-primary ui-icon');\n",
- " icon_img.addClass(image);\n",
- " icon_img.addClass('ui-corner-all');\n",
- "\n",
- " var tooltip_span = $('');\n",
- " tooltip_span.addClass('ui-button-text');\n",
- " tooltip_span.html(tooltip);\n",
- "\n",
- " button.append(icon_img);\n",
- " button.append(tooltip_span);\n",
- "\n",
- " nav_element.append(button);\n",
- " }\n",
- "\n",
- " var fmt_picker_span = $('');\n",
- "\n",
- " var fmt_picker = $('');\n",
- " fmt_picker.addClass('mpl-toolbar-option ui-widget ui-widget-content');\n",
- " fmt_picker_span.append(fmt_picker);\n",
- " nav_element.append(fmt_picker_span);\n",
- " this.format_dropdown = fmt_picker[0];\n",
- "\n",
- " for (var ind in mpl.extensions) {\n",
- " var fmt = mpl.extensions[ind];\n",
- " var option = $(\n",
- " '', {selected: fmt === mpl.default_extension}).html(fmt);\n",
- " fmt_picker.append(option);\n",
- " }\n",
- "\n",
- " // Add hover states to the ui-buttons\n",
- " $( \".ui-button\" ).hover(\n",
- " function() { $(this).addClass(\"ui-state-hover\");},\n",
- " function() { $(this).removeClass(\"ui-state-hover\");}\n",
- " );\n",
- "\n",
- " var status_bar = $('');\n",
- " nav_element.append(status_bar);\n",
- " this.message = status_bar[0];\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.request_resize = function(x_pixels, y_pixels) {\n",
- " // Request matplotlib to resize the figure. Matplotlib will then trigger a resize in the client,\n",
- " // which will in turn request a refresh of the image.\n",
- " this.send_message('resize', {'width': x_pixels, 'height': y_pixels});\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.send_message = function(type, properties) {\n",
- " properties['type'] = type;\n",
- " properties['figure_id'] = this.id;\n",
- " this.ws.send(JSON.stringify(properties));\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.send_draw_message = function() {\n",
- " if (!this.waiting) {\n",
- " this.waiting = true;\n",
- " this.ws.send(JSON.stringify({type: \"draw\", figure_id: this.id}));\n",
- " }\n",
- "}\n",
- "\n",
- "\n",
- "mpl.figure.prototype.handle_save = function(fig, msg) {\n",
- " var format_dropdown = fig.format_dropdown;\n",
- " var format = format_dropdown.options[format_dropdown.selectedIndex].value;\n",
- " fig.ondownload(fig, format);\n",
- "}\n",
- "\n",
- "\n",
- "mpl.figure.prototype.handle_resize = function(fig, msg) {\n",
- " var size = msg['size'];\n",
- " if (size[0] != fig.canvas.width || size[1] != fig.canvas.height) {\n",
- " fig._resize_canvas(size[0], size[1]);\n",
- " fig.send_message(\"refresh\", {});\n",
- " };\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.handle_rubberband = function(fig, msg) {\n",
- " var x0 = msg['x0'] / mpl.ratio;\n",
- " var y0 = (fig.canvas.height - msg['y0']) / mpl.ratio;\n",
- " var x1 = msg['x1'] / mpl.ratio;\n",
- " var y1 = (fig.canvas.height - msg['y1']) / mpl.ratio;\n",
- " x0 = Math.floor(x0) + 0.5;\n",
- " y0 = Math.floor(y0) + 0.5;\n",
- " x1 = Math.floor(x1) + 0.5;\n",
- " y1 = Math.floor(y1) + 0.5;\n",
- " var min_x = Math.min(x0, x1);\n",
- " var min_y = Math.min(y0, y1);\n",
- " var width = Math.abs(x1 - x0);\n",
- " var height = Math.abs(y1 - y0);\n",
- "\n",
- " fig.rubberband_context.clearRect(\n",
- " 0, 0, fig.canvas.width / mpl.ratio, fig.canvas.height / mpl.ratio);\n",
- "\n",
- " fig.rubberband_context.strokeRect(min_x, min_y, width, height);\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.handle_figure_label = function(fig, msg) {\n",
- " // Updates the figure title.\n",
- " fig.header.textContent = msg['label'];\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.handle_cursor = function(fig, msg) {\n",
- " var cursor = msg['cursor'];\n",
- " switch(cursor)\n",
- " {\n",
- " case 0:\n",
- " cursor = 'pointer';\n",
- " break;\n",
- " case 1:\n",
- " cursor = 'default';\n",
- " break;\n",
- " case 2:\n",
- " cursor = 'crosshair';\n",
- " break;\n",
- " case 3:\n",
- " cursor = 'move';\n",
- " break;\n",
- " }\n",
- " fig.rubberband_canvas.style.cursor = cursor;\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.handle_message = function(fig, msg) {\n",
- " fig.message.textContent = msg['message'];\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.handle_draw = function(fig, msg) {\n",
- " // Request the server to send over a new figure.\n",
- " fig.send_draw_message();\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.handle_image_mode = function(fig, msg) {\n",
- " fig.image_mode = msg['mode'];\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.updated_canvas_event = function() {\n",
- " // Called whenever the canvas gets updated.\n",
- " this.send_message(\"ack\", {});\n",
- "}\n",
- "\n",
- "// A function to construct a web socket function for onmessage handling.\n",
- "// Called in the figure constructor.\n",
- "mpl.figure.prototype._make_on_message_function = function(fig) {\n",
- " return function socket_on_message(evt) {\n",
- " if (evt.data instanceof Blob) {\n",
- " /* FIXME: We get \"Resource interpreted as Image but\n",
- " * transferred with MIME type text/plain:\" errors on\n",
- " * Chrome. But how to set the MIME type? It doesn't seem\n",
- " * to be part of the websocket stream */\n",
- " evt.data.type = \"image/png\";\n",
- "\n",
- " /* Free the memory for the previous frames */\n",
- " if (fig.imageObj.src) {\n",
- " (window.URL || window.webkitURL).revokeObjectURL(\n",
- " fig.imageObj.src);\n",
- " }\n",
- "\n",
- " fig.imageObj.src = (window.URL || window.webkitURL).createObjectURL(\n",
- " evt.data);\n",
- " fig.updated_canvas_event();\n",
- " fig.waiting = false;\n",
- " return;\n",
- " }\n",
- " else if (typeof evt.data === 'string' && evt.data.slice(0, 21) == \"data:image/png;base64\") {\n",
- " fig.imageObj.src = evt.data;\n",
- " fig.updated_canvas_event();\n",
- " fig.waiting = false;\n",
- " return;\n",
- " }\n",
- "\n",
- " var msg = JSON.parse(evt.data);\n",
- " var msg_type = msg['type'];\n",
- "\n",
- " // Call the \"handle_{type}\" callback, which takes\n",
- " // the figure and JSON message as its only arguments.\n",
- " try {\n",
- " var callback = fig[\"handle_\" + msg_type];\n",
- " } catch (e) {\n",
- " console.log(\"No handler for the '\" + msg_type + \"' message type: \", msg);\n",
- " return;\n",
- " }\n",
- "\n",
- " if (callback) {\n",
- " try {\n",
- " // console.log(\"Handling '\" + msg_type + \"' message: \", msg);\n",
- " callback(fig, msg);\n",
- " } catch (e) {\n",
- " console.log(\"Exception inside the 'handler_\" + msg_type + \"' callback:\", e, e.stack, msg);\n",
- " }\n",
- " }\n",
- " };\n",
- "}\n",
- "\n",
- "// from http://stackoverflow.com/questions/1114465/getting-mouse-location-in-canvas\n",
- "mpl.findpos = function(e) {\n",
- " //this section is from http://www.quirksmode.org/js/events_properties.html\n",
- " var targ;\n",
- " if (!e)\n",
- " e = window.event;\n",
- " if (e.target)\n",
- " targ = e.target;\n",
- " else if (e.srcElement)\n",
- " targ = e.srcElement;\n",
- " if (targ.nodeType == 3) // defeat Safari bug\n",
- " targ = targ.parentNode;\n",
- "\n",
- " // jQuery normalizes the pageX and pageY\n",
- " // pageX,Y are the mouse positions relative to the document\n",
- " // offset() returns the position of the element relative to the document\n",
- " var x = e.pageX - $(targ).offset().left;\n",
- " var y = e.pageY - $(targ).offset().top;\n",
- "\n",
- " return {\"x\": x, \"y\": y};\n",
- "};\n",
- "\n",
- "/*\n",
- " * return a copy of an object with only non-object keys\n",
- " * we need this to avoid circular references\n",
- " * http://stackoverflow.com/a/24161582/3208463\n",
- " */\n",
- "function simpleKeys (original) {\n",
- " return Object.keys(original).reduce(function (obj, key) {\n",
- " if (typeof original[key] !== 'object')\n",
- " obj[key] = original[key]\n",
- " return obj;\n",
- " }, {});\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.mouse_event = function(event, name) {\n",
- " var canvas_pos = mpl.findpos(event)\n",
- "\n",
- " if (name === 'button_press')\n",
- " {\n",
- " this.canvas.focus();\n",
- " this.canvas_div.focus();\n",
- " }\n",
- "\n",
- " var x = canvas_pos.x * mpl.ratio;\n",
- " var y = canvas_pos.y * mpl.ratio;\n",
- "\n",
- " this.send_message(name, {x: x, y: y, button: event.button,\n",
- " step: event.step,\n",
- " guiEvent: simpleKeys(event)});\n",
- "\n",
- " /* This prevents the web browser from automatically changing to\n",
- " * the text insertion cursor when the button is pressed. We want\n",
- " * to control all of the cursor setting manually through the\n",
- " * 'cursor' event from matplotlib */\n",
- " event.preventDefault();\n",
- " return false;\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._key_event_extra = function(event, name) {\n",
- " // Handle any extra behaviour associated with a key event\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.key_event = function(event, name) {\n",
- "\n",
- " // Prevent repeat events\n",
- " if (name == 'key_press')\n",
- " {\n",
- " if (event.which === this._key)\n",
- " return;\n",
- " else\n",
- " this._key = event.which;\n",
- " }\n",
- " if (name == 'key_release')\n",
- " this._key = null;\n",
- "\n",
- " var value = '';\n",
- " if (event.ctrlKey && event.which != 17)\n",
- " value += \"ctrl+\";\n",
- " if (event.altKey && event.which != 18)\n",
- " value += \"alt+\";\n",
- " if (event.shiftKey && event.which != 16)\n",
- " value += \"shift+\";\n",
- "\n",
- " value += 'k';\n",
- " value += event.which.toString();\n",
- "\n",
- " this._key_event_extra(event, name);\n",
- "\n",
- " this.send_message(name, {key: value,\n",
- " guiEvent: simpleKeys(event)});\n",
- " return false;\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.toolbar_button_onclick = function(name) {\n",
- " if (name == 'download') {\n",
- " this.handle_save(this, null);\n",
- " } else {\n",
- " this.send_message(\"toolbar_button\", {name: name});\n",
- " }\n",
- "};\n",
- "\n",
- "mpl.figure.prototype.toolbar_button_onmouseover = function(tooltip) {\n",
- " this.message.textContent = tooltip;\n",
- "};\n",
- "mpl.toolbar_items = [[\"Home\", \"Reset original view\", \"fa fa-home icon-home\", \"home\"], [\"Back\", \"Back to previous view\", \"fa fa-arrow-left icon-arrow-left\", \"back\"], [\"Forward\", \"Forward to next view\", \"fa fa-arrow-right icon-arrow-right\", \"forward\"], [\"\", \"\", \"\", \"\"], [\"Pan\", \"Pan axes with left mouse, zoom with right\", \"fa fa-arrows icon-move\", \"pan\"], [\"Zoom\", \"Zoom to rectangle\", \"fa fa-square-o icon-check-empty\", \"zoom\"], [\"\", \"\", \"\", \"\"], [\"Download\", \"Download plot\", \"fa fa-floppy-o icon-save\", \"download\"]];\n",
- "\n",
- "mpl.extensions = [\"eps\", \"pdf\", \"png\", \"ps\", \"raw\", \"svg\"];\n",
- "\n",
- "mpl.default_extension = \"png\";var comm_websocket_adapter = function(comm) {\n",
- " // Create a \"websocket\"-like object which calls the given IPython comm\n",
- " // object with the appropriate methods. Currently this is a non binary\n",
- " // socket, so there is still some room for performance tuning.\n",
- " var ws = {};\n",
- "\n",
- " ws.close = function() {\n",
- " comm.close()\n",
- " };\n",
- " ws.send = function(m) {\n",
- " //console.log('sending', m);\n",
- " comm.send(m);\n",
- " };\n",
- " // Register the callback with on_msg.\n",
- " comm.on_msg(function(msg) {\n",
- " //console.log('receiving', msg['content']['data'], msg);\n",
- " // Pass the mpl event to the overridden (by mpl) onmessage function.\n",
- " ws.onmessage(msg['content']['data'])\n",
- " });\n",
- " return ws;\n",
- "}\n",
- "\n",
- "mpl.mpl_figure_comm = function(comm, msg) {\n",
- " // This is the function which gets called when the mpl process\n",
- " // starts-up an IPython Comm through the \"matplotlib\" channel.\n",
- "\n",
- " var id = msg.content.data.id;\n",
- " // Get hold of the div created by the display call when the Comm\n",
- " // socket was opened in Python.\n",
- " var element = $(\"#\" + id);\n",
- " var ws_proxy = comm_websocket_adapter(comm)\n",
- "\n",
- " function ondownload(figure, format) {\n",
- " window.open(figure.imageObj.src);\n",
- " }\n",
- "\n",
- " var fig = new mpl.figure(id, ws_proxy,\n",
- " ondownload,\n",
- " element.get(0));\n",
- "\n",
- " // Call onopen now - mpl needs it, as it is assuming we've passed it a real\n",
- " // web socket which is closed, not our websocket->open comm proxy.\n",
- " ws_proxy.onopen();\n",
- "\n",
- " fig.parent_element = element.get(0);\n",
- " fig.cell_info = mpl.find_output_cell(\"\");\n",
- " if (!fig.cell_info) {\n",
- " console.error(\"Failed to find cell for figure\", id, fig);\n",
- " return;\n",
- " }\n",
- "\n",
- " var output_index = fig.cell_info[2]\n",
- " var cell = fig.cell_info[0];\n",
- "\n",
- "};\n",
- "\n",
- "mpl.figure.prototype.handle_close = function(fig, msg) {\n",
- " var width = fig.canvas.width/mpl.ratio\n",
- " fig.root.unbind('remove')\n",
- "\n",
- " // Update the output cell to use the data from the current canvas.\n",
- " fig.push_to_output();\n",
- " var dataURL = fig.canvas.toDataURL();\n",
- " // Re-enable the keyboard manager in IPython - without this line, in FF,\n",
- " // the notebook keyboard shortcuts fail.\n",
- " IPython.keyboard_manager.enable()\n",
- " $(fig.parent_element).html('');\n",
- " fig.close_ws(fig, msg);\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.close_ws = function(fig, msg){\n",
- " fig.send_message('closing', msg);\n",
- " // fig.ws.close()\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.push_to_output = function(remove_interactive) {\n",
- " // Turn the data on the canvas into data in the output cell.\n",
- " var width = this.canvas.width/mpl.ratio\n",
- " var dataURL = this.canvas.toDataURL();\n",
- " this.cell_info[1]['text/html'] = '';\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.updated_canvas_event = function() {\n",
- " // Tell IPython that the notebook contents must change.\n",
- " IPython.notebook.set_dirty(true);\n",
- " this.send_message(\"ack\", {});\n",
- " var fig = this;\n",
- " // Wait a second, then push the new image to the DOM so\n",
- " // that it is saved nicely (might be nice to debounce this).\n",
- " setTimeout(function () { fig.push_to_output() }, 1000);\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._init_toolbar = function() {\n",
- " var fig = this;\n",
- "\n",
- " var nav_element = $('');\n",
- " nav_element.attr('style', 'width: 100%');\n",
- " this.root.append(nav_element);\n",
- "\n",
- " // Define a callback function for later on.\n",
- " function toolbar_event(event) {\n",
- " return fig.toolbar_button_onclick(event['data']);\n",
- " }\n",
- " function toolbar_mouse_event(event) {\n",
- " return fig.toolbar_button_onmouseover(event['data']);\n",
- " }\n",
- "\n",
- " for(var toolbar_ind in mpl.toolbar_items){\n",
- " var name = mpl.toolbar_items[toolbar_ind][0];\n",
- " var tooltip = mpl.toolbar_items[toolbar_ind][1];\n",
- " var image = mpl.toolbar_items[toolbar_ind][2];\n",
- " var method_name = mpl.toolbar_items[toolbar_ind][3];\n",
- "\n",
- " if (!name) { continue; };\n",
- "\n",
- " var button = $('');\n",
- " button.click(method_name, toolbar_event);\n",
- " button.mouseover(tooltip, toolbar_mouse_event);\n",
- " nav_element.append(button);\n",
- " }\n",
- "\n",
- " // Add the status bar.\n",
- " var status_bar = $('');\n",
- " nav_element.append(status_bar);\n",
- " this.message = status_bar[0];\n",
- "\n",
- " // Add the close button to the window.\n",
- " var buttongrp = $('');\n",
- " var button = $('');\n",
- " button.click(function (evt) { fig.handle_close(fig, {}); } );\n",
- " button.mouseover('Stop Interaction', toolbar_mouse_event);\n",
- " buttongrp.append(button);\n",
- " var titlebar = this.root.find($('.ui-dialog-titlebar'));\n",
- " titlebar.prepend(buttongrp);\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._root_extra_style = function(el){\n",
- " var fig = this\n",
- " el.on(\"remove\", function(){\n",
- "\tfig.close_ws(fig, {});\n",
- " });\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._canvas_extra_style = function(el){\n",
- " // this is important to make the div 'focusable\n",
- " el.attr('tabindex', 0)\n",
- " // reach out to IPython and tell the keyboard manager to turn it's self\n",
- " // off when our div gets focus\n",
- "\n",
- " // location in version 3\n",
- " if (IPython.notebook.keyboard_manager) {\n",
- " IPython.notebook.keyboard_manager.register_events(el);\n",
- " }\n",
- " else {\n",
- " // location in version 2\n",
- " IPython.keyboard_manager.register_events(el);\n",
- " }\n",
- "\n",
- "}\n",
- "\n",
- "mpl.figure.prototype._key_event_extra = function(event, name) {\n",
- " var manager = IPython.notebook.keyboard_manager;\n",
- " if (!manager)\n",
- " manager = IPython.keyboard_manager;\n",
- "\n",
- " // Check for shift+enter\n",
- " if (event.shiftKey && event.which == 13) {\n",
- " this.canvas_div.blur();\n",
- " // select the cell after this one\n",
- " var index = IPython.notebook.find_cell_index(this.cell_info[0]);\n",
- " IPython.notebook.select(index + 1);\n",
- " }\n",
- "}\n",
- "\n",
- "mpl.figure.prototype.handle_save = function(fig, msg) {\n",
- " fig.ondownload(fig, null);\n",
- "}\n",
- "\n",
- "\n",
- "mpl.find_output_cell = function(html_output) {\n",
- " // Return the cell and output element which can be found *uniquely* in the notebook.\n",
- " // Note - this is a bit hacky, but it is done because the \"notebook_saving.Notebook\"\n",
- " // IPython event is triggered only after the cells have been serialised, which for\n",
- " // our purposes (turning an active figure into a static one), is too late.\n",
- " var cells = IPython.notebook.get_cells();\n",
- " var ncells = cells.length;\n",
- " for (var i=0; i= 3 moved mimebundle to data attribute of output\n",
- " data = data.data;\n",
- " }\n",
- " if (data['text/html'] == html_output) {\n",
- " return [cell, data, j];\n",
- " }\n",
- " }\n",
- " }\n",
- " }\n",
- "}\n",
- "\n",
- "// Register the function which deals with the matplotlib target/channel.\n",
- "// The kernel may be null if the page has been refreshed.\n",
- "if (IPython.notebook.kernel != null) {\n",
- " IPython.notebook.kernel.comm_manager.register_target('matplotlib', mpl.mpl_figure_comm);\n",
- "}\n"
- ],
- "text/plain": [
- ""
- ]
- },
- "metadata": {},
- "output_type": "display_data"
- },
- {
- "data": {
- "text/html": [
- ""
- ],
- "text/plain": [
- ""
- ]
- },
- "metadata": {},
- "output_type": "display_data"
- }
- ],
- "source": [
- "%matplotlib nbagg\n",
- "\n",
- "import matplotlib\n",
- "import matplotlib.pyplot as plt\n",
- "import numpy as np\n",
- "\n",
- "# create area for graphic\n",
- "%timeit x = np.linspace(0, np.pi, 100)\n",
- "\n",
- "# create function f(x) = sin(x)\n",
- "%timeit y = (lambda x: np.sin(x))\n",
- "\n",
- "# draw graphic\n",
- "# Function to plot\n",
- "plt.plot(x,y)\n",
- "\n",
- "# function to show the plot\n",
- "plt.show()"
- ]
- },
- {
- "cell_type": "code",
- "execution_count": 15,
- "id": "93d9ddfd",
- "metadata": {},
- "outputs": [
- {
- "ename": "ValueError",
- "evalue": "Input of type could not be converted to numpy.ndarray",
- "output_type": "error",
- "traceback": [
- "\u001b[1;31m---------------------------------------------------------------------------\u001b[0m",
- "\u001b[1;31mRuntimeError\u001b[0m Traceback (most recent call last)",
- "File \u001b[1;32mdpctl\\tensor\\_usmarray.pyx:936\u001b[0m, in \u001b[0;36mdpctl.tensor._usmarray.usm_ndarray.__setitem__\u001b[1;34m()\u001b[0m\n",
- "File \u001b[1;32m~\\Anaconda3\\envs\\my_env\\lib\\site-packages\\dpctl\\tensor\\_copy_utils.py:84\u001b[0m, in \u001b[0;36m_copy_from_numpy_into\u001b[1;34m(dst, np_ary)\u001b[0m\n\u001b[0;32m 83\u001b[0m src_ary \u001b[38;5;241m=\u001b[39m np\u001b[38;5;241m.\u001b[39mbroadcast_to(np_ary, dst\u001b[38;5;241m.\u001b[39mshape)\n\u001b[1;32m---> 84\u001b[0m \u001b[43mti\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43m_copy_numpy_ndarray_into_usm_ndarray\u001b[49m\u001b[43m(\u001b[49m\n\u001b[0;32m 85\u001b[0m \u001b[43m \u001b[49m\u001b[43msrc\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43msrc_ary\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43mdst\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43mdst\u001b[49m\u001b[43m,\u001b[49m\u001b[43m \u001b[49m\u001b[43msycl_queue\u001b[49m\u001b[38;5;241;43m=\u001b[39;49m\u001b[43mdst\u001b[49m\u001b[38;5;241;43m.\u001b[39;49m\u001b[43msycl_queue\u001b[49m\n\u001b[0;32m 86\u001b[0m \u001b[43m\u001b[49m\u001b[43m)\u001b[49m\n",
- "\u001b[1;31mRuntimeError\u001b[0m: The program was built for 1 devices\nBuild program log for 'Intel(R) Iris(R) Xe Graphics':\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, bool> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, bool> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, bool> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, bool> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, bool>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, bool>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, bool>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, bool>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, signed char> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, signed char> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, signed char>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, signed char>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned char> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned char> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned char>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned char>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, short> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, short> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, short>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, short>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned short> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned short> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned short>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned short>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, int> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, int> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, int>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, int>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned int> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned int> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned int>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned int>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, long long> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, long long> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, long long>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, long long>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned long long> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, unsigned long long> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned long long>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, unsigned long long>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, cl::sycl::detail::half_impl::half> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, cl::sycl::detail::half_impl::half> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, cl::sycl::detail::half_impl::half>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, cl::sycl::detail::half_impl::half>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, float> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, float> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, float>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, float>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, double> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, double> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, double>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, double>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, double> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, double> >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, double>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, double>'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, std::complex > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, std::complex > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, std::complex > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper, std::complex > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, std::complex >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, std::complex >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, std::complex >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel, std::complex >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for copy_cast_generic_kernel >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper > >'\nerror: backend compiler failed build.\n\nerror: Double type is not supported on this platform.\nin kernel: 'typeinfo name for cl::sycl::detail::__pf_kernel_wrapper