[OpenMP] Remove the `LibC` subdirectory from the DeviceRTL

[jhuber6]

Summary

This is redundant now that these functions are being implemented in the
GPU libc project https://libc.llvm.org/gpu/. Remove this to keep it
separate.

We do not need to do this immediately as it may regress for some people
until we do some more tests on including the libc in the upcoming
release.

[tianshilei1992]

Let's defer the land of this patch until libc is enabled by default for OpenMP target offloading.

[jhuber6]

I don't know if we need to wait until it's the default, since this is probably a
niche usage and doesn't really belong here anyway. We can probably document how
to set this up, but we probably should wait until after the fork probably.

openmp/libomptarget/DeviceRTL/CMakeLists.txt

set(LIBOMPTARGET_DEVICE_ARCHITECTURES
       "${LIBOMPTARGET_NVPTX_DETECTED_ARCH_LIST};${LIBOMPTARGET_AMDGPU_DETECTED_ARCH_LIST}")
 endif()
 list(REMOVE_DUPLICATES LIBOMPTARGET_DEVICE_ARCHITECTURES)
 
 set(include_files
   ${include_directory}/Configuration.h
   ${include_directory}/Debug.h
   ${include_directory}/Interface.h
-  ${include_directory}/LibC.h
   ${include_directory}/Mapping.h
   ${include_directory}/State.h
   ${include_directory}/Synchronization.h
   ${include_directory}/Types.h
   ${include_directory}/Utils.h
 )
 
 set(src_files
   ${source_directory}/Configuration.cpp
   ${source_directory}/Debug.cpp
   ${source_directory}/Kernel.cpp
-  ${source_directory}/LibC.cpp
   ${source_directory}/Mapping.cpp
   ${source_directory}/Misc.cpp
   ${source_directory}/Parallelism.cpp
   ${source_directory}/Reduction.cpp
   ${source_directory}/State.cpp
   ${source_directory}/Synchronization.cpp
   ${source_directory}/Tasking.cpp
   ${source_directory}/Utils.cpp

openmp/libomptarget/DeviceRTL/include/Debug.h

 //===-------- Debug.h ---- Debug utilities ------------------------ C++ -*-===//
 //
 // 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
 //
 //===----------------------------------------------------------------------===//
 //
 //
 //===----------------------------------------------------------------------===//
 
 #ifndef OMPTARGET_DEVICERTL_DEBUG_H
 #define OMPTARGET_DEVICERTL_DEBUG_H
 
 #include "Configuration.h"
-#include "LibC.h"
 
 /// Assertion
 ///
 /// {
 extern "C" {
+int printf(const char *format, ...);
 void __assert_assume(bool condition);
 void __assert_fail(const char *assertion, const char *file, unsigned line,
                    const char *function);
 }
 
 #define ASSERT(expr)                                                           \
   {                                                                            \
     if (config::isDebugMode(config::DebugKind::Assertion) && !(expr))          \

openmp/libomptarget/DeviceRTL/include/LibC.h

-//===--------- LibC.h - Simple implementation of libc functions --- C++ -*-===//
-//
-// 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
-//
-//===----------------------------------------------------------------------===//
-//
-//
-//===----------------------------------------------------------------------===//
-
-#ifndef OMPTARGET_LIBC_H
-#define OMPTARGET_LIBC_H
-
-#include "Types.h"
-
-extern "C" {
-
-int memcmp(const void *lhs, const void *rhs, size_t count);
-
-int printf(const char *format, ...);
-}
-
-#endif

openmp/libomptarget/DeviceRTL/src/LibC.cpp

-//===------- LibC.cpp - Simple implementation of libc functions --- C++ ---===//
-//
-// 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
-//
-//===----------------------------------------------------------------------===//
-
-#include "LibC.h"
-
-#pragma omp begin declare target device_type(nohost)
-
-namespace impl {
-int32_t omp_vprintf(const char *Format, void *Arguments, uint32_t);
-}
-
-#pragma omp begin declare variant match(                                       \
-        device = {arch(nvptx, nvptx64)},                                       \
-            implementation = {extension(match_any)})
-extern "C" int32_t vprintf(const char *, void *);
-namespace impl {
-int32_t omp_vprintf(const char *Format, void *Arguments, uint32_t) {
-  return vprintf(Format, Arguments);
-}
-} // namespace impl
-#pragma omp end declare variant
-
-// We do not have a vprintf implementation for AMD GPU yet so we use a stub.
-#pragma omp begin declare variant match(device = {arch(amdgcn)})
-namespace impl {
-int32_t omp_vprintf(const char *Format, void *Arguments, uint32_t) {
-  return -1;
-}
-} // namespace impl
-#pragma omp end declare variant
-
-extern "C" {
-
-int memcmp(const void *lhs, const void *rhs, size_t count) {
-  auto *L = reinterpret_cast<const unsigned char *>(lhs);
-  auto *R = reinterpret_cast<const unsigned char *>(rhs);
-
-  for (size_t I = 0; I < count; ++I)
-    if (L[I] != R[I])
-      return (int)L[I] - (int)R[I];
-
-  return 0;
-}
-
-/// printf() calls are rewritten by CGGPUBuiltin to __llvm_omp_vprintf
-int32_t __llvm_omp_vprintf(const char *Format, void *Arguments, uint32_t Size) {
-  return impl::omp_vprintf(Format, Arguments, Size);
-}
-}
-
-#pragma omp end declare target

openmp/libomptarget/DeviceRTL/src/Misc.cpp

int32_t __kmpc_cancel(IdentTy *, int32_t, int32_t) {
   return 0;
 }
 
 double omp_get_wtick(void) { return ompx::impl::getWTick(); }
 
 double omp_get_wtime(void) { return ompx::impl::getWTime(); }
 }
 
+namespace impl {
+int32_t omp_vprintf(const char *Format, void *Arguments, uint32_t);
+}
+
+#pragma omp begin declare variant match(                                       \
+        device = {arch(nvptx, nvptx64)},                                       \
+            implementation = {extension(match_any)})
+extern "C" int32_t vprintf(const char *, void *);
+namespace impl {
+int32_t omp_vprintf(const char *Format, void *Arguments, uint32_t) {
+  return vprintf(Format, Arguments);
+}
+} // namespace impl
+#pragma omp end declare variant
+
+// We do not have a vprintf implementation for AMD GPU yet so we use a stub.
+#pragma omp begin declare variant match(device = {arch(amdgcn)})
+namespace impl {
+int32_t omp_vprintf(const char *Format, void *Arguments, uint32_t) {
+  return -1;
+}
+} // namespace impl
+#pragma omp end declare variant
+
+extern "C" {
+
+/// printf() calls are rewritten by CGGPUBuiltin to __llvm_omp_vprintf.
+int32_t __llvm_omp_vprintf(const char *Format, void *Arguments, uint32_t Size) {
+  return impl::omp_vprintf(Format, Arguments, Size);
+}
+}
+
 ///}
 #pragma omp end declare target

openmp/libomptarget/test/libc/equal.cpp

+// RUN: %libomptarget-compilexx-run-and-check-generic
+
+// REQUIRES: libc
+
+#include <algorithm>
+
+extern "C" int printf(const char *, ...);
+
+// std::equal is lowered to the libc function memcmp.
+void test_memcpy() {
+  int r = 0;
+#pragma omp target map(from : r)
+  {
+    int x[2] = {0, 0};
+    int y[2] = {0, 0};
+    int z[2] = {0, 1};
+    bool eq1 = std::equal(x, x + 2, y);
+    bool eq2 = std::equal(x, x + 2, z);
+    r = eq1 && !eq2;
+  }
+  // CHECK: memcmp: PASS
+  printf("memcmp: %s\n", r ? "PASS" : "FAIL");
+}
+
+int main(int argc, char *argv[]) { test_memcpy(); }