[SYCL] Add (raw|decorated)_generic_ptr aliases

sycl/include/sycl/pointers.hpp

@@ -47,6 +47,11 @@ using private_ptr =
 // The interface exposes non-decorated pointer while keeping the
 // address space information internally.
 
+template <typename ElementType>
+using raw_generic_ptr =
+    multi_ptr<ElementType, access::address_space::generic_space,
+              access::decorated::no>;
+
 template <typename ElementType>
 using raw_global_ptr =
     multi_ptr<ElementType, access::address_space::global_space,
@@ -64,6 +69,11 @@ using raw_private_ptr =
 // Template specialization aliases for different pointer address spaces.
 // The interface exposes decorated pointer.
 
+template <typename ElementType>
+using decorated_generic_ptr =
+    multi_ptr<ElementType, access::address_space::generic_space,
+              access::decorated::yes>;
+
 template <typename ElementType>
 using decorated_global_ptr =
     multi_ptr<ElementType, access::address_space::global_space,

sycl/test/check_device_code/extensions/address_cast.cpp

@@ -11,12 +11,6 @@
 using namespace sycl;
 using namespace sycl::ext::oneapi::experimental;
 
-// FIXME: should be removed when https://github.com/intel/llvm/pull/15389 is merged in.
-template <typename ElementType>
-using decorated_generic_ptr =
-    multi_ptr<ElementType, access::address_space::generic_space,
-              access::decorated::yes>;
-
 namespace static_as_cast {
 // CHECK-LABEL: define dso_local spir_func void @_ZN14static_as_cast19to_global_decoratedEN4sycl3_V19multi_ptrIiLNS1_6access13address_spaceE6ELNS3_9decoratedE1EEE(
 // CHECK-SAME: ptr addrspace(4) dead_on_unwind noalias nocapture writable writeonly sret(%"class.sycl::_V1::multi_ptr") align 8 [[AGG_RESULT:%.*]], ptr nocapture noundef readonly byval(%"class.sycl::_V1::multi_ptr.0") align 8 [[P:%.*]]) local_unnamed_addr #[[ATTR0:[0-9]+]] !srcloc [[META6:![0-9]+]] !sycl_fixed_targets [[META7:![0-9]+]] {

sycl/test/multi_ptr/aliases.cpp

@@ -0,0 +1,87 @@
+//==--------------- aliases.cpp - SYCL multi_ptr aliases test --------------==//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s -Xclang -verify-ignore-unexpected=note,warning
+
+// expected-no-diagnostics
+
+#include <sycl/detail/core.hpp>
+
+#include <type_traits>
+
+template <typename T, sycl::access::decorated IsDecorated>
+void test_address_space_aliases() {
+  static_assert(std::is_same_v<
+                sycl::generic_ptr<T, IsDecorated>,
+                sycl::multi_ptr<T, sycl::access::address_space::generic_space,
+                                IsDecorated>>);
+  static_assert(std::is_same_v<
+                sycl::global_ptr<T, IsDecorated>,
+                sycl::multi_ptr<T, sycl::access ::address_space::global_space,
+                                IsDecorated>>);
+  static_assert(std::is_same_v<
+                sycl::local_ptr<T, IsDecorated>,
+                sycl::multi_ptr<T, sycl::access::address_space::local_space,
+                                IsDecorated>>);
+  static_assert(std::is_same_v<
+                sycl::private_ptr<T, IsDecorated>,
+                sycl::multi_ptr<T, sycl::access::address_space::private_space,
+                                IsDecorated>>);
+}
+
+template <typename T> void test_aliases() {
+  // Template specialization aliases for different pointer address spaces
+  test_address_space_aliases<T, sycl::access::decorated::yes>();
+  test_address_space_aliases<T, sycl::access::decorated::no>();
+  test_address_space_aliases<T, sycl::access::decorated::legacy>();
+
+  // Template specialization aliases for different pointer address spaces.
+  // The interface exposes non-decorated pointer while keeping the
+  // address space information internally.
+  static_assert(std::is_same_v<
+                sycl::raw_generic_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::generic_space,
+                                sycl::access::decorated::no>>);
+  static_assert(std::is_same_v<
+                sycl::raw_global_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::global_space,
+                                sycl::access::decorated::no>>);
+  static_assert(std::is_same_v<
+                sycl::raw_local_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::local_space,
+                                sycl::access::decorated::no>>);
+  static_assert(std::is_same_v<
+                sycl::raw_private_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::private_space,
+                                sycl::access::decorated::no>>);
+
+  // Template specialization aliases for different pointer address spaces.
+  // The interface exposes decorated pointer.
+  static_assert(std::is_same_v<
+                sycl::decorated_generic_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::generic_space,
+                                sycl::access::decorated::yes>>);
+  static_assert(std::is_same_v<
+                sycl::decorated_global_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::global_space,
+                                sycl::access::decorated::yes>>);
+  static_assert(std::is_same_v<
+                sycl::decorated_local_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::local_space,
+                                sycl::access::decorated::yes>>);
+  static_assert(std::is_same_v<
+                sycl::decorated_private_ptr<T>,
+                sycl::multi_ptr<T, sycl::access::address_space::private_space,
+                                sycl::access::decorated::yes>>);
+}
+
+// Test "minimal set of types" in the CTS. As we are just testing aliases are
+// present in this test, this should work for any type.
+
+template void test_aliases<int>();
+template void test_aliases<float>();