From 75d15ba401f77584f14a34679d5c05d831303bd0 Mon Sep 17 00:00:00 2001 From: Zixuan Fan Date: Fri, 12 Jun 2026 11:01:57 -0400 Subject: [PATCH 1/3] refactor(hip-tests): remove redundant title strings from hip_tests_config.hh macros The generated hip_tests_config.hh macros included a redundant title string as the first element that was never used and complicated macro expansion. The current implementation generates macros like #define {name} "{name}", "{tags}" which requires extra preprocessor indirection (SECOND_ARG and GET_TAGS macros) to extract the tags. This change simplifies to #define {name} "{tags}" and eliminates the unnecessary preprocessor complexity. Changes: - Modified parse_config.py (line 60) to generate macros in the simplified format: #define {name} "{tags}" instead of #define {name} "{name}", "{tags}" - Updated hip_test_common.hh (lines 24-32) to eliminate SECOND_ARG and GET_TAGS indirection, reducing from 4 macros to 2 macros: * Before: #define HIP_TEST_CASE(name) TEST_CASE(#name, GET_TAGS(name)) * After: #define HIP_TEST_CASE(name) TEST_CASE(#name, name) Benefits: - Reduces generated header size - Removes unnecessary preprocessor complexity - Simplifies macro expansion JIRA ID: AIRUNTIME-1933 (cherry picked from commit 86341708855d796af11b4ce48513202665c324fd) --- projects/hip-tests/catch/config/parse_config.py | 2 +- projects/hip-tests/catch/include/hip_test_common.hh | 7 ++----- 2 files changed, 3 insertions(+), 6 deletions(-) diff --git a/projects/hip-tests/catch/config/parse_config.py b/projects/hip-tests/catch/config/parse_config.py index 0651ba0940c..3de7dc7a3ae 100644 --- a/projects/hip-tests/catch/config/parse_config.py +++ b/projects/hip-tests/catch/config/parse_config.py @@ -81,7 +81,7 @@ def create_test_definition( for entry in disabled: tags_str += f"[exclude_{entry}]" - return f'#define {case_name} "{case_name}", "{tags_str}"' + return f'#define {case_name} "{tags_str}"' def generate_parameter_header(cmd_options, output_path): diff --git a/projects/hip-tests/catch/include/hip_test_common.hh b/projects/hip-tests/catch/include/hip_test_common.hh index cbc8f0f7b85..9f5dd5fb29e 100644 --- a/projects/hip-tests/catch/include/hip_test_common.hh +++ b/projects/hip-tests/catch/include/hip_test_common.hh @@ -26,13 +26,10 @@ #ifdef ENABLE_YAML_TAGS #include "hip_test_config.hh" -#define SECOND_ARG(a, b, ...) b -#define GET_TAGS(...) SECOND_ARG(__VA_ARGS__) -#define HIP_TEST_CASE(name) TEST_CASE(#name, GET_TAGS(name)) -#define HIP_TEMPLATE_TEST_CASE(name, ...) TEMPLATE_TEST_CASE(#name, GET_TAGS(name), __VA_ARGS__) +#define HIP_TEST_CASE(name) TEST_CASE(#name, name) +#define HIP_TEMPLATE_TEST_CASE(name, ...) TEMPLATE_TEST_CASE(#name, name, __VA_ARGS__) #else -#define GET_TAGS(...) #define HIP_TEST_CASE(name) TEST_CASE(#name, "") #define HIP_TEMPLATE_TEST_CASE(name, ...) TEMPLATE_TEST_CASE(#name, "", __VA_ARGS__) #endif From 01b1f3086af70f1c749a0778ee1a808355334a2a Mon Sep 17 00:00:00 2001 From: Zixuan Fan Date: Thu, 27 Aug 2026 16:27:09 -0400 Subject: [PATCH 2/3] Update hip-tests wrapper macros for single-string HIP_TEST_CASE format Fix raw TEST_CASE/TEMPLATE_TEST_CASE call sites and hip_test_common.hh wrapper macros broken by parse_config.py now emitting a single concatenated string literal instead of a name+tags pair. (cherry picked from commit 3db634fe58e57dc9e5e389882a2ed6edb99c7408) --- .../catch/include/hip_test_common.hh | 2 +- .../catch/unit/atomics/unsafeAtomicAdd.cc | 2 +- .../unit/cooperativeGrps/thread_block_tile.cc | 20 +++++++++---------- .../unit/deviceLib/AtomicAdd_Coherent.cc | 2 +- .../unit/deviceLib/AtomicAdd_NonCoherent.cc | 2 +- .../deviceLib/unsafeAtomicAdd_Coherent.cc | 2 +- .../deviceLib/unsafeAtomicAdd_NonCoherent.cc | 2 +- projects/hip-tests/catch/unit/rtc/rtc_coop.cc | 2 +- 8 files changed, 17 insertions(+), 17 deletions(-) diff --git a/projects/hip-tests/catch/include/hip_test_common.hh b/projects/hip-tests/catch/include/hip_test_common.hh index 9f5dd5fb29e..d257e98cb42 100644 --- a/projects/hip-tests/catch/include/hip_test_common.hh +++ b/projects/hip-tests/catch/include/hip_test_common.hh @@ -35,7 +35,7 @@ #endif /** - * @brief Check if running at quick level (level_0). + * @brief Check if r1ning at quick level (level_0). * Use this to reduce test parameters for faster execution. */ inline bool isQuickLevel() { diff --git a/projects/hip-tests/catch/unit/atomics/unsafeAtomicAdd.cc b/projects/hip-tests/catch/unit/atomics/unsafeAtomicAdd.cc index 6eb2c40bfe5..74f141d20fe 100644 --- a/projects/hip-tests/catch/unit/atomics/unsafeAtomicAdd.cc +++ b/projects/hip-tests/catch/unit/atomics/unsafeAtomicAdd.cc @@ -177,7 +177,7 @@ static void runUnsafeAtomicAddHalfAndBfloatTest() { HIP_CHECK(hipFree(out)); } -TEST_CASE(Unit_unsafe_atomic_add_half_and_bfloat) { +HIP_TEST_CASE(Unit_unsafe_atomic_add_half_and_bfloat) { SECTION("__half2") { runUnsafeAtomicAddHalfAndBfloatTest<__half2>(); } SECTION("__hip_bfloat162") { runUnsafeAtomicAddHalfAndBfloatTest<__hip_bfloat162>(); } SECTION("__half") { runUnsafeAtomicAddHalfAndBfloatTest<__half>(); } diff --git a/projects/hip-tests/catch/unit/cooperativeGrps/thread_block_tile.cc b/projects/hip-tests/catch/unit/cooperativeGrps/thread_block_tile.cc index e52a7a5cb3c..ba68fd6648e 100644 --- a/projects/hip-tests/catch/unit/cooperativeGrps/thread_block_tile.cc +++ b/projects/hip-tests/catch/unit/cooperativeGrps/thread_block_tile.cc @@ -992,7 +992,7 @@ HIP_TEST_CASE(Unit_Thread_Block_Tile_Reduce_Trivially_Copyable_Parameters) } } -TEST_CASE(Unit_Thread_Block_Tile_Scan_Trivially_Copyable_Parameters) +HIP_TEST_CASE(Unit_Thread_Block_Tile_Scan_Trivially_Copyable_Parameters) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1278,7 +1278,7 @@ HIP_TEST_CASE(Unit_Thread_Block_Tile_Reduce_All_Parameter_Sizes) } } -TEST_CASE(Unit_Thread_Block_Tile_Scan_All_Parameter_Sizes) +HIP_TEST_CASE(Unit_Thread_Block_Tile_Scan_All_Parameter_Sizes) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1475,7 +1475,7 @@ void testScanForTileSize() } } -TEST_CASE(Unit_Thread_Block_Tile_Inclusive_Scan_Basic) +HIP_TEST_CASE(Unit_Thread_Block_Tile_Inclusive_Scan_Basic) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1495,7 +1495,7 @@ TEST_CASE(Unit_Thread_Block_Tile_Inclusive_Scan_Basic) } } -TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Exclusive_Scan_Basic, int, half) +HIP_TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Exclusive_Scan_Basic, int, half) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1533,7 +1533,7 @@ TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Exclusive_Scan_Basic, int, half) // for all the tile sizes and all input types, using random input values, calculates the scan // values. Additionally, randomly make some threads not participate for the coalesced_threads case -TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Scan_Random_arithmetic, int, unsigned int, long long, +HIP_TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Scan_Random_arithmetic, int, unsigned int, long long, unsigned long long, float, half, double) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1559,7 +1559,7 @@ TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Scan_Random_arithmetic, int, unsigned } } -TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Scan_Random_boolean, int, unsigned int, long long, +HIP_TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Scan_Random_boolean, int, unsigned int, long long, unsigned long long) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1586,7 +1586,7 @@ TEMPLATE_TEST_CASE(Unit_Thread_Block_Tile_Scan_Random_boolean, int, unsigned int } // make sures that tiled blocks that use the y or z dimension work correctly -TEST_CASE(Unit_Thread_Block_Tile_2D_3D_Blocks) +HIP_TEST_CASE(Unit_Thread_Block_Tile_2D_3D_Blocks) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1621,7 +1621,7 @@ TEST_CASE(Unit_Thread_Block_Tile_2D_3D_Blocks) } } -TEMPLATE_TEST_CASE(Unit_Thread_Block_Coalesced_Scan_arithmetic, int, unsigned int, long long, +HIP_TEMPLATE_TEST_CASE(Unit_Thread_Block_Coalesced_Scan_arithmetic, int, unsigned int, long long, unsigned long long, float, half, double) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1647,7 +1647,7 @@ TEMPLATE_TEST_CASE(Unit_Thread_Block_Coalesced_Scan_arithmetic, int, unsigned in } } -TEMPLATE_TEST_CASE(Unit_Thread_Block_Coalesced_Scan_boolean, int, unsigned int, long long, +HIP_TEMPLATE_TEST_CASE(Unit_Thread_Block_Coalesced_Scan_boolean, int, unsigned int, long long, unsigned long long) { CHECK_COOPERATIVE_LAUNCH_SUPPORT @@ -1706,7 +1706,7 @@ void __global__ binaryPartitionTiled(int* out, int* ranks) } } -TEST_CASE(Unit_Thread_Block_Scan_partition) +HIP_TEST_CASE(Unit_Thread_Block_Scan_partition) { CHECK_COOPERATIVE_LAUNCH_SUPPORT diff --git a/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_Coherent.cc b/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_Coherent.cc index d6387fc537d..a2ff2a607fa 100644 --- a/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_Coherent.cc +++ b/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_Coherent.cc @@ -140,7 +140,7 @@ static void runAtomicAddCoherentUnsafeFlagTest(const std::string& gfxName) { HIP_CHECK(hipHostFree(result)); } -TEST_CASE(Unit_AtomicAdd_Coherent) { +HIP_TEST_CASE(Unit_AtomicAdd_Coherent) { hipDeviceProp_t prop; int device; HIP_CHECK(hipGetDevice(&device)); diff --git a/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_NonCoherent.cc b/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_NonCoherent.cc index d148669f7e8..39ca2e0f891 100644 --- a/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_NonCoherent.cc +++ b/projects/hip-tests/catch/unit/deviceLib/AtomicAdd_NonCoherent.cc @@ -134,7 +134,7 @@ static void runAtomicAddNonCoherentUnsafeFlagTest() { HIP_CHECK(hipHostFree(result)); } -TEST_CASE(Unit_AtomicAdd_NonCoherent) { +HIP_TEST_CASE(Unit_AtomicAdd_NonCoherent) { hipDeviceProp_t prop; int device; HIP_CHECK(hipGetDevice(&device)); diff --git a/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_Coherent.cc b/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_Coherent.cc index e07bee2dda1..ad94ec45aff 100644 --- a/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_Coherent.cc +++ b/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_Coherent.cc @@ -167,7 +167,7 @@ static void runUnsafeAtomicAddCoherentUnsafeFlagTest(const std::string& gfxName) HIP_CHECK(hipHostFree(result)); } -TEST_CASE(Unit_unsafeAtomicAdd_Coherent) { +HIP_TEST_CASE(Unit_unsafeAtomicAdd_Coherent) { hipDeviceProp_t prop; int device; HIP_CHECK(hipGetDevice(&device)); diff --git a/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_NonCoherent.cc b/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_NonCoherent.cc index bdbfb6b6eb4..adac88bfcc4 100644 --- a/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_NonCoherent.cc +++ b/projects/hip-tests/catch/unit/deviceLib/unsafeAtomicAdd_NonCoherent.cc @@ -146,7 +146,7 @@ static void runUnsafeAtomicAddNonCoherentUnsafeFlagTest() { HIP_CHECK(hipHostFree(result)); } -TEST_CASE(Unit_unsafeAtomicAdd_NonCoherent) { +HIP_TEST_CASE(Unit_unsafeAtomicAdd_NonCoherent) { hipDeviceProp_t prop; int device; HIP_CHECK(hipGetDevice(&device)); diff --git a/projects/hip-tests/catch/unit/rtc/rtc_coop.cc b/projects/hip-tests/catch/unit/rtc/rtc_coop.cc index 7ee7351dc62..556a5e6982d 100644 --- a/projects/hip-tests/catch/unit/rtc/rtc_coop.cc +++ b/projects/hip-tests/catch/unit/rtc/rtc_coop.cc @@ -261,7 +261,7 @@ HIP_TEST_CASE(Unit_Rtc_CoopReduce) } } -TEST_CASE(Unit_Rtc_CoopScan) +HIP_TEST_CASE(Unit_Rtc_CoopScan) { const std::tuple allTypes; const std::tuple integralTypes; From a9a591ed33c151f0d6d282f00e4aee9788e94b49 Mon Sep 17 00:00:00 2001 From: Zixuan Fan Date: Fri, 28 Aug 2026 18:05:20 -0400 Subject: [PATCH 3/3] docs(hip-tests): update AUTHORING.md for single-string HIP_TEST_CASE The guide described the removed GET_TAGS/SECOND_ARG indirection and the old compile error. Update it to the current expansion and the error a missing YAML entry now produces. Co-Authored-By: Claude Opus 5 (cherry picked from commit 468973dabab08f604a8736095bcbf963845d92b1) --- projects/hip-tests/catch/contract/AUTHORING.md | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/projects/hip-tests/catch/contract/AUTHORING.md b/projects/hip-tests/catch/contract/AUTHORING.md index 65d8f16c20c..0a8cee3cc3a 100644 --- a/projects/hip-tests/catch/contract/AUTHORING.md +++ b/projects/hip-tests/catch/contract/AUTHORING.md @@ -75,11 +75,11 @@ per-API notes on why each gap exists. tags: [] ``` - This is **required, not optional**. `HIP_TEST_CASE(name)` expands through - `GET_TAGS(name)` → `SECOND_ARG(name)`, which needs a `name` macro that the - build **generates from this YAML** into `hip_test_config.hh`. A test-case name + This is **required, not optional**. `HIP_TEST_CASE(name)` expands to + `TEST_CASE(#name, name)`, which needs a `name` macro that the build + **generates from this YAML** into `hip_test_config.hh`. A test-case name with no YAML entry fails to compile with - *"too few arguments to function-like macro SECOND_ARG"*. After editing the + *"use of undeclared identifier ''"*. After editing the YAML you must re-run CMake configure so the header regenerates. 6. **Regenerate the test plan and update the docs**: run