diff --git a/libc/config/gpu/api.td b/libc/config/gpu/api.td --- a/libc/config/gpu/api.td +++ b/libc/config/gpu/api.td @@ -1,6 +1,7 @@ include "config/public_api.td" include "spec/stdc.td" +include "spec/posix.td" include "spec/gpu_ext.td" def StringAPI : PublicAPI<"string.h"> { @@ -38,5 +39,7 @@ def TimeAPI : PublicAPI<"time.h"> { let Types = [ "clock_t", + "time_t", + "struct timespec", ]; } diff --git a/libc/config/gpu/entrypoints.txt b/libc/config/gpu/entrypoints.txt --- a/libc/config/gpu/entrypoints.txt +++ b/libc/config/gpu/entrypoints.txt @@ -98,6 +98,7 @@ # time.h entrypoints libc.src.time.clock + libc.src.time.nanosleep # gpu/rpc.h entrypoints libc.src.gpu.rpc_reset diff --git a/libc/docs/gpu/support.rst b/libc/docs/gpu/support.rst --- a/libc/docs/gpu/support.rst +++ b/libc/docs/gpu/support.rst @@ -136,4 +136,5 @@ Function Name Available RPC Required ============= ========= ============ clock |check| +nanosleep |check| ============= ========= ============ diff --git a/libc/src/time/CMakeLists.txt b/libc/src/time/CMakeLists.txt --- a/libc/src/time/CMakeLists.txt +++ b/libc/src/time/CMakeLists.txt @@ -106,19 +106,6 @@ libc.src.errno.errno ) -add_entrypoint_object( - nanosleep - SRCS - nanosleep.cpp - HDRS - nanosleep.h - DEPENDS - libc.include.time - libc.include.sys_syscall - libc.src.__support.OSUtil.osutil - libc.src.errno.errno -) - add_entrypoint_object( time ALIAS @@ -132,3 +119,10 @@ DEPENDS .${LIBC_TARGET_OS}.clock ) + +add_entrypoint_object( + nanosleep + ALIAS + DEPENDS + .${LIBC_TARGET_OS}.nanosleep +) diff --git a/libc/src/time/gpu/CMakeLists.txt b/libc/src/time/gpu/CMakeLists.txt --- a/libc/src/time/gpu/CMakeLists.txt +++ b/libc/src/time/gpu/CMakeLists.txt @@ -17,3 +17,15 @@ libc.src.__support.GPU.utils .time_utils ) + +add_entrypoint_object( + nanosleep + SRCS + nanosleep.cpp + HDRS + ../nanosleep.h + DEPENDS + libc.include.time + libc.src.__support.GPU.utils + .time_utils +) diff --git a/libc/src/time/gpu/nanosleep.cpp b/libc/src/time/gpu/nanosleep.cpp new file mode 100644 --- /dev/null +++ b/libc/src/time/gpu/nanosleep.cpp @@ -0,0 +1,55 @@ +//===-- GPU implementation of the nanosleep function ----------------------===// +// +// 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 "src/time/nanosleep.h" + +#include "time_utils.h" + +namespace __llvm_libc { + +constexpr uint64_t TICKS_PER_NS = 1000000000UL; + +LLVM_LIBC_FUNCTION(int, nanosleep, + (const struct timespec *req, struct timespec *rem)) { + if (!GPU_CLOCKS_PER_SEC || !req) + return -1; + + uint64_t nsecs = req->tv_nsec + req->tv_sec * TICKS_PER_NS; + + uint64_t start = gpu::fixed_frequency_clock(); +#if defined(LIBC_TARGET_ARCH_IS_NVPTX) + LIBC_INLINE_ASM("nanosleep.u32 %0;" ::"r"(nsecs)); +#elif defined(LIBC_TARGET_ARCH_IS_AMDGPU) + uint64_t end = start + nsecs / (TICKS_PER_NS / GPU_CLOCKS_PER_SEC); + uint64_t cur = gpu::fixed_frequency_clock(); + // The AMDGPU architecture does not provide a sleep implementation with a + // known delay so we simply repeatedly sleep with a large value of ~960 clock + // cycles and check until we've passed the time using the known frequency. + __builtin_amdgcn_s_sleep(2); + while (cur < end) { + __builtin_amdgcn_s_sleep(15); + cur = gpu::fixed_frequency_clock(); + } +#endif + uint64_t stop = gpu::fixed_frequency_clock(); + + // Check to make sure we slept for at least the desired duration and set the + // remaining time if not. + uint64_t elapsed = (stop - start) * (TICKS_PER_NS / GPU_CLOCKS_PER_SEC); + if (elapsed < nsecs) { + if (rem) { + rem->tv_sec = (nsecs - elapsed) / TICKS_PER_NS; + rem->tv_nsec = (nsecs - elapsed) % TICKS_PER_NS; + } + return -1; + } + + return 0; +} + +} // namespace __llvm_libc diff --git a/libc/src/time/linux/CMakeLists.txt b/libc/src/time/linux/CMakeLists.txt --- a/libc/src/time/linux/CMakeLists.txt +++ b/libc/src/time/linux/CMakeLists.txt @@ -24,3 +24,17 @@ libc.src.__support.OSUtil.osutil libc.src.errno.errno ) + +add_entrypoint_object( + nanosleep + SRCS + nanosleep.cpp + HDRS + ../nanosleep.h + DEPENDS + libc.include.time + libc.include.sys_syscall + libc.src.__support.CPP.limits + libc.src.__support.OSUtil.osutil + libc.src.errno.errno +) diff --git a/libc/src/time/nanosleep.cpp b/libc/src/time/linux/nanosleep.cpp rename from libc/src/time/nanosleep.cpp rename to libc/src/time/linux/nanosleep.cpp --- a/libc/src/time/nanosleep.cpp +++ b/libc/src/time/linux/nanosleep.cpp @@ -1,4 +1,4 @@ -//===-- Implementation of nanosleep function ------------------------------===// +//===-- Linux implementation of nanosleep function ------------------------===// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. // See https://llvm.org/LICENSE.txt for license information. @@ -16,7 +16,6 @@ namespace __llvm_libc { -// TODO(michaelrj): Move this into time/linux with the other syscalls. LLVM_LIBC_FUNCTION(int, nanosleep, (const struct timespec *req, struct timespec *rem)) { #if SYS_nanosleep