[libomptarget][nfc] Move non-freestanding headers out of common

[JonChesterfield]

Summary

[libomptarget][nfc] Move non-freestanding headers out of common

Lowers the bar for building deviceRTL.
Drops math.h entirely as it wasn't used and libm is a big dependency.

[JonChesterfield]

openmp/libomptarget/deviceRTLs/amdgcn/src/target_impl.h

21

No stdio in amdgcn's target_impl as the only function used from it is printf, declared at the end of this header.

Can reduce amdgcn to wholly ffreestanding by providing an implementation of assert and two macros from inttypes, which may be a good idea as I think HIP is currently using the glibc versions of these headers for the device code.

[jdoerfert]

openmp/libomptarget/deviceRTLs/amdgcn/src/target_impl.h

21

I think this seems reasonable, depending on the macros.

[JonChesterfield]

openmp/libomptarget/deviceRTLs/amdgcn/src/target_impl.h

21

The macros are PRId64 and PRIu64 iirc, which expand to "ld"/"lu". Pretty minor. For a later patch anyway.

This reduces the requirements on common code to freestanding, target_impl functions, printf, assert, two print format macros. It's still quite a lot, but much less than stdlib + math suggests.

[jdoerfert]

LGTM. Let's look at freestanding later.

openmp/libomptarget/deviceRTLs/amdgcn/src/target_impl.h

 #define OMPTARGET_AMDGCN_TARGET_IMPL_H
 
 #ifndef __AMDGCN__
 #error "amdgcn target_impl.h expects to be compiled under __AMDGCN__"
 #endif
 
 #include "amdgcn_interface.h"
 
+#include <assert.h>
+#include <inttypes.h>
 #include <stddef.h>
 #include <stdint.h>
 
 #define DEVICE __attribute__((device))
 #define INLINE inline DEVICE
 #define NOINLINE __attribute__((noinline)) DEVICE
 #define SHARED __attribute__((shared))
 #define ALIGN(N) __attribute__((aligned(N)))

openmp/libomptarget/deviceRTLs/common/debug.h

 #define OMPTARGET_NVPTX_WARNING (LW_SET_FUSSY)
 #endif
 
 ////////////////////////////////////////////////////////////////////////////////
 // implementation for debug
 ////////////////////////////////////////////////////////////////////////////////
 
 #if OMPTARGET_NVPTX_DEBUG || OMPTARGET_NVPTX_TEST || OMPTARGET_NVPTX_WARNING
-#include <stdio.h>
 #include "common/support.h"
 
 template <typename... Arguments>
 NOINLINE static void log(const char *fmt, Arguments... parameters) {
   printf(fmt, (int)GetBlockIdInKernel(), (int)GetThreadIdInBlock(),
          (int)GetWarpId(), (int)GetLaneId(), parameters...);
 }
 
 #endif
 #if OMPTARGET_NVPTX_TEST
-#include <assert.h>
 
 template <typename... Arguments>
 NOINLINE static void check(bool cond, const char *fmt,
                            Arguments... parameters) {
   if (!cond)
     printf(fmt, (int)GetBlockIdInKernel(), (int)GetThreadIdInBlock(),
            (int)GetWarpId(), (int)GetLaneId(), parameters...);
   assert(cond);

openmp/libomptarget/deviceRTLs/common/omptarget.h

 //===---- omptarget.h - OpenMP GPU initialization ---------------- CUDA -*-===//
 //
 // 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
 //
 //===----------------------------------------------------------------------===//
 //
 // This file contains the declarations of all library macros, types,
 // and functions.
 //
 //===----------------------------------------------------------------------===//
 
 #ifndef OMPTARGET_H
 #define OMPTARGET_H
 
-// std includes
-#include <inttypes.h>
-#include <math.h>
-
-// local includes
 #include "target_impl.h"
 #include "common/debug.h"     // debug
 #include "interface.h" // interfaces with omp, compiler, and user
 #include "common/state-queue.h"
 #include "common/support.h"
 
 #define OMPTARGET_NVPTX_VERSION 1.1
 

openmp/libomptarget/deviceRTLs/common/src/data_sharing.cu

 //===----- data_sharing.cu - OpenMP GPU data sharing ------------- CUDA -*-===//
 //
 // 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
 //
 //===----------------------------------------------------------------------===//
 //
 // This file contains the implementation of data sharing environments
 //
 //===----------------------------------------------------------------------===//
 #include "common/omptarget.h"
 #include "target_impl.h"
-#include <stdio.h>
 
 // Return true if this is the master thread.
 INLINE static bool IsMasterThread(bool isSPMDExecutionMode) {
   return !isSPMDExecutionMode && GetMasterThreadID() == GetThreadIdInBlock();
 }
 
 /// Return the provided size aligned to the size of a pointer.
 INLINE static size_t AlignVal(size_t Val) {

openmp/libomptarget/deviceRTLs/nvptx/src/target_impl.h

 //===------------ target_impl.h - NVPTX OpenMP GPU options ------- CUDA -*-===//
 //
 // 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
 //
 //===----------------------------------------------------------------------===//
 //
 // Definitions of target specific functions
 //
 //===----------------------------------------------------------------------===//
 #ifndef _TARGET_IMPL_H_
 #define _TARGET_IMPL_H_
 
+#include <assert.h>
 #include <cuda.h>
+#include <inttypes.h>
+#include <stdio.h>
 #include <stdlib.h>
 
 #include "nvptx_interface.h"
 
 #define DEVICE __device__
 #define INLINE __forceinline__ DEVICE
 #define NOINLINE __noinline__ DEVICE
 #define SHARED __shared__