[libc] Fix accidental LIBC_NAMESPACE_clock_freq

libc/src/time/gpu/time_utils.cpp

@@ -16,7 +16,7 @@ namespace LIBC_NAMESPACE {
 // TODO: Once we have another use-case for this we should put it in a common
 // device environment struct.
 extern "C" [[gnu::visibility("protected")]] uint64_t
-    [[clang::address_space(4)]] LIBC_NAMESPACE_clock_freq = clock_freq;
+    [[clang::address_space(4)]] __llvm_libc_clock_freq = clock_freq;
 #endif
 
 } // namespace LIBC_NAMESPACE

libc/src/time/gpu/time_utils.h

@@ -39,8 +39,8 @@ constexpr uint64_t clock_freq = 0;
 // We provide an externally visible symbol such that the runtime can set this to
 // the correct value. If it is not set we try to default to the known values.
 extern "C" [[gnu::visibility("protected")]] uint64_t
-    [[clang::address_space(4)]] LIBC_NAMESPACE_clock_freq;
-#define GPU_CLOCKS_PER_SEC static_cast<clock_t>(LIBC_NAMESPACE_clock_freq)
+    [[clang::address_space(4)]] __llvm_libc_clock_freq;
+#define GPU_CLOCKS_PER_SEC static_cast<clock_t>(__llvm_libc_clock_freq)
 
 #elif defined(LIBC_TARGET_ARCH_IS_NVPTX)
 // NPVTX uses a single 1 GHz fixed frequency clock for all target architectures.

libc/startup/gpu/amdgpu/start.cpp

@@ -19,7 +19,7 @@ namespace LIBC_NAMESPACE {
 // real time. However, the frequency of this clock varies between cards and can
 // only be obtained via the driver. The loader will set this so we can use it.
 extern "C" [[gnu::visibility("protected")]] uint64_t
-    [[clang::address_space(4)]] LIBC_NAMESPACE_clock_freq = 0;
+    [[clang::address_space(4)]] __llvm_libc_clock_freq = 0;
 
 extern "C" uintptr_t __init_array_start[];
 extern "C" uintptr_t __init_array_end[];

libc/test/UnitTest/LibcTest.cpp

@@ -23,8 +23,8 @@ static long clock() { return LIBC_NAMESPACE::gpu::fixed_frequency_clock(); }
 #else
 // The AMDGPU loader needs to initialize this at runtime by querying the driver.
 extern "C" [[gnu::visibility("protected")]] uint64_t
-    [[clang::address_space(4)]] LIBC_NAMESPACE_clock_freq;
-#define CLOCKS_PER_SEC LIBC_NAMESPACE_clock_freq
+    [[clang::address_space(4)]] __llvm_libc_clock_freq;
+#define CLOCKS_PER_SEC __llvm_libc_clock_freq
 #endif
 #else
 static long clock() { return 0; }

libc/utils/gpu/loader/amdgpu/Loader.cpp

@@ -477,7 +477,7 @@ int load(int argc, char **argv, char **envp, void *image, size_t size,
   // If the clock_freq symbol is missing, no work to do.
   hsa_executable_symbol_t freq_sym;
   if (HSA_STATUS_SUCCESS ==
-      hsa_executable_get_symbol_by_name(executable, "LIBC_NAMESPACE_clock_freq",
+      hsa_executable_get_symbol_by_name(executable, "__llvm_libc_clock_freq",
                                         &dev_agent, &freq_sym)) {
 
     void *host_clock_freq;