8 files changed, 152 insertions, 0 deletions
diff --git a/Tests/HIP/MixedLanguage/CMakeLists.txt b/Tests/HIP/MixedLanguage/CMakeLists.txt
new file mode 100644
index 000000000..4f6dd3b0f
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/CMakeLists.txt
@@ -0,0 +1,19 @@
+cmake_minimum_required(VERSION 3.18)
+project (MixedLanguage C CXX HIP)
+
+set(CMAKE_HIP_STANDARD 14)
+set(CMAKE_CXX_STANDARD 14)
+
+#Goal for this example:
+#make sure that we can build multiple languages into targets
+#and have the link language always be HIP
+add_library(MixedSharedLib SHARED shared.c)
+add_library(MixedObjectLib OBJECT shared.cxx shared.hip)
+set_target_properties(MixedObjectLib PROPERTIES POSITION_INDEPENDENT_CODE ON)
+target_link_libraries(MixedSharedLib PRIVATE MixedObjectLib)
+
+add_library(MixedStaticLib STATIC static.c static.cxx static.hip)
+set_target_properties(MixedStaticLib PROPERTIES POSITION_INDEPENDENT_CODE ON)
+
+add_executable(HIPMixedLanguage main.cxx)
+target_link_libraries(HIPMixedLanguage PRIVATE MixedStaticLib MixedSharedLib)
diff --git a/Tests/HIP/MixedLanguage/main.cxx b/Tests/HIP/MixedLanguage/main.cxx
new file mode 100644
index 000000000..003d18ac4
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/main.cxx
@@ -0,0 +1,40 @@
+
+#include <iostream>
+
+#ifdef __HIP_PLATFORM_HCC__
+#  error "__HIP_PLATFORM_HCC__ propagated to C++ compilation!"
+#endif
+
+#ifdef __HIP_ROCclr__
+#  error "__HIP_ROCclr__ propagated to C++ compilation!"
+#endif
+
+#ifdef _WIN32
+#  define IMPORT __declspec(dllimport)
+#else
+#  define IMPORT
+#endif
+
+extern "C" {
+IMPORT int shared_c_func(int);
+int static_c_func(int);
+}
+
+IMPORT int shared_cxx_func(int);
+IMPORT int shared_hip_func(int);
+
+int static_cxx_func(int);
+int static_hip_func(int);
+
+int main(int argc, char** argv)
+{
+  static_c_func(int(42));
+  static_cxx_func(int(42));
+  static_hip_func(int(42));
+
+  shared_c_func(int(42));
+  shared_cxx_func(int(42));
+  shared_hip_func(int(42));
+
+  return 0;
+}
diff --git a/Tests/HIP/MixedLanguage/shared.c b/Tests/HIP/MixedLanguage/shared.c
new file mode 100644
index 000000000..347fba925
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/shared.c
@@ -0,0 +1,12 @@
+
+
+#ifdef _WIN32
+#  define EXPORT __declspec(dllexport)
+#else
+#  define EXPORT
+#endif
+
+EXPORT int shared_c_func(int x)
+{
+  return -x;
+}
diff --git a/Tests/HIP/MixedLanguage/shared.cxx b/Tests/HIP/MixedLanguage/shared.cxx
new file mode 100644
index 000000000..8e6c1d35f
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/shared.cxx
@@ -0,0 +1,21 @@
+
+#include <type_traits>
+
+#ifdef __HIP_PLATFORM_HCC__
+#  error "__HIP_PLATFORM_HCC__ propagated to C++ compilation!"
+#endif
+
+#ifdef __HIP_ROCclr__
+#  error "__HIP_ROCclr__ propagated to C++ compilation!"
+#endif
+
+#ifdef _WIN32
+#  define EXPORT __declspec(dllexport)
+#else
+#  define EXPORT
+#endif
+
+EXPORT int shared_cxx_func(int x)
+{
+  return x * x + std::integral_constant<int, 14>::value;
+}
diff --git a/Tests/HIP/MixedLanguage/shared.hip b/Tests/HIP/MixedLanguage/shared.hip
new file mode 100644
index 000000000..e6fea9fe2
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/shared.hip
@@ -0,0 +1,26 @@
+#include <system_error>
+#include <type_traits>
+#include <hip/hip_runtime_api.h>
+
+#ifdef _WIN32
+#  define EXPORT __declspec(dllexport)
+#else
+#  define EXPORT
+#endif
+
+
+static __global__ void fake_hip_kernel()
+{
+}
+
+
+int __host__ shared_hip_func(int x)
+{
+
+  fake_hip_kernel<<<1, 1>>>();
+  bool valid = (hipSuccess == hipGetLastError());
+  if (!valid) {
+    throw std::system_error(ENODEV, std::generic_category(), "no hip device");
+  }
+  return x * x + std::integral_constant<int, 17>::value;
+}
diff --git a/Tests/HIP/MixedLanguage/static.c b/Tests/HIP/MixedLanguage/static.c
new file mode 100644
index 000000000..06c33b4cc
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/static.c
@@ -0,0 +1,6 @@
+
+
+int static_c_func(int x)
+{
+  return -x;
+}
diff --git a/Tests/HIP/MixedLanguage/static.cxx b/Tests/HIP/MixedLanguage/static.cxx
new file mode 100644
index 000000000..2c14fb101
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/static.cxx
@@ -0,0 +1,7 @@
+
+#include <type_traits>
+
+int static_cxx_func(int x)
+{
+  return x * x + std::integral_constant<int, 14>::value;
+}
diff --git a/Tests/HIP/MixedLanguage/static.hip b/Tests/HIP/MixedLanguage/static.hip
new file mode 100644
index 000000000..359b9facd
--- /dev/null
+++ b/Tests/HIP/MixedLanguage/static.hip
@@ -0,0 +1,21 @@
+
+#include <type_traits>
+#include <system_error>
+#include <hip/hip_runtime_api.h>
+
+
+static __global__ void fake_hip_kernel()
+{
+}
+
+
+int __host__ static_hip_func(int x)
+{
+
+  fake_hip_kernel<<<1, 1>>>();
+  bool valid = (hipSuccess == hipGetLastError());
+  if (!valid) {
+    throw std::system_error(ENODEV, std::generic_category(), "no hip device");
+  }
+  return x * x + std::integral_constant<int, 17>::value;
+}