-
Notifications
You must be signed in to change notification settings - Fork 790
[SYCL][DOC] Move sycl::complex's extension from proposed to experimental #9867
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
Closed
jle-quel
wants to merge
14
commits into
intel:sycl
from
jle-quel:jle-quel/move-complex-extension-to-experimental
Closed
Changes from all commits
Commits
Show all changes
14 commits
Select commit
Hold shift + click to select a range
7cba1fd
move sycl::complex's extension from proposed to experimental
jle-quel aed3d3f
update the status section
jle-quel 410be92
Update the extension with the reviewers's feedback
jle-quel f315009
Add more clarification to the extension
jle-quel 68bb448
Merge branch 'sycl' of github.com:jle-quel/llvm into jle-quel/move-co…
jle-quel fdf9028
Bring latest modifications made to the proposed extension to the expe…
jle-quel a681828
Update Math's pow section
jle-quel 1622a26
Fix typo in the math's pow section
jle-quel 5fbe411
shorten the namespace declarations
jle-quel 7c18b40
remove all references of the literal operators
jle-quel e6ee73e
re-introduce proposed/sycl_ext_oneapi_complex.asciidoc
jle-quel a4e255c
Merge branch 'sycl' of github.com:jle-quel/llvm into jle-quel/move-co…
jle-quel 1be71af
delete proposed extension
jle-quel 5d0837f
Merge branch 'sycl' of github.com:jle-quel/llvm into jle-quel/move-co…
jle-quel File filter
Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
There are no files selected for viewing
340 changes: 340 additions & 0 deletions
340
sycl/doc/extensions/experimental/sycl_ext_oneapi_complex.asciidoc
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,340 @@ | ||
= sycl_ext_oneapi_complex | ||
|
||
:source-highlighter: coderay | ||
:coderay-linenums-mode: table | ||
|
||
// This section needs to be after the document title. | ||
:doctype: book | ||
:toc2: | ||
:toc: left | ||
:encoding: utf-8 | ||
:lang: en | ||
:dpcpp: pass:[DPC++] | ||
|
||
// Set the default source code type in this document to C++, | ||
// for syntax highlighting purposes. This is needed because | ||
// docbook uses c++ and html5 uses cpp. | ||
:language: {basebackend@docbook:c++:cpp} | ||
|
||
|
||
== Notice | ||
|
||
[%hardbreaks] | ||
Copyright (C) 2022-2022 Codeplay Ltd. All rights reserved. | ||
|
||
Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are trademarks | ||
of The Khronos Group Inc. OpenCL(TM) is a trademark of Apple Inc. used by | ||
permission by Khronos. | ||
|
||
|
||
== Contact | ||
|
||
To report problems with this extension, please open a new issue at: | ||
|
||
https://github.com/intel/llvm/issues | ||
|
||
|
||
== Dependencies | ||
|
||
This extension is written against the SYCL 2020 revision 5 specification. All | ||
references below to the "core SYCL specification" or to section numbers in the | ||
SYCL specification refer to that revision. | ||
|
||
== Status | ||
|
||
This is an experimental extension specification, intended to provide early | ||
access to features and gather community feedback. Interfaces defined in this | ||
specification are implemented in {dpcpp}, but they are not finalized and may | ||
change incompatibly in future versions of {dpcpp} without prior notice. | ||
*Shipping software products should not rely on APIs defined in this | ||
specification.* | ||
|
||
== Overview | ||
|
||
While {dpcpp} has support for `std::complex` in device code, it limits the | ||
complex interface and operations to the existing C++ standard. This proposal | ||
defines a SYCL complex extension based on but independent of the `std::complex` | ||
interface. This framework would allow for further development of complex math | ||
within oneAPI. Possible areas for deviation with `std::complex` include adding | ||
complex support for `marray` and `vec` and overloading mathematical | ||
functions to handle the element-wise operations. | ||
|
||
== Specification | ||
|
||
=== Feature test macro | ||
|
||
This extension provides a feature-test macro as described in the core SYCL | ||
specification. An implementation supporting this extension must predefine the | ||
macro `SYCL_EXT_ONEAPI_COMPLEX` to one of the values defined in the table | ||
below. Applications can test for the existence of this macro to determine if | ||
the implementation supports this feature, or applications can test the macro's | ||
value to determine which of the extension's features the implementation | ||
supports. | ||
|
||
[%header,cols="1,5"] | ||
|=== | ||
|Value | ||
|Description | ||
|
||
|1 | ||
|The APIs of this experimental extension are not versioned, so the feature-test macro always has this value. | ||
|=== | ||
|
||
=== Complex Class | ||
|
||
The core of this extension is the complex math class. This class contains a real | ||
and imaginary component and enables mathematical operations between complex | ||
numbers and decimals. The complex class interface and operations (shown below) | ||
are available in both host code and device code. Some operations, however, are | ||
available only in host code as noted below. | ||
|
||
The complex type is trivially copyable and type trait `is_device_copyable` | ||
should resolve to `std::true_type`. | ||
|
||
The `T` template parameter must be one of the types float, double, or | ||
sycl::half. | ||
|
||
Note: When performing operations between complex numbers and decimals, | ||
the decimal is treated as a complex number with a real component equal to | ||
the decimal and an imaginary component equal to 0. | ||
|
||
```C++ | ||
namespace sycl::ext::oneapi::experimental { | ||
|
||
template <typename T> | ||
class complex { | ||
public: | ||
using value_type = T; | ||
|
||
/// Constructs the complex number from real and imaginary parts. | ||
constexpr complex(value_type __re = value_type(), value_type __im = value_type()); | ||
|
||
/// Converting constructor. Constructs the object from a complex number of a different type. | ||
template <typename X> | ||
constexpr complex(const complex<X> &); | ||
|
||
/// Converting constructor. Constructs the object from a std::complex number. | ||
template <class X> | ||
constexpr complex(const std::complex<X> &); | ||
|
||
/// Constructs a std::complex number from a sycl::complex. | ||
template <class X> | ||
constexpr operator std::complex<X>() const; | ||
|
||
/// Returns the real part. | ||
constexpr value_type real() const; | ||
/// Returns the imaginary part. | ||
constexpr value_type imag() const; | ||
|
||
/// Sets the real part from value. | ||
void real(value_type value); | ||
/// Sets the imaginary part from value. | ||
void imag(value_type value); | ||
|
||
/// Assigns x to the real part of the complex number. Imaginary part is set to zero. | ||
complex<value_type> &operator=(value_type x); | ||
/// Adds and assigns real number y to complex number z. | ||
friend complex<value_type> &operator+=(complex<value_type> &z, value_type y); | ||
/// Subtracts and assigns real number y to complex number z. | ||
friend complex<value_type> &operator-=(complex<value_type> &z, value_type y); | ||
/// Multiplies and assigns real number y to complex number z. | ||
friend complex<value_type> &operator*=(complex<value_type> &z, value_type y); | ||
/// Divides and assigns real number y to complex number z. | ||
friend complex<value_type> &operator/=(complex<value_type> &z, value_type y); | ||
|
||
/// Assigns cx.real() and cx.imag() to the real and the imaginary parts of the complex number respectively. | ||
complex<value_type> &operator=(const complex<value_type> &cx); | ||
/// Adds and assigns complex number w to complex number z. | ||
template <class X> friend complex<value_type> &operator+=(complex<value_type> &z, const complex<X> &w); | ||
/// Subtracts and assigns complex number w to complex number z. | ||
template <class X> friend complex<value_type> &operator-=(complex<value_type> &z, const complex<X> &w); | ||
/// Multiplies and assigns complex number w to complex number z. | ||
template <class X> friend complex<value_type> &operator*=(complex<value_type> &z, const complex<X> &w); | ||
/// Divides and assigns complex number w to complex number z. | ||
template <class X> friend complex<value_type> &operator/=(complex<value_type> &z, const complex<X> &w); | ||
|
||
/// Adds complex numbers z and w and returns the value. | ||
friend complex<value_type> operator+(const complex<value_type> &z, const complex<value_type> &w); | ||
/// Adds complex number z and real y and returns the value. | ||
friend complex<value_type> operator+(const complex<value_type> &z, value_type y); | ||
/// Adds real x and complex number w and returns the value. | ||
friend complex<value_type> operator+(value_type x, const complex<value_type> &w); | ||
/// Returns the value of its argument. | ||
friend complex<value_type> operator+(const complex<value_type> &); | ||
|
||
/// Subtracts complex numbers z and w and returns the value. | ||
friend complex<value_type> operator-(const complex<value_type> &z, const complex<value_type> &w); | ||
/// Subtracts complex number z and real y and returns the value. | ||
friend complex<value_type> operator-(const complex<value_type> &z, value_type y); | ||
/// Subtracts real x and complex number w and returns the value. | ||
friend complex<value_type> operator-(value_type x, const complex<value_type> &w); | ||
/// Negates the argument. | ||
friend complex<value_type> operator-(const complex<value_type> &); | ||
|
||
/// Multiplies complex numbers z and w and returns the value. | ||
friend complex<value_type> operator*(const complex<value_type> &z, const complex<value_type> &w); | ||
/// Multiplies complex number z and real y and returns the value. | ||
friend complex<value_type> operator*(const complex<value_type> &z, value_type y); | ||
/// Multiplies real x and complex number w and returns the value. | ||
friend complex<value_type> operator*(value_type x, const complex<value_type> &w); | ||
|
||
/// Divides complex numbers z and w and returns the value. | ||
friend complex<value_type> operator/(const complex<value_type> &z, const complex<value_type> &w); | ||
/// Divides complex number z and real y and returns the value. | ||
friend complex<value_type> operator/(const complex<value_type> &z, value_type y); | ||
/// Divides real x and complex number w and returns the value. | ||
friend complex<value_type> operator/(value_type x, const complex<value_type> &w); | ||
|
||
/// Compares complex numbers z and w and returns true if they are the same, otherwise false. | ||
friend constexpr bool operator==(const complex<value_type> &z, const complex<value_type> &w); | ||
/// Compares complex number z and real y and returns true if they are the same, otherwise false. | ||
friend constexpr bool operator==(const complex<value_type> &z, value_type y); | ||
/// Compares real x and complex number w and returns true if they are the same, otherwise false. | ||
friend constexpr bool operator==(value_type x, const complex<value_type> &w); | ||
|
||
/// Compares complex numbers z and w and returns true if they are different, otherwise false. | ||
friend constexpr bool operator!=(const complex<value_type> &z, const complex<value_type> &w); | ||
///Compares complex number z and real y and returns true if they are different, otherwise false. | ||
friend constexpr bool operator!=(const complex<value_type> &z, value_type y); | ||
/// Compares real x and complex number w and returns true if they are different, otherwise false. | ||
friend constexpr bool operator!=(value_type x, const complex<value_type> &w); | ||
|
||
/// Reads a complex number from is. | ||
/// Not allowed in device code. | ||
template <class C, class T> friend std::basic_istream<C, T> &operator>>(std::basic_istream<C, T> &is, complex<value_type> &); | ||
/// Writes to os the complex number z in the form (real,imaginary). | ||
/// Not allowed in device code. | ||
template <class C, class T> friend std::basic_ostream<C, T> &operator<<(std::basic_ostream<C, T> &os, const complex<value_type> &); | ||
gmlueck marked this conversation as resolved.
Show resolved
Hide resolved
|
||
/// Streams the complex number z in the format "(real,imaginary)" into `sycl::stream` x and return the result. | ||
friend const sycl::stream &operator<<(const sycl::stream &x, const complex<value_type> &z); | ||
|
||
} // namespace sycl::ext::oneapi::experimental | ||
``` | ||
|
||
=== Mathematical operations | ||
|
||
This proposal adds to the `sycl::ext::oneapi::experimental` namespace, math | ||
functions accepting the complex types `complex<sycl::half>`, `complex<float>`, | ||
`complex<double>` as well as the scalar types `sycl::half`, `float` and `double` | ||
for the SYCL math functions, `abs`, `acos`, `asin`, `atan`, `acosh`, `asinh`, | ||
`atanh`, `arg`, `conj`, `cos`, `cosh`, `exp`, `log`, `log10`, `norm`, `polar`, | ||
`pow`, `proj`, `sin`, `sinh`, `sqrt`, `tan`, and `tanh`. | ||
|
||
gmlueck marked this conversation as resolved.
Show resolved
Hide resolved
|
||
These functions are available in both host and device code, and each math | ||
function should follow the C++ standard for handling NaN's and Inf values. | ||
|
||
Note: In the case of the `pow` function, additional overloads have been added | ||
to ensure that for their first argument `base` and second argument `exponent`: | ||
|
||
* If `base` and/or `exponent` has type `complex<double>` or `double`, | ||
then `pow(base, exponent)` has the same effect as | ||
`pow(complex<double>(base), complex<double>(exponent))`. | ||
|
||
* Otherwise, if `base` and/or `exponent` has type `complex<float>` or `float`, | ||
then `pow(base, exponent)` has the same effect as | ||
`pow(complex<float>(base), complex<float>(exponent))`. | ||
|
||
* Otherwise, if `base` and/or `exponent` has type `complex<sycl::half>` or `sycl::half`, | ||
then `pow(base, exponent)` has the same effect as | ||
`pow(complex<sycl::half>(base), complex<sycl::half>(exponent))`. | ||
|
||
```C++ | ||
namespace sycl::ext::oneapi::experimental { | ||
|
||
/// VALUES: | ||
/// Returns the real component of the complex number z. | ||
template <class T> constexpr T real(const complex<T> &); | ||
/// Returns the real component of the number y, treated as complex numbers with zero imaginary component. | ||
template <class T> constexpr T real(T); | ||
/// Returns the imaginary component of the complex number z. | ||
template <class T> constexpr T imag(const complex<T> &); | ||
/// Returns the imaginary component of the number y, treated as complex numbers with zero imaginary component. | ||
template <class T> constexpr T imag(T); | ||
|
||
/// Compute the magnitude of complex number x. | ||
template <class T> T abs(const complex<T> &); | ||
/// Compute phase angle in radians of complex number x. | ||
template <class T> T arg(const complex<T> &); | ||
/// Compute phase angle in radians of complex number x, treated as complex number with positive zero imaginary component. | ||
template <class T> T arg(T); | ||
/// Compute the squared magnitude of complex number x. | ||
template <class T> T norm(const complex<T> &); | ||
/// Compute the squared magnitude of number x, treated as complex number with positive zero imaginary component. | ||
template <class T> T norm(T); | ||
/// Compute the conjugate of complex number x. | ||
template <class T> complex<T> conj(const complex<T> &); | ||
/// Compute the conjugate of number y, treated as complex number with positive zero imaginary component. | ||
template <class T> complex<T> conj(T); | ||
/// Compute the projection of complex number x. | ||
template <class T> complex<T> proj(const complex<T> &); | ||
/// Compute the projection of number y, treated as complex number with positive zero imaginary component. | ||
template <class T> complex<T> proj(T); | ||
/// Construct a complex number from polar coordinates with mangitude rho and angle theta. | ||
template <class T> complex<T> polar(const T &rho, const T &theta = T()); | ||
|
||
/// TRANSCENDENTALS: | ||
/// Compute the natural log of complex number x. | ||
template <class T> complex<T> log(const complex<T> &); | ||
/// Compute the base-10 log of complex number x. | ||
template <class T> complex<T> log10(const complex<T> &); | ||
/// Compute the square root of complex number x. | ||
template <class T> complex<T> sqrt(const complex<T> &); | ||
/// Compute the base-e exponent of complex number x. | ||
template <class T> complex<T> exp(const complex<T> &); | ||
|
||
/// Compute complex number z raised to the power of complex number y. | ||
template <class T> complex<T> pow(const complex<T> &, const complex<T> &); | ||
/// Compute complex number z raised to the power of complex number y. | ||
template <class T, class U> complex</*Promoted*/> pow(const complex<T> &, const complex<U> &); | ||
/// Compute complex number z raised to the power of real number y. | ||
template <class T, class U> complex</*Promoted*/> pow(const complex<T> &, const U &); | ||
/// Compute real number x raised to the power of complex number y. | ||
template <class T, class U> complex</*Promoted*/> pow(const T &, const complex<U> &); | ||
gmlueck marked this conversation as resolved.
Show resolved
Hide resolved
|
||
|
||
/// Compute the inverse hyperbolic sine of complex number x. | ||
template <class T> complex<T> asinh(const complex<T> &); | ||
/// Compute the inverse hyperbolic cosine of complex number x. | ||
template <class T> complex<T> acosh(const complex<T> &); | ||
/// Compute the inverse hyperbolic tangent of complex number x. | ||
template <class T> complex<T> atanh(const complex<T> &); | ||
/// Compute the hyperbolic sine of complex number x. | ||
template <class T> complex<T> sinh(const complex<T> &); | ||
/// Compute the hyperbolic cosine of complex number x. | ||
template <class T> complex<T> cosh(const complex<T> &); | ||
/// Compute the hyperbolic tangent of complex number x. | ||
template <class T> complex<T> tanh(const complex<T> &); | ||
/// Compute the inverse sine of complex number x. | ||
template <class T> complex<T> asin(const complex<T> &); | ||
/// Compute the inverse cosine of complex number x. | ||
template <class T> complex<T> acos(const complex<T> &); | ||
/// Compute the inverse tangent of complex number x. | ||
template <class T> complex<T> atan(const complex<T> &); | ||
/// Compute the sine of complex number x. | ||
template <class T> complex<T> sin(const complex<T> &); | ||
/// Compute the cosine of complex number x. | ||
template <class T> complex<T> cos(const complex<T> &); | ||
// Compute the tangent of complex number x. | ||
template <class T> complex<T> tan(const complex<T> &); | ||
|
||
} // namespace sycl::ext::oneapi::experimental | ||
``` | ||
jle-quel marked this conversation as resolved.
Show resolved
Hide resolved
|
||
|
||
== Implementation notes | ||
|
||
The complex mathematical operations can all be defined using SYCL built-ins. | ||
Therefore, implementing complex with SYCL built-ins would allow any backend | ||
with SYCL built-ins to support `sycl::ext::oneapi::experimental::complex`. | ||
The current implementation of `std::complex` relies on `libdevice`, which | ||
requires adjusting and altering the clang driver. This additional work would not | ||
be necessary for adding complex support with this extension. | ||
|
||
== Issues | ||
|
||
The motivation for adding this extension is to allow for complex support of | ||
`marray` and `vec`. This raises the issue of if this should be represented as | ||
an array of structs or a struct of arrays. The advantage of having an array | ||
of structs is that this is the most intuitive format for the user. As the | ||
user is likely thinking about the problem as a vector of complex numbers. | ||
However, this would cause the real and imaginary vectors to be non-contiguous. | ||
Conversely, having a struct of arrays would be less intuitive but would keep | ||
the vector's memory contiguous. |
Oops, something went wrong.
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
Uh oh!
There was an error while loading. Please reload this page.