Skip to content

Commit

Permalink
Remove underspecified vector_t
Browse files Browse the repository at this point in the history
This is change 9 of 9 that fix problems with the specification of the
`vec` class.  An implementation that follows the existing specification
would not accept common code patterns and would not pass the CTS.  None
of the existing implementations actually follow the existing
specification.

This change removes the `vector_t` type alias from the `vec` class and
also removes the conversions to and from this type.  The intention of
this type was to provide a conversion between `vec` and a
compiler-specific builtin vector type.  However, these conversion also
lead to ambiguity errors in common code patterns such as:

```
sycl::vec<sycl::half, 1> h1;
sycl::half h;
h = h1;
```

Prior to this change, the last line could result in an ambiguity error
from the compiler between:

* Convert `vec<half,1>` -> `half`, call `half::operator=(const half&)`
* Convert `vec<half,1>` -> `_Float16`, convert `_Float16` -> `float`,
  call `half::operator=(float)`

What's worse is that the exact type alias of `vector_t` is
implementation-defined, so these ambiguity errors could be different
for each SYCL implementation.

During the WG meeting, we learned that the two major SYCL
implementations (DPC++ and AdaptiveCpp) define `vector_t` differently
even though they are both based on clang.  Given this lack of
commonality, the WG decided to remove `vector_t` entirely from the spec.
Implementations can decide for themselves whether to provide a
compiler-specific conversion like this.

These changes correspond to slides 29 - 31 of the presentation that was
discussed in the WG meetings.
  • Loading branch information
gmlueck committed Dec 3, 2024
1 parent 94eb558 commit a57f1a7
Show file tree
Hide file tree
Showing 3 changed files with 6 additions and 63 deletions.
4 changes: 2 additions & 2 deletions adoc/chapters/opencl_backend.adoc
Original file line number Diff line number Diff line change
Expand Up @@ -1063,8 +1063,8 @@ OpenCL C functions that are linked with, using either offline or online
compilation, must be declared as [code]#extern "C"# function declarations.
The function parameters of these function declarations must be defined as the
OpenCL C interoperability aliases; [code]#pointer# of the [code]#multi_ptr#
class template, [code]#vector_t# of the [code]#vec# class template and scalar
data type aliases described in <<table.types.aliases>>.
class template, and scalar data type aliases described in
<<table.types.aliases>>.

// \include{opencl_extensions}
// %%%%%%%%%%%%%%%%%%%%%%%%%%%% begin opencl_extensions %%%%%%%%%%%%%%%%%%%%%%%%%%%%
Expand Down
55 changes: 4 additions & 51 deletions adoc/chapters/programming_interface.adoc
Original file line number Diff line number Diff line change
Expand Up @@ -17763,18 +17763,10 @@ Any other value shall produce a compilation failure.
The element type parameter, _DataT_, must be one of the basic scalar types
supported in device code.

The SYCL [code]#vec# class template provides interoperability with the
underlying vector type defined by [code]#vector_t# which is available only when
compiled for the device.
The SYCL [code]#vec# class can be constructed from an instance of
[code]#vector_t# and can implicitly convert to an instance of [code]#vector_t#
in order to support interoperability with native <<backend>> functions from a
SYCL kernel function.

An instance of the SYCL [code]#vec# class template can also be implicitly
converted to an instance of the data type when the number of elements is
[code]#1# in order to allow single element vectors and scalars to be convertible
with each other.
An instance of the SYCL [code]#vec# class template can be implicitly converted
to an instance of the data type when the number of elements is [code]#1# in
order to allow single element vectors and scalars to be convertible with each
other.

==== Vec interface

Expand Down Expand Up @@ -17826,15 +17818,6 @@ constexpr vec(const vec<DataT, NumElements>& rhs)
----
a@ Construct a vector of element type [code]#DataT# and number of elements [code]#NumElements# by copy from another similar vector.

a@
[source]
----
vec(vector_t nativeVector)
----
a@ Available only when: compiled for the device.

Constructs a SYCL [code]#vec# instance from an instance of the underlying backend-native vector type defined by [code]#vector_t#.

|====


Expand All @@ -17844,16 +17827,6 @@ Constructs a SYCL [code]#vec# instance from an instance of the underlying backen
[width="100%",options="header",separator="@",cols="65%,35%"]
|====
@ Member function @ Description
a@
[source]
----
operator vector_t() const
----
a@ Available only when: compiled for the device.

Converts this SYCL [code]#vec# instance to the underlying backend-native vector type
defined by [code]#vector_t#.

a@
[source]
----
Expand Down Expand Up @@ -18598,22 +18571,6 @@ using value_type = DataT
Each of these type aliases tells the type of an element in the underlying
[code]#vec#.

'''

[frame=all,grid=none,separator="@"]
|====
a@
[source]
----
#ifdef __SYCL_DEVICE_ONLY__
using vector_t = /*unspecified*/
#endif
----
|====
This type alias is available only in device code.
The [code]#vector_t# type represents a native vector type of [code]#NumElements#
elements where each element's type is [code]#DataT#.

[[sec:constructors.swizzled-vec]]
===== Constructors for the swizzled vector class templates

Expand Down Expand Up @@ -18656,10 +18613,6 @@ The destructors have no visible effect.
a@
[source]
----
#ifdef __SYCL_DEVICE_ONLY__
operator vector_t() const
#endif

operator DataT() const
static constexpr size_t byte_size() noexcept
static constexpr size_t size() noexcept
Expand Down
10 changes: 0 additions & 10 deletions adoc/headers/vec.h
Original file line number Diff line number Diff line change
Expand Up @@ -37,10 +37,6 @@ template <typename DataT, int NumElements> class vec {
using element_type = DataT;
using value_type = DataT;

#ifdef __SYCL_DEVICE_ONLY__
using vector_t = __unspecified__;
#endif

vec();

explicit constexpr vec(const DataT& arg);
Expand All @@ -49,12 +45,6 @@ template <typename DataT, int NumElements> class vec {

constexpr vec(const vec<DataT, NumElements>& rhs);

#ifdef __SYCL_DEVICE_ONLY__
vec(vector_t nativeVector);

operator vector_t() const;
#endif

// Available only when: NumElements == 1
operator DataT() const;

Expand Down

0 comments on commit a57f1a7

Please sign in to comment.