From 3b1c0c3464b93ff40e2a1f7452b9028f76cc49ce Mon Sep 17 00:00:00 2001 From: MachineTom Date: Wed, 5 Nov 2025 09:33:41 -0500 Subject: [PATCH] SWDEV-558845 - Support image in rocr on Windows (#1582) Enable image build in Windows. Remove some useless codes that fail building in Windows. Some minor improvement. Temporarily exclude mipmap test files. Prevent negative tests affect some tests. Move some catch info log codes into failed cases. --- .../hip-tests/catch/unit/surface/surf1D.cc | 1 + .../catch/unit/texture/CMakeLists.txt | 2 +- .../catch/unit/texture/hipBindTexture2D.cc | 1 + .../unit/texture/hipTex1DFetchCheckModes.cc | 1 + .../catch/unit/texture/hipTexObjPitch.cc | 7 +- .../catch/unit/texture/hipTexObjectCreate.cc | 2 + .../unit/texture/hipTextureObj1DFetch.cc | 1 + .../texture/hipTextureObj2DCheckSRGBModes.cc | 1 + .../hip-tests/catch/unit/texture/tex1D.cc | 32 ++++----- .../hip-tests/catch/unit/texture/tex1DGrad.cc | 32 ++++----- .../catch/unit/texture/tex1DLayered.cc | 36 +++++----- .../catch/unit/texture/tex1DLayeredGrad.cc | 38 +++++----- .../catch/unit/texture/tex1DLayeredLod.cc | 36 +++++----- .../hip-tests/catch/unit/texture/tex1DLod.cc | 32 ++++----- .../catch/unit/texture/tex1Dfetch.cc | 12 ++-- .../hip-tests/catch/unit/texture/tex2D.cc | 36 +++++----- .../hip-tests/catch/unit/texture/tex2DGrad.cc | 36 +++++----- .../catch/unit/texture/tex2DLayered.cc | 40 +++++------ .../catch/unit/texture/tex2DLayeredGrad.cc | 40 +++++------ .../catch/unit/texture/tex2DLayeredLod.cc | 40 +++++------ .../hip-tests/catch/unit/texture/tex2DLod.cc | 36 +++++----- .../catch/unit/texture/tex2Dgather.cc | 20 +++--- .../hip-tests/catch/unit/texture/tex3D.cc | 44 ++++++------ .../hip-tests/catch/unit/texture/tex3DGrad.cc | 44 ++++++------ .../hip-tests/catch/unit/texture/tex3DLod.cc | 44 ++++++------ .../catch/unit/texture/texCubemap.cc | 58 ++++++++------- .../catch/unit/texture/texCubemapGrad.cc | 58 ++++++++------- .../catch/unit/texture/texCubemapLayered.cc | 58 ++++++++------- .../unit/texture/texCubemapLayeredGrad.cc | 62 ++++++++-------- .../unit/texture/texCubemapLayeredLod.cc | 62 ++++++++-------- .../catch/unit/texture/texCubemapLod.cc | 58 ++++++++------- .../runtime/hsa-runtime/CMakeLists.txt | 11 +-- .../hsa-runtime/core/runtime/amd_topology.cpp | 3 +- .../runtime/hsa-runtime/image/blit_kernel.cpp | 5 -- .../hsa-runtime/image/blit_src/CMakeLists.txt | 17 ++++- .../image/blit_src/create_hsaco_ascii_file.py | 71 +++++++++++++++++++ .../hsa-runtime/image/image_manager.cpp | 5 -- .../hsa-runtime/image/image_manager_ai.cpp | 1 - .../hsa-runtime/image/image_manager_gfx11.cpp | 1 - .../hsa-runtime/image/image_manager_gfx12.cpp | 1 - .../hsa-runtime/image/image_manager_kv.cpp | 5 +- .../hsa-runtime/image/image_manager_nv.cpp | 1 - .../hsa-runtime/image/image_runtime.cpp | 1 - .../runtime/hsa-runtime/image/util.h | 16 ----- 44 files changed, 577 insertions(+), 531 deletions(-) create mode 100644 projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/create_hsaco_ascii_file.py diff --git a/projects/hip-tests/catch/unit/surface/surf1D.cc b/projects/hip-tests/catch/unit/surface/surf1D.cc index ddb9c137f8..7646e3e99a 100644 --- a/projects/hip-tests/catch/unit/surface/surf1D.cc +++ b/projects/hip-tests/catch/unit/surface/surf1D.cc @@ -66,6 +66,7 @@ template __global__ void surf1DKernelRW(hipSurfaceObject_t surfaceO } template static void runTestR(const int width) { + (void) hipGetLastError(); // Prevent negative tests affecting this unsigned int size = width * sizeof(T); T* hData = (T*)malloc(size); memset(hData, 0, size); diff --git a/projects/hip-tests/catch/unit/texture/CMakeLists.txt b/projects/hip-tests/catch/unit/texture/CMakeLists.txt index 53716114d6..8e28936c30 100644 --- a/projects/hip-tests/catch/unit/texture/CMakeLists.txt +++ b/projects/hip-tests/catch/unit/texture/CMakeLists.txt @@ -129,7 +129,7 @@ else() endif() # Mipmap APIs are not supported on Linux -if(WIN32) +if(0) set(TEST_SRC ${TEST_SRC} hipBindTextureToMipmappedArray.cc diff --git a/projects/hip-tests/catch/unit/texture/hipBindTexture2D.cc b/projects/hip-tests/catch/unit/texture/hipBindTexture2D.cc index 15ec24c869..edd7cfa684 100644 --- a/projects/hip-tests/catch/unit/texture/hipBindTexture2D.cc +++ b/projects/hip-tests/catch/unit/texture/hipBindTexture2D.cc @@ -57,6 +57,7 @@ TEST_CASE("Unit_hipBindTexture2D_Pitch") { HipTest::HIP_SKIP_TEST("__HIP_NO_IMAGE_SUPPORT is set"); return; #endif + (void) hipGetLastError(); // Prevent negative tests affecting this float* b; float* a; diff --git a/projects/hip-tests/catch/unit/texture/hipTex1DFetchCheckModes.cc b/projects/hip-tests/catch/unit/texture/hipTex1DFetchCheckModes.cc index e152ee4515..bda4e17dab 100644 --- a/projects/hip-tests/catch/unit/texture/hipTex1DFetchCheckModes.cc +++ b/projects/hip-tests/catch/unit/texture/hipTex1DFetchCheckModes.cc @@ -110,6 +110,7 @@ TEST_CASE("Unit_tex1Dfetch_CheckModes") { HipTest::HIP_SKIP_TEST("__HIP_NO_IMAGE_SUPPORT is set"); return; #endif + (void) hipGetLastError(); // Prevent negative tests affecting this SECTION("hipAddressModeClamp AND hipFilterModePoint") { runTest(hipAddressModeClamp, hipFilterModePoint); diff --git a/projects/hip-tests/catch/unit/texture/hipTexObjPitch.cc b/projects/hip-tests/catch/unit/texture/hipTexObjPitch.cc index abdac1341d..ec713a961a 100644 --- a/projects/hip-tests/catch/unit/texture/hipTexObjPitch.cc +++ b/projects/hip-tests/catch/unit/texture/hipTexObjPitch.cc @@ -58,14 +58,13 @@ static __global__ void texture2dCopyKernel(hipTextureObject_t texObj, TYPE_t* ds TEMPLATE_TEST_CASE("Unit_hipTexObjPitch_texture2D", "", char, unsigned char, short, unsigned short, int, unsigned int, float) { CHECK_IMAGE_SUPPORT -#if HT_NVIDIA - (void) - hipGetLastError(); // Prevent negative tests affecting this -#endif + #if __HIP_NO_IMAGE_SUPPORT HipTest::HIP_SKIP_TEST("__HIP_NO_IMAGE_SUPPORT is set"); return; #endif + (void) hipGetLastError(); // Prevent negative tests affecting this + TestType* B; TestType* A; TestType* devPtrB; diff --git a/projects/hip-tests/catch/unit/texture/hipTexObjectCreate.cc b/projects/hip-tests/catch/unit/texture/hipTexObjectCreate.cc index eaac1d9b9c..4ca8e5531b 100644 --- a/projects/hip-tests/catch/unit/texture/hipTexObjectCreate.cc +++ b/projects/hip-tests/catch/unit/texture/hipTexObjectCreate.cc @@ -244,6 +244,7 @@ TEST_CASE("Unit_TexObjectCreate_TypeArray_NullptrArray") { CTX_DESTROY(); } +#if 0 TEST_CASE("Unit_TexObjectCreate_TypeMipmapped") { #if __linux__ HipTest::HIP_SKIP_TEST("Mipmap APIs are not supported on Linux"); @@ -311,6 +312,7 @@ TEST_CASE("Unit_TexObjectCreate_TypeMipmaped_IncompleteInit") { HIP_CHECK(hipFreeMipmappedArray(mipmapped_array)); CTX_DESTROY(); } +#endif TEST_CASE("Unit_TexObjectCreate_TypePitch2D") { CHECK_IMAGE_SUPPORT diff --git a/projects/hip-tests/catch/unit/texture/hipTextureObj1DFetch.cc b/projects/hip-tests/catch/unit/texture/hipTextureObj1DFetch.cc index d53bfbe9d3..d2e29df3ab 100644 --- a/projects/hip-tests/catch/unit/texture/hipTextureObj1DFetch.cc +++ b/projects/hip-tests/catch/unit/texture/hipTextureObj1DFetch.cc @@ -39,6 +39,7 @@ TEST_CASE("Unit_hipCreateTextureObject_tex1DfetchVerification") { HipTest::HIP_SKIP_TEST("__HIP_NO_IMAGE_SUPPORT is set"); return; #endif + (void) hipGetLastError(); // Prevent negative tests affecting this // Allocating the required buffer on gpu device float *texBuf, *texBufOut; diff --git a/projects/hip-tests/catch/unit/texture/hipTextureObj2DCheckSRGBModes.cc b/projects/hip-tests/catch/unit/texture/hipTextureObj2DCheckSRGBModes.cc index f8becb2318..0005918c59 100644 --- a/projects/hip-tests/catch/unit/texture/hipTextureObj2DCheckSRGBModes.cc +++ b/projects/hip-tests/catch/unit/texture/hipTextureObj2DCheckSRGBModes.cc @@ -159,6 +159,7 @@ TEST_CASE("Unit_hipTextureObj2DCheckRGBAModes") { HipTest::HIP_SKIP_TEST("__HIP_NO_IMAGE_SUPPORT is set"); return; #endif + (void) hipGetLastError(); // Prevent negative tests affecting this SECTION("RGBA 2D hipAddressModeClamp, hipFilterModePoint, regularCoords") { runTest(256, 256, -3.9, 6.1); diff --git a/projects/hip-tests/catch/unit/texture/tex1D.cc b/projects/hip-tests/catch/unit/texture/tex1D.cc index f09338c2b1..527ce8474c 100644 --- a/projects/hip-tests/catch/unit/texture/tex1D.cc +++ b/projects/hip-tests/catch/unit/texture/tex1D.cc @@ -69,15 +69,15 @@ TEMPLATE_TEST_CASE("Unit_tex1D_Positive_ReadModeElementType", "", char, unsigned for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Index: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1D(x, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Index: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } @@ -119,15 +119,15 @@ TEMPLATE_TEST_CASE("Unit_tex1D_Positive_ReadModeNormalizedFloat", "", char, unsi for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1D(x, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("i: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex1DGrad.cc b/projects/hip-tests/catch/unit/texture/tex1DGrad.cc index ea7dae3db3..0d7400b93d 100644 --- a/projects/hip-tests/catch/unit/texture/tex1DGrad.cc +++ b/projects/hip-tests/catch/unit/texture/tex1DGrad.cc @@ -69,15 +69,15 @@ TEMPLATE_TEST_CASE("Unit_tex1DGrad_Positive_ReadModeElementType", "", char, unsi for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Index: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1D(x, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Index: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } @@ -119,15 +119,15 @@ TEMPLATE_TEST_CASE("Unit_tex1DGrad_Positive_ReadModeNormalizedFloat", "", char, for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1D(x, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("i: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex1DLayered.cc b/projects/hip-tests/catch/unit/texture/tex1DLayered.cc index e202b0e663..d2b7a1f4a5 100644 --- a/projects/hip-tests/catch/unit/texture/tex1DLayered.cc +++ b/projects/hip-tests/catch/unit/texture/tex1DLayered.cc @@ -72,16 +72,16 @@ TEMPLATE_TEST_CASE("Unit_tex1DLayered_Positive_ReadModeElementType", "", char, u for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - const auto ref_val = fixture.tex_h.Tex1DLayered(x, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } } @@ -127,16 +127,16 @@ TEMPLATE_TEST_CASE("Unit_tex1DLayered_Positive_ReadModeNormalizedFloat", "", cha for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Index: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1DLayered(x, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Index: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/tex1DLayeredGrad.cc b/projects/hip-tests/catch/unit/texture/tex1DLayeredGrad.cc index 504d52fc91..d6155f0ede 100644 --- a/projects/hip-tests/catch/unit/texture/tex1DLayeredGrad.cc +++ b/projects/hip-tests/catch/unit/texture/tex1DLayeredGrad.cc @@ -72,16 +72,16 @@ TEMPLATE_TEST_CASE("Unit_tex1DLayeredGrad_Positive_ReadModeElementType", "", cha for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Index: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1DLayered(x, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Index: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } } @@ -127,17 +127,17 @@ TEMPLATE_TEST_CASE("Unit_tex1DLayeredGrad_Positive_ReadModeNormalizedFloat", "", for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Filter mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1DLayered(x, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Filter mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/tex1DLayeredLod.cc b/projects/hip-tests/catch/unit/texture/tex1DLayeredLod.cc index 0e0517352a..5d58da3e8f 100644 --- a/projects/hip-tests/catch/unit/texture/tex1DLayeredLod.cc +++ b/projects/hip-tests/catch/unit/texture/tex1DLayeredLod.cc @@ -72,16 +72,16 @@ TEMPLATE_TEST_CASE("Unit_tex1DLayeredLod_Positive_ReadModeElementType", "", char for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Index: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1DLayered(x, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Index: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } } @@ -127,16 +127,16 @@ TEMPLATE_TEST_CASE("Unit_tex1DLayeredLod_Positive_ReadModeNormalizedFloat", "", for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1DLayered(x, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/tex1DLod.cc b/projects/hip-tests/catch/unit/texture/tex1DLod.cc index 3880c0549f..8e2bf84ea1 100644 --- a/projects/hip-tests/catch/unit/texture/tex1DLod.cc +++ b/projects/hip-tests/catch/unit/texture/tex1DLod.cc @@ -69,15 +69,15 @@ TEMPLATE_TEST_CASE("Unit_tex1DLod_Positive_ReadModeElementType", "", char, unsig for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Index: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1D(x, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Index: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } @@ -119,15 +119,15 @@ TEMPLATE_TEST_CASE("Unit_tex1DLod_Positive_ReadModeNormalizedFloat", "", char, u for (auto i = 0u; i < params.NumItersX(); ++i) { float x = GetCoordinate(i, params.NumItersX(), params.Width(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - auto ref_val = fixture.tex_h.Tex1D(x, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("i: " << i); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex1Dfetch.cc b/projects/hip-tests/catch/unit/texture/tex1Dfetch.cc index f9211d9550..82c65c1b9e 100644 --- a/projects/hip-tests/catch/unit/texture/tex1Dfetch.cc +++ b/projects/hip-tests/catch/unit/texture/tex1Dfetch.cc @@ -89,9 +89,11 @@ TEMPLATE_TEST_CASE("Unit_tex1Dfetch_Positive_ReadModeElementType", "", char, uns HIP_CHECK(hipDeviceSynchronize()); for (auto i = 0u; i < out_alloc_h.size(); ++i) { - INFO("Index: " << i); const auto ref_val = tex_h[i]; - REQUIRE(out_alloc_h[i] == ref_val); + if (out_alloc_h[i] != ref_val) { + INFO("Index: " << i); + REQUIRE(false); + } } } @@ -151,9 +153,11 @@ TEMPLATE_TEST_CASE("Unit_tex1Dfetch_Positive_ReadModeNormalizedFloat", "", char, HIP_CHECK(hipDeviceSynchronize()); for (auto i = 0u; i < out_alloc_h.size(); ++i) { - INFO("Index: " << i); const auto ref_val = Vec4Map(tex_h[i]); - REQUIRE(out_alloc_h[i] == ref_val); + if (out_alloc_h[i] != ref_val) { + INFO("Index: " << i); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex2D.cc b/projects/hip-tests/catch/unit/texture/tex2D.cc index 3daa7f6eac..bb1ecb547a 100644 --- a/projects/hip-tests/catch/unit/texture/tex2D.cc +++ b/projects/hip-tests/catch/unit/texture/tex2D.cc @@ -51,6 +51,7 @@ THE SOFTWARE. TEMPLATE_TEST_CASE("Unit_tex2D_Positive_ReadModeElementType", "", char, unsigned char, short, unsigned short, int, unsigned int, float) { CHECK_IMAGE_SUPPORT; + (void) hipGetLastError(); // Prevent negative tests affecting this TextureTestParams params = {}; params.extent = make_hipExtent(16, 4, 0); @@ -85,16 +86,16 @@ TEMPLATE_TEST_CASE("Unit_tex2D_Positive_ReadModeElementType", "", char, unsigned params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - const auto ref_val = fixture.tex_h.Tex2D(x, y, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } @@ -152,15 +153,16 @@ TEMPLATE_TEST_CASE("Unit_tex2D_Positive_ReadModeNormalizedFloat", "", char, unsi params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); auto ref_val = fixture.tex_h.Tex2D(x, y, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex2DGrad.cc b/projects/hip-tests/catch/unit/texture/tex2DGrad.cc index 6750f7c425..80b8def7b4 100644 --- a/projects/hip-tests/catch/unit/texture/tex2DGrad.cc +++ b/projects/hip-tests/catch/unit/texture/tex2DGrad.cc @@ -86,16 +86,16 @@ TEMPLATE_TEST_CASE("Unit_tex2DGrad_Positive_ReadModeElementType", "", char, unsi params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - const auto ref_val = fixture.tex_h.Tex2D(x, y, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } @@ -154,16 +154,16 @@ TEMPLATE_TEST_CASE("Unit_tex2DGrad_Positive_ReadModeNormalizedFloat", "", char, params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - auto ref_val = fixture.tex_h.Tex2D(x, y, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex2DLayered.cc b/projects/hip-tests/catch/unit/texture/tex2DLayered.cc index 826305b72d..a20fd647dc 100644 --- a/projects/hip-tests/catch/unit/texture/tex2DLayered.cc +++ b/projects/hip-tests/catch/unit/texture/tex2DLayered.cc @@ -88,17 +88,17 @@ TEMPLATE_TEST_CASE("Unit_tex2DLayered_Positive_ReadModeElementType", "", char, u params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - const auto ref_val = fixture.tex_h.Tex2DLayered(x, y, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } } @@ -160,17 +160,17 @@ TEMPLATE_TEST_CASE("Unit_tex2DLayered_Positive_ReadModeNormalizedFloat", "", cha params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - auto ref_val = fixture.tex_h.Tex2DLayered(x, y, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/tex2DLayeredGrad.cc b/projects/hip-tests/catch/unit/texture/tex2DLayeredGrad.cc index 7e45c9e021..9094eba9fc 100644 --- a/projects/hip-tests/catch/unit/texture/tex2DLayeredGrad.cc +++ b/projects/hip-tests/catch/unit/texture/tex2DLayeredGrad.cc @@ -88,17 +88,17 @@ TEMPLATE_TEST_CASE("Unit_tex2DLayeredGrad_Positive_ReadModeElementType", "", cha params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - const auto ref_val = fixture.tex_h.Tex2DLayered(x, y, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } } @@ -160,17 +160,17 @@ TEMPLATE_TEST_CASE("Unit_tex2DLayeredGrad_Positive_ReadModeNormalizedFloat", "", params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - auto ref_val = fixture.tex_h.Tex2DLayered(x, y, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/tex2DLayeredLod.cc b/projects/hip-tests/catch/unit/texture/tex2DLayeredLod.cc index b14ccfe4f6..0384405e2b 100644 --- a/projects/hip-tests/catch/unit/texture/tex2DLayeredLod.cc +++ b/projects/hip-tests/catch/unit/texture/tex2DLayeredLod.cc @@ -88,17 +88,17 @@ TEMPLATE_TEST_CASE("Unit_tex2DLayeredLod_Positive_ReadModeElementType", "", char params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - const auto ref_val = fixture.tex_h.Tex2DLayered(x, y, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Layer: " << layer); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } } @@ -160,17 +160,17 @@ TEMPLATE_TEST_CASE("Unit_tex2DLayeredLod_Positive_ReadModeNormalizedFloat", "", params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - auto ref_val = fixture.tex_h.Tex2DLayered(x, y, layer, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val) { + INFO("Layer: " << layer); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/tex2DLod.cc b/projects/hip-tests/catch/unit/texture/tex2DLod.cc index 017456d0ed..9b1880b63c 100644 --- a/projects/hip-tests/catch/unit/texture/tex2DLod.cc +++ b/projects/hip-tests/catch/unit/texture/tex2DLod.cc @@ -86,16 +86,16 @@ TEMPLATE_TEST_CASE("Unit_tex2DLod_Positive_ReadModeElementType", "", char, unsig params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - const auto ref_val = fixture.tex_h.Tex2D(x, y, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } @@ -154,16 +154,16 @@ TEMPLATE_TEST_CASE("Unit_tex2DLod_Positive_ReadModeNormalizedFloat", "", char, u params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - auto ref_val = fixture.tex_h.Tex2D(x, y, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex2Dgather.cc b/projects/hip-tests/catch/unit/texture/tex2Dgather.cc index 6896c53657..050a0a75b9 100644 --- a/projects/hip-tests/catch/unit/texture/tex2Dgather.cc +++ b/projects/hip-tests/catch/unit/texture/tex2Dgather.cc @@ -88,17 +88,17 @@ TEMPLATE_TEST_CASE("Unit_tex2Dgather_Positive_ReadModeElementType", "", char, un params.tex_desc.normalizedCoords); y = GetCoordinate(y, params.NumItersY(), params.Height(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("comp: " << comp); - const auto ref_val = fixture.tex_h.Tex2DGather(x, y, comp, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("comp: " << comp); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex3D.cc b/projects/hip-tests/catch/unit/texture/tex3D.cc index bce788ce69..f1203f945b 100644 --- a/projects/hip-tests/catch/unit/texture/tex3D.cc +++ b/projects/hip-tests/catch/unit/texture/tex3D.cc @@ -95,18 +95,18 @@ TEMPLATE_TEST_CASE("Unit_tex3D_Positive_ReadModeElementType", "", char, unsigned params.tex_desc.normalizedCoords); z = GetCoordinate(z, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - const auto ref_val = fixture.tex_h.Tex3D(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } @@ -172,18 +172,18 @@ TEMPLATE_TEST_CASE("Unit_tex3D_Positive_ReadModeNormalizedFloat", "", char, unsi params.tex_desc.normalizedCoords); z = GetCoordinate(z, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto ref_val = fixture.tex_h.Tex3D(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex3DGrad.cc b/projects/hip-tests/catch/unit/texture/tex3DGrad.cc index 624aa13d8a..dfbc2b5429 100644 --- a/projects/hip-tests/catch/unit/texture/tex3DGrad.cc +++ b/projects/hip-tests/catch/unit/texture/tex3DGrad.cc @@ -93,18 +93,18 @@ TEMPLATE_TEST_CASE("Unit_tex3DGrad_Positive_ReadModeElementType", "", char, unsi params.tex_desc.normalizedCoords); z = GetCoordinate(z, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - const auto ref_val = fixture.tex_h.Tex3D(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } @@ -170,18 +170,18 @@ TEMPLATE_TEST_CASE("Unit_tex3DGrad_Positive_ReadModeNormalizedFloat", "", char, params.tex_desc.normalizedCoords); z = GetCoordinate(z, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto ref_val = fixture.tex_h.Tex3D(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/tex3DLod.cc b/projects/hip-tests/catch/unit/texture/tex3DLod.cc index cb418e1998..5aa31198e9 100644 --- a/projects/hip-tests/catch/unit/texture/tex3DLod.cc +++ b/projects/hip-tests/catch/unit/texture/tex3DLod.cc @@ -93,18 +93,18 @@ TEMPLATE_TEST_CASE("Unit_tex3DLod_Positive_ReadModeElementType", "", char, unsig params.tex_desc.normalizedCoords); z = GetCoordinate(z, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - const auto ref_val = fixture.tex_h.Tex3D(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } @@ -170,18 +170,18 @@ TEMPLATE_TEST_CASE("Unit_tex3DLod_Positive_ReadModeNormalizedFloat", "", char, u params.tex_desc.normalizedCoords); z = GetCoordinate(z, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto ref_val = fixture.tex_h.Tex3D(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[i], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[i], ref_val)) { + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } diff --git a/projects/hip-tests/catch/unit/texture/texCubemap.cc b/projects/hip-tests/catch/unit/texture/texCubemap.cc index 3c30028269..5c67381754 100644 --- a/projects/hip-tests/catch/unit/texture/texCubemap.cc +++ b/projects/hip-tests/catch/unit/texture/texCubemap.cc @@ -92,23 +92,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemap_Positive_ReadModeElementType", "", char, uns params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - const auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } @@ -175,23 +174,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemap_Positive_ReadModeNormalizedFloat", "", char, params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/texCubemapGrad.cc b/projects/hip-tests/catch/unit/texture/texCubemapGrad.cc index 6151bbdcc9..da72490d9d 100644 --- a/projects/hip-tests/catch/unit/texture/texCubemapGrad.cc +++ b/projects/hip-tests/catch/unit/texture/texCubemapGrad.cc @@ -92,23 +92,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemapGrad_Positive_ReadModeElementType", "", char, params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - const auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } @@ -175,23 +174,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemapGrad_Positive_ReadModeNormalizedFloat", "", c params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/texCubemapLayered.cc b/projects/hip-tests/catch/unit/texture/texCubemapLayered.cc index 1ae0cd462a..cef3de8e04 100644 --- a/projects/hip-tests/catch/unit/texture/texCubemapLayered.cc +++ b/projects/hip-tests/catch/unit/texture/texCubemapLayered.cc @@ -94,23 +94,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLayered_Positive_ReadModeElementType", "", ch params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - const auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } @@ -180,23 +179,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLayered_Positive_ReadModeNormalizedFloat", "" params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/texCubemapLayeredGrad.cc b/projects/hip-tests/catch/unit/texture/texCubemapLayeredGrad.cc index fe7943ba9a..765c5f8be9 100644 --- a/projects/hip-tests/catch/unit/texture/texCubemapLayeredGrad.cc +++ b/projects/hip-tests/catch/unit/texture/texCubemapLayeredGrad.cc @@ -94,24 +94,23 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLayeredGrad_Positive_ReadModeElementType", "" params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - const auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } @@ -181,24 +180,23 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLayeredGrad_Positive_ReadModeNormalizedFloat" params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false;) + } } } } diff --git a/projects/hip-tests/catch/unit/texture/texCubemapLayeredLod.cc b/projects/hip-tests/catch/unit/texture/texCubemapLayeredLod.cc index c64e0d5933..330f47bf2b 100644 --- a/projects/hip-tests/catch/unit/texture/texCubemapLayeredLod.cc +++ b/projects/hip-tests/catch/unit/texture/texCubemapLayeredLod.cc @@ -94,24 +94,23 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLayeredLod_Positive_ReadModeElementType", "", params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - const auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } @@ -181,24 +180,23 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLayeredLod_Positive_ReadModeNormalizedFloat", params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("Layer: " << layer); - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("Layer: " << layer); + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } diff --git a/projects/hip-tests/catch/unit/texture/texCubemapLod.cc b/projects/hip-tests/catch/unit/texture/texCubemapLod.cc index 40d2bd8a12..96c0cfce8e 100644 --- a/projects/hip-tests/catch/unit/texture/texCubemapLod.cc +++ b/projects/hip-tests/catch/unit/texture/texCubemapLod.cc @@ -92,23 +92,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLod_Positive_ReadModeElementType", "", char, params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - const auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } @@ -175,23 +174,22 @@ TEMPLATE_TEST_CASE("Unit_texCubemapLod_Positive_ReadModeNormalizedFloat", "", ch params.tex_desc.normalizedCoords); float z = GetCoordinate(k, params.NumItersZ(), params.Depth(), params.num_subdivisions, params.tex_desc.normalizedCoords); - - INFO("i: " << i); - INFO("j: " << j); - INFO("k: " << k); - INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); - INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); - INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); - INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); - INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); - INFO("x: " << std::fixed << std::setprecision(16) << x); - INFO("y: " << std::fixed << std::setprecision(16) << y); - INFO("z: " << std::fixed << std::setprecision(16) << z); - auto index = k * params.NumItersX() * params.NumItersY() + j * params.NumItersX() + i; - auto ref_val = fixture.tex_h.TexCubemap(x, y, z, params.tex_desc); - REQUIRE(fixture.Verify(fixture.out_alloc_h[index], ref_val)); + if (!fixture.Verify(fixture.out_alloc_h[index], ref_val)) { + INFO("i: " << i); + INFO("j: " << j); + INFO("k: " << k); + INFO("Filtering mode: " << FilteringModeToString(params.tex_desc.filterMode)); + INFO("Normalized coordinates: " << std::boolalpha << params.tex_desc.normalizedCoords); + INFO("Address mode X: " << AddressModeToString(params.tex_desc.addressMode[0])); + INFO("Address mode Y: " << AddressModeToString(params.tex_desc.addressMode[1])); + INFO("Address mode Z: " << AddressModeToString(params.tex_desc.addressMode[2])); + INFO("x: " << std::fixed << std::setprecision(16) << x); + INFO("y: " << std::fixed << std::setprecision(16) << y); + INFO("z: " << std::fixed << std::setprecision(16) << z); + REQUIRE(false); + } } } } diff --git a/projects/rocr-runtime/runtime/hsa-runtime/CMakeLists.txt b/projects/rocr-runtime/runtime/hsa-runtime/CMakeLists.txt index a9237129b5..83b9e749b1 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/CMakeLists.txt +++ b/projects/rocr-runtime/runtime/hsa-runtime/CMakeLists.txt @@ -261,15 +261,10 @@ if (${PC_SAMPLING_SUPPORT}) target_sources( ${CORE_RUNTIME_TARGET} PRIVATE ${PCS_SRCS} ) endif() -if (UNIX) - if ( NOT DEFINED IMAGE_SUPPORT AND CMAKE_SYSTEM_PROCESSOR MATCHES "i?86|x86_64|amd64|AMD64|loongarch64" ) - set ( IMAGE_SUPPORT ON ) - endif() -set ( IMAGE_SUPPORT ${IMAGE_SUPPORT} CACHE BOOL "Build with image support (default: ON for x86, OFF elsewise)." ) -else() - # Force IMAGE_SUPPORT to be OFF - set(IMAGE_SUPPORT OFF CACHE BOOL "Build with image support (forced to OFF)" FORCE) +if (NOT DEFINED IMAGE_SUPPORT AND CMAKE_SYSTEM_PROCESSOR MATCHES "i?86|x86_64|amd64|AMD64|loongarch64") + set (IMAGE_SUPPORT ON) endif() +set (IMAGE_SUPPORT ${IMAGE_SUPPORT} CACHE BOOL "Build with image support (default: ON for x86, OFF elsewise).") ## Optional image module defintions. if(${IMAGE_SUPPORT}) diff --git a/projects/rocr-runtime/runtime/hsa-runtime/core/runtime/amd_topology.cpp b/projects/rocr-runtime/runtime/hsa-runtime/core/runtime/amd_topology.cpp index d141a73793..ba69a84075 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/core/runtime/amd_topology.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/core/runtime/amd_topology.cpp @@ -289,7 +289,9 @@ void SurfaceGpuList(std::vector& gpu_list, bool xnack_mode, bool enable core::g_use_interrupt_wait = false; if (core::Runtime::runtime_singleton_->thunkLoader()->IsDXG()) { + core::Runtime::runtime_singleton_->flag().disable_image(true); #if defined(_WIN32) + core::Runtime::runtime_singleton_->flag().disable_image(false); if (node_prop.Capability2.ui32.AqlEmulationPm4_) #endif { @@ -297,7 +299,6 @@ void SurfaceGpuList(std::vector& gpu_list, bool xnack_mode, bool enable core::Runtime::runtime_singleton_->flag().disable_scratch(); } core::Runtime::runtime_singleton_->flag().set_sdma(false, false); - core::Runtime::runtime_singleton_->flag().disable_image(true); core::Runtime::runtime_singleton_->flag().disable_xnack(); core::Runtime::runtime_singleton_->flag().disable_fine_grain_pcie(); core::Runtime::runtime_singleton_->flag().set_ipc_mode_legacy(false); diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/blit_kernel.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/blit_kernel.cpp index 01b54fafa9..c34888d6d9 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/blit_kernel.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/blit_kernel.cpp @@ -41,11 +41,6 @@ //////////////////////////////////////////////////////////////////////////////// #include "blit_kernel.h" - -#if (defined(WIN32) || defined(_WIN32)) -#define NOMINMAX -#endif - #include #include #include diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/CMakeLists.txt b/projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/CMakeLists.txt index e7a5a22a48..93cb816de7 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/CMakeLists.txt +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/CMakeLists.txt @@ -67,10 +67,12 @@ endif() ## Add custom command to generate a kernel code object file ##========================================== function(gen_kernel_bc TARGET_ID INPUT_FILE OUTPUT_FILE) - + if(WIN32) + set (WIN_CL_COMPILE_OPTION "-fshort-wchar") + endif() separate_arguments(CLANG_ARG_LIST UNIX_COMMAND "-O2 -x cl -Xclang -finclude-default-header -cl-denorms-are-zero -cl-std=CL2.0 - -target amdgcn-amd-amdhsa -mcpu=${TARGET_ID} -mcode-object-version=4 + -target amdgcn-amd-amdhsa -mcpu=${TARGET_ID} -mcode-object-version=4 ${WIN_CL_COMPILE_OPTION} -o ${OUTPUT_FILE} ${INPUT_FILE}") ## Add custom command to produce a code object file. @@ -128,9 +130,18 @@ function(generate_blit_file BFILE) ## Add a custom command that generates opencl_blit_objects.cpp ## This depends on all the generated code object files and the C++ generator script. - add_custom_command(OUTPUT ${BFILE}.cpp + if (UNIX) + add_custom_command(OUTPUT ${BFILE}.cpp COMMAND ${CMAKE_CURRENT_SOURCE_DIR}/create_hsaco_ascii_file.sh ${CMAKE_CURRENT_BINARY_DIR}/${BFILE}.cpp DEPENDS ${HSACO_TARG_LIST} create_hsaco_ascii_file.sh ) + else() + find_package(Python3 COMPONENTS Interpreter REQUIRED) + add_custom_command( + OUTPUT ${BFILE}.cpp + COMMAND ${Python3_EXECUTABLE} ${CMAKE_CURRENT_SOURCE_DIR}/create_hsaco_ascii_file.py ${CMAKE_CURRENT_BINARY_DIR}/${BFILE}.cpp ${HSACO_TARG_LIST} + COMMENT "Collating blit shaders..." + DEPENDS ${HSACO_TARG_LIST} create_hsaco_ascii_file.py) + endif() ## Export a target that builds (and depends on) opencl_blit_objects.cpp add_custom_target( ${BFILE} DEPENDS ${CMAKE_CURRENT_BINARY_DIR}/${BFILE}.cpp ) diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/create_hsaco_ascii_file.py b/projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/create_hsaco_ascii_file.py new file mode 100644 index 0000000000..91fac67d42 --- /dev/null +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/blit_src/create_hsaco_ascii_file.py @@ -0,0 +1,71 @@ +################################################################################ +## +## Copyright (c) Advanced Micro Devices, Inc., or its affiliates. +## +## SPDX-License-Identifier: MIT +## +################################################################################ +import sys + +def GetSize(fileobject): + fileobject.seek(0,2) # move the cursor to the end of the file + size = fileobject.tell() + return size + +def DumpFile(header, input_name): + try: + with open(input_name, "rb") as binary_file: + # Read the entire content of the file as bytes + binary_data = binary_file.read() + file_size = GetSize(binary_file) + #print(f"Binary size: {file_size}") + # Reset file pointer + binary_file.seek(0) + parts = input_name.split('.') + file_name = parts[0] + content = f"unsigned char {file_name}""[] = {\n " + + header.write(content) + line = 0 + count = 0 + for byte_value in binary_data: + count += 1 + padded_hex = '{:02x}'.format(byte_value) + if (count != file_size): + header.write(f"0x{padded_hex},") + else: + header.write(f"0x{padded_hex}") + line += 1 + if (line == 12): + header.write(f"\n ") + line = 0 + else: + header.write(f" ") + + header.write("\n};\nunsigned int "f"{file_name}_len = {file_size};\n") + + except FileNotFoundError: + print(f"Error: The file {input_name} was not found.") + except Exception as e: + print(f"An error occurred: {e}") + + +if len(sys.argv) > 1: + header_name = sys.argv[1]; + with open(header_name, 'w') as header: + header.write("//==============================================================================\n") + header.write("// This file is automatically generated during build process, don't modify it\n") + header.write("//==============================================================================\n\n") + header.write("namespace rocr {\n") + header.write("namespace image {\n\n") + + for i, arg in enumerate(sys.argv): + if (i > 1): + #print(f"File {i}: {arg}\n") + DumpFile(header, arg) + header.write("} // namespace image\n") + header.write("} // namespace rocr\n\n") + +else: + print("Empty arguments!") + \ No newline at end of file diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager.cpp index f3090eaccf..3750a64532 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager.cpp @@ -52,11 +52,6 @@ #include #include -#if (defined(WIN32) || defined(_WIN32)) -#define NOMINMAX -__inline long int lrintf(float f) { return _mm_cvtss_si32(_mm_load_ss(&f)); } -#endif - namespace rocr { namespace image { diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_ai.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_ai.cpp index 8a6ea1d6ea..bfe0cb729b 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_ai.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_ai.cpp @@ -40,7 +40,6 @@ // //////////////////////////////////////////////////////////////////////////////// -#define NOMINMAX #include "image_manager_ai.h" #include diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx11.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx11.cpp index 1c1727885c..3f993d3150 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx11.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx11.cpp @@ -40,7 +40,6 @@ // //////////////////////////////////////////////////////////////////////////////// -#define NOMINMAX #include "image_manager_gfx11.h" #include diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx12.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx12.cpp index 2ac9640c56..0f0211d05d 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx12.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_gfx12.cpp @@ -40,7 +40,6 @@ // //////////////////////////////////////////////////////////////////////////////// -#define NOMINMAX #include "image_manager_gfx12.h" #include diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_kv.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_kv.cpp index b7e142fe3a..1bf4cc592a 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_kv.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_kv.cpp @@ -40,7 +40,6 @@ // //////////////////////////////////////////////////////////////////////////////// -#define NOMINMAX #include "image_manager_kv.h" #include @@ -100,8 +99,8 @@ hsa_status_t ImageManagerKv::Initialize(hsa_agent_t agent_handle) { assert(status == HSA_STATUS_SUCCESS); HsaGpuTileConfig tileConfig = {0}; - unsigned int tc[40]; - unsigned int mtc[40]; + unsigned int tc[40] = {0}; + unsigned int mtc[40] = {0}; tileConfig.TileConfig = &tc[0]; tileConfig.NumTileConfigs = 40; tileConfig.MacroTileConfig = &mtc[0]; diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_nv.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_nv.cpp index 64a73dd6c2..81205e181c 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_nv.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/image_manager_nv.cpp @@ -40,7 +40,6 @@ // //////////////////////////////////////////////////////////////////////////////// -#define NOMINMAX #include "image_manager_nv.h" #include diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/image_runtime.cpp b/projects/rocr-runtime/runtime/hsa-runtime/image/image_runtime.cpp index 178e306852..5a0450849c 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/image_runtime.cpp +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/image_runtime.cpp @@ -40,7 +40,6 @@ // //////////////////////////////////////////////////////////////////////////////// -#define NOMINMAX #include "image_runtime.h" #include diff --git a/projects/rocr-runtime/runtime/hsa-runtime/image/util.h b/projects/rocr-runtime/runtime/hsa-runtime/image/util.h index 88cdf4ccc9..18f4a11ba9 100644 --- a/projects/rocr-runtime/runtime/hsa-runtime/image/util.h +++ b/projects/rocr-runtime/runtime/hsa-runtime/image/util.h @@ -280,22 +280,6 @@ static __forceinline uint64_t NextPow2(uint64_t value) { static __forceinline bool strIsEmpty(const char* str) noexcept { return str[0] == '\0'; } -static __forceinline std::string& ltrim(std::string& s) { - auto it = std::find_if(s.begin(), s.end(), - [](char c) { return !std::isspace(c, std::locale::classic()); }); - s.erase(s.begin(), it); - return s; -} - -static __forceinline std::string& rtrim(std::string& s) { - auto it = std::find_if(s.rbegin(), s.rend(), - [](char c) { return !std::isspace(c, std::locale::classic()); }); - s.erase(it.base(), s.end()); - return s; -} - -static __forceinline std::string& trim(std::string& s) { return ltrim(rtrim(s)); } - template static __forceinline uint32_t BitSelect(T p) { static_assert(sizeof(T) <= sizeof(uintptr_t), "Type out of range.");