From b05040b9198e6ea1b2002b7ba6240cfde070e534 Mon Sep 17 00:00:00 2001 From: Illia Silin <98187287+illsilin@users.noreply.github.com> Date: Fri, 8 May 2026 14:15:31 +0000 Subject: [PATCH] [rocm-libraries] ROCm/rocm-libraries#7111 (commit 651947f) [CK] Fix latest batch of staging compiler warnings ## Motivation Suppress the new batch of clang lifetimebound and invalidation warnings with the latest staging compiler. ## Technical Details ## Test Plan ## Test Result ## Submission Checklist - [ ] Look over the contributing guidelines at https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests. --- Jenkinsfile | 16 ++++++++++------ cmake/gtest.cmake | 4 +++- ..._conv2d_fwd_perlayer_quantization_example.inc | 5 +---- .../ck_tile/01_fmha/fmha_fwd_head_grouping.hpp | 3 +++ example/ck_tile/01_fmha/fmha_fwd_runner.hpp | 3 +++ .../grouped_convolution_utils.hpp | 4 ++++ include/ck/library/utility/gpu_verification.hpp | 3 +++ ...d_contraction_multiple_d_wmma_cshuffle_v3.hpp | 3 +++ ...tched_contraction_multiple_d_xdl_cshuffle.hpp | 3 +++ .../impl/device_batched_gemm_multi_d_xdl.hpp | 3 +++ ...emm_multiple_d_layernorm_wmma_cshuffle_v3.hpp | 3 +++ ...e_gemm_multiple_d_multiple_r_xdl_cshuffle.hpp | 3 +++ .../device_gemm_multiple_d_wmma_cshuffle_v3.hpp | 3 +++ .../impl/device_gemm_multiple_d_xdl_cshuffle.hpp | 3 +++ ...ed_conv_fwd_multiple_abd_wmma_cshuffle_v3.hpp | 3 +++ ...rouped_conv_fwd_multiple_abd_xdl_cshuffle.hpp | 3 +++ ...ped_conv_fwd_multiple_abd_xdl_cshuffle_v3.hpp | 3 +++ ...nv_fwd_multiple_d_multiple_r_xdl_cshuffle.hpp | 3 +++ ...grouped_conv_fwd_multiple_d_wmma_cshuffle.hpp | 3 +++ .../impl/device_multiple_reduce_multiblock.hpp | 3 +++ .../impl/device_multiple_reduce_threadwise.hpp | 3 +++ .../gpu/device/impl/device_permute_impl.hpp | 3 +++ ...plitk_contraction_multiple_d_xdl_cshuffle.hpp | 3 +++ .../gridwise_gemm_xdl_cshuffle_v3_multi_abd.hpp | 3 +++ include/ck/utility/type_convert.hpp | 3 +++ include/ck_tile/core/numeric/type_convert.hpp | 3 +++ include/ck_tile/ref/naive_attention.hpp | 3 +++ .../gpu/reference_gemm.hpp | 3 +++ .../profiler/profile_contraction_utils.hpp | 3 +++ test/common/csv_test_loader.hpp | 3 +++ .../test_grouped_convnd_bwd_data_dataset_xdl.cpp | 3 +++ ...est_grouped_convnd_bwd_weight_dataset_xdl.cpp | 3 +++ .../test_grouped_convnd_fwd_dataset_xdl.cpp | 3 +++ .../test_grouped_gemm_fixed_nk_bias.cpp | 3 +++ .../test_grouped_gemm_multi_abd_fixed_nk.cpp | 3 +++ test/grouped_gemm/test_grouped_gemm_util.hpp | 3 +++ .../test_grouped_gemm_tile_loop_util.hpp | 3 +++ 37 files changed, 117 insertions(+), 11 deletions(-) diff --git a/Jenkinsfile b/Jenkinsfile index 73f7c9630d..69dcab60d3 100644 --- a/Jenkinsfile +++ b/Jenkinsfile @@ -2119,9 +2119,11 @@ pipeline { } success { script { - // Report the parent stage build ck and run tests status - setGithubStatus("${env.STAGE_NAME}", 'success', "Stage ${env.STAGE_NAME} passed") - echo "Reporting success status for build ck and run tests" + node(rocmnode("nogpu")) { + // Report the parent stage build ck and run tests status + setGithubStatus("${env.STAGE_NAME}", 'success', "Stage ${env.STAGE_NAME} passed") + echo "Reporting success status for build ck and run tests" + } } } } @@ -2145,9 +2147,11 @@ pipeline { post { success { script { - // Report the skipped parent's stage status - setGithubStatus("${env.STAGE_NAME}", 'success', "Stage ${env.STAGE_NAME} passed") - echo "Process Performance Test Results stage skipped." + node(rocmnode("nogpu")) { + // Report the skipped parent's stage status + setGithubStatus("${env.STAGE_NAME}", 'success', "Stage ${env.STAGE_NAME} passed") + echo "Process Performance Test Results stage skipped." + } } } } diff --git a/cmake/gtest.cmake b/cmake/gtest.cmake index 33f0b4d548..46f2d8990c 100644 --- a/cmake/gtest.cmake +++ b/cmake/gtest.cmake @@ -71,6 +71,7 @@ set(GTEST_CXX_FLAGS -Wno-lifetime-safety-intra-tu-suggestions -Wno-lifetime-safety-cross-tu-suggestions -Wno-character-conversion + -Wno-lifetime-safety-invalidation ) if(WIN32) @@ -78,7 +79,8 @@ if(WIN32) -Wno-suggest-destructor-override -Wno-suggest-override -Wno-nonportable-system-include-path - -Wno-language-extension-token) + -Wno-language-extension-token + -Wno-lifetime-safety-invalidation) endif() target_compile_options(gtest PRIVATE ${GTEST_CXX_FLAGS}) diff --git a/example/40_conv2d_fwd_quantization/run_conv2d_fwd_perlayer_quantization_example.inc b/example/40_conv2d_fwd_quantization/run_conv2d_fwd_perlayer_quantization_example.inc index 26c3165446..19e365a3d8 100644 --- a/example/40_conv2d_fwd_quantization/run_conv2d_fwd_perlayer_quantization_example.inc +++ b/example/40_conv2d_fwd_quantization/run_conv2d_fwd_perlayer_quantization_example.inc @@ -132,10 +132,7 @@ bool run_grouped_conv_fwd(bool do_verification, ref_invoker.Run(ref_argument); - out_host.ForEach([&](auto&, auto idx) - { - out_element_op(out_host(idx), c_host(idx)); - }); + out_host.ForEach([&](auto&, auto idx) { out_element_op(out_host(idx), c_host(idx)); }); out_device_buf.FromDevice(out_device.mData.data()); diff --git a/example/ck_tile/01_fmha/fmha_fwd_head_grouping.hpp b/example/ck_tile/01_fmha/fmha_fwd_head_grouping.hpp index 9dad951d41..d0d85a2660 100644 --- a/example/ck_tile/01_fmha/fmha_fwd_head_grouping.hpp +++ b/example/ck_tile/01_fmha/fmha_fwd_head_grouping.hpp @@ -22,6 +22,8 @@ #define CK_TILE_FMHA_ENABLE_HEAD_GROUPING 1 #endif +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" #if CK_TILE_FMHA_ENABLE_HEAD_GROUPING CK_TILE_DECLARE_ENV_VAR_BOOL(CK_TILE_FMHA_HEAD_GROUP_LOG) CK_TILE_DECLARE_ENV_VAR_BOOL(CK_TILE_FMHA_DISABLE_HEAD_GROUPING) @@ -427,3 +429,4 @@ float run_fwd_head_grouped(const ck_tile::stream_config& sc, } // namespace fmha_fwd_head_grouping #endif // CK_TILE_FMHA_ENABLE_HEAD_GROUPING +#pragma clang diagnostic pop diff --git a/example/ck_tile/01_fmha/fmha_fwd_runner.hpp b/example/ck_tile/01_fmha/fmha_fwd_runner.hpp index e01555193f..1c99dffcda 100644 --- a/example/ck_tile/01_fmha/fmha_fwd_runner.hpp +++ b/example/ck_tile/01_fmha/fmha_fwd_runner.hpp @@ -27,6 +27,8 @@ #error "we should enable fmha_fwd_splitkv() api in order to cooperate with fmha_fwd_appendkv()" #endif +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" enum class fwd_result { success, @@ -2484,3 +2486,4 @@ fwd_result fmha_fwd_run(mode_enum mode, return pass ? fwd_result::success : fwd_result::failure; } +#pragma clang diagnostic pop diff --git a/example/ck_tile/20_grouped_convolution/grouped_convolution_utils.hpp b/example/ck_tile/20_grouped_convolution/grouped_convolution_utils.hpp index 5a714f6da7..ac04e6e2e4 100644 --- a/example/ck_tile/20_grouped_convolution/grouped_convolution_utils.hpp +++ b/example/ck_tile/20_grouped_convolution/grouped_convolution_utils.hpp @@ -15,6 +15,9 @@ #include "ck_tile/ops/elementwise/unary_element_wise_operation.hpp" #include "conv_configs.hpp" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" + template auto calculate_rtol_atol(const ck_tile::index_t GemmK, const ck_tile::index_t kbatch, @@ -144,3 +147,4 @@ struct InvokerResult float ave_time; ck_tile::index_t split_k; }; +#pragma clang diagnostic pop diff --git a/include/ck/library/utility/gpu_verification.hpp b/include/ck/library/utility/gpu_verification.hpp index 6be7e1886b..6fcc9ff391 100644 --- a/include/ck/library/utility/gpu_verification.hpp +++ b/include/ck/library/utility/gpu_verification.hpp @@ -18,6 +18,8 @@ #include "ck/host_utility/hip_check_error.hpp" #include "ck/library/utility/check_err.hpp" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" namespace ck { namespace profiler { @@ -517,3 +519,4 @@ float gpu_reduce_max(Iterator device_buffer, std::size_t size, hipStream_t strea } // namespace profiler } // namespace ck +#pragma clang diagnostic pop diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_wmma_cshuffle_v3.hpp index b59357ffe9..14e63bc8aa 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_wmma_cshuffle_v3.hpp @@ -19,6 +19,8 @@ #include "ck/host_utility/kernel_launch.hpp" #include "ck/utility/scheduler_enum.hpp" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" namespace ck { template & y, const Array } } // namespace ck +#pragma clang diagnostic pop diff --git a/include/ck_tile/core/numeric/type_convert.hpp b/include/ck_tile/core/numeric/type_convert.hpp index da5579f5f0..54d26444d5 100644 --- a/include/ck_tile/core/numeric/type_convert.hpp +++ b/include/ck_tile/core/numeric/type_convert.hpp @@ -13,6 +13,8 @@ #include "ck_tile/core/numeric/int8.hpp" #include "ck_tile/core/numeric/mxfp_convert.hpp" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" namespace ck_tile { #if CK_TILE_USE_CUSTOM_DATA_TYPE @@ -147,3 +149,4 @@ CK_TILE_SCALED_TYPE_CONVERT(fp16_t, fp16, pk_fp4_t, pk_fp4) #endif } // namespace ck_tile +#pragma clang diagnostic pop diff --git a/include/ck_tile/ref/naive_attention.hpp b/include/ck_tile/ref/naive_attention.hpp index fd7a4b31cb..e5e2c0a125 100644 --- a/include/ck_tile/ref/naive_attention.hpp +++ b/include/ck_tile/ref/naive_attention.hpp @@ -9,6 +9,8 @@ #include #include +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" namespace ck_tile { enum class naive_attention_layout_enum @@ -824,3 +826,4 @@ CK_TILE_HOST float naive_attention_fwd(naive_attention_fwd_traits t, #undef CK_TILE_DISPATCH_NAIVE_ATTEN_FWD_INTERNAL_ } // namespace ck_tile +#pragma clang diagnostic pop diff --git a/library/include/ck/library/reference_tensor_operation/gpu/reference_gemm.hpp b/library/include/ck/library/reference_tensor_operation/gpu/reference_gemm.hpp index c05c0605bf..37588fb2c3 100644 --- a/library/include/ck/library/reference_tensor_operation/gpu/reference_gemm.hpp +++ b/library/include/ck/library/reference_tensor_operation/gpu/reference_gemm.hpp @@ -10,6 +10,8 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/host_utility/kernel_launch.hpp" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" namespace ck { template & strides, std::vector& csv_pat } // namespace test } // namespace ck +#pragma clang diagnostic pop diff --git a/test/grouped_convnd_bwd_data/test_grouped_convnd_bwd_data_dataset_xdl.cpp b/test/grouped_convnd_bwd_data/test_grouped_convnd_bwd_data_dataset_xdl.cpp index a30f5d349d..777a83ea54 100644 --- a/test/grouped_convnd_bwd_data/test_grouped_convnd_bwd_data_dataset_xdl.cpp +++ b/test/grouped_convnd_bwd_data/test_grouped_convnd_bwd_data_dataset_xdl.cpp @@ -13,6 +13,8 @@ using namespace ck::tensor_layout::convolution; // Import tensor layout names (GNHWK, GKYXC, etc.) +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" // Load CSV data for 2D tests static std::vector Get2DTestCases() { @@ -315,3 +317,4 @@ TEST_P(TestGroupedConvndBwdData3dNDHWGKBFloat16SplitK2, ConvTest) INSTANTIATE_TEST_SUITE_P(Dataset, TestGroupedConvndBwdData3dNDHWGKBFloat16SplitK2, ::testing::ValuesIn(Get3DTestCases())); +#pragma clang diagnostic pop diff --git a/test/grouped_convnd_bwd_weight/test_grouped_convnd_bwd_weight_dataset_xdl.cpp b/test/grouped_convnd_bwd_weight/test_grouped_convnd_bwd_weight_dataset_xdl.cpp index 07d80dfad2..1de3434890 100644 --- a/test/grouped_convnd_bwd_weight/test_grouped_convnd_bwd_weight_dataset_xdl.cpp +++ b/test/grouped_convnd_bwd_weight/test_grouped_convnd_bwd_weight_dataset_xdl.cpp @@ -15,6 +15,8 @@ #include "profiler/profile_grouped_conv_bwd_weight_impl.hpp" // The actual GPU profiler that does convolution work #include "../common/csv_test_loader.hpp" // Shared CSV test case loader +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" using namespace ck::tensor_layout::convolution; // Load CSV data for 2D tests @@ -256,3 +258,4 @@ TEST_P(TestGroupedConvndBwdWeight3dNDHWGKBFloat16SplitK2, ConvTest) INSTANTIATE_TEST_SUITE_P(Dataset, TestGroupedConvndBwdWeight3dNDHWGKBFloat16SplitK2, ::testing::ValuesIn(Get3DTestCases())); +#pragma clang diagnostic pop diff --git a/test/grouped_convnd_fwd/test_grouped_convnd_fwd_dataset_xdl.cpp b/test/grouped_convnd_fwd/test_grouped_convnd_fwd_dataset_xdl.cpp index 8bfdbabd54..73da3713b6 100644 --- a/test/grouped_convnd_fwd/test_grouped_convnd_fwd_dataset_xdl.cpp +++ b/test/grouped_convnd_fwd/test_grouped_convnd_fwd_dataset_xdl.cpp @@ -11,6 +11,8 @@ #include "profiler/profile_grouped_conv_fwd_impl.hpp" // The actual GPU profiler that does convolution work #include "../common/csv_test_loader.hpp" // Shared CSV test case loader +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" using namespace ck::tensor_layout::convolution; // Import tensor layout names (NHWGC, GKYXC, etc.) // Load CSV data for 2D tests @@ -167,3 +169,4 @@ TEST_P(TestGroupedConvndFwd3dBFloat16, ConvTest) INSTANTIATE_TEST_SUITE_P(Dataset, TestGroupedConvndFwd3dBFloat16, ::testing::ValuesIn(Get3DTestCases())); +#pragma clang diagnostic pop diff --git a/test/grouped_gemm/test_grouped_gemm_fixed_nk_bias.cpp b/test/grouped_gemm/test_grouped_gemm_fixed_nk_bias.cpp index 63acfbb2e3..f247329f32 100644 --- a/test/grouped_gemm/test_grouped_gemm_fixed_nk_bias.cpp +++ b/test/grouped_gemm/test_grouped_gemm_fixed_nk_bias.cpp @@ -18,6 +18,8 @@ #include #include +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" static ck::index_t param_mask = 0xffffff; using FP32 = float; @@ -302,3 +304,4 @@ int main(int argc, char** argv) } return RUN_ALL_TESTS(); } +#pragma clang diagnostic pop diff --git a/test/grouped_gemm/test_grouped_gemm_multi_abd_fixed_nk.cpp b/test/grouped_gemm/test_grouped_gemm_multi_abd_fixed_nk.cpp index 05caa6ed6d..ab441b0156 100644 --- a/test/grouped_gemm/test_grouped_gemm_multi_abd_fixed_nk.cpp +++ b/test/grouped_gemm/test_grouped_gemm_multi_abd_fixed_nk.cpp @@ -15,6 +15,8 @@ #include "gtest/gtest.h" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" using FP32 = float; using FP16 = ck::half_t; using BF16 = ck::bhalf_t; @@ -237,3 +239,4 @@ int main(int argc, char** argv) testing::InitGoogleTest(&argc, argv); return RUN_ALL_TESTS(); } +#pragma clang diagnostic pop diff --git a/test/grouped_gemm/test_grouped_gemm_util.hpp b/test/grouped_gemm/test_grouped_gemm_util.hpp index 4f9b739351..c5d5994579 100644 --- a/test/grouped_gemm/test_grouped_gemm_util.hpp +++ b/test/grouped_gemm/test_grouped_gemm_util.hpp @@ -18,6 +18,8 @@ #include "profiler/profile_grouped_gemm_impl.hpp" #include "profiler/profile_grouped_gemm_fixed_nk_impl.hpp" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" extern ck::index_t param_mask; extern ck::index_t instance_index; @@ -325,3 +327,4 @@ class TestGroupedGemm : public testing::Test } // namespace test } // namespace ck +#pragma clang diagnostic pop diff --git a/test/grouped_gemm_tile_loop/test_grouped_gemm_tile_loop_util.hpp b/test/grouped_gemm_tile_loop/test_grouped_gemm_tile_loop_util.hpp index 66c291fb58..0f3cfea6f9 100644 --- a/test/grouped_gemm_tile_loop/test_grouped_gemm_tile_loop_util.hpp +++ b/test/grouped_gemm_tile_loop/test_grouped_gemm_tile_loop_util.hpp @@ -19,6 +19,8 @@ #include "example/ck_tile/17_grouped_gemm/grouped_gemm_multi_d.hpp" #include "profiler/profile_grouped_gemm_tile_loop_generic_impl.hpp" +#pragma clang diagnostic push +#pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" extern ck::index_t param_mask; extern ck::index_t instance_index; @@ -171,3 +173,4 @@ class TestGroupedGemmTileLoop : public testing::Test } // namespace test } // namespace ck +#pragma clang diagnostic pop