[SYCL] Implement get_multi_ptr for accessor and local_accessor

sycl/include/sycl/accessor.hpp

@@ -2082,6 +2082,11 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
     return constant_ptr<DataT>(getPointerAdjusted());
   }
 
+  template <access::decorated IsDecorated>
+  accessor_ptr<IsDecorated> get_multi_ptr() const noexcept {
+    return accessor_ptr<IsDecorated>(getPointerAdjusted());
+  }
+
   // accessor::has_property for runtime properties is only available in host
   // code. This restriction is not listed in the core spec and will be added in
   // future versions.
@@ -2765,6 +2770,11 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(local_accessor) local_accessor
     return const_reverse_iterator(begin());
   }
 
+  template <access::decorated IsDecorated>
+  accessor_ptr<IsDecorated> get_multi_ptr() const noexcept {
+    return accessor_ptr<IsDecorated>(local_acc::getQualifiedPtr());
+  }
+
   template <typename Property> bool has_property() const noexcept {
 #ifndef __SYCL_DEVICE_ONLY__
     return this->getPropList().template has_property<Property>();

sycl/test/basic_tests/accessor/accessor_default_ctor.cpp

@@ -11,9 +11,9 @@ int main() {
   assert(B.byte_size() == 0);
   // The return values of get_pointer() and get_multi_ptr() are unspecified.
   assert(B.get_pointer() == nullptr);
-  // TODO: uncomment check with get_multi_ptr() when SYCL 2020 mupti_ptr feature
-  // will be merged
-  // assert(B.get_multi_ptr() == nullptr);
+  assert(B.get_multi_ptr<sycl::access::decorated::yes>() == nullptr);
+  assert(B.get_multi_ptr<sycl::access::decorated::no>() == nullptr);
+  assert(B.get_multi_ptr<sycl::access::decorated::legacy>() == nullptr);
 
   return 0;
 }

sycl/test/basic_tests/accessor/accessor_multi_ptr.cpp

@@ -0,0 +1,28 @@
+// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -fsyntax-only
+
+#include <cassert>
+#include <sycl/sycl.hpp>
+#include <type_traits>
+
+using namespace sycl;
+
+constexpr static int size = 1;
+
+void test_get_multi_ptr(handler &cgh, buffer<int, size> &buffer) {
+  using accessor_t =
+      accessor<int, size, access::mode::read_write, access::target::device,
+               access::placeholder::true_t>;
+  using local_accessor_t = local_accessor<int, size>;
+
+  auto ptr = buffer.get_access<access_mode::read_write, target::device>(cgh);
+  auto local_ptr = local_accessor<int, size>({size}, cgh);
+  auto acc_multi_ptr = ptr.get_multi_ptr<access::decorated::yes>();
+  auto local_acc_multi_ptr = local_ptr.get_multi_ptr<access::decorated::yes>();
+  static_assert(
+      std::is_same_v<
+          decltype(acc_multi_ptr),
+          typename accessor_t::template accessor_ptr<access::decorated::yes>>);
+  static_assert(std::is_same_v<decltype(local_acc_multi_ptr),
+                               typename local_accessor_t::template accessor_ptr<
+                                   access::decorated::yes>>);
+}

sycl/unittests/accessor/LocalAccessorDefaultCtor.cpp

@@ -37,8 +37,10 @@ TEST(LocalAccessorDefaultCtorTest, LocalAcessorDefaultCtorPtrQueries) {
   // unspecified. Just check they can run without any issue.
   auto ptr = acc.get_pointer();
   (void)ptr;
-  // TODO: uncomment check with get_multi_ptr() when SYCL 2020 multi_ptr feature
-  // will be merged
-  // auto multi_ptr = acc.get_multi_ptr();
-  // (void)multi_ptr;
+  auto multi_ptr = acc.get_multi_ptr<access::decorated::yes>();
+  (void)multi_ptr;
+  auto multi_ptr_no_decorated = acc.get_multi_ptr<access::decorated::no>();
+  (void)multi_ptr_no_decorated;
+  auto multi_ptr_legacy = acc.get_multi_ptr<access::decorated::legacy>();
+  (void)multi_ptr_legacy;
 }