From c9962c0113f016a1b82a5fe6c046b103c407b3b5 Mon Sep 17 00:00:00 2001
From: Nicolas Miller <nicolas.miller@codeplay.com>
Date: Fri, 21 Oct 2022 11:50:20 -0400
Subject: [PATCH] [SYCL][HIP] Add interop header and devcie specialization

---
 sycl/include/sycl/ext/oneapi/backend/hip.hpp | 31 ++++++++++
 sycl/test/basic_tests/interop-hip.cpp        | 60 ++++++++++++++++++++
 2 files changed, 91 insertions(+)
 create mode 100644 sycl/include/sycl/ext/oneapi/backend/hip.hpp
 create mode 100644 sycl/test/basic_tests/interop-hip.cpp

diff --git a/sycl/include/sycl/ext/oneapi/backend/hip.hpp b/sycl/include/sycl/ext/oneapi/backend/hip.hpp
new file mode 100644
index 0000000000000..35a8ae078e2a8
--- /dev/null
+++ b/sycl/include/sycl/ext/oneapi/backend/hip.hpp
@@ -0,0 +1,31 @@
+//==--------- hip.hpp - SYCL HIP backend -----------------------------------==//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+#pragma once
+
+#include <sycl/backend.hpp>
+
+namespace sycl {
+__SYCL_INLINE_VER_NAMESPACE(_V1) {
+
+template <>
+inline backend_return_t<backend::ext_oneapi_hip, device>
+get_native<backend::ext_oneapi_hip, device>(const device &Obj) {
+  // TODO use SYCL 2020 exception when implemented
+  if (Obj.get_backend() != backend::ext_oneapi_hip) {
+    throw sycl::runtime_error(errc::backend_mismatch, "Backends mismatch",
+                              PI_ERROR_INVALID_OPERATION);
+  }
+  // HIP uses a 32-bit int instead of an opaque pointer like other backends,
+  // so we need a specialization with static_cast instead of reinterpret_cast.
+  return static_cast<backend_return_t<backend::ext_oneapi_hip, device>>(
+      Obj.getNative());
+}
+
+} // __SYCL_INLINE_VER_NAMESPACE(_V1)
+} // namespace sycl
diff --git a/sycl/test/basic_tests/interop-hip.cpp b/sycl/test/basic_tests/interop-hip.cpp
new file mode 100644
index 0000000000000..32019bcff147d
--- /dev/null
+++ b/sycl/test/basic_tests/interop-hip.cpp
@@ -0,0 +1,60 @@
+// REQUIRES: hip_be
+// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out
+// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note -D__SYCL_INTERNAL_API %s -o %t.out
+// expected-no-diagnostics
+
+// Test for HIP interop API
+
+#include <sycl/sycl.hpp>
+#include <sycl/ext/oneapi/backend/hip.hpp>
+
+using namespace sycl;
+
+//
+// 4.5.1 SYCL application interoperability may be provided for
+// platform,
+// device,
+// context,
+// queue,
+// event,
+// buffer,
+// device_image,
+// sampled_image,
+// unsampled_image.
+
+int main() {
+
+  // Create SYCL objects
+  device Device;
+  context Context(Device);
+  queue Queue(Device);
+  event Event;
+
+  // 4.5.1.1 For each SYCL runtime class T which supports SYCL application
+  // interoperability with the SYCL backend, a specialization of return_type
+  // must be defined as the type of SYCL application interoperability native
+  // backend object associated with T for the SYCL backend, specified in the
+  // SYCL backend specification.
+  //
+  // return_type is used when retrieving the backend specific native object from
+  // a SYCL object. See the relevant backend specification for details.
+
+  backend_traits<backend::ext_oneapi_hip>::return_type<device> hip_device;
+  backend_traits<backend::ext_oneapi_hip>::return_type<context> hip_context;
+  backend_traits<backend::ext_oneapi_hip>::return_type<event> hip_event;
+  backend_traits<backend::ext_oneapi_hip>::return_type<queue> hip_queue;
+
+  // 4.5.1.2 For each SYCL runtime class T which supports SYCL application
+  // interoperability, a specialization of get_native must be defined, which
+  // takes an instance of T and returns a SYCL application interoperability
+  // native backend object associated with syclObject which can be used for SYCL
+  // application interoperability. The lifetime of the object returned are
+  // backend-defined and specified in the backend specification.
+
+  hip_device = get_native<backend::ext_oneapi_hip>(Device);
+  hip_context = get_native<backend::ext_oneapi_hip>(Context);
+  hip_event = get_native<backend::ext_oneapi_hip>(Event);
+  hip_queue = get_native<backend::ext_oneapi_hip>(Queue);
+
+  return 0;
+}