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