From 41b6754cc4957829324cc9295441515cace25ee3 Mon Sep 17 00:00:00 2001
From: John Pennycook <john.pennycook@intel.com>
Date: Mon, 10 Mar 2025 09:44:45 +0000
Subject: [PATCH 1/2] Clarify that certain types are trivially copyable

id, range, nd_range, vec and marray are permitted as kernel arguments,
but the specification does not say that they are trivially copyable
nor that they are device copyable.
---
 adoc/chapters/programming_interface.adoc | 15 +++++++++++++++
 1 file changed, 15 insertions(+)

diff --git a/adoc/chapters/programming_interface.adoc b/adoc/chapters/programming_interface.adoc
index e420c3922..10fd7319d 100644
--- a/adoc/chapters/programming_interface.adoc
+++ b/adoc/chapters/programming_interface.adoc
@@ -11965,6 +11965,9 @@ It can be constructed from integers.
 The SYCL [code]#range# class template provides the common by-value semantics
 (see <<sec:byval-semantics>>).
 
+Every instantiation of the SYCL [code]#range# class template is a trivially
+copyable type.
+
 A synopsis of the SYCL [code]#range# class is provided below.
 The constructors, member functions and non-member functions of the SYCL
 [code]#range# class are listed in <<table.constructors.range>>,
@@ -12212,6 +12215,9 @@ which the kernel is to be executed, and the range of each work group.
 The SYCL [code]#nd_range# class template provides the common by-value semantics
 (see <<sec:byval-semantics>>).
 
+Every instantiation of the SYCL [code]#nd_range# class template is a trivially
+copyable type.
+
 A synopsis of the SYCL [code]#nd_range# class is provided below.
 The constructors and member functions of the SYCL [code]#nd_range# class are
 listed in <<table.constructors.ndrange>> and <<table.members.ndrange>>
@@ -12296,6 +12302,9 @@ as a [code]#size_t#.
 The SYCL [code]#id# class template provides the common by-value semantics (see
 <<sec:byval-semantics>>).
 
+Every instantiation of the SYCL [code]#id# class template is a trivially
+copyable type.
+
 A synopsis of the SYCL [code]#id# class is provided below.
 The constructors, member functions and non-member functions of the SYCL
 [code]#id# class are listed in <<table.constructors.id>>, <<table.members.id>>
@@ -17901,6 +17910,9 @@ The element type parameter, [code]#DataT#, must be the cv-unqualified version of
 one of the following: one of the built-in scalar data types listed in
 <<subsec:scalartypes>>, [code]#half#, or [code]#sycl::byte#.
 
+Every instantiation of the SYCL [code]#vec# class template is a trivially
+copyable type.
+
 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.
@@ -19558,6 +19570,9 @@ the [code]#std::size_t# type.
 The element type parameter, [code]#DataT#, must be a _numeric type_ as it is
 defined by {cpp} standard.
 
+If [code]#DataT# is a trivially copyable type, then [code]#marray<DataT,
+NumElements># is a trivially copyable type.
+
 An instance of the [code]#marray# 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 arrays and scalars to be convertible

From ae34962f1cd1b5466c40330e04ef68b5293a1d0e Mon Sep 17 00:00:00 2001
From: John Pennycook <john.pennycook@intel.com>
Date: Wed, 23 Apr 2025 13:35:28 +0100
Subject: [PATCH 2/2] Simplify rules for parameter passing to kernels

id, range, marray and vec are now guaranteed to be trivially copyable, so do
not need to be listed explicitly as SYCL types that can be passed to kernels.
---
 adoc/chapters/programming_interface.adoc | 8 ++------
 1 file changed, 2 insertions(+), 6 deletions(-)

diff --git a/adoc/chapters/programming_interface.adoc b/adoc/chapters/programming_interface.adoc
index 10fd7319d..5acf15f11 100644
--- a/adoc/chapters/programming_interface.adoc
+++ b/adoc/chapters/programming_interface.adoc
@@ -17328,12 +17328,8 @@ allowable types for a kernel parameter:
   - [code]#unsampled_image_accessor# when templated with
     [code]#image_target::device#;
   - [code]#sampled_image_accessor# when templated with
-    [code]#image_target::device#;
-  - [code]#stream#;
-  - [code]#id#;
-  - [code]#range#;
-  - [code]#marray<T, NumElements># when [code]#T# is <<device-copyable>>;
-  - [code]#vec<T, NumElements>#.
+    [code]#image_target::device#; and
+  - [code]#stream#.
 
 * An array of element types [code]#T# is a legal parameter type if [code]#T# is
   a legal parameter type.