diff --git a/libc/src/time/gpu/time_utils.cpp b/libc/src/time/gpu/time_utils.cpp index 935a5394e5fdb..300a72b102360 100644 --- a/libc/src/time/gpu/time_utils.cpp +++ b/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 diff --git a/libc/src/time/gpu/time_utils.h b/libc/src/time/gpu/time_utils.h index 075fcb61768ec..53548181e17e2 100644 --- a/libc/src/time/gpu/time_utils.h +++ b/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(LIBC_NAMESPACE_clock_freq) + [[clang::address_space(4)]] __llvm_libc_clock_freq; +#define GPU_CLOCKS_PER_SEC static_cast(__llvm_libc_clock_freq) #elif defined(LIBC_TARGET_ARCH_IS_NVPTX) // NPVTX uses a single 1 GHz fixed frequency clock for all target architectures. diff --git a/libc/startup/gpu/amdgpu/start.cpp b/libc/startup/gpu/amdgpu/start.cpp index 9e642dbe3fc7d..89b0be208d2e7 100644 --- a/libc/startup/gpu/amdgpu/start.cpp +++ b/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[]; diff --git a/libc/test/UnitTest/LibcTest.cpp b/libc/test/UnitTest/LibcTest.cpp index b01f40f052646..5c27e2e69bead 100644 --- a/libc/test/UnitTest/LibcTest.cpp +++ b/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; } diff --git a/libc/utils/gpu/loader/amdgpu/Loader.cpp b/libc/utils/gpu/loader/amdgpu/Loader.cpp index 1fb67e841cfec..1d0247a6dc5dc 100644 --- a/libc/utils/gpu/loader/amdgpu/Loader.cpp +++ b/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;