diff --git a/runtime/hsa-runtime/core/runtime/runtime.cpp b/runtime/hsa-runtime/core/runtime/runtime.cpp index 37af826d4f..2604973c08 100644 --- a/runtime/hsa-runtime/core/runtime/runtime.cpp +++ b/runtime/hsa-runtime/core/runtime/runtime.cpp @@ -809,14 +809,17 @@ hsa_status_t Runtime::SetAsyncSignalHandler(hsa_signal_t signal, void* arg) { struct AsyncEventsInfo* asyncInfo = &asyncSignals_; + int priority = runtime_singleton_->flag().async_events_thread_priority(); if (signal.handle != 0) { // Indicate that this signal is in use. hsa_signal_handle(signal)->Retain(); core::Signal* coreSignal = core::Signal::Convert(signal); - if (coreSignal->EopEvent() && coreSignal->EopEvent()->EventData.EventType != HSA_EVENTTYPE_SIGNAL) + if (coreSignal->EopEvent() && coreSignal->EopEvent()->EventData.EventType != HSA_EVENTTYPE_SIGNAL) { + priority = os::OS_THREAD_PRIORITY_DEFAULT; asyncInfo = &asyncExceptions_; + } } ScopedAcquire scope_lock(&asyncInfo->control.lock); @@ -835,7 +838,7 @@ hsa_status_t Runtime::SetAsyncSignalHandler(hsa_signal_t signal, // Start event monitoring thread asyncInfo->control.exit = false; asyncInfo->control.async_events_thread_ = - os::CreateThread(AsyncEventsLoop, asyncInfo); + os::CreateThread(AsyncEventsLoop, asyncInfo, 0, priority); if (asyncInfo->control.async_events_thread_ == NULL) { assert(false && "Asyncronous events thread creation error."); return HSA_STATUS_ERROR_OUT_OF_RESOURCES; diff --git a/runtime/hsa-runtime/core/util/flag.h b/runtime/hsa-runtime/core/util/flag.h index 4a23e7a256..fcba3536cd 100644 --- a/runtime/hsa-runtime/core/util/flag.h +++ b/runtime/hsa-runtime/core/util/flag.h @@ -257,6 +257,22 @@ class Flag { /* hsa_signal_wait_relaxed abort timeout */ var = os::GetEnvVar("HSA_SIGNAL_WAIT_ABORT_TIMEOUT"); signal_abort_timeout_ = var.empty() ? 0 : atoi(var.c_str()); + + /* Valid inputs are 0-99, HIGH, MAX */ + var = os::GetEnvVar("HSA_ASYNCEVENTS_THREAD_PRIORITY"); + async_events_thread_priority_ = os::OS_THREAD_PRIORITY_DEFAULT; + if (var == "MAX") { + async_events_thread_priority_ = os::OS_THREAD_PRIORITY_MAX; + } else if (var == "HIGH") { + async_events_thread_priority_ = os::OS_THREAD_PRIORITY_HIGH; + } else if (var != "") { + char* end; + int input = strtol(var.c_str(), &end, 10); + if (input >= 0 && input <= 99) + async_events_thread_priority_ = input; + else + fprintf(stderr, "Failed to parse HSA_ASYNCEVENTS_THREAD_PRIORITY"); + } } void parse_masks(uint32_t maxGpu, uint32_t maxCU) { @@ -373,6 +389,8 @@ class Flag { uint32_t signal_abort_timeout() const { return signal_abort_timeout_; } + int async_events_thread_priority() const { return async_events_thread_priority_; } + private: bool check_flat_scratch_; bool enable_vm_fault_message_; @@ -404,6 +422,7 @@ class Flag { bool wait_any_; bool dev_mem_queue_; uint32_t signal_abort_timeout_; + int async_events_thread_priority_; SDMA_OVERRIDE enable_sdma_; SDMA_OVERRIDE enable_peer_sdma_; diff --git a/runtime/hsa-runtime/core/util/lnx/os_linux.cpp b/runtime/hsa-runtime/core/util/lnx/os_linux.cpp index 72e189be2f..d1e9d0e805 100644 --- a/runtime/hsa-runtime/core/util/lnx/os_linux.cpp +++ b/runtime/hsa-runtime/core/util/lnx/os_linux.cpp @@ -90,7 +90,10 @@ void* __stdcall ThreadTrampoline(void* arg) { // Thread container allows multiple waits and separate close (destroy). class os_thread { public: - explicit os_thread(ThreadEntry function, void* threadArgument, uint stackSize) + explicit os_thread(ThreadEntry function, + void* threadArgument, + uint stackSize, + int priority) : thread(0), lock(nullptr), state(RUNNING) { int err; lock = CreateMutex(); @@ -162,6 +165,38 @@ class os_thread { return; } } while (stackSize < 20 * 1024 * 1024); + + struct sched_param param = {}; + if (priority != OS_THREAD_PRIORITY_DEFAULT) { + int set_priority; + int max_priority = sched_get_priority_max(SCHED_FIFO); + + if (priority == OS_THREAD_PRIORITY_MAX) + set_priority = max_priority; + else if (priority == OS_THREAD_PRIORITY_HIGH) + set_priority = max_priority - 1; + else if (priority > max_priority) + set_priority = max_priority; + else + set_priority = priority; + + param.sched_priority = set_priority; + if (pthread_setschedparam(thread, SCHED_FIFO, ¶m)) { + fprintf(stderr, "pthread_setschedparam failed\n"); + return; + } + + int policy = 0; + if (pthread_getschedparam(thread, &policy, ¶m)) + fprintf(stderr, "pthread_getschedparam failed: %s\n", strerror(err)); + + if (policy != SCHED_FIFO || param.sched_priority != set_priority) + fprintf(stderr, "Failed to adjust thread priority (policy:%s requested:%d current:%d)\n", + policy == SCHED_FIFO ? "FIFO" : + policy == SCHED_OTHER ? "OTHER" : + policy == SCHED_RR ? "RR" : "Unknown", + set_priority, param.sched_priority); + } } os_thread(os_thread&& rhs) { @@ -413,8 +448,8 @@ void uSleep(int delayInUs) { usleep(delayInUs); } void YieldThread() { sched_yield(); } -Thread CreateThread(ThreadEntry function, void* threadArgument, uint stackSize) { - os_thread* result = new os_thread(function, threadArgument, stackSize); +Thread CreateThread(ThreadEntry function, void* threadArgument, uint stackSize, int priority) { + os_thread* result = new os_thread(function, threadArgument, stackSize, priority); if (!result->Valid()) { delete result; return nullptr; diff --git a/runtime/hsa-runtime/core/util/os.h b/runtime/hsa-runtime/core/util/os.h index 2eec51a34e..6973c2533c 100644 --- a/runtime/hsa-runtime/core/util/os.h +++ b/runtime/hsa-runtime/core/util/os.h @@ -58,6 +58,12 @@ typedef void* SharedMutex; typedef void* Thread; typedef void* EventHandle; +typedef enum { + OS_THREAD_PRIORITY_DEFAULT = -1, + OS_THREAD_PRIORITY_HIGH = 254, + OS_THREAD_PRIORITY_MAX = 255, +} ThreadPriority; + enum class os_t { OS_WIN = 0, OS_LINUX, COUNT }; static __forceinline std::underlying_type::type os_index(os_t val) { return std::underlying_type::type(val); @@ -209,9 +215,10 @@ typedef void (*ThreadEntry)(void*); /// @param: entry_argument(Input), a pointer to the argument of the thread /// function. /// @param: stack_size(Input), size of the thread's stack, 0 by default. +/// @param: priority(Input), thread priority. /// @return: Thread, a handle to thread created. Thread CreateThread(ThreadEntry entry_function, void* entry_argument, - uint stack_size = 0); + uint stack_size = 0, int priority = OS_THREAD_PRIORITY_DEFAULT); /// @brief: Destroys the thread. /// @param: thread(Input), thread handle to what will be destroyed. diff --git a/runtime/hsa-runtime/core/util/win/os_win.cpp b/runtime/hsa-runtime/core/util/win/os_win.cpp index 81c90cd266..648a732899 100644 --- a/runtime/hsa-runtime/core/util/win/os_win.cpp +++ b/runtime/hsa-runtime/core/util/win/os_win.cpp @@ -151,7 +151,7 @@ unsigned __stdcall ThreadTrampoline(void* arg) { } Thread CreateThread(ThreadEntry entry_function, void* entry_argument, - uint stack_size) { + uint stack_size, int priority_unused) { ThreadArgs* thread_args = new ThreadArgs(); thread_args->entry_args = entry_argument; thread_args->entry_function = entry_function;