[SYCL][NFC] Remove some mentions of host device.

sycl/include/sycl/group.hpp

@@ -127,12 +127,8 @@ template <int Dimensions = 1> class __SYCL_TYPE(group) group {
 #ifdef __SYCL_DEVICE_ONLY__
     return __spirv::initLocalInvocationId<Dimensions, id<Dimensions>>();
 #else
-    throw runtime_error("get_local_id() is not implemented on host device",
+    throw runtime_error("get_local_id() is not implemented on host",
                         PI_ERROR_INVALID_DEVICE);
-    // Implementing get_local_id() on host device requires ABI breaking change.
-    // It requires extending class group with local item which represents
-    // local_id. Currently this local id is only used in nd_item and group
-    // cannot access it.
 #endif
   }
 
@@ -545,7 +541,7 @@ group<Dims> this_group() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 
@@ -557,7 +553,7 @@ template <int Dims> group<Dims> this_group() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 } // namespace ext::oneapi::experimental

sycl/include/sycl/group_algorithm.hpp

@@ -159,7 +159,7 @@ Function for_each(Group g, Ptr first, Ptr last, Function f) {
   (void)first;
   (void)last;
   (void)f;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -186,7 +186,7 @@ reduce_over_group(Group, T x, BinaryOperation binary_op) {
                             sycl::detail::spirv::group_scope<Group>::value>(
       typename sycl::detail::GroupOpTag<T>::type(), x, binary_op);
 #else
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -209,7 +209,7 @@ reduce_over_group(Group g, T x, BinaryOperation binary_op) {
   (void)g;
   (void)x;
   (void)binary_op;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -258,7 +258,7 @@ reduce_over_group(Group g, V x, T init, BinaryOperation binary_op) {
   return binary_op(init, reduce_over_group(g, x, binary_op));
 #else
   (void)g;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -286,7 +286,7 @@ reduce_over_group(Group g, V x, T init, BinaryOperation binary_op) {
   return result;
 #else
   (void)g;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -310,7 +310,7 @@ joint_reduce(Group g, Ptr first, Ptr last, BinaryOperation binary_op) {
   (void)first;
   (void)last;
   (void)binary_op;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -345,7 +345,7 @@ joint_reduce(Group g, Ptr first, Ptr last, T init, BinaryOperation binary_op) {
 #else
   (void)g;
   (void)last;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -358,7 +358,7 @@ any_of_group(Group, bool pred) {
   return sycl::detail::spirv::GroupAny<Group>(pred);
 #else
   (void)pred;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -384,7 +384,7 @@ joint_any_of(Group g, Ptr first, Ptr last, Predicate pred) {
   (void)first;
   (void)last;
   (void)pred;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -397,7 +397,7 @@ all_of_group(Group, bool pred) {
   return sycl::detail::spirv::GroupAll<Group>(pred);
 #else
   (void)pred;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -423,7 +423,7 @@ joint_all_of(Group g, Ptr first, Ptr last, Predicate pred) {
   (void)first;
   (void)last;
   (void)pred;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -436,7 +436,7 @@ none_of_group(Group, bool pred) {
   return sycl::detail::spirv::GroupAll<Group>(!pred);
 #else
   (void)pred;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -459,7 +459,7 @@ joint_none_of(Group g, Ptr first, Ptr last, Predicate pred) {
   (void)first;
   (void)last;
   (void)pred;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -478,7 +478,7 @@ shift_group_left(Group, T x, typename Group::linear_id_type delta = 1) {
 #else
   (void)x;
   (void)delta;
-  throw runtime_error("Sub-groups are not supported on host device.",
+  throw runtime_error("Sub-groups are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -497,7 +497,7 @@ shift_group_right(Group, T x, typename Group::linear_id_type delta = 1) {
 #else
   (void)x;
   (void)delta;
-  throw runtime_error("Sub-groups are not supported on host device.",
+  throw runtime_error("Sub-groups are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -516,7 +516,7 @@ permute_group_by_xor(Group, T x, typename Group::linear_id_type mask) {
 #else
   (void)x;
   (void)mask;
-  throw runtime_error("Sub-groups are not supported on host device.",
+  throw runtime_error("Sub-groups are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -535,7 +535,7 @@ select_from_group(Group, T x, typename Group::id_type local_id) {
 #else
   (void)x;
   (void)local_id;
-  throw runtime_error("Sub-groups are not supported on host device.",
+  throw runtime_error("Sub-groups are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -554,7 +554,7 @@ group_broadcast(Group, T x, typename Group::id_type local_id) {
 #else
   (void)x;
   (void)local_id;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -573,7 +573,7 @@ group_broadcast(Group g, T x, typename Group::linear_id_type linear_local_id) {
   (void)g;
   (void)x;
   (void)linear_local_id;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -589,7 +589,7 @@ group_broadcast(Group g, T x) {
 #else
   (void)g;
   (void)x;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -614,7 +614,7 @@ exclusive_scan_over_group(Group, T x, BinaryOperation binary_op) {
                             sycl::detail::spirv::group_scope<Group>::value>(
       typename sycl::detail::GroupOpTag<T>::type(), x, binary_op);
 #else
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -637,7 +637,7 @@ exclusive_scan_over_group(Group g, T x, BinaryOperation binary_op) {
   (void)g;
   (void)x;
   (void)binary_op;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -715,7 +715,7 @@ exclusive_scan_over_group(Group g, V x, T init, BinaryOperation binary_op) {
   return scan;
 #else
   (void)g;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -773,7 +773,7 @@ joint_exclusive_scan(Group g, InPtr first, InPtr last, OutPtr result, T init,
   (void)last;
   (void)result;
   (void)init;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -845,7 +845,7 @@ inclusive_scan_over_group(Group, T x, BinaryOperation binary_op) {
                             sycl::detail::spirv::group_scope<Group>::value>(
       typename sycl::detail::GroupOpTag<T>::type(), x, binary_op);
 #else
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -867,7 +867,7 @@ inclusive_scan_over_group(Group g, T x, BinaryOperation binary_op) {
   (void)g;
   (void)x;
   (void)binary_op;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -897,7 +897,7 @@ inclusive_scan_over_group(Group g, V x, BinaryOperation binary_op, T init) {
   return inclusive_scan_over_group(g, x, binary_op);
 #else
   (void)g;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }
@@ -975,7 +975,7 @@ joint_inclusive_scan(Group g, InPtr first, InPtr last, OutPtr result,
   (void)g;
   (void)last;
   (void)result;
-  throw runtime_error("Group algorithms are not supported on host device.",
+  throw runtime_error("Group algorithms are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }

sycl/include/sycl/id.hpp

@@ -356,7 +356,7 @@ id<Dims> this_id() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 
@@ -367,7 +367,7 @@ template <int Dims> id<Dims> this_id() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 } // namespace ext::oneapi::experimental

sycl/include/sycl/item.hpp

@@ -148,7 +148,7 @@ item<Dims> this_item() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 
@@ -159,7 +159,7 @@ template <int Dims> item<Dims> this_item() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 } // namespace ext::oneapi::experimental

sycl/include/sycl/nd_item.hpp

@@ -198,7 +198,7 @@ nd_item<Dims> this_nd_item() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 
@@ -210,7 +210,7 @@ template <int Dims> nd_item<Dims> this_nd_item() {
 #else
   throw sycl::exception(
       sycl::make_error_code(sycl::errc::feature_not_supported),
-      "Free function calls are not supported on host device");
+      "Free function calls are not supported on host");
 #endif
 }
 } // namespace ext::oneapi::experimental

sycl/include/sycl/sub_group.hpp

@@ -22,7 +22,7 @@ inline sub_group this_sub_group() {
 #ifdef __SYCL_DEVICE_ONLY__
   return sub_group();
 #else
-  throw runtime_error("Sub-groups are not supported on host device.",
+  throw runtime_error("Sub-groups are not supported on host.",
                       PI_ERROR_INVALID_DEVICE);
 #endif
 }