From 9e390a140857355b3524924075a261b9d06ae850 Mon Sep 17 00:00:00 2001 From: Joseph Huber Date: Mon, 30 Oct 2023 14:58:09 -0500 Subject: [PATCH] [libc][Obvious] Fix missing semicolon in AMDGPU loader implementation Summary: Title --- libc/utils/gpu/loader/amdgpu/Loader.cpp | 5 ++--- 1 file changed, 2 insertions(+), 3 deletions(-) diff --git a/libc/utils/gpu/loader/amdgpu/Loader.cpp b/libc/utils/gpu/loader/amdgpu/Loader.cpp index c2a11fd8aab7..2f99076a720e 100644 --- a/libc/utils/gpu/loader/amdgpu/Loader.cpp +++ b/libc/utils/gpu/loader/amdgpu/Loader.cpp @@ -248,9 +248,8 @@ hsa_status_t launch_kernel(hsa_agent_t dev_agent, hsa_executable_t executable, (HSA_PACKET_TYPE_KERNEL_DISPATCH << HSA_PACKET_HEADER_TYPE) | (HSA_FENCE_SCOPE_SYSTEM << HSA_PACKET_HEADER_SCACQUIRE_FENCE_SCOPE) | (HSA_FENCE_SCOPE_SYSTEM << HSA_PACKET_HEADER_SCRELEASE_FENCE_SCOPE); - uint32_t header_word = - header | (setup << 16u) __atomic_store_n((uint32_t *)&packet->header, - header_word, __ATOMIC_RELEASE); + uint32_t header_word = header | (setup << 16u); + __atomic_store_n((uint32_t *)&packet->header, header_word, __ATOMIC_RELEASE); hsa_signal_store_relaxed(queue->doorbell_signal, packet_id); // Wait until the kernel has completed execution on the device. Periodically