diff --git a/CMakeLists.txt b/CMakeLists.txt index e2f3c46619..2056f23a7f 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -135,6 +135,11 @@ add_compile_options(-Wno-bit-int-extension) add_compile_options(-Wno-pass-failed) add_compile_options(-Wno-switch-default) add_compile_options(-Wno-unique-object-duplication) +if(CMAKE_CXX_COMPILER_ID STREQUAL "Clang") + add_compile_options(-Wno-unknown-warning-option) +else() + add_compile_options(-Wno-unknown-pragmas) +endif() # Increase the number of max elements in fold expressions add_compile_options(-fbracket-depth=1024) diff --git a/codegen/include/ck/host/types.hpp b/codegen/include/ck/host/types.hpp index 4db3941414..91dd3013a1 100644 --- a/codegen/include/ck/host/types.hpp +++ b/codegen/include/ck/host/types.hpp @@ -90,12 +90,15 @@ std::string SequenceStr(const std::vector& v); std::string MakeTuple(const std::vector& v); +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wglobal-constructors" +#endif template const std::string S = SequenceStr({xs...}); +#ifdef __clang__ #pragma clang diagnostic pop +#endif constexpr const char* PassThrough = "ck::tensor_operation::element_wise::PassThrough"; constexpr const char* Bilinear = "ck::tensor_operation::element_wise::Bilinear"; diff --git a/codegen/src/headers.cpp b/codegen/src/headers.cpp index 9d59176c92..0929879e6a 100644 --- a/codegen/src/headers.cpp +++ b/codegen/src/headers.cpp @@ -7,11 +7,14 @@ namespace ck { namespace host { +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wglobal-constructors" +#endif const std::string config_header = ""; +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif std::unordered_map GetHeaders() { diff --git a/codegen/test/include/test.hpp b/codegen/test/include/test.hpp index b799df372d..eb9d29187a 100644 --- a/codegen/test/include/test.hpp +++ b/codegen/test/include/test.hpp @@ -843,9 +843,8 @@ inline void run(int argc, const char* argv[]) TEST_CASE_REGISTER(__VA_ARGS__) \ void __VA_ARGS__() -#ifdef __clang__ +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wglobal-constructors" #endif diff --git a/example/15_grouped_gemm/grouped_gemm_multiple_d_xdl_fp16.cpp b/example/15_grouped_gemm/grouped_gemm_multiple_d_xdl_fp16.cpp index ae097865f6..b24e9c089d 100644 --- a/example/15_grouped_gemm/grouped_gemm_multiple_d_xdl_fp16.cpp +++ b/example/15_grouped_gemm/grouped_gemm_multiple_d_xdl_fp16.cpp @@ -23,10 +23,10 @@ #include "ck/library/utility/literals.hpp" #include "ck/library/reference_tensor_operation/cpu/reference_gemm_multiple_d.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif using ::ck::DeviceMem; using ::ck::hip_check_error; using ::ck::HostTensorDescriptor; @@ -79,4 +79,6 @@ using DeviceGemmInstance = int main(int argc, char* argv[]) { return !run_grouped_gemm_example(argc, argv); } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/28_grouped_gemm_bias_e_permute/grouped_gemm_bias_e_permute_xdl_fp16.cpp b/example/28_grouped_gemm_bias_e_permute/grouped_gemm_bias_e_permute_xdl_fp16.cpp index 63417dfd95..2e0d9c2299 100644 --- a/example/28_grouped_gemm_bias_e_permute/grouped_gemm_bias_e_permute_xdl_fp16.cpp +++ b/example/28_grouped_gemm_bias_e_permute/grouped_gemm_bias_e_permute_xdl_fp16.cpp @@ -18,10 +18,10 @@ #include "ck/library/utility/host_tensor_generator.hpp" #include "ck/library/utility/numeric.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif using ::ck::DeviceMem; using ::ck::HostTensorDescriptor; using ::ck::make_ParallelTensorFunctor; @@ -482,4 +482,6 @@ int main(int argc, char* argv[]) return pass ? 0 : 1; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/29_batched_gemm_bias_e_permute/run_batched_gemm_bias_e_permute_example.inc b/example/29_batched_gemm_bias_e_permute/run_batched_gemm_bias_e_permute_example.inc index 081e27c843..f6134fab5a 100644 --- a/example/29_batched_gemm_bias_e_permute/run_batched_gemm_bias_e_permute_example.inc +++ b/example/29_batched_gemm_bias_e_permute/run_batched_gemm_bias_e_permute_example.inc @@ -1,7 +1,7 @@ +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // hardcoded for NumDimM == NumDimN == NumDimK == 2 template = 23 #pragma clang diagnostic pop +#endif diff --git a/example/36_sparse_embedding/sparse_embedding3_forward_layernorm.cpp b/example/36_sparse_embedding/sparse_embedding3_forward_layernorm.cpp index 8c6121c402..3f1dafe3a7 100644 --- a/example/36_sparse_embedding/sparse_embedding3_forward_layernorm.cpp +++ b/example/36_sparse_embedding/sparse_embedding3_forward_layernorm.cpp @@ -19,10 +19,10 @@ #include "ck/library/utility/host_tensor_generator.hpp" #include "ck/library/reference_tensor_operation/cpu/reference_sparse_embedding3_forward_layernorm.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif using ::ck::DeviceMem; using ::ck::HostTensorDescriptor; using ::ck::Tensor; @@ -222,4 +222,6 @@ int main(int argc, char* argv[]) return 0; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/39_permute/common.hpp b/example/39_permute/common.hpp index 4478118f34..c6c1087ebe 100644 --- a/example/39_permute/common.hpp +++ b/example/39_permute/common.hpp @@ -26,10 +26,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif using ::ck::DeviceMem; using ::ck::HostTensorDescriptor; using ::ck::Tensor; @@ -463,4 +463,6 @@ auto host_permute(const Tensor& src, const Axes& axes, Functor functor, Ten return true; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp16.cpp b/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp16.cpp index d14e4e5c32..b537c59835 100644 --- a/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp16.cpp +++ b/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp16.cpp @@ -16,10 +16,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif using ::ck::DeviceMem; using ::ck::HostTensorDescriptor; using ::ck::make_ParallelTensorFunctor; @@ -423,4 +423,6 @@ int main(int argc, char* argv[]) return 0; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp32.cpp b/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp32.cpp index 7181ffebc0..4cad866132 100644 --- a/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp32.cpp +++ b/example/43_splitk_gemm_bias_e_permute/splitk_gemm_bias_e_permute_xdl_fp32.cpp @@ -16,10 +16,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif using ::ck::DeviceMem; using ::ck::HostTensorDescriptor; using ::ck::make_ParallelTensorFunctor; @@ -423,4 +423,6 @@ int main(int argc, char* argv[]) return 0; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/ck_tile/01_fmha/codegen/ops/fmha_fwd_splitkv.py b/example/ck_tile/01_fmha/codegen/ops/fmha_fwd_splitkv.py index 875a655976..849f463afa 100644 --- a/example/ck_tile/01_fmha/codegen/ops/fmha_fwd_splitkv.py +++ b/example/ck_tile/01_fmha/codegen/ops/fmha_fwd_splitkv.py @@ -121,10 +121,10 @@ using trait_{F_idx} = fmha_fwd_splitkv_traits_<{F_hdim}, {F_dtype}, {F_mode}, {F {F_pipeline_enum}, {F_logits}, fmha_mask_{F_idx}, {F_bias}, {F_lse}, {F_squant}, {F_pagedkv}, {F_sink}, {F_spad}, {F_skpad}, {F_dpad}, {F_dvpad}>; +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wtautological-compare" - +#endif namespace {{ template void run_instance(const ck_tile::stream_config& s, fmha_fwd_splitkv_args a) {{ @@ -142,7 +142,9 @@ void run_instance(const ck_tile::stream_config& s, fmha_fwd_splitkv_args a) {{ }} }} // anonymous namespace +#ifdef __clang__ #pragma clang diagnostic pop +#endif template<> void fmha_fwd_splitkv_oneshot_(const ck_tile::stream_config& s, fmha_fwd_splitkv_args a) 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 1565fe8566..a4e48ff124 100644 --- a/example/ck_tile/01_fmha/fmha_fwd_head_grouping.hpp +++ b/example/ck_tile/01_fmha/fmha_fwd_head_grouping.hpp @@ -22,9 +22,10 @@ #define CK_TILE_FMHA_ENABLE_HEAD_GROUPING 1 #endif +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif #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) @@ -430,4 +431,6 @@ float run_fwd_head_grouped(const ck_tile::stream_config& sc, } // namespace fmha_fwd_head_grouping #endif // CK_TILE_FMHA_ENABLE_HEAD_GROUPING +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/ck_tile/01_fmha/fmha_fwd_runner.hpp b/example/ck_tile/01_fmha/fmha_fwd_runner.hpp index 2b0033524b..0b51dffa46 100644 --- a/example/ck_tile/01_fmha/fmha_fwd_runner.hpp +++ b/example/ck_tile/01_fmha/fmha_fwd_runner.hpp @@ -27,9 +27,10 @@ #error "we should enable fmha_fwd_splitkv() api in order to cooperate with fmha_fwd_appendkv()" #endif +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif enum class fwd_result { success, @@ -2492,4 +2493,6 @@ fwd_result fmha_fwd_run(mode_enum mode, return pass ? fwd_result::success : fwd_result::failure; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/ck_tile/01_fmha/quant.hpp b/example/ck_tile/01_fmha/quant.hpp index 885d014695..4f854fde1d 100644 --- a/example/ck_tile/01_fmha/quant.hpp +++ b/example/ck_tile/01_fmha/quant.hpp @@ -8,10 +8,10 @@ #include "ck_tile/core.hpp" #include "ck_tile/ops/fmha.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // keep sync with BlockAttentionQuantScaleEnum enum class quant_scale_enum { @@ -76,4 +76,6 @@ struct quant_scale_info return os; } }; +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/example/ck_tile/03_gemm/gemm_splitk_two_stage_reduce.cpp b/example/ck_tile/03_gemm/gemm_splitk_two_stage_reduce.cpp index 1b339c39bb..6faca63c17 100644 --- a/example/ck_tile/03_gemm/gemm_splitk_two_stage_reduce.cpp +++ b/example/ck_tile/03_gemm/gemm_splitk_two_stage_reduce.cpp @@ -16,10 +16,10 @@ #include "gemm_utils.hpp" #include "run_gemm_example.inc" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif /** * @brief Tile partitioner with output offset support. * @@ -979,4 +979,6 @@ int main(int argc, char* argv[]) return EXIT_SUCCESS; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 bb403e6062..fcd7004f35 100644 --- a/example/ck_tile/20_grouped_convolution/grouped_convolution_utils.hpp +++ b/example/ck_tile/20_grouped_convolution/grouped_convolution_utils.hpp @@ -15,10 +15,10 @@ #include "ck_tile/ops/elementwise/unary_element_wise_operation.hpp" #include "conv_configs.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" - +#endif struct GemmWarpConfig_Mfma { static constexpr ck_tile::index_t M_Warp_Tile = 32; @@ -163,4 +163,6 @@ struct InvokerResult float ave_time; ck_tile::index_t split_k; }; +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/experimental/builder/test/unit_device_buffer.cpp b/experimental/builder/test/unit_device_buffer.cpp index 1691088e1d..1f512720b3 100644 --- a/experimental/builder/test/unit_device_buffer.cpp +++ b/experimental/builder/test/unit_device_buffer.cpp @@ -50,9 +50,10 @@ TEST(DeviceBuffer, AutoFree) // In this test we are explicitly testing a pointer that is out of scope, so // we have to disable the clang compiler's lifestime safety checks. +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-permissive" +#endif { auto buffer = ckt::alloc_buffer(size); ptr = buffer.get(); @@ -60,7 +61,9 @@ TEST(DeviceBuffer, AutoFree) // Trying to use a pointer after freeing should return en error in HIP. EXPECT_THAT(hipMemset(ptr, 0xFF, size), HipError(hipErrorInvalidValue)); +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif // Reset internal HIP error state. // Otherwise, the error may leak into other tests, triggering anything that diff --git a/include/ck/library/utility/gpu_verification.hpp b/include/ck/library/utility/gpu_verification.hpp index b69fb33691..cd16354541 100644 --- a/include/ck/library/utility/gpu_verification.hpp +++ b/include/ck/library/utility/gpu_verification.hpp @@ -18,9 +18,10 @@ #include "ck/host_utility/hip_check_error.hpp" #include "ck/library/utility/check_err.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace profiler { @@ -520,4 +521,6 @@ float gpu_reduce_max(Iterator device_buffer, std::size_t size, hipStream_t strea } // namespace profiler } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/library/utility/host_tensor.hpp b/include/ck/library/utility/host_tensor.hpp index c3865820d1..c6028e5351 100644 --- a/include/ck/library/utility/host_tensor.hpp +++ b/include/ck/library/utility/host_tensor.hpp @@ -23,11 +23,11 @@ #include "ck/tensor_operation/gpu/device/tensor_layout.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" #pragma clang diagnostic ignored "-Wlifetime-safety-cross-tu-suggestions" - +#endif namespace ck { template @@ -1202,4 +1202,6 @@ struct Tensor }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor/static_tensor.hpp b/include/ck/tensor/static_tensor.hpp index c29c1a4a66..5543ab6271 100644 --- a/include/ck/tensor/static_tensor.hpp +++ b/include/ck/tensor/static_tensor.hpp @@ -4,9 +4,10 @@ #ifndef CK_STATIC_TENSOR_HPP #define CK_STATIC_TENSOR_HPP +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { // StaticTensor for Scalar @@ -273,5 +274,7 @@ __host__ __device__ constexpr auto make_static_tensor(TensorDesc, X invalid_elem } } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop #endif +#endif diff --git a/include/ck/tensor_description/multi_index_transform.hpp b/include/ck/tensor_description/multi_index_transform.hpp index ef1d6177fa..ab806b1bb2 100644 --- a/include/ck/tensor_description/multi_index_transform.hpp +++ b/include/ck/tensor_description/multi_index_transform.hpp @@ -6,10 +6,10 @@ #include "ck/utility/common_header.hpp" #include "ck/utility/multi_index.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -2281,4 +2281,6 @@ struct Xor } }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_description/tensor_descriptor.hpp b/include/ck/tensor_description/tensor_descriptor.hpp index 1958c2b981..3a5d258fae 100644 --- a/include/ck/tensor_description/tensor_descriptor.hpp +++ b/include/ck/tensor_description/tensor_descriptor.hpp @@ -7,9 +7,10 @@ #include "ck/utility/sequence_helper.hpp" #include "ck/tensor_description/multi_index_transform.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template @@ -642,4 +643,6 @@ using TensorCoordinateStep_t = decltype(make_tensor_coordinate_step( TensorDesc{}, MultiIndex::GetNumOfDimension()>{})); } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_description/tensor_space_filling_curve.hpp b/include/ck/tensor_description/tensor_space_filling_curve.hpp index 353ce65c61..673485cba1 100644 --- a/include/ck/tensor_description/tensor_space_filling_curve.hpp +++ b/include/ck/tensor_description/tensor_space_filling_curve.hpp @@ -10,10 +10,10 @@ #include "ck/utility/tuple_helper.hpp" #include "ck/tensor_description/tensor_adaptor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace detail { @@ -197,4 +197,6 @@ struct SpaceFillingCurve } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_wmmaops_base.hpp b/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_wmmaops_base.hpp index 6d684696b4..9b30fd5580 100644 --- a/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_wmmaops_base.hpp +++ b/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_wmmaops_base.hpp @@ -10,9 +10,10 @@ #include "ck/tensor_operation/gpu/warp/wmma_gemm.hpp" #include "ck/tensor_description/tensor_adaptor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops.hpp b/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops.hpp index 4611aa6d6d..c2e9e2edac 100644 --- a/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops.hpp +++ b/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops.hpp @@ -13,9 +13,10 @@ // Prefetech 2 stage // Local prefetch 1 stage +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops_base.hpp b/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops_base.hpp index 40a2aa28b6..ba710781b5 100644 --- a/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops_base.hpp +++ b/include/ck/tensor_operation/gpu/block/blockwise_gemm_pipeline_xdlops_base.hpp @@ -9,10 +9,10 @@ #include "ck/tensor_operation/gpu/warp/xdlops_gemm.hpp" #include "ck/tensor_description/tensor_adaptor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/block/blockwise_gemm_wmma.hpp b/include/ck/tensor_operation/gpu/block/blockwise_gemm_wmma.hpp index bb3377930d..827ea29c07 100644 --- a/include/ck/tensor_operation/gpu/block/blockwise_gemm_wmma.hpp +++ b/include/ck/tensor_operation/gpu/block/blockwise_gemm_wmma.hpp @@ -11,10 +11,10 @@ #define CK_MNK_LOOP +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { #ifdef __gfx12__ @@ -1014,4 +1014,6 @@ struct BlockwiseGemmWMMA #endif } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops.hpp b/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops.hpp index 25a29a277b..4f73be7773 100644 --- a/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops.hpp +++ b/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops.hpp @@ -9,9 +9,10 @@ #include "ck/tensor_operation/gpu/warp/xdlops_gemm.hpp" #include "ck/tensor_description/tensor_adaptor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template @@ -1033,4 +1034,6 @@ struct BlockwiseGemmXdlops_v2 }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops_skip_b_lds.hpp b/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops_skip_b_lds.hpp index 4fe0b77f27..e80562ed8e 100644 --- a/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops_skip_b_lds.hpp +++ b/include/ck/tensor_operation/gpu/block/blockwise_gemm_xdlops_skip_b_lds.hpp @@ -8,10 +8,10 @@ #include "ck/tensor_operation/gpu/warp/xdlops_gemm.hpp" #include "ck/tensor_description/tensor_adaptor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/device_grouped_gemm.hpp b/include/ck/tensor_operation/gpu/device/device_grouped_gemm.hpp index 426b008257..7b6dbd6f94 100644 --- a/include/ck/tensor_operation/gpu/device/device_grouped_gemm.hpp +++ b/include/ck/tensor_operation/gpu/device/device_grouped_gemm.hpp @@ -12,10 +12,10 @@ #include "device_base.hpp" #include "ck/utility/ignore.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -183,4 +183,6 @@ struct DeviceGroupedGemm : public BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 2f4926d452..14fc62f4a9 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 @@ -20,9 +20,10 @@ #include "ck/host_utility/kernel_launch.hpp" #include "ck/utility/scheduler_enum.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_xdl_cshuffle.hpp index 419b8b3a1f..af6d12e5ad 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_contraction_multiple_d_xdl_cshuffle.hpp @@ -18,9 +18,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_wmma_cshuffle_v3_common.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_wmma_cshuffle_v3_common.hpp index 0346f90b72..9055f9a02a 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_wmma_cshuffle_v3_common.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_wmma_cshuffle_v3_common.hpp @@ -18,10 +18,10 @@ #include "ck/utility/scheduler_enum.hpp" #include "ck/utility/integral_constant.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -705,16 +705,15 @@ struct DeviceGemmGemm_Wmma_CShuffleV3_Common_Invoker_Arg auto print = [&curFunc](const char* format, ...) -> void { if(ck::EnvIsEnabled(CK_ENV(CK_LOGGING))) { -#if defined(__clang__) +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wformat-nonliteral" #endif va_list args; va_start(args, format); std::vfprintf(stdout, format, args); va_end(args); -#if defined(__clang__) +#ifdef __clang__ #pragma clang diagnostic pop #endif std::cout << "In file: " << __FILE__ << ", function: " << curFunc << "\n"; @@ -905,4 +904,6 @@ struct DeviceGemmGemm_Wmma_CShuffleV3_Common_Invoker_Arg } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_xdl_cshuffle.hpp index 97af5fd88d..5392055556 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_gemm_xdl_cshuffle.hpp @@ -18,10 +18,10 @@ #include "ck/host_utility/kernel_launch.hpp" #include "ck/host_utility/io.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -809,4 +809,6 @@ struct DeviceBatchedGemmGemm_Xdl_CShuffle : public DeviceBatchedGemmGemm= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_multi_d_xdl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_multi_d_xdl.hpp index 7e414eeed7..c6d18d57d5 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_multi_d_xdl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_multi_d_xdl.hpp @@ -18,9 +18,10 @@ #include "ck/host_utility/kernel_launch.hpp" #include "ck/host_utility/io.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -794,4 +795,6 @@ struct DeviceBatchedGemmMultiD_Xdl : public DeviceBatchedGemmMultiD= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_wmma_cshuffle_v3.hpp index 26891b4367..febc1b079e 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_wmma_cshuffle_v3.hpp @@ -19,10 +19,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_xdl_cshuffle.hpp index f197ee3157..3748c6fa7d 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_reduce_xdl_cshuffle.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1048,4 +1048,6 @@ struct DeviceBatchedGemmReduce_Xdl_CShuffle : public DeviceGemmReduce<0, ReduceO } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_wmma_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_wmma_cshuffle.hpp index adcfe9c5b6..069d77dd1b 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_wmma_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_wmma_cshuffle.hpp @@ -22,10 +22,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1599,4 +1599,6 @@ struct DeviceBatchedGemmSoftmaxGemmPermute_Wmma_CShuffle } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_xdl_cshuffle.hpp index 187ed9ad72..c840c1217b 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_permute_xdl_cshuffle.hpp @@ -19,10 +19,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1010,4 +1010,6 @@ struct DeviceBatchedGemmSoftmaxGemmPermute_Xdl_CShuffle } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_xdl_cshuffle.hpp index 63d1125fc9..9b84f665fc 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_softmax_gemm_xdl_cshuffle.hpp @@ -22,10 +22,10 @@ #include "ck/tensor_operation/gpu/device/matrix_padder.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_batched_gemm_softmax_gemm_xdl_cshuffle_v1.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1209,4 +1209,6 @@ struct DeviceBatchedGemmSoftmaxGemm_Xdl_CShuffle } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl.hpp index 14f89a49bb..f87c41bff8 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batched_gemm_xdl.hpp @@ -16,10 +16,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -493,4 +493,6 @@ struct DeviceBatchedGemmXdl : public DeviceBatchedGemm= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_backward_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_backward_impl.hpp index 202ed1088d..634cfec213 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_backward_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_backward_impl.hpp @@ -18,10 +18,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -879,4 +879,6 @@ struct DeviceBatchNormBwdImpl : public DeviceBatchNormBwd= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl.hpp index a4995ccc62..8b7afbce00 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl.hpp @@ -19,10 +19,10 @@ #include "ck/host_utility/kernel_launch.hpp" #include "ck/host_utility/hip_check_error.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -827,4 +827,6 @@ struct DeviceBatchNormFwdImpl : public DeviceBatchNormFwd= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl_obsolete.hpp b/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl_obsolete.hpp index 01f2b60cc8..2c2f494467 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl_obsolete.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_batchnorm_forward_impl_obsolete.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -719,4 +719,6 @@ struct DeviceBatchNormFwdImpl : public DeviceBatchNormFwd= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_cgemm_4gemm_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_cgemm_4gemm_xdl_cshuffle.hpp index e43babb783..46d940f17a 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_cgemm_4gemm_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_cgemm_4gemm_xdl_cshuffle.hpp @@ -19,10 +19,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -679,4 +679,6 @@ struct DeviceCGemm_4Gemm_Xdl_CShuffle } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_add_nhwc_kyxc_nhwk.hpp b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_add_nhwc_kyxc_nhwk.hpp index 3aae690e61..43fa9fb8ad 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_add_nhwc_kyxc_nhwk.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_add_nhwc_kyxc_nhwk.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1014,4 +1014,6 @@ struct } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_nhwc_kyxc_nhwk.hpp b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_nhwc_kyxc_nhwk.hpp index b5b0fa7674..d76e1505fd 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_nhwc_kyxc_nhwk.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_bias_activation_nhwc_kyxc_nhwk.hpp @@ -18,10 +18,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -974,4 +974,6 @@ struct DeviceConv2dFwdXdl_C_Shuffle_Bias_Activation_Input_N_Hi_Wi_C_Weight_K_Y_X } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk.hpp b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk.hpp index e8995de746..dc1fe4da0e 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -941,4 +941,6 @@ struct DeviceConv2dFwdXdl_C_Shuffle_Input_N_Hi_Wi_C_Weight_K_Y_X_C_Output_N_Ho_W } // namespace tensor_operation } // namespace ck // +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_nhwc_kyxc_nhwk.hpp b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_nhwc_kyxc_nhwk.hpp index 894dd7c3ac..dc84f2aaca 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_nhwc_kyxc_nhwk.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_conv2d_fwd_xdl_nhwc_kyxc_nhwk.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -728,4 +728,6 @@ struct DeviceConv2dFwdXdl_Input_N_Hi_Wi_C_Weight_K_Y_X_C_Output_N_Ho_Wo_K } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_conv3d_fwd_naive_ndhwc_kzyxc_ndhwk.hpp b/include/ck/tensor_operation/gpu/device/impl/device_conv3d_fwd_naive_ndhwc_kzyxc_ndhwk.hpp index 19fb1afbe8..d44458620a 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_conv3d_fwd_naive_ndhwc_kzyxc_ndhwk.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_conv3d_fwd_naive_ndhwc_kzyxc_ndhwk.hpp @@ -14,10 +14,10 @@ #include "ck/library/utility/convolution_parameter.hpp" #include "ck/library/reference_tensor_operation/gpu/naive_conv_fwd_gpu.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -266,4 +266,6 @@ struct DeviceConv3dFwdNaive_Input_N_Di_Hi_Wi_C_Weight_K_Z_Y_X_C_Output_N_Do_Ho_W } // namespace ck #endif +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_dl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_dl.hpp index 3980c055cf..cbd97bfb33 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_dl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_dl.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1589,4 +1589,6 @@ struct DeviceConvNdBwdDataNwcKxcNwk_Dl } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_xdl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_xdl.hpp index f037f00b28..a21d6c633c 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_xdl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_convnd_bwd_data_nwc_kxc_nwk_xdl.hpp @@ -18,10 +18,10 @@ #include "ck/host_utility/kernel_launch.hpp" #include "ck/library/utility/numeric.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1589,5 +1589,7 @@ struct DeviceConvNdBwdDataNwcKxcNwk_Xdl } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_wmma_cshuffle_v3.hpp index c84fa51a40..6f3715378c 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_wmma_cshuffle_v3.hpp @@ -18,10 +18,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_xdl_cshuffle.hpp index ac70ef0dbc..6caa070543 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_bias_add_reduce_xdl_cshuffle.hpp @@ -16,10 +16,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -919,4 +919,6 @@ struct DeviceGemmBiasAddReduce_Xdl_CShuffle : public DeviceGemmReduce<1, ReduceO } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_layernorm_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_layernorm_wmma_cshuffle_v3.hpp index c1e9048d1b..4e0eaec3b2 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_layernorm_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_layernorm_wmma_cshuffle_v3.hpp @@ -19,9 +19,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_multiple_r_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_multiple_r_xdl_cshuffle.hpp index bf22814a07..d077dc5438 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_multiple_r_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_multiple_r_xdl_cshuffle.hpp @@ -17,9 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_wmma_cshuffle_v3.hpp index 53e8964bfb..18a46b3bb8 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_wmma_cshuffle_v3.hpp @@ -18,9 +18,10 @@ #include "ck/host_utility/flush_cache.hpp" #include "ck/tensor_operation/gpu/device/impl/device_gemm_wmma_cshuffle_v3_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -411,4 +412,6 @@ struct DeviceGemmMultipleD_Wmma_CShuffleV3 } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle.hpp index 42ed344702..b21aa08f69 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_multiple_d_xdl_cshuffle.hpp @@ -19,9 +19,10 @@ #include "ck/tensor_operation/gpu/device/matrix_padder.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_multiple_d_xdl_cshuffle.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_wmma_cshuffle_v3.hpp index e223cb5fbe..618fd8abb7 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_wmma_cshuffle_v3.hpp @@ -18,10 +18,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_xdl_cshuffle.hpp index 4cf8456117..f4e97a4bd9 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_reduce_xdl_cshuffle.hpp @@ -16,10 +16,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -883,4 +883,6 @@ struct DeviceGemmReduce_Xdl_CShuffle : public DeviceGemmReduce<0, ReduceOperatio } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma_cshuffle_v3r1.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma_cshuffle_v3r1.hpp index 2e8d32ddb4..3112dea36d 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma_cshuffle_v3r1.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_wmma_cshuffle_v3r1.hpp @@ -26,10 +26,10 @@ #include "ck/tensor_operation/gpu/device/reduction_operator_mapping.hpp" #include "ck/tensor_operation/gpu/device/impl/device_reduce_threadwise_multi_d.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -584,4 +584,6 @@ struct DeviceGemm_Wmma_CShuffleV3R1 : public DeviceGemmV2R1= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_layernorm_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_layernorm_cshuffle.hpp index 1c0d00b8f7..02ad9079ea 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_layernorm_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_layernorm_cshuffle.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -816,4 +816,6 @@ struct DeviceGemmLayerNorm_Xdl_CShuffle : public BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_skip_b_lds.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_skip_b_lds.hpp index 92f6d4370f..fb5fea27f1 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_skip_b_lds.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_skip_b_lds.hpp @@ -18,10 +18,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -560,4 +560,6 @@ struct DeviceGemmXdlSkipBLds : public DeviceGemm= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_waveletmodel_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_waveletmodel_cshuffle.hpp index 9c1ada3b68..075f3db489 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_waveletmodel_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_gemm_xdl_waveletmodel_cshuffle.hpp @@ -17,10 +17,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_bwd_weight_dl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_bwd_weight_dl.hpp index 35bc2d1b11..136d59a160 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_bwd_weight_dl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_bwd_weight_dl.hpp @@ -27,10 +27,10 @@ #endif #include "ck/tensor_operation/gpu/device/tensor_size_check.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1396,4 +1396,6 @@ struct DeviceGroupedConvBwdWeight_Dl : public DeviceGroupedConvBwdWeight= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_wmma_cshuffle_v3.hpp index 862ddb50d5..9b0358e5d2 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_wmma_cshuffle_v3.hpp @@ -36,9 +36,10 @@ #endif #include "ck/tensor_operation/gpu/device/tensor_size_check.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -2410,4 +2411,6 @@ struct DeviceGroupedConvFwdMultipleABD_Wmma_CShuffle_V3 #ifdef CK_EXPERIMENTAL_BUILDER #include "ck_tile/builder/reflect/reflect_device_grouped_conv_fwd_multiple_abd_wmma_cshuffle_v3.inc" #endif +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle.hpp index 4d2351a348..21eebd4481 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle.hpp @@ -34,9 +34,10 @@ #endif #include "ck/tensor_operation/gpu/device/tensor_size_check.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -2264,4 +2265,6 @@ struct DeviceGroupedConvFwdMultipleABD_Xdl_CShuffle } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle_v3.hpp index ee578971e2..10785a17d0 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_abd_xdl_cshuffle_v3.hpp @@ -33,9 +33,10 @@ #endif #include "ck/tensor_operation/gpu/device/tensor_size_check.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -2424,4 +2425,6 @@ struct DeviceGroupedConvFwdMultipleABD_Xdl_CShuffle_V3 } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_multiple_r_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_multiple_r_xdl_cshuffle.hpp index 572314c023..371e59988d 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_multiple_r_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_multiple_r_xdl_cshuffle.hpp @@ -24,9 +24,10 @@ #include "ck/host_utility/io.hpp" #include "ck/library/utility/numeric.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1160,4 +1161,6 @@ struct DeviceGroupedConvFwdMultipleDMultipleR_Xdl_CShuffle } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_wmma_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_wmma_cshuffle.hpp index 68845f7fa9..737c019edd 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_wmma_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_conv_fwd_multiple_d_wmma_cshuffle.hpp @@ -30,9 +30,10 @@ #endif #include "ck/tensor_operation/gpu/device/tensor_size_check.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1059,4 +1060,6 @@ struct DeviceGroupedConvFwdMultipleD_Wmma_CShuffle } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_grouped_query_attention_forward_wmma.hpp b/include/ck/tensor_operation/gpu/device/impl/device_grouped_query_attention_forward_wmma.hpp index 72b1fb6b5f..d878cfc1b5 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_grouped_query_attention_forward_wmma.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_grouped_query_attention_forward_wmma.hpp @@ -22,10 +22,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1124,4 +1124,6 @@ struct DeviceGroupedQueryAttentionForward_Wmma } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_image_to_column_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_image_to_column_impl.hpp index 7f7f669602..1fac0160c5 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_image_to_column_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_image_to_column_impl.hpp @@ -19,10 +19,10 @@ #include "ck/tensor_operation/gpu/device/impl/device_grouped_conv_utils.hpp" #include "ck/host_utility/io.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -411,4 +411,6 @@ struct DeviceImageToColumnImpl } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_multi_query_attention_forward_wmma.hpp b/include/ck/tensor_operation/gpu/device/impl/device_multi_query_attention_forward_wmma.hpp index 0a8922de8a..91dfd504b2 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_multi_query_attention_forward_wmma.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_multi_query_attention_forward_wmma.hpp @@ -22,10 +22,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -1114,4 +1114,6 @@ struct DeviceMultiQueryAttentionForward_Wmma } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_multiblock.hpp b/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_multiblock.hpp index 0e31488487..671894546b 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_multiblock.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_multiblock.hpp @@ -17,9 +17,10 @@ #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -596,4 +597,6 @@ struct DeviceMultipleReduceMultiBlock : public DeviceMultipleReduce= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_threadwise.hpp b/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_threadwise.hpp index 389754f533..62198843e6 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_threadwise.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_multiple_reduce_threadwise.hpp @@ -16,9 +16,10 @@ #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -423,4 +424,6 @@ struct DeviceMultipleReduceThreadWise : public DeviceMultipleReduce= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_data_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_data_impl.hpp index 887322379c..8ae6a305c6 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_data_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_data_impl.hpp @@ -14,10 +14,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // M is Invariant dimension, K is reduced dimension namespace ck { namespace tensor_operation { @@ -468,4 +468,6 @@ struct DeviceNormalizationBwdDataImpl : public DeviceNormalizationBwdData= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_gamma_beta_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_gamma_beta_impl.hpp index 153685af13..b2d50c4c41 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_gamma_beta_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_normalization_bwd_gamma_beta_impl.hpp @@ -14,10 +14,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // M is Invariant dimension, K is reduced dimension namespace ck { namespace tensor_operation { @@ -481,4 +481,6 @@ struct DeviceNormalizationBwdGammaBetaImpl } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_normalization_fwd_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_normalization_fwd_impl.hpp index 429e9cff79..cbf670b0eb 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_normalization_fwd_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_normalization_fwd_impl.hpp @@ -14,10 +14,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -478,4 +478,6 @@ struct DeviceNormalizationFwdImpl : public DeviceNormalizationFwd= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_permute_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_permute_impl.hpp index ea52a442ae..12ca511a5e 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_permute_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_permute_impl.hpp @@ -17,9 +17,10 @@ #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { namespace tensor_operation { namespace device { @@ -283,4 +284,6 @@ struct DevicePermuteImpl : DevicePermute= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_pool2d_fwd_nhwc_nhwc.hpp b/include/ck/tensor_operation/gpu/device/impl/device_pool2d_fwd_nhwc_nhwc.hpp index 3c0d110540..9ee5e9b07c 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_pool2d_fwd_nhwc_nhwc.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_pool2d_fwd_nhwc_nhwc.hpp @@ -16,10 +16,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -391,4 +391,6 @@ struct DevicePool2dFwd_NHWC_NHWC : public DevicePoolFwd<4, } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_pool3d_fwd_ndhwc_ndhwc.hpp b/include/ck/tensor_operation/gpu/device/impl/device_pool3d_fwd_ndhwc_ndhwc.hpp index d53ab6cadb..30c6b51103 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_pool3d_fwd_ndhwc_ndhwc.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_pool3d_fwd_ndhwc_ndhwc.hpp @@ -16,10 +16,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -414,4 +414,6 @@ struct DevicePool3dFwd_NDHWC_NDHWC : public DevicePoolFwd<5, } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_put_element_impl.hpp b/include/ck/tensor_operation/gpu/device/impl/device_put_element_impl.hpp index cef984d581..10bdd780fb 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_put_element_impl.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_put_element_impl.hpp @@ -14,10 +14,10 @@ #include "ck/host_utility/kernel_launch.hpp" #include "ck/host_utility/stream_utility.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -158,4 +158,6 @@ struct DevicePutElementImpl } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise.hpp b/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise.hpp index d711e4705b..13c16a4bcb 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise.hpp @@ -13,10 +13,10 @@ #include "ck/tensor_operation/gpu/device/impl/device_reduce_common.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_2d_reduction_threadwise.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -395,4 +395,6 @@ struct DeviceReduceThreadWise : public DeviceReduce= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise_multi_d.hpp b/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise_multi_d.hpp index d6c7afd8b2..dc9db637aa 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise_multi_d.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_reduce_threadwise_multi_d.hpp @@ -14,10 +14,10 @@ #include "ck/tensor_operation/gpu/grid/gridwise_2d_reduction_threadwise_multi_d.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -415,4 +415,6 @@ struct DeviceReduceThreadWiseMultiD : public DeviceReduceMultiD= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_sparse_embeddings_forward_layernorm.hpp b/include/ck/tensor_operation/gpu/device/impl/device_sparse_embeddings_forward_layernorm.hpp index 10a4d74ddc..1b5e401319 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_sparse_embeddings_forward_layernorm.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_sparse_embeddings_forward_layernorm.hpp @@ -18,10 +18,10 @@ #include "ck/tensor_operation/gpu/grid/gridwise_sparse_embeddings_forward_layernorm.hpp" #endif +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -200,4 +200,6 @@ struct DeviceSparseEmbeddingsForwardLayernorm : public BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/device/impl/device_splitk_contraction_multiple_d_xdl_cshuffle.hpp b/include/ck/tensor_operation/gpu/device/impl/device_splitk_contraction_multiple_d_xdl_cshuffle.hpp index 6213b44db4..29db59fe32 100644 --- a/include/ck/tensor_operation/gpu/device/impl/device_splitk_contraction_multiple_d_xdl_cshuffle.hpp +++ b/include/ck/tensor_operation/gpu/device/impl/device_splitk_contraction_multiple_d_xdl_cshuffle.hpp @@ -18,9 +18,10 @@ #include "ck/host_utility/device_prop.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/epilogue_cshuffle_v3_welford_wmma.hpp b/include/ck/tensor_operation/gpu/grid/epilogue_cshuffle_v3_welford_wmma.hpp index 4d92dbdc0c..aaeaf96338 100644 --- a/include/ck/tensor_operation/gpu/grid/epilogue_cshuffle_v3_welford_wmma.hpp +++ b/include/ck/tensor_operation/gpu/grid/epilogue_cshuffle_v3_welford_wmma.hpp @@ -8,10 +8,10 @@ #include "ck/tensor_operation/gpu/block/blockwise_welford.hpp" #include "ck/tensor_operation/gpu/thread/threadwise_welford.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_batched_gemm_gemm_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_batched_gemm_gemm_wmma_cshuffle_v3.hpp index 36ea80f472..c276be5dbe 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_batched_gemm_gemm_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_batched_gemm_gemm_wmma_cshuffle_v3.hpp @@ -510,16 +510,17 @@ struct GridwiseBatchedGemmGemm_wmma_cshuffle_v3 auto print = [&curFunc](const char* format, ...) -> void { if(ck::EnvIsEnabled(CK_ENV(CK_LOGGING))) { -#if defined(__clang__) + +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wformat-nonliteral" #endif va_list args; va_start(args, format); std::vfprintf(stdout, format, args); va_end(args); -#if defined(__clang__) + +#ifdef __clang__ #pragma clang diagnostic pop #endif std::cout << "In file: " << __FILE__ << ", function: " << curFunc << "\n"; diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_dpp.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_dpp.hpp index 787a957219..52bd7e33fb 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_dpp.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_dpp.hpp @@ -16,10 +16,10 @@ #include "ck/tensor_operation/gpu/thread/threadwise_tensor_slice_transfer.hpp" #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -697,4 +697,6 @@ struct GridwiseGemm_ak0mak1_bk0nbk1_mn_dpp } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3.hpp index 844b1cc7a2..897891ea4c 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3.hpp @@ -20,10 +20,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { /// @brief \"Universal\" GEMM kernel with SplitK support. @@ -1189,4 +1189,6 @@ struct GridwiseGemm_wmma_cshuffle_v3 }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3_ab_scale.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3_ab_scale.hpp index 2c7a8e2eaf..1d0eb4f2d0 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3_ab_scale.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3_ab_scale.hpp @@ -16,10 +16,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_wmma_cshuffle_v3_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_conv_v3.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_conv_v3.hpp index 1673be0dcd..c9434ebfa2 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_conv_v3.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_conv_v3.hpp @@ -15,10 +15,10 @@ #include "ck/tensor_operation/gpu/block/thread_group_tensor_slice_transfer_direct_load.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_streamk_v3.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_streamk_v3.hpp index f2abdefff0..ad3c31e467 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_streamk_v3.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_streamk_v3.hpp @@ -17,10 +17,10 @@ #include "ck/utility/reduction_functions_accumulate.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -2164,4 +2164,6 @@ struct GridwiseGemm_xdl_cshuffle_streamk_v3 } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v1.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v1.hpp index 2ff3f04821..ebace17416 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v1.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v1.hpp @@ -17,10 +17,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -1034,4 +1034,6 @@ struct GridwiseGemm_k0mk1_k0nk1_mn_xdl_cshuffle_v1 } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v2.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v2.hpp index e49fb5be40..adaf3209a3 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v2.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v2.hpp @@ -15,10 +15,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -1028,4 +1028,6 @@ struct GridwiseGemm_xdl_cshuffle_v2 } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3.hpp index 3bdcbc336c..0c8707476e 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3.hpp @@ -17,10 +17,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1739,4 +1739,6 @@ struct GridwiseGemm_xdl_cshuffle_v3 } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_preshuffle.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_preshuffle.hpp index 1c39f62e4d..a8eb85cf7b 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_preshuffle.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_preshuffle.hpp @@ -15,10 +15,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1476,4 +1476,6 @@ struct GridwiseGemm_xdl_cshuffle_v3_b_preshuffle }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_scale.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_scale.hpp index 7a6de68579..19e98469d0 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_scale.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_b_scale.hpp @@ -15,10 +15,10 @@ #include "ck/utility/env.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1607,4 +1607,6 @@ struct GridwiseGemm_xdl_cshuffle_v3 } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_abd.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_abd.hpp index 73f54ea93b..753d801a98 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_abd.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_abd.hpp @@ -15,9 +15,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1696,4 +1697,6 @@ struct GridwiseGemm_xdl_cshuffle_v3 }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d.hpp index 38385711ba..99e5828661 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d.hpp @@ -20,10 +20,10 @@ #define DEBUG_LOG 0 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1961,4 +1961,6 @@ struct GridwiseGemmMultiD_xdl_cshuffle_v3 } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_ab_scale.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_ab_scale.hpp index 7d2c6d8f28..27d5c4be61 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_ab_scale.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_ab_scale.hpp @@ -17,10 +17,10 @@ #define DEBUG_LOG 0 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1288,4 +1288,6 @@ struct GridwiseGemmMultiD_ABScale_xdl_cshuffle_v3 }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_b_preshuffle.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_b_preshuffle.hpp index e2a2c0145e..f0b750aef2 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_b_preshuffle.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_b_preshuffle.hpp @@ -17,10 +17,10 @@ #define DEBUG_LOG 0 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1474,4 +1474,6 @@ struct GridwiseGemmMultiD_xdl_cshuffle_v3_b_preshuffle } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_blockscale_b_preshuffle.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_blockscale_b_preshuffle.hpp index 6c4c279fa4..c0abfb107c 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_blockscale_b_preshuffle.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_multi_d_blockscale_b_preshuffle.hpp @@ -17,10 +17,10 @@ #define DEBUG_LOG 0 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1549,4 +1549,6 @@ struct GridwiseGemmMultiD_blockscale_xdl_cshuffle_v3_b_preshuffle }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx.hpp index 25e68bcb10..061aaf8c39 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx.hpp @@ -17,10 +17,10 @@ #include "ck/tensor_operation/gpu/block/thread_group_tensor_slice_transfer_direct_load.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1766,4 +1766,6 @@ struct GridwiseGemmMX_xdl_cshuffle_v3 }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx_bpreshuffle.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx_bpreshuffle.hpp index 7206a5766b..bad66d58a6 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx_bpreshuffle.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_v3_mx_bpreshuffle.hpp @@ -16,10 +16,10 @@ #include "ck/tensor_operation/gpu/block/thread_group_tensor_slice_transfer_direct_load.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -1502,4 +1502,6 @@ struct GridwiseGemmMX_xdl_cshuffle_v3_bpreshuffle }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_bwd_weight.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_bwd_weight.hpp index 0539f5a2af..181572d3f9 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_bwd_weight.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_bwd_weight.hpp @@ -17,10 +17,10 @@ #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" #include "ck/tensor_operation/gpu/device/device_base.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Implementation of "Merge" transformation primitive that uses division and mod. It is supposed to @@ -1147,4 +1147,6 @@ struct GridwiseGemm_bk0mk1_bk0nk1_mn_xdlops_bwd_weight }; // namespace ck } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_splitk_lds_direct_load.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_splitk_lds_direct_load.hpp index 3bdc593ca8..1270220dbe 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_splitk_lds_direct_load.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_splitk_lds_direct_load.hpp @@ -21,10 +21,10 @@ #include "ck/tensor_operation/gpu/device/matrix_padder.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_streamk.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_streamk.hpp index 66188babee..84809dbeee 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_streamk.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_streamk.hpp @@ -20,10 +20,10 @@ #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" #include "ck/host_utility/device_prop.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -1212,4 +1212,6 @@ struct GridwiseGemm_bk0mk1_bk0nk1_mn_xdlops_streamk } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r3.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r3.hpp index 10020c166b..ed84774cae 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r3.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r3.hpp @@ -17,10 +17,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r4r2.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r4r2.hpp index 7b97c0f5b7..f99190ee45 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r4r2.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_gemm_xdlops_v2r4r2.hpp @@ -18,10 +18,10 @@ #include "ck/tensor_operation/gpu/device/matrix_padder.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm.hpp index 7ea7439f22..049388766c 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm.hpp @@ -18,10 +18,10 @@ #define DEBUG_LOG 0 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -2126,4 +2126,6 @@ struct GridwiseMoeGemm : public GridwiseGemm_xdl_cshuffle_base< }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm_blockscale.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm_blockscale.hpp index ce89c94c04..5b791d3668 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm_blockscale.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_moe_gemm_blockscale.hpp @@ -15,10 +15,10 @@ #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" #define DEBUG_LOG 0 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -2224,4 +2224,6 @@ struct GridwiseMoeGemmBlockScale }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm.hpp index 75b4b22465..ab71fb2f77 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm.hpp @@ -19,10 +19,10 @@ #include "ck/tensor_operation/gpu/block/thread_group_tensor_slice_transfer_gather_direct_load.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif #define DEBUG_LOG 0 namespace ck { @@ -2512,4 +2512,6 @@ struct GridwiseMoeGemmMX }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bns.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bns.hpp index 3dd1b6557b..5c7c37272f 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bns.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bns.hpp @@ -21,10 +21,10 @@ #define DEBUG_LOG 0 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Currently we do not have a elegant way to put single lds buffer & double lds buffer pipe in same @@ -2394,4 +2394,6 @@ struct GridwiseMoeGemmMXBNS }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bpreshuffle.hpp b/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bpreshuffle.hpp index 29bac1d4a9..04042530a2 100644 --- a/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bpreshuffle.hpp +++ b/include/ck/tensor_operation/gpu/grid/gridwise_moe_mx_gemm_bpreshuffle.hpp @@ -19,10 +19,10 @@ #include "ck/tensor_operation/gpu/block/thread_group_tensor_slice_transfer_gather_direct_load.hpp" #include "ck/tensor_operation/gpu/grid/gridwise_gemm_xdl_cshuffle_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif #define DEBUG_LOG 0 namespace ck { @@ -2222,4 +2222,6 @@ struct GridwiseMoeGemmMX_BPreshuffle }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/tensor_operation/gpu/thread/threadwise_tensor_slice_transfer.hpp b/include/ck/tensor_operation/gpu/thread/threadwise_tensor_slice_transfer.hpp index 142a90ae3a..765b395cfe 100644 --- a/include/ck/tensor_operation/gpu/thread/threadwise_tensor_slice_transfer.hpp +++ b/include/ck/tensor_operation/gpu/thread/threadwise_tensor_slice_transfer.hpp @@ -12,10 +12,10 @@ #include "ck/tensor_operation/gpu/thread/threadwise_tensor_slice_transfer_util.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // Assume: // 1. src: @@ -2100,4 +2100,6 @@ struct ThreadwiseTensorSliceTransfer_StaticToStatic_IntraRow } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/amd_address_space.hpp b/include/ck/utility/amd_address_space.hpp index 46557b3d41..cfafed05fe 100644 --- a/include/ck/utility/amd_address_space.hpp +++ b/include/ck/utility/amd_address_space.hpp @@ -25,11 +25,14 @@ __device__ T* cast_pointer_to_generic_address_space(T CK_CONSTANT_ADDRESS_SPACE* { // cast a pointer in "Constant" address space (4) to "Generic" address space (0) // only c-style pointer cast seems be able to be compiled +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" +#endif return (T*)p; // NOLINT(old-style-cast) +#ifdef __clang__ #pragma clang diagnostic pop +#endif } template @@ -37,11 +40,14 @@ __host__ __device__ T CK_CONSTANT_ADDRESS_SPACE* cast_pointer_to_constant_addres { // cast a pointer in "Generic" address space (0) to "Constant" address space (4) // only c-style pointer cast seems be able to be compiled +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" +#endif return (T CK_CONSTANT_ADDRESS_SPACE*)p; // NOLINT(old-style-cast) +#ifdef __clang__ #pragma clang diagnostic pop +#endif } } // namespace ck diff --git a/include/ck/utility/amd_cluster_load.hpp b/include/ck/utility/amd_cluster_load.hpp index dcd811e8aa..6a70281dbd 100644 --- a/include/ck/utility/amd_cluster_load.hpp +++ b/include/ck/utility/amd_cluster_load.hpp @@ -19,11 +19,15 @@ namespace detail { template __device__ __attribute__((address_space(1))) T* to_global(const T* ptr) { +#ifdef __clang__ #pragma clang diagnostic push #pragma clang diagnostic ignored "-Wold-style-cast" #pragma clang diagnostic ignored "-Wcast-qual" +#endif return (__attribute__((address_space(1))) T*)(ptr); +#ifdef __clang__ #pragma clang diagnostic pop +#endif } } // namespace detail diff --git a/include/ck/utility/array.hpp b/include/ck/utility/array.hpp index b0e90b2cc0..4f1147feae 100644 --- a/include/ck/utility/array.hpp +++ b/include/ck/utility/array.hpp @@ -9,10 +9,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -98,6 +98,8 @@ __host__ __device__ constexpr auto make_array() } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif #endif diff --git a/include/ck/utility/c_style_pointer_cast.hpp b/include/ck/utility/c_style_pointer_cast.hpp index 023e0c38ad..dc973ca0fb 100644 --- a/include/ck/utility/c_style_pointer_cast.hpp +++ b/include/ck/utility/c_style_pointer_cast.hpp @@ -14,12 +14,15 @@ template && is_pointer_v, bool>::type = false> __host__ __device__ PY c_style_pointer_cast([[clang::lifetimebound]] PX p_x) { +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" #pragma clang diagnostic ignored "-Wcast-align" +#endif return (PY)p_x; // NOLINT(old-style-cast, cast-align) +#ifdef __clang__ #pragma clang diagnostic pop +#endif } } // namespace ck diff --git a/include/ck/utility/dtype_vector.hpp b/include/ck/utility/dtype_vector.hpp index 1e44d3bb28..017bd814b6 100644 --- a/include/ck/utility/dtype_vector.hpp +++ b/include/ck/utility/dtype_vector.hpp @@ -5,9 +5,10 @@ #include "ck/utility/data_type.hpp" #include "ck/utility/math.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { __device__ int static err = 0; @@ -855,4 +856,6 @@ using pk_i4x4_t = typename vector_type::type; using pk_i4x8_t = typename vector_type::type; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/dynamic_buffer.hpp b/include/ck/utility/dynamic_buffer.hpp index 2f1ed3b121..07fb93e9e2 100644 --- a/include/ck/utility/dynamic_buffer.hpp +++ b/include/ck/utility/dynamic_buffer.hpp @@ -16,10 +16,10 @@ #include "generic_memory_space_atomic.hpp" #include "data_cache_prefetch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { // T may be scalar or vector @@ -542,4 +542,6 @@ make_dynamic_buffer(T* p, ElementSpaceSize element_space_size, X invalid_element } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/env.hpp b/include/ck/utility/env.hpp index 1460f02b62..812b17a807 100644 --- a/include/ck/utility/env.hpp +++ b/include/ck/utility/env.hpp @@ -9,10 +9,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace internal { template @@ -87,8 +87,10 @@ struct EnvVar is_unset = false; value = val; } +#if __clang_major__ >= 23 #pragma clang diagnostic push #pragma clang diagnostic ignored "-Wdeprecated-declarations" +#endif explicit EnvVar(const char* const name, const T& def_val) { // NOLINTNEXTLINE (concurrency-mt-unsafe) @@ -103,7 +105,9 @@ struct EnvVar value = def_val; } } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif }; } // end namespace internal @@ -194,5 +198,7 @@ void UpdateEnvVar(EnvVar, const std::string_view& val) // environment variable to enable logging: // export CK_LOGGING=ON or CK_LOGGING=1 or CK_LOGGING=ENABLED CK_DECLARE_ENV_VAR_BOOL(CK_LOGGING) +#if __clang_major__ >= 23 #pragma clang diagnostic pop #endif +#endif diff --git a/include/ck/utility/functional2.hpp b/include/ck/utility/functional2.hpp index b704f2881f..827b26c7cf 100644 --- a/include/ck/utility/functional2.hpp +++ b/include/ck/utility/functional2.hpp @@ -8,10 +8,10 @@ #include "ck/utility/tuple.hpp" #include "ck/utility/type.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace detail { @@ -117,4 +117,6 @@ struct identity }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/sequence.hpp b/include/ck/utility/sequence.hpp index 8987f9fbf8..7f15d5bb3b 100644 --- a/include/ck/utility/sequence.hpp +++ b/include/ck/utility/sequence.hpp @@ -12,10 +12,10 @@ #include "ck/utility/functional.hpp" #include "ck/utility/math.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -1027,7 +1027,9 @@ using uniform_sequence_gen_t = typename uniform_sequence_gen::type; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif #if !defined(__HIPCC_RTC__) || !defined(CK_CODE_GEN_RTC) template diff --git a/include/ck/utility/span.hpp b/include/ck/utility/span.hpp index 72efd79711..d6f40a2a5e 100644 --- a/include/ck/utility/span.hpp +++ b/include/ck/utility/span.hpp @@ -7,10 +7,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { template @@ -70,4 +70,6 @@ class span } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/static_buffer.hpp b/include/ck/utility/static_buffer.hpp index bde35b8a0c..ecfc2b8ebe 100644 --- a/include/ck/utility/static_buffer.hpp +++ b/include/ck/utility/static_buffer.hpp @@ -5,9 +5,10 @@ #include "statically_indexed_array.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { // static buffer for scalar @@ -198,4 +199,6 @@ __host__ __device__ constexpr auto make_static_buffer(LongNumber) } } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/statically_indexed_array.hpp b/include/ck/utility/statically_indexed_array.hpp index 0d3be06c6b..b7660b7178 100644 --- a/include/ck/utility/statically_indexed_array.hpp +++ b/include/ck/utility/statically_indexed_array.hpp @@ -7,10 +7,10 @@ #include "functional2.hpp" #include "tuple.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace detail { @@ -107,6 +107,8 @@ struct StaticallyIndexedArray_v2 } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif #endif diff --git a/include/ck/utility/tuple.hpp b/include/ck/utility/tuple.hpp index 0067433839..685e4aa160 100644 --- a/include/ck/utility/tuple.hpp +++ b/include/ck/utility/tuple.hpp @@ -9,11 +9,11 @@ #include "ck/utility/enable_if.hpp" #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" #pragma clang diagnostic ignored "-Wlifetime-safety-lifetimebound-violation" - +#endif namespace ck { namespace detail { @@ -275,4 +275,6 @@ using tuple_element_or_t = typename detail::tuple_element_or_impl= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/type_convert.hpp b/include/ck/utility/type_convert.hpp index 2e7317a378..5d035e10a0 100644 --- a/include/ck/utility/type_convert.hpp +++ b/include/ck/utility/type_convert.hpp @@ -13,9 +13,10 @@ #include "ck/utility/amd_inline_asm.hpp" #include "ck/utility/type.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { // Define the common macro for MI300 models #if defined(__gfx942__) || defined(__gfx950__) @@ -2948,4 +2949,6 @@ inline __host__ __device__ void array_convert(Array& y, const Array } } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/utility/workgroup_barrier.hpp b/include/ck/utility/workgroup_barrier.hpp index 791357b8fa..75127075f6 100644 --- a/include/ck/utility/workgroup_barrier.hpp +++ b/include/ck/utility/workgroup_barrier.hpp @@ -5,10 +5,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { struct workgroup_barrier { @@ -65,4 +65,6 @@ struct workgroup_barrier }; } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/wrapper/layout.hpp b/include/ck/wrapper/layout.hpp index 11d16ab144..1d289f93f1 100644 --- a/include/ck/wrapper/layout.hpp +++ b/include/ck/wrapper/layout.hpp @@ -5,10 +5,10 @@ #include "ck/wrapper/utils/layout_utils.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // Disable from doxygen docs generation /// @cond INTERNAL namespace ck { @@ -485,4 +485,6 @@ struct Layout } // namespace wrapper } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck/wrapper/tensor.hpp b/include/ck/wrapper/tensor.hpp index 3b88228db4..38cd450f76 100644 --- a/include/ck/wrapper/tensor.hpp +++ b/include/ck/wrapper/tensor.hpp @@ -7,10 +7,10 @@ #include "utils/tensor_partition.hpp" #include "utils/layout_utils.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // Disable from doxygen docs generation /// @cond INTERNAL namespace ck { @@ -444,4 +444,6 @@ struct Tensor } // namespace wrapper } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/algorithm/coordinate_transform.hpp b/include/ck_tile/core/algorithm/coordinate_transform.hpp index 71a1a14b05..5969440dd5 100644 --- a/include/ck_tile/core/algorithm/coordinate_transform.hpp +++ b/include/ck_tile/core/algorithm/coordinate_transform.hpp @@ -11,10 +11,10 @@ #include "ck_tile/core/utility/magic_div.hpp" #include "ck_tile/core/utility/print.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { enum struct coord_transform_enum @@ -1808,4 +1808,6 @@ make_indexing_transform_with_adaptor(const UpLength& up_lengths, const IndexingA } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/arch/amd_buffer_addressing.hpp b/include/ck_tile/core/arch/amd_buffer_addressing.hpp index 86c52078f4..4134b077db 100644 --- a/include/ck_tile/core/arch/amd_buffer_addressing.hpp +++ b/include/ck_tile/core/arch/amd_buffer_addressing.hpp @@ -106,9 +106,10 @@ struct buffer_store; template struct buffer_store_if; +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wundefined-reinterpret-cast" +#endif // TODO: strict aliasing rule seems fail when reinterpret_cast between vector type // (exp_vector_type(xxx)) @@ -487,7 +488,9 @@ struct buffer_load_if<1, pre_nop> }; #endif -#pragma clang diagnostic pop // "-Wundefined-reinterpret-cast" +#ifdef __clang__ +#pragma clang diagnostic pop +#endif // "-Wundefined-reinterpret-cast" template <> struct buffer_store<16> @@ -1917,9 +1920,10 @@ CK_TILE_DEVICE void amd_async_buffer_load(CK_TILE_LDS_ADDR T* smem, if constexpr(oob_conditional_check) v_offset = flag ? v_offset : src_wave_buffer_resource[2]; +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" +#endif // Use C-style cast to change address space without dropping llvm noalias attribute llvm_amdgcn_raw_buffer_load_lds(src_wave_buffer_resource, (as3_uint32_ptr)(smem), @@ -1928,7 +1932,9 @@ CK_TILE_DEVICE void amd_async_buffer_load(CK_TILE_LDS_ADDR T* smem, src_wave_addr_offset, /*src_immediate_addr_offset*/ 0, static_cast(coherence)); +#ifdef __clang__ #pragma clang diagnostic pop +#endif } template (in_ptr)); +#ifdef __clang__ #pragma clang diagnostic pop +#endif if constexpr(std::is_same_v, ck_tile::half_t>) { typedef __attribute__((__vector_size__(4 * sizeof(__fp16)))) __fp16 llvm_fp16x4_t; diff --git a/include/ck_tile/core/arch/amd_buffer_addressing_builtins.hpp b/include/ck_tile/core/arch/amd_buffer_addressing_builtins.hpp index 7f1c264bbf..ac2b81cbfb 100644 --- a/include/ck_tile/core/arch/amd_buffer_addressing_builtins.hpp +++ b/include/ck_tile/core/arch/amd_buffer_addressing_builtins.hpp @@ -113,9 +113,10 @@ struct buffer_store; template struct buffer_store_if; +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wundefined-reinterpret-cast" +#endif // TODO: strict aliasing rule seems fail when reinterpret_cast between vector type // (exp_vector_type(xxx)) @@ -500,7 +501,9 @@ struct buffer_load_if<1, pre_nop> } }; #endif -#pragma clang diagnostic pop // "-Wundefined-reinterpret-cast" +#ifdef __clang__ +#pragma clang diagnostic pop +#endif // "-Wundefined-reinterpret-cast" template struct buffer_store; @@ -1984,9 +1987,10 @@ CK_TILE_DEVICE void amd_async_buffer_load(CK_TILE_LDS_ADDR T* smem, if constexpr(oob_conditional_check) v_offset = flag ? v_offset : 0x7fffffff; // large offset to cause OOB access +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" +#endif // Use C-style cast to change address space without dropping llvm noalias attribute __builtin_amdgcn_raw_ptr_buffer_load_lds(rsrc, smem, @@ -1995,7 +1999,9 @@ CK_TILE_DEVICE void amd_async_buffer_load(CK_TILE_LDS_ADDR T* smem, src_wave_addr_offset, /*imm*/ IMM, static_cast(coherence)); +#ifdef __clang__ #pragma clang diagnostic pop +#endif } template (in_ptr)); +#ifdef __clang__ #pragma clang diagnostic pop +#endif if constexpr(std::is_same_v, ck_tile::half_t>) { #if defined(__gfx950__) diff --git a/include/ck_tile/core/arch/amd_cluster_load.hpp b/include/ck_tile/core/arch/amd_cluster_load.hpp index 19ca7690a6..893d2d7334 100644 --- a/include/ck_tile/core/arch/amd_cluster_load.hpp +++ b/include/ck_tile/core/arch/amd_cluster_load.hpp @@ -14,20 +14,28 @@ namespace ck_tile { template CK_TILE_DEVICE __attribute__((address_space(1))) T* to_global(const T* ptr) { +#ifdef __clang__ #pragma clang diagnostic push #pragma clang diagnostic ignored "-Wold-style-cast" #pragma clang diagnostic ignored "-Wcast-qual" +#endif return (__attribute__((address_space(1))) T*)(ptr); +#ifdef __clang__ #pragma clang diagnostic pop +#endif } template CK_TILE_DEVICE __attribute__((address_space(3))) T* to_lds(T* ptr) { +#ifdef __clang__ #pragma clang diagnostic push #pragma clang diagnostic ignored "-Wold-style-cast" +#endif return (__attribute__((address_space(3))) T*)(ptr); +#ifdef __clang__ #pragma clang diagnostic pop +#endif } #endif // __gfx1250__ diff --git a/include/ck_tile/core/arch/arch.hpp b/include/ck_tile/core/arch/arch.hpp index d51a02df0b..6c8dcd7290 100644 --- a/include/ck_tile/core/arch/arch.hpp +++ b/include/ck_tile/core/arch/arch.hpp @@ -1299,11 +1299,14 @@ __device__ T* cast_pointer_to_generic_address_space(T CK_TILE_CONSTANT_ADDRESS_S { // cast a pointer in "Constant" address space (4) to "Generic" address space (0) // only c-style pointer cast seems be able to be compiled +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" +#endif return (T*)(p); // NOLINT(old-style-cast) +#ifdef __clang__ #pragma clang diagnostic pop +#endif } template @@ -1311,11 +1314,14 @@ __host__ __device__ T CK_TILE_CONSTANT_ADDRESS_SPACE* cast_pointer_to_constant_a { // cast a pointer in "Generic" address space (0) to "Constant" address space (4) // only c-style pointer cast seems be able to be compiled; +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" +#endif return (T CK_TILE_CONSTANT_ADDRESS_SPACE*)p; // NOLINT(old-style-cast) +#ifdef __clang__ #pragma clang diagnostic pop +#endif } CK_TILE_HOST_DEVICE constexpr index_t get_smem_capacity() diff --git a/include/ck_tile/core/arch/mma/amdgcn_mma.hpp b/include/ck_tile/core/arch/mma/amdgcn_mma.hpp index c2a31e92c8..5985f63440 100644 --- a/include/ck_tile/core/arch/mma/amdgcn_mma.hpp +++ b/include/ck_tile/core/arch/mma/amdgcn_mma.hpp @@ -23,10 +23,10 @@ #include #endif +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile::core::arch::mma { /**--------------------------------------------------- @@ -390,7 +390,9 @@ CK_TILE_HOST_DEVICE void print(amdgcn_mma= 23 #pragma clang diagnostic pop +#endif // Include the implementations #include "wmma/wmma.hpp" // should be included before the below headers diff --git a/include/ck_tile/core/arch/mma/mma_pipeline.hpp b/include/ck_tile/core/arch/mma/mma_pipeline.hpp index 418cad282d..de01760620 100644 --- a/include/ck_tile/core/arch/mma/mma_pipeline.hpp +++ b/include/ck_tile/core/arch/mma/mma_pipeline.hpp @@ -9,10 +9,10 @@ #include "mma_traits.hpp" #include "mma_transforms.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile::core::arch::mma { /*! @enum MmaPipelineOptionFlag @@ -346,4 +346,6 @@ concept MmaPipelineInterface = std::derived_from= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/container/array.hpp b/include/ck_tile/core/container/array.hpp index ab1d64c0ea..69f460bfe9 100644 --- a/include/ck_tile/core/container/array.hpp +++ b/include/ck_tile/core/container/array.hpp @@ -12,10 +12,10 @@ #include "ck_tile/core/utility/type_traits.hpp" #include "ck_tile/core/utility/functional.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { // use aggregate initialization for this type @@ -289,4 +289,6 @@ CK_TILE_HOST_DEVICE constexpr auto to_array(const X& x) } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/container/map.hpp b/include/ck_tile/core/container/map.hpp index ef91fb6af9..5b7ea9c406 100644 --- a/include/ck_tile/core/container/map.hpp +++ b/include/ck_tile/core/container/map.hpp @@ -8,10 +8,10 @@ #include "ck_tile/core/container/sequence.hpp" #include "ck_tile/core/container/tuple.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { // naive map @@ -161,4 +161,6 @@ CK_TILE_HOST_DEVICE static void print(const map& m) } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/container/span.hpp b/include/ck_tile/core/container/span.hpp index 9839cb533f..7835bfd90f 100644 --- a/include/ck_tile/core/container/span.hpp +++ b/include/ck_tile/core/container/span.hpp @@ -8,10 +8,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { // implement the c++20 std::span, lightweight, non-owning reference to a sequence @@ -80,4 +80,6 @@ class span }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/container/static_array.hpp b/include/ck_tile/core/container/static_array.hpp index 00c352c6f6..2d24ae8bfb 100644 --- a/include/ck_tile/core/container/static_array.hpp +++ b/include/ck_tile/core/container/static_array.hpp @@ -6,10 +6,10 @@ #include "ck_tile/core/config.hpp" #include "ck_tile/core/numeric/integer.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /** @@ -41,4 +41,6 @@ struct static_array }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/container/thread_buffer.hpp b/include/ck_tile/core/container/thread_buffer.hpp index 37a9b0c3e0..2976b878ee 100644 --- a/include/ck_tile/core/container/thread_buffer.hpp +++ b/include/ck_tile/core/container/thread_buffer.hpp @@ -7,10 +7,10 @@ #include "ck_tile/core/container/array.hpp" #include "ck_tile/core/container/tuple.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { #if CK_TILE_THREAD_BUFFER_DEFAULT == CK_TILE_THREAD_BUFFER_USE_TUPLE @@ -148,4 +148,6 @@ struct vector_traits, std::enable_if_t>> } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/container/tuple.hpp b/include/ck_tile/core/container/tuple.hpp index c147208f78..996ea3f52f 100644 --- a/include/ck_tile/core/container/tuple.hpp +++ b/include/ck_tile/core/container/tuple.hpp @@ -13,11 +13,11 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" #pragma clang diagnostic ignored "-Wlifetime-safety-lifetimebound-violation" - +#endif #ifndef CK_TILE_TUPLE_IMPL #define CK_TILE_TUPLE_IMPL 1 #endif @@ -861,4 +861,6 @@ struct tuple_element> } \ }() #endif +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/numeric/e4m3.hpp b/include/ck_tile/core/numeric/e4m3.hpp index 1010c01517..715cb80f26 100644 --- a/include/ck_tile/core/numeric/e4m3.hpp +++ b/include/ck_tile/core/numeric/e4m3.hpp @@ -5,10 +5,10 @@ #include "ck_tile/core/config.hpp" #include "ck_tile/core/numeric/scale_util.hpp" - +#if __clang_major__ >= 23 #pragma clang diagnostic push #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { struct e4m3_bexp_t @@ -89,4 +89,6 @@ CK_TILE_HOST_DEVICE e4m3_bexp_t::operator float() const } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/numeric/e5m3.hpp b/include/ck_tile/core/numeric/e5m3.hpp index 07f0868dec..daf799daf6 100644 --- a/include/ck_tile/core/numeric/e5m3.hpp +++ b/include/ck_tile/core/numeric/e5m3.hpp @@ -5,10 +5,10 @@ #include "ck_tile/core/config.hpp" #include "ck_tile/core/numeric/scale_util.hpp" - +#if __clang_major__ >= 23 #pragma clang diagnostic push #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { struct e5m3_bexp_t @@ -95,4 +95,6 @@ CK_TILE_HOST_DEVICE e5m3_bexp_t::operator float() const } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/numeric/e8m0.hpp b/include/ck_tile/core/numeric/e8m0.hpp index c3f85bf1f2..79b5e485db 100644 --- a/include/ck_tile/core/numeric/e8m0.hpp +++ b/include/ck_tile/core/numeric/e8m0.hpp @@ -6,10 +6,10 @@ #include "ck_tile/core/config.hpp" #include "ck_tile/core/numeric/mxfp_convert.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /** @@ -104,4 +104,6 @@ CK_TILE_HOST_DEVICE constexpr e8m0_bexp_t::operator float() const } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/numeric/pk_fp4.hpp b/include/ck_tile/core/numeric/pk_fp4.hpp index 3a63a374b3..1ee8a5557c 100644 --- a/include/ck_tile/core/numeric/pk_fp4.hpp +++ b/include/ck_tile/core/numeric/pk_fp4.hpp @@ -12,10 +12,10 @@ #include "ck_tile/core/numeric/mxfp_scale.hpp" #include "ck_tile/core/numeric/vector_type.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif #if defined(__gfx950__) || defined(__gfx125__) #define CK_TILE_FP4_CVT_DEVICE 1 #else @@ -736,4 +736,6 @@ CK_TILE_HOST_DEVICE constexpr bf8x2_t pk_fp4_t::to_bf8x2(float scale) const } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/numeric/type_convert.hpp b/include/ck_tile/core/numeric/type_convert.hpp index 2c582f67fd..af21d6e655 100644 --- a/include/ck_tile/core/numeric/type_convert.hpp +++ b/include/ck_tile/core/numeric/type_convert.hpp @@ -13,9 +13,10 @@ #include "ck_tile/core/numeric/int8.hpp" #include "ck_tile/core/numeric/mxfp_convert.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck_tile { #if CK_TILE_USE_CUSTOM_DATA_TYPE @@ -318,4 +319,6 @@ struct pk4scaled_type_convert_impl #endif } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/tensor/buffer_view.hpp b/include/ck_tile/core/tensor/buffer_view.hpp index bd69cedbb6..029c390157 100644 --- a/include/ck_tile/core/tensor/buffer_view.hpp +++ b/include/ck_tile/core/tensor/buffer_view.hpp @@ -19,11 +19,11 @@ #include "ck_tile/core/utility/type_traits.hpp" #include "ck_tile/core/utility/ignore.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" #pragma clang diagnostic ignored "-Wlifetime-safety-lifetimebound-violation" - +#endif namespace ck_tile { // T may be scalar or vector @@ -1588,4 +1588,6 @@ CK_TILE_HOST_DEVICE void print(const buffer_view= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/tensor/static_distributed_tensor.hpp b/include/ck_tile/core/tensor/static_distributed_tensor.hpp index e9eb51f1a5..9207fd2be1 100644 --- a/include/ck_tile/core/tensor/static_distributed_tensor.hpp +++ b/include/ck_tile/core/tensor/static_distributed_tensor.hpp @@ -14,10 +14,10 @@ #include "ck_tile/core/tensor/tile_distribution.hpp" #include "ck_tile/core/container/thread_buffer.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { template @@ -271,4 +271,6 @@ inline constexpr bool is_similiar_distributed_tensor_v = } // namespace detail } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/tensor/tensor_adaptor.hpp b/include/ck_tile/core/tensor/tensor_adaptor.hpp index 5ca7f468b3..43937f9b8f 100644 --- a/include/ck_tile/core/tensor/tensor_adaptor.hpp +++ b/include/ck_tile/core/tensor/tensor_adaptor.hpp @@ -12,10 +12,10 @@ #include "ck_tile/core/utility/type_traits.hpp" #include "ck_tile/core/numeric/numeric.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { // Transforms: Tuple @@ -955,4 +955,6 @@ CK_TILE_HOST_DEVICE constexpr auto chain_tensor_adaptors(const X& x, const Xs&.. remove_cvref_t, \ remove_cvref_t>{trans}; \ }() +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/tensor/tensor_adaptor_coordinate.hpp b/include/ck_tile/core/tensor/tensor_adaptor_coordinate.hpp index a517a99dbd..fb0b627314 100644 --- a/include/ck_tile/core/tensor/tensor_adaptor_coordinate.hpp +++ b/include/ck_tile/core/tensor/tensor_adaptor_coordinate.hpp @@ -14,10 +14,10 @@ #include "ck_tile/core/utility/type_traits.hpp" #include "ck_tile/core/utility/print.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { template @@ -348,11 +348,14 @@ struct CK_PRINT_X_, str_literal> Args&&... args) const { constexpr auto fmt_wrap_v = get_prefix() + str_literal{} + get_suffix(); +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wformat-nonliteral" +#endif printf(fmt_wrap_v.data, get_thread_id(), args..., targs.at(number())...); +#ifdef __clang__ #pragma clang diagnostic pop +#endif } template CK_TILE_HOST_DEVICE void operator()(T&& x, Args&&... args) const @@ -372,4 +375,6 @@ CK_TILE_HOST_DEVICE void print(const tensor_adaptor_coordinate& coord) detail::CK_PRINT_X_<>{}(coord); } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/tensor/tensor_view.hpp b/include/ck_tile/core/tensor/tensor_view.hpp index 6096b47e00..dd7ca94559 100644 --- a/include/ck_tile/core/tensor/tensor_view.hpp +++ b/include/ck_tile/core/tensor/tensor_view.hpp @@ -14,10 +14,10 @@ #include "ck_tile/core/utility/functional.hpp" #include "ck_tile/core/utility/type_traits.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { // placeholder type if we want to opt-out a tensor view parameter @@ -704,4 +704,6 @@ pad_tensor_view(const TensorView& tensor_view, const TileLengths& tile_lengths, } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/tensor/tile_distribution.hpp b/include/ck_tile/core/tensor/tile_distribution.hpp index 8b63b26964..606b346e19 100644 --- a/include/ck_tile/core/tensor/tile_distribution.hpp +++ b/include/ck_tile/core/tensor/tile_distribution.hpp @@ -15,10 +15,10 @@ #include "ck_tile/core/utility/functional.hpp" #include "ck_tile/core/utility/type_traits.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { template @@ -708,4 +708,6 @@ CK_TILE_HOST_DEVICE void print(const tile_distribution= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/tensor/tile_elementwise.hpp b/include/ck_tile/core/tensor/tile_elementwise.hpp index 12437095bd..914306a9ab 100644 --- a/include/ck_tile/core/tensor/tile_elementwise.hpp +++ b/include/ck_tile/core/tensor/tile_elementwise.hpp @@ -193,9 +193,10 @@ CK_TILE_DEVICE auto cast_tile_pk_fp8_fp32(const InTensor& in_dstr_tensors) constexpr index_t thread_buffer_size_pk = thread_buffer_size / 4; auto out_dstr_tensor = make_static_distributed_tensor(in_tile_dstr); +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wuninitialized" +#endif // __builtin_amdgcn_cvt_pk_fp8_f32() this builtin requires the old value, and // will generate a v_mov_b32 vxxx [old] before cvt, which result in unwanted ISA // so we prepare an uninitialized variable purposely, and turn off the warning @@ -218,7 +219,9 @@ CK_TILE_DEVICE auto cast_tile_pk_fp8_fp32(const InTensor& in_dstr_tensors) vec_t d = bit_cast(y); out_dstr_tensor.get_thread_buffer().template set_as(number{}, d); }); +#ifdef __clang__ #pragma clang diagnostic pop +#endif return out_dstr_tensor; #else diff --git a/include/ck_tile/core/utility/debug.hpp b/include/ck_tile/core/utility/debug.hpp index 23c6215a5e..111bca09bc 100644 --- a/include/ck_tile/core/utility/debug.hpp +++ b/include/ck_tile/core/utility/debug.hpp @@ -90,15 +90,18 @@ struct CK_PRINTF(make_str_literal(" ")); constexpr auto fmt_wrap_v = get_prefix() + fmt_v + get_suffix(); +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wformat-nonliteral" +#endif printf(fmt_wrap_v.data, get_thread_id(), N, args..., bit_cast>(type_convert(buf[Is]))...); +#ifdef __clang__ #pragma clang diagnostic pop +#endif } template diff --git a/include/ck_tile/core/utility/env.hpp b/include/ck_tile/core/utility/env.hpp index 01a6b39a74..1800467caa 100644 --- a/include/ck_tile/core/utility/env.hpp +++ b/include/ck_tile/core/utility/env.hpp @@ -6,10 +6,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { template @@ -112,8 +112,10 @@ struct EnvVar is_unset = false; value = val; } +#if __clang_major__ >= 23 #pragma clang diagnostic push #pragma clang diagnostic ignored "-Wdeprecated-declarations" +#endif explicit EnvVar(const char* const name, const T& def_val) { // NOLINTNEXTLINE (concurrency-mt-unsafe) @@ -128,7 +130,9 @@ struct EnvVar value = def_val; } } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif }; } // end namespace internal @@ -220,4 +224,6 @@ void UpdateEnvVar(EnvVar, const std::string_view& val) // environment variable to enable logging: // export CK_TILE_LOGGING=ON or CK_TILE_LOGGING=1 or CK_TILE_LOGGING=ENABLED CK_TILE_DECLARE_ENV_VAR_BOOL(CK_TILE_LOGGING) +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/utility/functional.hpp b/include/ck_tile/core/utility/functional.hpp index cceb87fb98..1125b2098d 100644 --- a/include/ck_tile/core/utility/functional.hpp +++ b/include/ck_tile/core/utility/functional.hpp @@ -10,9 +10,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck_tile { namespace detail { @@ -355,4 +356,6 @@ constexpr auto conditional_expr(X&& x, Y&& y) } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/core/utility/type_traits.hpp b/include/ck_tile/core/utility/type_traits.hpp index 1273062612..a8bedd964f 100644 --- a/include/ck_tile/core/utility/type_traits.hpp +++ b/include/ck_tile/core/utility/type_traits.hpp @@ -119,12 +119,15 @@ template < typename std::enable_if && std::is_pointer_v, bool>::type = false> CK_TILE_HOST_DEVICE PY c_style_pointer_cast([[clang::lifetimebound]] PX p_x) { +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" #pragma clang diagnostic ignored "-Wcast-align" +#endif return (PY)p_x; // NOLINT(old-style-cast, cast-align) +#ifdef __clang__ #pragma clang diagnostic pop +#endif } // Template ternary: if Cond == Match, use TrueType, else FalseType diff --git a/include/ck_tile/host/arg_parser.hpp b/include/ck_tile/host/arg_parser.hpp index dee28f9263..1a8f220bc4 100644 --- a/include/ck_tile/host/arg_parser.hpp +++ b/include/ck_tile/host/arg_parser.hpp @@ -13,10 +13,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /* * a host side utility, arg parser for, either @@ -238,4 +238,6 @@ class ArgParser std::vector keys; }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/host/host_tensor.hpp b/include/ck_tile/host/host_tensor.hpp index ca7ae452f8..e6cd9f6e77 100644 --- a/include/ck_tile/host/host_tensor.hpp +++ b/include/ck_tile/host/host_tensor.hpp @@ -17,10 +17,10 @@ #include "ck_tile/host/joinable_thread.hpp" #include "ck_tile/host/ranges.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { template @@ -785,4 +785,6 @@ auto get_default_stride(std::size_t row, return stride; } } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/host/joinable_thread.hpp b/include/ck_tile/host/joinable_thread.hpp index 586524221e..474c7ba5f4 100644 --- a/include/ck_tile/host/joinable_thread.hpp +++ b/include/ck_tile/host/joinable_thread.hpp @@ -59,11 +59,14 @@ class cpu_core_guard CPU_ZERO(&new_cpu_set); for(unsigned int i = 0; i < num_cores; ++i) { +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wold-style-cast" +#endif CPU_SET(i, &new_cpu_set); // NOLINT(old-style-cast) +#ifdef __clang__ #pragma clang diagnostic pop +#endif } sched_setaffinity(0, sizeof(cpu_set_t), &new_cpu_set); } diff --git a/include/ck_tile/host/kernel_launch.hpp b/include/ck_tile/host/kernel_launch.hpp index 6213ccfdb9..06cd45c1ea 100644 --- a/include/ck_tile/host/kernel_launch.hpp +++ b/include/ck_tile/host/kernel_launch.hpp @@ -15,11 +15,11 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" #pragma clang diagnostic ignored "-Wlifetime-safety-lifetimebound-violation" - +#endif namespace ck_tile { template @@ -375,4 +375,6 @@ CK_TILE_HOST float launch_kernel_time_mask_flush_cache(const stream_config& s, } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/batched_contraction/kernel/batched_contraction_kernel.hpp b/include/ck_tile/ops/batched_contraction/kernel/batched_contraction_kernel.hpp index bb0d1f7bb3..47a358600f 100644 --- a/include/ck_tile/ops/batched_contraction/kernel/batched_contraction_kernel.hpp +++ b/include/ck_tile/ops/batched_contraction/kernel/batched_contraction_kernel.hpp @@ -8,10 +8,10 @@ #include "ck_tile/ops/batched_contraction/utils/tensor_descriptor_utils.hpp" #include "ck_tile/ops/gemm/kernel/universal_gemm_kernel.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif /** * @file batched_contraction_kernel.hpp * @brief Batched Tensor Contraction Operations @@ -689,4 +689,6 @@ struct BatchedContractionKernel } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/flatmm/block/flatmm_32x512x128_1x4x1_16x16x32.hpp b/include/ck_tile/ops/flatmm/block/flatmm_32x512x128_1x4x1_16x16x32.hpp index c3e105e3b2..922393c861 100644 --- a/include/ck_tile/ops/flatmm/block/flatmm_32x512x128_1x4x1_16x16x32.hpp +++ b/include/ck_tile/ops/flatmm/block/flatmm_32x512x128_1x4x1_16x16x32.hpp @@ -467,9 +467,10 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_BF16 : public Flatmm_32x512x128_1x4x1_16 fp32x4_t v_acc[32]{.0f}; // B nr->kr +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif // clang-format off asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_BF16 @@ -484,7 +485,9 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_BF16 : public Flatmm_32x512x128_1x4x1_16 : _EXPAND_ASM_ARGS_CLOBBER, "s24", "s25", "s26", "s27" ); // clang-format on +#ifdef __clang__ #pragma clang diagnostic pop +#endif // return local scratch auto c = make_tuple(MakeCBlockTile(), MakeCBlockTile()); @@ -510,9 +513,10 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_BF16 : public Flatmm_32x512x128_1x4x1_16 fp32x4_t v_acc[16]{.0f}; // B nr->kr +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif // clang-format off asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_BF16 @@ -522,7 +526,9 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_BF16 : public Flatmm_32x512x128_1x4x1_16 : _EXPAND_ASM_ARGS_CLOBBER ); // clang-format on +#ifdef __clang__ #pragma clang diagnostic pop +#endif // return local scratch auto c = MakeCBlockTile(); @@ -606,9 +612,10 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_FP16 : public Flatmm_32x512x128_1x4x1_16 fp32x4_t v_acc[32]{.0f}; // B nr->kr +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif // clang-format off asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_FP16 @@ -623,7 +630,9 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_FP16 : public Flatmm_32x512x128_1x4x1_16 : _EXPAND_ASM_ARGS_CLOBBER, "s24", "s25", "s26", "s27" ); // clang-format on +#ifdef __clang__ #pragma clang diagnostic pop +#endif // return local scratch auto c = make_tuple(MakeCBlockTile(), MakeCBlockTile()); @@ -649,9 +658,10 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_FP16 : public Flatmm_32x512x128_1x4x1_16 fp32x4_t v_acc[16]{.0f}; // B nr->kr +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif // clang-format off asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_FP16 @@ -661,7 +671,9 @@ struct Flatmm_32x512x128_1x4x1_16x16x32_FP16 : public Flatmm_32x512x128_1x4x1_16 : _EXPAND_ASM_ARGS_CLOBBER ); // clang-format on +#ifdef __clang__ #pragma clang diagnostic pop +#endif // return local scratch auto c = MakeCBlockTile(); diff --git a/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32.hpp b/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32.hpp index 968c79471e..15886fcf85 100644 --- a/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32.hpp +++ b/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32.hpp @@ -169,9 +169,10 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_BF16 : public FlatmmSn_32x128x512_1x4x // B nr->kr // clang-format off +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_BF16 #include "uk/flatmm_sn_uk_gfx9_32x128x512_1x4x1_16x16x16.inc" @@ -310,7 +311,9 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_BF16 : public FlatmmSn_32x128x512_1x4x "v244", "v245", "v246", "v247", "v248", "v249", "v250", "v251", "v252", "v253", "v254", "v255" ); +#ifdef __clang__ #pragma clang diagnostic pop +#endif // clang-format on } }; @@ -414,9 +417,10 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_FP16 : public FlatmmSn_32x128x512_1x4x // B nr->kr // clang-format off +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_FP16 #include "uk/flatmm_sn_uk_gfx9_32x128x512_1x4x1_16x16x16.inc" @@ -555,7 +559,9 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_FP16 : public FlatmmSn_32x128x512_1x4x "v244", "v245", "v246", "v247", "v248", "v249", "v250", "v251", "v252", "v253", "v254", "v255" ); +#ifdef __clang__ #pragma clang diagnostic pop +#endif // clang-format on } }; diff --git a/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32_itl.hpp b/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32_itl.hpp index 6384db685a..066ccc4eaf 100644 --- a/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32_itl.hpp +++ b/include/ck_tile/ops/flatmm/block/flatmm_sn_32x128x512_1x4x1_16x16x32_itl.hpp @@ -113,9 +113,10 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_BF16_itl : public FlatmmSn_32x128x512_ // B nr->kr // clang-format off +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_BF16 #include "uk/flatmm_sn_uk_gfx9_32x128x512_1x4x1_16x16x16_itl.inc" @@ -257,7 +258,9 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_BF16_itl : public FlatmmSn_32x128x512_ "v244", "v245", "v246", "v247", "v248", "v249", "v250", "v251", "v252", "v253", "v254", "v255" ); +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif // clang-format on } }; @@ -361,9 +364,10 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_FP16_itl : public FlatmmSn_32x128x512_ // B nr->kr // clang-format off +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Winline-asm" +#endif asm volatile( #define CK_TILE_FLATMM_UK_MFMA CK_TILE_FLATMM_UK_MFMA_FP16 #include "uk/flatmm_sn_uk_gfx9_32x128x512_1x4x1_16x16x16_itl.inc" @@ -504,7 +508,9 @@ struct FlatmmSn_32x128x512_1x4x1_16x16x32_FP16_itl : public FlatmmSn_32x128x512_ "v244", "v245", "v246", "v247", "v248", "v249", "v250", "v251", "v252", "v253", "v254", "v255" ); +#ifdef __clang__ #pragma clang diagnostic pop +#endif // clang-format on } }; diff --git a/include/ck_tile/ops/flatmm/kernel/flatmm_kernel.hpp b/include/ck_tile/ops/flatmm/kernel/flatmm_kernel.hpp index 8ad4a24ff5..5a4b194edb 100644 --- a/include/ck_tile/ops/flatmm/kernel/flatmm_kernel.hpp +++ b/include/ck_tile/ops/flatmm/kernel/flatmm_kernel.hpp @@ -10,10 +10,10 @@ #include "ck_tile/ops/common.hpp" #include "ck_tile/ops/gemm/pipeline/gemm_pipeline_ag_bg_cr_scheduler.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { struct FlatmmProblem { @@ -982,4 +982,6 @@ struct FlatmmKernel } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/flatmm/kernel/grouped_flatmm_kernel.hpp b/include/ck_tile/ops/flatmm/kernel/grouped_flatmm_kernel.hpp index d1feda700f..161ff8e0d0 100644 --- a/include/ck_tile/ops/flatmm/kernel/grouped_flatmm_kernel.hpp +++ b/include/ck_tile/ops/flatmm/kernel/grouped_flatmm_kernel.hpp @@ -10,10 +10,10 @@ #include "ck_tile/ops/common.hpp" #include "ck_tile/ops/flatmm/kernel/flatmm_kernel.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { template , @@ -473,4 +473,6 @@ struct GroupedFlatmmKernel : FlatmmKernel= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/flatmm/kernel/moe_flatmm_kernel.hpp b/include/ck_tile/ops/flatmm/kernel/moe_flatmm_kernel.hpp index 284ecb22f5..cc6385865c 100644 --- a/include/ck_tile/ops/flatmm/kernel/moe_flatmm_kernel.hpp +++ b/include/ck_tile/ops/flatmm/kernel/moe_flatmm_kernel.hpp @@ -10,10 +10,10 @@ #include "ck_tile/ops/gemm/kernel/gemm_tile_partitioner.hpp" #include "ck_tile/host.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // #define disable_tile_gs namespace ck_tile { @@ -1499,4 +1499,6 @@ struct MoeFlatmmKernel } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/fmha/block/cast_tile_mx.hpp b/include/ck_tile/ops/fmha/block/cast_tile_mx.hpp index 4ed82f4ab4..a6925efa9a 100644 --- a/include/ck_tile/ops/fmha/block/cast_tile_mx.hpp +++ b/include/ck_tile/ops/fmha/block/cast_tile_mx.hpp @@ -56,9 +56,10 @@ cast_tile_mx(DstTensor& dst_tensor, DstScaleTensor& dst_scale_tensor, const SrcT // These builtins require the old value, and will generate a v_mov_b32 // vxxx [old] before cvt, which result in unwanted ISA so we prepare an // uninitialized variable x purposely, and turn off the warning +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wuninitialized" +#endif vec_t x; x = __builtin_amdgcn_cvt_scalef32_pk_fp4_f32( x, @@ -86,7 +87,9 @@ cast_tile_mx(DstTensor& dst_tensor, DstScaleTensor& dst_scale_tensor, const SrcT 3); // byte 3 dst_tensor.get_thread_buffer().template set_as( number{}, x); +#ifdef __clang__ #pragma clang diagnostic pop +#endif }); // Save scale for the corresponding lane @@ -184,9 +187,10 @@ cast_tile_mx(DstTensor& dst_tensor, DstScaleTensor& dst_scale_tensor, const SrcT // These builtins require the old value, and will generate a v_mov_b32 // vxxx [old] before cvt, which result in unwanted ISA so we prepare an // uninitialized variable x purposely, and turn off the warning +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wuninitialized" +#endif vec_t x; if constexpr(std::is_same_v) { @@ -220,7 +224,9 @@ cast_tile_mx(DstTensor& dst_tensor, DstScaleTensor& dst_scale_tensor, const SrcT } dst_tensor.get_thread_buffer().template set_as( number{}, x); +#ifdef __clang__ #pragma clang diagnostic pop +#endif }); // Save scale for the corresponding lane diff --git a/include/ck_tile/ops/fmha/block/page_block_navigator.hpp b/include/ck_tile/ops/fmha/block/page_block_navigator.hpp index 28d5796f9a..412d8bbd04 100644 --- a/include/ck_tile/ops/fmha/block/page_block_navigator.hpp +++ b/include/ck_tile/ops/fmha/block/page_block_navigator.hpp @@ -6,10 +6,10 @@ #include "ck_tile/core.hpp" #include "ck_tile/core/tensor/tile_window.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { // assume that we have only 1 page-block/tensor view @@ -361,4 +361,6 @@ CK_TILE_HOST_DEVICE auto make_page_block_navigator(copy_const_t* } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/fmha/block/variants.hpp b/include/ck_tile/ops/fmha/block/variants.hpp index f287ec8382..0eb4ae28b4 100644 --- a/include/ck_tile/ops/fmha/block/variants.hpp +++ b/include/ck_tile/ops/fmha/block/variants.hpp @@ -20,10 +20,10 @@ #define CK_TILE_ATTENTION_USE_SOFTSIGN_ASM 0 #endif +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { namespace internal { __device__ inline float @@ -339,4 +339,6 @@ struct ComposedAttention } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/fused_moe/kernel/moe_sorting_kernel.hpp b/include/ck_tile/ops/fused_moe/kernel/moe_sorting_kernel.hpp index 9e610d803f..2223b6bb32 100644 --- a/include/ck_tile/ops/fused_moe/kernel/moe_sorting_kernel.hpp +++ b/include/ck_tile/ops/fused_moe/kernel/moe_sorting_kernel.hpp @@ -10,10 +10,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif #if !defined(CK_TILE_HAS_ROW_NEWBCAST) // row_newbcast (DPP modifier 0x157) support by architecture: // - Not supported: gfx908 (MI100) and older @@ -3130,4 +3130,6 @@ struct MoeSortingMultiPhaseKernel_P23 } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm/kernel/gemm_kernel.hpp b/include/ck_tile/ops/gemm/kernel/gemm_kernel.hpp index df77349ace..ba8188047b 100644 --- a/include/ck_tile/ops/gemm/kernel/gemm_kernel.hpp +++ b/include/ck_tile/ops/gemm/kernel/gemm_kernel.hpp @@ -15,10 +15,10 @@ #include "ck_tile/ops/gemm/kernel/universal_gemm_kernel.hpp" #include "ck_tile/core/utility/type_traits.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /// @brief The GEMM kernel host arguments. @@ -175,4 +175,6 @@ struct GemmKernel }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm/kernel/gemm_multi_abd_kernel.hpp b/include/ck_tile/ops/gemm/kernel/gemm_multi_abd_kernel.hpp index 356377fba6..02a27968ec 100644 --- a/include/ck_tile/ops/gemm/kernel/gemm_multi_abd_kernel.hpp +++ b/include/ck_tile/ops/gemm/kernel/gemm_multi_abd_kernel.hpp @@ -15,10 +15,10 @@ #include "ck_tile/ops/gemm/kernel/universal_gemm_kernel.hpp" #include "ck_tile/core/utility/type_traits.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /// @brief The MultiABD GEMM kernel host arguments. @@ -200,4 +200,6 @@ struct GemmKernelMultiABD }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm/kernel/gemm_multi_d_kernel.hpp b/include/ck_tile/ops/gemm/kernel/gemm_multi_d_kernel.hpp index 1a6bcba7e8..9a03428d53 100644 --- a/include/ck_tile/ops/gemm/kernel/gemm_multi_d_kernel.hpp +++ b/include/ck_tile/ops/gemm/kernel/gemm_multi_d_kernel.hpp @@ -15,10 +15,10 @@ #include "ck_tile/ops/gemm/kernel/universal_gemm_kernel.hpp" #include "ck_tile/core/utility/type_traits.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /// @brief The MultiD GEMM kernel host arguments. @@ -195,4 +195,6 @@ struct GemmKernelMultiD }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm/kernel/grouped_gemm_kernel.hpp b/include/ck_tile/ops/gemm/kernel/grouped_gemm_kernel.hpp index 1e33ba0bcb..d2c5e4598c 100644 --- a/include/ck_tile/ops/gemm/kernel/grouped_gemm_kernel.hpp +++ b/include/ck_tile/ops/gemm/kernel/grouped_gemm_kernel.hpp @@ -14,10 +14,10 @@ #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /// @brief The Grouped GEMM kernel host arguments. @@ -583,4 +583,6 @@ struct GroupedGemmKernel } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm_mx/kernel/scale_pointer.hpp b/include/ck_tile/ops/gemm_mx/kernel/scale_pointer.hpp index c0a95fca05..0214f26a9c 100644 --- a/include/ck_tile/ops/gemm_mx/kernel/scale_pointer.hpp +++ b/include/ck_tile/ops/gemm_mx/kernel/scale_pointer.hpp @@ -5,10 +5,10 @@ #include "ck_tile/core.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { template @@ -115,4 +115,6 @@ struct MXScalePointer }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm_quant/block/block_gemm_quant_common.hpp b/include/ck_tile/ops/gemm_quant/block/block_gemm_quant_common.hpp index 4730439501..d4ff15a50e 100644 --- a/include/ck_tile/ops/gemm_quant/block/block_gemm_quant_common.hpp +++ b/include/ck_tile/ops/gemm_quant/block/block_gemm_quant_common.hpp @@ -5,10 +5,10 @@ #include "ck_tile/core.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { // Common utilities for quantized GEMM block operations @@ -228,4 +228,6 @@ struct AQPickerCommon : public BlockGemmQuantBase float scale_reg_f = 0.0f; }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm_quant/kernel/gemm_quant_kernel.hpp b/include/ck_tile/ops/gemm_quant/kernel/gemm_quant_kernel.hpp index b277780d6a..4c81718468 100644 --- a/include/ck_tile/ops/gemm_quant/kernel/gemm_quant_kernel.hpp +++ b/include/ck_tile/ops/gemm_quant/kernel/gemm_quant_kernel.hpp @@ -14,10 +14,10 @@ #include "ck_tile/host/concat.hpp" #include "ck_tile/ops/gemm_quant/pipeline/tile_gemm_quant_traits.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { namespace detail { @@ -1578,4 +1578,6 @@ struct QuantGemmKernel }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/gemm_quant/kernel/grouped_gemm_quant_kernel.hpp b/include/ck_tile/ops/gemm_quant/kernel/grouped_gemm_quant_kernel.hpp index 1e43836d57..46284c62ae 100644 --- a/include/ck_tile/ops/gemm_quant/kernel/grouped_gemm_quant_kernel.hpp +++ b/include/ck_tile/ops/gemm_quant/kernel/grouped_gemm_quant_kernel.hpp @@ -15,10 +15,10 @@ #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /// @brief The Grouped GEMM kernel host arguments. @@ -650,4 +650,6 @@ struct QuantGroupedGemmKernel }; } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/grouped_convolution/utils/grouped_convolution_utils.hpp b/include/ck_tile/ops/grouped_convolution/utils/grouped_convolution_utils.hpp index e526918ff7..b513f4fd75 100644 --- a/include/ck_tile/ops/grouped_convolution/utils/grouped_convolution_utils.hpp +++ b/include/ck_tile/ops/grouped_convolution/utils/grouped_convolution_utils.hpp @@ -7,10 +7,10 @@ #include "ck_tile/host/convolution_parameter.hpp" #include "ck_tile/ops/elementwise/unary_element_wise_operation.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { enum class GroupedConvDirection @@ -376,4 +376,6 @@ CK_TILE_HOST SplitImagePieceInfo calculate_spatial_piece(ck_tile::index_t piece_ } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ops/pooling/kernel/pool_kernel.hpp b/include/ck_tile/ops/pooling/kernel/pool_kernel.hpp index 297e397b11..8f6ee14663 100644 --- a/include/ck_tile/ops/pooling/kernel/pool_kernel.hpp +++ b/include/ck_tile/ops/pooling/kernel/pool_kernel.hpp @@ -8,10 +8,10 @@ #include "ck_tile/ops/common.hpp" #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck_tile { /// @brief Host arguments for pooling operations @@ -580,4 +580,6 @@ struct PoolKernel } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/ck_tile/ref/naive_attention.hpp b/include/ck_tile/ref/naive_attention.hpp index f7d1da36e2..6e401b128f 100644 --- a/include/ck_tile/ref/naive_attention.hpp +++ b/include/ck_tile/ref/naive_attention.hpp @@ -9,9 +9,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck_tile { enum class naive_attention_layout_enum @@ -827,4 +828,6 @@ CK_TILE_HOST float naive_attention_fwd(naive_attention_fwd_traits t, #undef CK_TILE_DISPATCH_NAIVE_ATTEN_FWD_INTERNAL_ } // namespace ck_tile +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/include/rapidjson/document.h b/include/rapidjson/document.h index 0b12550a00..e937a3ca99 100644 --- a/include/rapidjson/document.h +++ b/include/rapidjson/document.h @@ -29,7 +29,7 @@ #endif RAPIDJSON_DIAG_PUSH -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_OFF(padded) RAPIDJSON_DIAG_OFF(switch - enum) RAPIDJSON_DIAG_OFF(c++ 98 - compat) diff --git a/include/rapidjson/encodedstream.h b/include/rapidjson/encodedstream.h index 4b96e79b7b..075dfc8691 100644 --- a/include/rapidjson/encodedstream.h +++ b/include/rapidjson/encodedstream.h @@ -23,7 +23,7 @@ RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(effc++) #endif -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) #endif @@ -416,7 +416,7 @@ class AutoUTFOutputStream RAPIDJSON_NAMESPACE_END -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP #endif diff --git a/include/rapidjson/error/en.h b/include/rapidjson/error/en.h index 9ea11e40ec..f11abe5cd7 100644 --- a/include/rapidjson/error/en.h +++ b/include/rapidjson/error/en.h @@ -17,7 +17,7 @@ #include "error.h" -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(switch - enum) RAPIDJSON_DIAG_OFF(covered - switch - default) @@ -266,7 +266,7 @@ GetPointerParseError_En(PointerParseErrorCode pointerParseErrorCode) RAPIDJSON_NAMESPACE_END -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP #endif diff --git a/include/rapidjson/error/error.h b/include/rapidjson/error/error.h index 12ca3e085c..fcd308dc99 100644 --- a/include/rapidjson/error/error.h +++ b/include/rapidjson/error/error.h @@ -17,7 +17,7 @@ #include "../rapidjson.h" -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) #endif @@ -298,7 +298,7 @@ typedef const RAPIDJSON_ERROR_CHARTYPE* (*GetPointerParseErrorFunc)(PointerParse RAPIDJSON_NAMESPACE_END -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP #endif diff --git a/include/rapidjson/filereadstream.h b/include/rapidjson/filereadstream.h index 8cdd792f44..a12d408ffb 100644 --- a/include/rapidjson/filereadstream.h +++ b/include/rapidjson/filereadstream.h @@ -18,7 +18,7 @@ #include "stream.h" #include -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) RAPIDJSON_DIAG_OFF(unreachable - code) @@ -116,7 +116,7 @@ class FileReadStream RAPIDJSON_NAMESPACE_END -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP #endif diff --git a/include/rapidjson/filewritestream.h b/include/rapidjson/filewritestream.h index fd805d38d0..5ab74e646d 100644 --- a/include/rapidjson/filewritestream.h +++ b/include/rapidjson/filewritestream.h @@ -18,7 +18,7 @@ #include "stream.h" #include -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(unreachable - code) #endif @@ -128,7 +128,7 @@ inline void PutN(FileWriteStream& stream, char c, size_t n) RAPIDJSON_NAMESPACE_END -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP #endif diff --git a/include/rapidjson/internal/diyfp.h b/include/rapidjson/internal/diyfp.h index 7dfe4aa550..354495088e 100644 --- a/include/rapidjson/internal/diyfp.h +++ b/include/rapidjson/internal/diyfp.h @@ -40,7 +40,7 @@ RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(effc++) #endif -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) #endif @@ -264,7 +264,7 @@ inline DiyFp GetCachedPower10(int exp, int* outExp) RAPIDJSON_DIAG_POP #endif -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP RAPIDJSON_DIAG_OFF(padded) #endif diff --git a/include/rapidjson/internal/regex.h b/include/rapidjson/internal/regex.h index 1a078f1ba0..6f86d673cc 100644 --- a/include/rapidjson/internal/regex.h +++ b/include/rapidjson/internal/regex.h @@ -19,7 +19,7 @@ #include "../stream.h" #include "stack.h" -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) RAPIDJSON_DIAG_OFF(switch - enum) diff --git a/include/rapidjson/istreamwrapper.h b/include/rapidjson/istreamwrapper.h index ad07e5ca3c..a2fe18a48b 100644 --- a/include/rapidjson/istreamwrapper.h +++ b/include/rapidjson/istreamwrapper.h @@ -19,7 +19,7 @@ #include #include -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) #elif defined(_MSC_VER) diff --git a/include/rapidjson/memorystream.h b/include/rapidjson/memorystream.h index 221b756d7f..7a6471ac3e 100644 --- a/include/rapidjson/memorystream.h +++ b/include/rapidjson/memorystream.h @@ -17,7 +17,7 @@ #include "stream.h" -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(unreachable - code) RAPIDJSON_DIAG_OFF(missing - noreturn) @@ -74,7 +74,7 @@ struct MemoryStream RAPIDJSON_NAMESPACE_END -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP #endif diff --git a/include/rapidjson/ostreamwrapper.h b/include/rapidjson/ostreamwrapper.h index 547bd09fc7..f3dc362b11 100644 --- a/include/rapidjson/ostreamwrapper.h +++ b/include/rapidjson/ostreamwrapper.h @@ -18,7 +18,7 @@ #include "stream.h" #include -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) #endif @@ -89,7 +89,7 @@ class BasicOStreamWrapper typedef BasicOStreamWrapper OStreamWrapper; typedef BasicOStreamWrapper WOStreamWrapper; -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_POP #endif diff --git a/include/rapidjson/pointer.h b/include/rapidjson/pointer.h index bba4519774..387837c88b 100644 --- a/include/rapidjson/pointer.h +++ b/include/rapidjson/pointer.h @@ -20,7 +20,7 @@ #include "internal/itoa.h" #include "error/error.h" // PointerParseErrorCode -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(switch - enum) #elif defined(_MSC_VER) diff --git a/include/rapidjson/reader.h b/include/rapidjson/reader.h index fe4d6e3ec9..f580dfaf18 100644 --- a/include/rapidjson/reader.h +++ b/include/rapidjson/reader.h @@ -38,7 +38,7 @@ #include #endif -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(old - style - cast) RAPIDJSON_DIAG_OFF(padded) diff --git a/include/rapidjson/schema.h b/include/rapidjson/schema.h index 8a542afc66..dca3f2d910 100644 --- a/include/rapidjson/schema.h +++ b/include/rapidjson/schema.h @@ -53,7 +53,7 @@ RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(effc++) #endif -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_OFF(weak - vtables) RAPIDJSON_DIAG_OFF(exit - time - destructors) RAPIDJSON_DIAG_OFF(c++ 98 - compat - pedantic) diff --git a/include/rapidjson/writer.h b/include/rapidjson/writer.h index 634060020d..4a325d5bdb 100644 --- a/include/rapidjson/writer.h +++ b/include/rapidjson/writer.h @@ -37,7 +37,7 @@ #include #endif -#ifdef __clang__ +#if defined(__clang__) RAPIDJSON_DIAG_PUSH RAPIDJSON_DIAG_OFF(padded) RAPIDJSON_DIAG_OFF(unreachable - code) diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_avgpool_bwd.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_avgpool_bwd.hpp index c61ab5eae7..76fbf9d922 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_avgpool_bwd.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_avgpool_bwd.hpp @@ -10,10 +10,10 @@ #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -357,4 +357,6 @@ struct ReferenceAvgPoolBwd : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_batched_gemm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_batched_gemm.hpp index 7e219f12d5..de23d18d15 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_batched_gemm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_batched_gemm.hpp @@ -10,10 +10,10 @@ #include "ck/library/utility/host_tensor.hpp" #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -426,4 +426,6 @@ struct ReferenceBatchedGemm_GQA : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_cgemm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_cgemm.hpp index 23b8d69ba7..1897271c5b 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_cgemm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_cgemm.hpp @@ -10,10 +10,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -187,4 +187,6 @@ struct ReferenceCGemm : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_column_to_image.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_column_to_image.hpp index d193f08a99..19f1f43e21 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_column_to_image.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_column_to_image.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/tensor_layout.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -371,4 +371,6 @@ struct ReferenceColumnToImage : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_contraction.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_contraction.hpp index c4164a3e18..3d2eb46649 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_contraction.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_contraction.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/element/element_wise_operation.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -512,4 +512,6 @@ struct ReferenceBatchedContraction_G1_M3_N2_K1 : public ck::tensor_operation::de } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_data.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_data.hpp index 572855d4d4..da26103ada 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_data.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_data.hpp @@ -10,10 +10,10 @@ #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -511,4 +511,6 @@ struct ReferenceConvBwdData : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_weight.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_weight.hpp index cc5ca29c6d..d96e7b16d3 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_weight.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_bwd_weight.hpp @@ -10,10 +10,10 @@ #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -456,4 +456,6 @@ struct ReferenceConvBwdWeight : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd.hpp index 6f6d0f464e..c42676b209 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd.hpp @@ -23,10 +23,10 @@ #include "ck/library/utility/convolution_parameter.hpp" #include "ck/library/utility/convolution_host_tensor_descriptor_helper.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -572,4 +572,6 @@ struct ReferenceConvFwd : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation.hpp index 6daa0d0ba1..475d6e1fc2 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation.hpp @@ -9,10 +9,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -195,4 +195,6 @@ struct ReferenceConvFwd_Bias_Activation : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation_add.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation_add.hpp index 61a44acc28..b6c86abf7c 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation_add.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_conv_fwd_bias_activation_add.hpp @@ -9,10 +9,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -203,4 +203,6 @@ struct ReferenceConvFwd_Bias_Activation_Add : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_elementwise.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_elementwise.hpp index fe2cb4be4a..d6be12e20f 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_elementwise.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_elementwise.hpp @@ -10,10 +10,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -113,4 +113,6 @@ struct ReferenceElementwise : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_fpAintB_gemm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_fpAintB_gemm.hpp index cf8fb45179..26cecfc0fa 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_fpAintB_gemm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_fpAintB_gemm.hpp @@ -10,10 +10,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -149,4 +149,6 @@ struct ReferencefpAintBGemm : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm.hpp index b9dc834e80..8073b41bd2 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -251,4 +251,6 @@ struct ReferenceGemm : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_layernorm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_layernorm.hpp index 33ef0b97cc..304a13644e 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_layernorm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_layernorm.hpp @@ -7,10 +7,10 @@ #include #include "ck/library/reference_tensor_operation/cpu/reference_gemm.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -239,4 +239,6 @@ struct ReferenceGemmLayernorm : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multi_abd.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multi_abd.hpp index b55b1dbf99..490b808399 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multi_abd.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multi_abd.hpp @@ -14,10 +14,10 @@ #include "ck/library/reference_tensor_operation/cpu/reference_gemm.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -197,4 +197,6 @@ struct ReferenceGemmMultiABD : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multiple_d.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multiple_d.hpp index 8aeb0069b7..1ec311c7df 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multiple_d.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_gemm_multiple_d.hpp @@ -10,10 +10,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -160,4 +160,6 @@ struct ReferenceGemmMultipleD : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm.hpp index 2e1e4e0261..3b4764721e 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm.hpp @@ -12,10 +12,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -215,4 +215,6 @@ struct ReferenceGroupnorm : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm_bwd.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm_bwd.hpp index 57c9c16c77..7d89ea6fd8 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm_bwd.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_groupnorm_bwd.hpp @@ -12,10 +12,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -235,4 +235,6 @@ struct ReferenceGroupnormBwd : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_image_to_column.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_image_to_column.hpp index d935c766e7..33910f5c91 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_image_to_column.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_image_to_column.hpp @@ -12,10 +12,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/numeric.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -335,4 +335,6 @@ struct ReferenceImageToColumn : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm.hpp index a9c372510b..aa4837aac6 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm.hpp @@ -12,10 +12,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -263,4 +263,6 @@ struct ReferenceLayernorm : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm_bwd.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm_bwd.hpp index 5564bb7249..428b944d15 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm_bwd.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_layernorm_bwd.hpp @@ -12,10 +12,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -204,4 +204,6 @@ struct ReferenceLayernormBwd : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_maxpool_bwd.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_maxpool_bwd.hpp index cd3fb7922e..157b6e56e5 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_maxpool_bwd.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_maxpool_bwd.hpp @@ -11,10 +11,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -115,4 +115,6 @@ struct ReferenceMaxPoolBwd : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm.hpp index 843128ce97..9ebadcd665 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -297,4 +297,6 @@ struct ReferenceMoeGemm : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale.hpp index aa5a22b305..0b2c9152ee 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -282,4 +282,6 @@ struct ReferenceMoeGemm1BlockScale : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale_splitk.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale_splitk.hpp index 1f152c6aae..0755cf3cd7 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale_splitk.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm1_blockscale_splitk.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -234,4 +234,6 @@ struct ReferenceMoeGemm1BlockScaleSplitK : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2.hpp index 451528fb50..074a0b6f82 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2.hpp @@ -12,10 +12,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -270,4 +270,6 @@ struct ReferenceMoeGemm2 : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2_blockscale.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2_blockscale.hpp index 8f2a35bd26..f672546a6b 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2_blockscale.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_gemm2_blockscale.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -250,4 +250,6 @@ struct ReferenceMoeGemm2BlockScale : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm1.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm1.hpp index 5911f539b1..4d50cb6af0 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm1.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm1.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -266,4 +266,6 @@ struct ReferenceMoeMXGemm1 : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm2.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm2.hpp index 88e5c6bc3e..0d3022e517 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm2.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_moe_mx_gemm2.hpp @@ -11,10 +11,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/library/utility/host_tensor.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -243,4 +243,6 @@ struct ReferenceMoeMXGemm2 : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_mx_gemm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_mx_gemm.hpp index 1732cef098..caa962e0e1 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_mx_gemm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_mx_gemm.hpp @@ -11,10 +11,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/reference_tensor_operation/cpu/reference_gemm.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -239,4 +239,6 @@ struct ReferenceMXGemm : public device::BaseOperator } // namespace host } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_pool_fwd.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_pool_fwd.hpp index 8bb217ff83..d14da641ab 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_pool_fwd.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_pool_fwd.hpp @@ -14,10 +14,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -363,4 +363,6 @@ struct ReferencePoolingFwd : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/library/include/ck/library/reference_tensor_operation/cpu/reference_sparse_embedding3_forward_layernorm.hpp b/library/include/ck/library/reference_tensor_operation/cpu/reference_sparse_embedding3_forward_layernorm.hpp index 9d4e34ed4e..b050a35337 100644 --- a/library/include/ck/library/reference_tensor_operation/cpu/reference_sparse_embedding3_forward_layernorm.hpp +++ b/library/include/ck/library/reference_tensor_operation/cpu/reference_sparse_embedding3_forward_layernorm.hpp @@ -12,10 +12,10 @@ #include "ck/library/utility/host_tensor.hpp" #include "ck/library/utility/host_tensor_generator.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace host { @@ -208,4 +208,6 @@ struct ReferenceSparseEmbedding3ForwardLayernorm : public device::BaseOperator } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 3155428e57..b366659e7f 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,9 +10,10 @@ #include "ck/tensor_operation/gpu/device/device_base.hpp" #include "ck/host_utility/kernel_launch.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif namespace ck { template = 23 #pragma clang diagnostic pop +#endif diff --git a/library/src/tensor_operation_instance/gpu/gemm_multi_abd/device_gemm_wmma_multi_abd_gelu_bf16_i8_bf16_mk_kn_mn_v1_instance.cpp b/library/src/tensor_operation_instance/gpu/gemm_multi_abd/device_gemm_wmma_multi_abd_gelu_bf16_i8_bf16_mk_kn_mn_v1_instance.cpp index d93eb1f89e..5893116f23 100644 --- a/library/src/tensor_operation_instance/gpu/gemm_multi_abd/device_gemm_wmma_multi_abd_gelu_bf16_i8_bf16_mk_kn_mn_v1_instance.cpp +++ b/library/src/tensor_operation_instance/gpu/gemm_multi_abd/device_gemm_wmma_multi_abd_gelu_bf16_i8_bf16_mk_kn_mn_v1_instance.cpp @@ -12,10 +12,10 @@ #include "ck/library/tensor_operation_instance/add_device_operation_instance.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif namespace ck { namespace tensor_operation { namespace device { @@ -61,4 +61,6 @@ void add_device_gemm_wmma_multi_abd_bf16_i8_bf16_mk_kn_mn_gelu_v1_instances( } // namespace device } // namespace tensor_operation } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/profiler/include/profiler/profile_contraction_utils.hpp b/profiler/include/profiler/profile_contraction_utils.hpp index e5cc521178..88d4af2f0b 100644 --- a/profiler/include/profiler/profile_contraction_utils.hpp +++ b/profiler/include/profiler/profile_contraction_utils.hpp @@ -7,9 +7,10 @@ #include "ck/ck.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif using Row = ck::tensor_layout::gemm::RowMajor; using Col = ck::tensor_layout::gemm::ColumnMajor; @@ -84,4 +85,6 @@ assign_default_strides(Col, std::vector& strides, std::vector= 23 #pragma clang diagnostic pop +#endif diff --git a/profiler/src/profiler_operation_registry.hpp b/profiler/src/profiler_operation_registry.hpp index 262793d09e..49e27f36e9 100644 --- a/profiler/src/profiler_operation_registry.hpp +++ b/profiler/src/profiler_operation_registry.hpp @@ -9,10 +9,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif class ProfilerOperationRegistry final { ProfilerOperationRegistry() = default; @@ -87,4 +87,6 @@ class ProfilerOperationRegistry final ::ProfilerOperationRegistry::GetInstance().Add(name, description, operation) \ _Pragma("clang diagnostic pop") // clang-format on +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/test/ck_tile/flatmm/test_mx_flatmm_base.hpp b/test/ck_tile/flatmm/test_mx_flatmm_base.hpp index 919a295d52..8ce3fb5a86 100644 --- a/test/ck_tile/flatmm/test_mx_flatmm_base.hpp +++ b/test/ck_tile/flatmm/test_mx_flatmm_base.hpp @@ -203,9 +203,10 @@ class TestMXFlatmmBase : public ::testing::Test // mx_flatmm_calc is explicitly instantiated in the linked object library; // suppress the -Wundefined-func-template warning that fires when the // compiler sees only the forward declaration in mx_flatmm.hpp. +#ifdef __clang__ #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wundefined-func-template" +#endif BaseFlatmmPipeline::template TailHandler( [&](auto has_hot_loop_, auto tail_num_) { constexpr auto has_hot_loop_v = has_hot_loop_.value; @@ -232,7 +233,9 @@ class TestMXFlatmmBase : public ::testing::Test }, has_hot_loop, tail_num); +#ifdef __clang__ #pragma clang diagnostic pop +#endif c_dev_buf.FromDevice(c_rslt_host.data()); diff --git a/test/common/csv_test_loader.hpp b/test/common/csv_test_loader.hpp index ac3ac5f97e..43d94049c0 100644 --- a/test/common/csv_test_loader.hpp +++ b/test/common/csv_test_loader.hpp @@ -12,9 +12,10 @@ #include "ck/library/utility/convolution_parameter.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif namespace ck { namespace test { @@ -247,4 +248,6 @@ inline bool load_and_populate_test_cases(const std::vector& csv_pat } // namespace test } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 28795d1415..a779ae1107 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 @@ -15,9 +15,10 @@ using namespace ck::tensor_layout::convolution; // Import tensor layout names (G static ck::index_t param_mask = 0xffff; static ck::index_t instance_index = -1; +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif // Load CSV data for 2D tests static std::vector Get2DTestCases() { @@ -343,4 +344,6 @@ int main(int argc, char** argv) } return RUN_ALL_TESTS(); } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 e78660325a..7dadcbd4e2 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,9 +15,10 @@ #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 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif using namespace ck::tensor_layout::convolution; static ck::index_t param_mask = 0xffff; static ck::index_t instance_index = -1; @@ -282,4 +283,6 @@ int main(int argc, char** argv) } return RUN_ALL_TESTS(); } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 ffc9067c12..3c4a77f4bd 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,9 +11,10 @@ #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 +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif using namespace ck::tensor_layout::convolution; // Import tensor layout names (NHWGC, GKYXC, etc.) static ck::index_t param_mask = 0xffff; static ck::index_t instance_index = -1; @@ -196,4 +197,6 @@ int main(int argc, char** argv) } return RUN_ALL_TESTS(); } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 07c107d6c8..265f6e33e6 100644 --- a/test/grouped_gemm/test_grouped_gemm_fixed_nk_bias.cpp +++ b/test/grouped_gemm/test_grouped_gemm_fixed_nk_bias.cpp @@ -18,9 +18,10 @@ #include #include +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif static ck::index_t param_mask = 0xffffff; using FP32 = float; @@ -305,4 +306,6 @@ int main(int argc, char** argv) } return RUN_ALL_TESTS(); } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 70ca4b7705..b6be6e6f69 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,9 +15,10 @@ #include "gtest/gtest.h" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif using FP32 = float; using FP16 = ck::half_t; using BF16 = ck::bhalf_t; @@ -240,4 +241,6 @@ int main(int argc, char** argv) testing::InitGoogleTest(&argc, argv); return RUN_ALL_TESTS(); } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/test/grouped_gemm/test_grouped_gemm_util.hpp b/test/grouped_gemm/test_grouped_gemm_util.hpp index fd835bb5aa..ca5932093e 100644 --- a/test/grouped_gemm/test_grouped_gemm_util.hpp +++ b/test/grouped_gemm/test_grouped_gemm_util.hpp @@ -18,10 +18,10 @@ #include "profiler/profile_grouped_gemm_impl.hpp" #include "profiler/profile_grouped_gemm_fixed_nk_impl.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" - +#endif static ck::index_t param_mask = 0xffffff; static ck::index_t instance_index = -1; @@ -346,4 +346,6 @@ int main(int argc, char** argv) } return RUN_ALL_TESTS(); } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif 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 dd31a02cec..fabd74daf6 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,9 +19,10 @@ #include "example/ck_tile/17_grouped_gemm/grouped_gemm_multi_d.hpp" #include "profiler/profile_grouped_gemm_tile_loop_generic_impl.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-invalidation" +#endif extern ck::index_t param_mask; extern ck::index_t instance_index; @@ -174,4 +175,6 @@ class TestGroupedGemmTileLoop : public testing::Test } // namespace test } // namespace ck +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/test/position_embedding/position_embedding.cpp b/test/position_embedding/position_embedding.cpp index 04b77be9cd..d72c0fff03 100644 --- a/test/position_embedding/position_embedding.cpp +++ b/test/position_embedding/position_embedding.cpp @@ -9,10 +9,10 @@ #include "ck_tile/core.hpp" #include "ck_tile/ops/fmha.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif #ifndef TEST_ALIBI_VERBOSE #define TEST_ALIBI_VERBOSE 0 #endif @@ -217,4 +217,6 @@ int main() // clang-format on return rtn ? 0 : -1; } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/tile_engine/ops/gemm/gemm_multi_d/gemm_multi_d_benchmark.hpp b/tile_engine/ops/gemm/gemm_multi_d/gemm_multi_d_benchmark.hpp index 24cfebdf65..d1345c527e 100644 --- a/tile_engine/ops/gemm/gemm_multi_d/gemm_multi_d_benchmark.hpp +++ b/tile_engine/ops/gemm/gemm_multi_d/gemm_multi_d_benchmark.hpp @@ -13,10 +13,10 @@ #include "ck_tile/host.hpp" #include "gemm/gemm_benchmark.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif // Data types and Layouts are defined by the generated kernel headers // No hardcoded type definitions here to avoid conflicts struct GemmMultiDProblem : GemmProblem @@ -74,4 +74,6 @@ void gemm_multi_d_host_reference(int verify, a_m_k, b_k_n, {d0_m_n, d1_m_n}, c_m_n_host_result); } } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/tile_engine/ops/gemm/gemm_preshuffle/gemm_preshuffle_benchmark.hpp b/tile_engine/ops/gemm/gemm_preshuffle/gemm_preshuffle_benchmark.hpp index 8b9d95a087..beebc0ffc6 100644 --- a/tile_engine/ops/gemm/gemm_preshuffle/gemm_preshuffle_benchmark.hpp +++ b/tile_engine/ops/gemm/gemm_preshuffle/gemm_preshuffle_benchmark.hpp @@ -2,10 +2,10 @@ // SPDX-License-Identifier: MIT #pragma once +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif #include "ck_tile/core.hpp" #include "ck_tile/host.hpp" #include "gemm_preshuffle_common.hpp" @@ -70,4 +70,6 @@ void gemm_host_reference(int verify, c_m_n_gpu_buf_ref.FromDevice(c_m_n_ref.data()); } } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/tile_engine/ops/gemm/gemm_universal/gemm_universal_benchmark.hpp b/tile_engine/ops/gemm/gemm_universal/gemm_universal_benchmark.hpp index 64e8b3e226..23338a6cd0 100644 --- a/tile_engine/ops/gemm/gemm_universal/gemm_universal_benchmark.hpp +++ b/tile_engine/ops/gemm/gemm_universal/gemm_universal_benchmark.hpp @@ -13,9 +13,10 @@ #include "ck_tile/host.hpp" #include "gemm/gemm_benchmark.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif // Data types and Layouts are defined by the generated kernel headers // No hardcoded type definitions here to avoid conflicts @@ -67,4 +68,6 @@ void gemm_host_reference(int verify, c_m_n_gpu_buf_ref.FromDevice(c_m_n_host_result.data()); } } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/tile_engine/ops/gemm/grouped_gemm/grouped_gemm_benchmark.hpp b/tile_engine/ops/gemm/grouped_gemm/grouped_gemm_benchmark.hpp index 521e0c1e21..3cbeb6caba 100644 --- a/tile_engine/ops/gemm/grouped_gemm/grouped_gemm_benchmark.hpp +++ b/tile_engine/ops/gemm/grouped_gemm/grouped_gemm_benchmark.hpp @@ -15,9 +15,10 @@ #include "ck_tile/host.hpp" #include "grouped_gemm_common.hpp" +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" +#endif // Data types and Layouts are defined by the generated kernel headers // No hardcoded type definitions here to avoid conflicts @@ -264,4 +265,6 @@ void gemm_host_reference_grouped(int verify, } } } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif diff --git a/tile_engine/ops/gemm_streamk/gemm_streamk_benchmark.hpp b/tile_engine/ops/gemm_streamk/gemm_streamk_benchmark.hpp index 9097dbac78..f92910b40a 100644 --- a/tile_engine/ops/gemm_streamk/gemm_streamk_benchmark.hpp +++ b/tile_engine/ops/gemm_streamk/gemm_streamk_benchmark.hpp @@ -17,10 +17,10 @@ // Data types and Layouts are defined by the generated kernel headers // No hardcoded type definitions here to avoid conflicts +#if __clang_major__ >= 23 #pragma clang diagnostic push -#pragma clang diagnostic ignored "-Wno-unknown-warning-option" #pragma clang diagnostic ignored "-Wlifetime-safety-intra-tu-suggestions" - +#endif enum class Metric { LATENCY = 0, @@ -203,4 +203,6 @@ void gemm_host_reference(int verify, c_m_n_gpu_buf_ref.FromDevice(c_m_n_host_result.data()); } } +#if __clang_major__ >= 23 #pragma clang diagnostic pop +#endif