[flang-commits] [flang] [flang][runtime] Fixed flang-runtime-cuda-gcc builder after af964c7. (PR #85144)

[Slava Zakharin via flang-commits]

https://github.com/vzakhari created https://github.com/llvm/llvm-project/pull/85144

None

>From 2d98654c6a0e26699c4694377076b5b50b7fde11 Mon Sep 17 00:00:00 2001
From: Slava Zakharin <szakharin at nvidia.com>
Date: Wed, 13 Mar 2024 15:04:49 -0700
Subject: [PATCH] [flang][runtime] Fixed flang-runtime-cuda-gcc builder after
 af964c7.

---
 flang/include/flang/Runtime/api-attrs.h | 12 ++++++++++++
 flang/runtime/environment.cpp           |  4 +++-
 flang/runtime/environment.h             | 12 ++++++++++--
 3 files changed, 25 insertions(+), 3 deletions(-)

diff --git a/flang/include/flang/Runtime/api-attrs.h b/flang/include/flang/Runtime/api-attrs.h
index 9c8a67ffc34a82..fc3eb42e1b73f5 100644
--- a/flang/include/flang/Runtime/api-attrs.h
+++ b/flang/include/flang/Runtime/api-attrs.h
@@ -109,6 +109,18 @@
 #endif
 #endif /* !defined(RT_CONST_VAR_ATTRS) */
 
+/*
+ * RT_VAR_ATTRS is marking non-const/constexpr module scope variables
+ * referenced by Flang runtime.
+ */
+#ifndef RT_VAR_ATTRS
+#if (defined(__CUDACC__) || defined(__CUDA__)) && defined(__CUDA_ARCH__)
+#define RT_VAR_ATTRS __device__
+#else
+#define RT_VAR_ATTRS
+#endif
+#endif /* !defined(RT_VAR_ATTRS) */
+
 /*
  * RT_DEVICE_COMPILATION is defined for any device compilation.
  * Note that it can only be used reliably with compilers that perform
diff --git a/flang/runtime/environment.cpp b/flang/runtime/environment.cpp
index 29196ae8f31051..fe6701d72c9ff5 100644
--- a/flang/runtime/environment.cpp
+++ b/flang/runtime/environment.cpp
@@ -23,7 +23,9 @@ extern char **environ;
 
 namespace Fortran::runtime {
 
-ExecutionEnvironment executionEnvironment;
+RT_OFFLOAD_VAR_GROUP_BEGIN
+RT_VAR_ATTRS ExecutionEnvironment executionEnvironment;
+RT_OFFLOAD_VAR_GROUP_END
 
 static void SetEnvironmentDefaults(const EnvironmentDefaultList *envDefaults) {
   if (!envDefaults) {
diff --git a/flang/runtime/environment.h b/flang/runtime/environment.h
index 6da2c7bb3cf78a..49c7dbd2940f8c 100644
--- a/flang/runtime/environment.h
+++ b/flang/runtime/environment.h
@@ -10,6 +10,7 @@
 #define FORTRAN_RUNTIME_ENVIRONMENT_H_
 
 #include "flang/Decimal/decimal.h"
+#include "flang/Runtime/api-attrs.h"
 #include <optional>
 
 struct EnvironmentDefaultList;
@@ -32,7 +33,11 @@ enum class Convert { Unknown, Native, LittleEndian, BigEndian, Swap };
 std::optional<Convert> GetConvertFromString(const char *, std::size_t);
 
 struct ExecutionEnvironment {
-  constexpr ExecutionEnvironment(){};
+#if !defined(_OPENMP)
+  // FIXME: https://github.com/llvm/llvm-project/issues/84942
+  constexpr
+#endif
+      ExecutionEnvironment(){};
   void Configure(int argc, const char *argv[], const char *envp[],
       const EnvironmentDefaultList *envDefaults);
   const char *GetEnv(
@@ -51,7 +56,10 @@ struct ExecutionEnvironment {
   bool checkPointerDeallocation{true}; // FORT_CHECK_POINTER_DEALLOCATION
 };
 
-extern ExecutionEnvironment executionEnvironment;
+RT_OFFLOAD_VAR_GROUP_BEGIN
+extern RT_VAR_ATTRS ExecutionEnvironment executionEnvironment;
+RT_OFFLOAD_VAR_GROUP_END
+
 } // namespace Fortran::runtime
 
 #endif // FORTRAN_RUNTIME_ENVIRONMENT_H_