From b8dd3b208a6ac3d764050d5b504226758974380f Mon Sep 17 00:00:00 2001 From: Brad King Date: Wed, 9 Apr 2025 12:08:48 -0400 Subject: clang-format: Apply formatting to HIP sources with .hip extension --- .gitattributes | 1 + Tests/HIP/InferHipLang1/interface.hip | 3 +-- Tests/HIP/InferHipLang2/interface.hip | 3 +-- Tests/HIP/MathFunctions/main.hip | 22 ++++++++++++---------- Tests/HIP/MixedLanguage/shared.hip | 4 +--- Tests/HIP/MixedLanguage/static.hip | 7 ++----- Tests/HIP/TryCompile/device_function.hip | 2 +- 7 files changed, 19 insertions(+), 23 deletions(-) diff --git a/.gitattributes b/.gitattributes index 0e7ab06..d414c68 100644 --- a/.gitattributes +++ b/.gitattributes @@ -60,6 +60,7 @@ coverage.xml.in tab-indent *.H our-c-style *.h our-c-style *.hh our-c-style +*.hip our-c-style *.hpp our-c-style *.hxx our-c-style *.notcu our-c-style diff --git a/Tests/HIP/InferHipLang1/interface.hip b/Tests/HIP/InferHipLang1/interface.hip index 6ac8641..21b7b0b 100644 --- a/Tests/HIP/InferHipLang1/interface.hip +++ b/Tests/HIP/InferHipLang1/interface.hip @@ -1,15 +1,14 @@ #include #include + #include static __global__ void fake_hip_kernel() { } - int __host__ interface_hip_func(int x) { - fake_hip_kernel<<<1, 1>>>(); bool valid = (hipSuccess == hipGetLastError()); if (!valid) { diff --git a/Tests/HIP/InferHipLang2/interface.hip b/Tests/HIP/InferHipLang2/interface.hip index 6ac8641..21b7b0b 100644 --- a/Tests/HIP/InferHipLang2/interface.hip +++ b/Tests/HIP/InferHipLang2/interface.hip @@ -1,15 +1,14 @@ #include #include + #include static __global__ void fake_hip_kernel() { } - int __host__ interface_hip_func(int x) { - fake_hip_kernel<<<1, 1>>>(); bool valid = (hipSuccess == hipGetLastError()); if (!valid) { diff --git a/Tests/HIP/MathFunctions/main.hip b/Tests/HIP/MathFunctions/main.hip index dae89fc..149cfc1 100644 --- a/Tests/HIP/MathFunctions/main.hip +++ b/Tests/HIP/MathFunctions/main.hip @@ -1,15 +1,15 @@ - -#include #include -#include #include +#include -#include #include +#include +#include namespace { -template -__global__ void global_entry_point(F f, T *out) { +template +__global__ void global_entry_point(F f, T* out) +{ *out = f(); } @@ -23,7 +23,9 @@ bool verify(F f, T expected) } bool result = true; hipLaunchKernelGGL(global_entry_point, 1, 1, 0, 0, f, gpu_T); - result = hipMemcpy(cpu_T.get(), gpu_T, sizeof(T), hipMemcpyDeviceToHost) == hipSuccess && result; + result = hipMemcpy(cpu_T.get(), gpu_T, sizeof(T), hipMemcpyDeviceToHost) == + hipSuccess && + result; result = hipFree(gpu_T) == hipSuccess && result; result = *cpu_T == expected && result; return result; @@ -32,9 +34,9 @@ bool verify(F f, T expected) int main(int argc, char** argv) { - bool valid = verify([]__device__(){ return std::round(1.4f); }, 1.0f); - valid &= verify([]__device__(){ return max<_Float16>(1.0f, 2.0f); }, 2.0f); - valid &= verify([]__device__(){ return min<_Float16>(1.0f, 2.0f); }, 1.0f); + bool valid = verify([] __device__() { return std::round(1.4f); }, 1.0f); + valid &= verify([] __device__() { return max<_Float16>(1.0f, 2.0f); }, 2.0f); + valid &= verify([] __device__() { return min<_Float16>(1.0f, 2.0f); }, 1.0f); if (valid) { return 0; diff --git a/Tests/HIP/MixedLanguage/shared.hip b/Tests/HIP/MixedLanguage/shared.hip index e6fea9f..4b0454d 100644 --- a/Tests/HIP/MixedLanguage/shared.hip +++ b/Tests/HIP/MixedLanguage/shared.hip @@ -1,5 +1,6 @@ #include #include + #include #ifdef _WIN32 @@ -8,15 +9,12 @@ # define EXPORT #endif - static __global__ void fake_hip_kernel() { } - int __host__ shared_hip_func(int x) { - fake_hip_kernel<<<1, 1>>>(); bool valid = (hipSuccess == hipGetLastError()); if (!valid) { diff --git a/Tests/HIP/MixedLanguage/static.hip b/Tests/HIP/MixedLanguage/static.hip index 359b9fa..8664a62 100644 --- a/Tests/HIP/MixedLanguage/static.hip +++ b/Tests/HIP/MixedLanguage/static.hip @@ -1,17 +1,14 @@ - -#include #include -#include +#include +#include static __global__ void fake_hip_kernel() { } - int __host__ static_hip_func(int x) { - fake_hip_kernel<<<1, 1>>>(); bool valid = (hipSuccess == hipGetLastError()); if (!valid) { diff --git a/Tests/HIP/TryCompile/device_function.hip b/Tests/HIP/TryCompile/device_function.hip index 7c1205e..f8b97a6 100644 --- a/Tests/HIP/TryCompile/device_function.hip +++ b/Tests/HIP/TryCompile/device_function.hip @@ -1,4 +1,5 @@ #include + #include static __global__ void fake_hip_kernel() @@ -7,7 +8,6 @@ static __global__ void fake_hip_kernel() int __host__ try_compile_hip_func(int x) { - fake_hip_kernel<<<1, 1>>>(); bool valid = (hipSuccess == hipGetLastError()); if (!valid) { -- cgit v0.12 From 37089ffe207e4af516e1cddcc5f4e75dab88ca1b Mon Sep 17 00:00:00 2001 From: Brad King Date: Wed, 9 Apr 2025 12:20:25 -0400 Subject: Tests/HIP: Improve cmake code formatting --- Tests/HIP/CMakeLists.txt | 3 +-- 1 file changed, 1 insertion(+), 2 deletions(-) diff --git a/Tests/HIP/CMakeLists.txt b/Tests/HIP/CMakeLists.txt index f1e2e51..fb6238f 100644 --- a/Tests/HIP/CMakeLists.txt +++ b/Tests/HIP/CMakeLists.txt @@ -1,7 +1,6 @@ macro (add_hip_test_macro name) add_test_macro("${name}" ${ARGN}) - set_property(TEST "${name}" APPEND - PROPERTY LABELS "HIP") + set_property(TEST "${name}" APPEND PROPERTY LABELS "HIP") endmacro () add_hip_test_macro(HIP.ArchitectureOff HIPOnlyArchitectureOff) -- cgit v0.12 From c7ae9400c9d9375d37a66fe5561a5b086db241b0 Mon Sep 17 00:00:00 2001 From: Brad King Date: Wed, 9 Apr 2025 12:24:32 -0400 Subject: Tests/HIP: Improve diagnostics on failure --- Tests/HIP/CMakeLists.txt | 1 + Tests/HIP/InferHipLang1/interface.hip | 8 ++++---- Tests/HIP/InferHipLang2/interface.hip | 8 ++++---- Tests/HIP/MathFunctions/main.hip | 22 ++++++++++++++++------ Tests/HIP/MixedLanguage/shared.hip | 8 ++++---- Tests/HIP/MixedLanguage/static.hip | 8 ++++---- Tests/HIP/TryCompile/device_function.hip | 8 ++++---- 7 files changed, 37 insertions(+), 26 deletions(-) diff --git a/Tests/HIP/CMakeLists.txt b/Tests/HIP/CMakeLists.txt index fb6238f..92d6ddb 100644 --- a/Tests/HIP/CMakeLists.txt +++ b/Tests/HIP/CMakeLists.txt @@ -1,6 +1,7 @@ macro (add_hip_test_macro name) add_test_macro("${name}" ${ARGN}) set_property(TEST "${name}" APPEND PROPERTY LABELS "HIP") + set_property(TEST "${name}" APPEND PROPERTY ENVIRONMENT "AMD_LOG_LEVEL=4") endmacro () add_hip_test_macro(HIP.ArchitectureOff HIPOnlyArchitectureOff) diff --git a/Tests/HIP/InferHipLang1/interface.hip b/Tests/HIP/InferHipLang1/interface.hip index 21b7b0b..0ee49c3 100644 --- a/Tests/HIP/InferHipLang1/interface.hip +++ b/Tests/HIP/InferHipLang1/interface.hip @@ -1,4 +1,4 @@ -#include +#include #include #include @@ -10,9 +10,9 @@ static __global__ void fake_hip_kernel() int __host__ interface_hip_func(int x) { fake_hip_kernel<<<1, 1>>>(); - bool valid = (hipSuccess == hipGetLastError()); - if (!valid) { - throw std::system_error(ENODEV, std::generic_category(), "no hip device"); + hipError_t err = hipGetLastError(); + if (err != hipSuccess) { + throw std::runtime_error(hipGetErrorString(err)); } return x * x + std::integral_constant::value; } diff --git a/Tests/HIP/InferHipLang2/interface.hip b/Tests/HIP/InferHipLang2/interface.hip index 21b7b0b..0ee49c3 100644 --- a/Tests/HIP/InferHipLang2/interface.hip +++ b/Tests/HIP/InferHipLang2/interface.hip @@ -1,4 +1,4 @@ -#include +#include #include #include @@ -10,9 +10,9 @@ static __global__ void fake_hip_kernel() int __host__ interface_hip_func(int x) { fake_hip_kernel<<<1, 1>>>(); - bool valid = (hipSuccess == hipGetLastError()); - if (!valid) { - throw std::system_error(ENODEV, std::generic_category(), "no hip device"); + hipError_t err = hipGetLastError(); + if (err != hipSuccess) { + throw std::runtime_error(hipGetErrorString(err)); } return x * x + std::integral_constant::value; } diff --git a/Tests/HIP/MathFunctions/main.hip b/Tests/HIP/MathFunctions/main.hip index 149cfc1..e97ad9a 100644 --- a/Tests/HIP/MathFunctions/main.hip +++ b/Tests/HIP/MathFunctions/main.hip @@ -1,4 +1,5 @@ #include +#include #include #include @@ -18,14 +19,23 @@ bool verify(F f, T expected) { std::unique_ptr cpu_T(new T); T* gpu_T = nullptr; - if (hipMalloc((void**)&gpu_T, sizeof(T)) != hipSuccess) { - return false; - } bool result = true; + { + hipError_t err = hipMalloc((void**)&gpu_T, sizeof(T)); + if (err != hipSuccess) { + std::cerr << "hipMalloc failed: " << hipGetErrorString(err) << std::endl; + result = false; + } + } hipLaunchKernelGGL(global_entry_point, 1, 1, 0, 0, f, gpu_T); - result = hipMemcpy(cpu_T.get(), gpu_T, sizeof(T), hipMemcpyDeviceToHost) == - hipSuccess && - result; + { + hipError_t err = + hipMemcpy(cpu_T.get(), gpu_T, sizeof(T), hipMemcpyDeviceToHost); + if (err != hipSuccess) { + std::cerr << "hipMemcpy failed: " << hipGetErrorString(err) << std::endl; + result = false; + } + } result = hipFree(gpu_T) == hipSuccess && result; result = *cpu_T == expected && result; return result; diff --git a/Tests/HIP/MixedLanguage/shared.hip b/Tests/HIP/MixedLanguage/shared.hip index 4b0454d..fc242a7 100644 --- a/Tests/HIP/MixedLanguage/shared.hip +++ b/Tests/HIP/MixedLanguage/shared.hip @@ -1,4 +1,4 @@ -#include +#include #include #include @@ -16,9 +16,9 @@ static __global__ void fake_hip_kernel() int __host__ shared_hip_func(int x) { fake_hip_kernel<<<1, 1>>>(); - bool valid = (hipSuccess == hipGetLastError()); - if (!valid) { - throw std::system_error(ENODEV, std::generic_category(), "no hip device"); + hipError_t err = hipGetLastError(); + if (err != hipSuccess) { + throw std::runtime_error(hipGetErrorString(err)); } return x * x + std::integral_constant::value; } diff --git a/Tests/HIP/MixedLanguage/static.hip b/Tests/HIP/MixedLanguage/static.hip index 8664a62..dbcd309 100644 --- a/Tests/HIP/MixedLanguage/static.hip +++ b/Tests/HIP/MixedLanguage/static.hip @@ -1,4 +1,4 @@ -#include +#include #include #include @@ -10,9 +10,9 @@ static __global__ void fake_hip_kernel() int __host__ static_hip_func(int x) { fake_hip_kernel<<<1, 1>>>(); - bool valid = (hipSuccess == hipGetLastError()); - if (!valid) { - throw std::system_error(ENODEV, std::generic_category(), "no hip device"); + hipError_t err = hipGetLastError(); + if (err != hipSuccess) { + throw std::runtime_error(hipGetErrorString(err)); } return x * x + std::integral_constant::value; } diff --git a/Tests/HIP/TryCompile/device_function.hip b/Tests/HIP/TryCompile/device_function.hip index f8b97a6..e8b684b 100644 --- a/Tests/HIP/TryCompile/device_function.hip +++ b/Tests/HIP/TryCompile/device_function.hip @@ -1,4 +1,4 @@ -#include +#include #include @@ -9,9 +9,9 @@ static __global__ void fake_hip_kernel() int __host__ try_compile_hip_func(int x) { fake_hip_kernel<<<1, 1>>>(); - bool valid = (hipSuccess == hipGetLastError()); - if (!valid) { - throw std::system_error(ENODEV, std::generic_category(), "no hip device"); + hipError_t err = hipGetLastError(); + if (err != hipSuccess) { + throw std::runtime_error(hipGetErrorString(err)); } return x * x; } -- cgit v0.12