From f4c11b6d22ec1e55ef8f74004fe1becb622151f0 Mon Sep 17 00:00:00 2001 From: John Shumway Date: Sun, 14 Sep 2025 14:16:40 +0000 Subject: [PATCH] Provide default value for API VERSION. Create a LATEST_API_VERSION value and use that as the default. This simplifies code that shouldn't depend the API version. --- .../builder/include/ck_tile/builder/conv_builder.hpp | 4 ++-- .../builder/include/ck_tile/builder/conv_factory.hpp | 2 +- .../ck_tile/builder/{versions.h => versions.hpp} | 2 ++ experimental/builder/test/test_conv_builder.cpp | 5 +---- experimental/builder/test/test_conv_grp_fwd_2d.cpp | 11 ++++------- experimental/builder/test/test_conv_grp_fwd_3d.cpp | 5 +---- experimental/builder/test/test_conv_instances.cpp | 5 +---- 7 files changed, 12 insertions(+), 22 deletions(-) rename experimental/builder/include/ck_tile/builder/{versions.h => versions.hpp} (85%) diff --git a/experimental/builder/include/ck_tile/builder/conv_builder.hpp b/experimental/builder/include/ck_tile/builder/conv_builder.hpp index 1b7a0c2130..910192c341 100644 --- a/experimental/builder/include/ck_tile/builder/conv_builder.hpp +++ b/experimental/builder/include/ck_tile/builder/conv_builder.hpp @@ -6,7 +6,7 @@ #include #include #include -#include +#include namespace ck_tile::builder { @@ -24,7 +24,7 @@ namespace ck_tile::builder { */ template + StringLiteral VERSION = LATEST_API_VERSION> requires SupportedVersion && ValidConvSignature struct ConvBuilder { diff --git a/experimental/builder/include/ck_tile/builder/conv_factory.hpp b/experimental/builder/include/ck_tile/builder/conv_factory.hpp index 7f683273e8..5efca46807 100644 --- a/experimental/builder/include/ck_tile/builder/conv_factory.hpp +++ b/experimental/builder/include/ck_tile/builder/conv_factory.hpp @@ -7,7 +7,7 @@ #include #include #include -#include +#include namespace ck_tile::builder { diff --git a/experimental/builder/include/ck_tile/builder/versions.h b/experimental/builder/include/ck_tile/builder/versions.hpp similarity index 85% rename from experimental/builder/include/ck_tile/builder/versions.h rename to experimental/builder/include/ck_tile/builder/versions.hpp index c14003502c..c7543f11ae 100644 --- a/experimental/builder/include/ck_tile/builder/versions.h +++ b/experimental/builder/include/ck_tile/builder/versions.hpp @@ -10,6 +10,8 @@ namespace ck_tile::builder { static constexpr StringLiteral V0_0_0 = "0.0.0"; static constexpr StringLiteral V0_1_0 = "0.1.0"; +static constexpr StringLiteral LATEST_API_VERSION = V0_1_0; + template concept SupportedVersion = (V == V0_0_0) || (V == V0_1_0); diff --git a/experimental/builder/test/test_conv_builder.cpp b/experimental/builder/test/test_conv_builder.cpp index 38612e53f5..9d72eca7e2 100644 --- a/experimental/builder/test/test_conv_builder.cpp +++ b/experimental/builder/test/test_conv_builder.cpp @@ -23,14 +23,11 @@ struct DefaultAlgorithm }; static_assert(ckb::ConvAlgorithmDescriptor); -constexpr char API_VERSION[] = "0.1.0"; -static_assert(ckb::SupportedVersion); - TEST(ConvBuilderTest, TestDefaultInstance) { static constexpr const ConvSignature SIGNATURE; static constexpr const DefaultAlgorithm ALGORITHM; - using Builder = ckb::ConvBuilder; + using Builder = ckb::ConvBuilder; EXPECT_EQ( Builder::Instance::TypeString(), "DeviceGroupedConvFwdMultipleABD_Xdl_CShuffle_V3<256, 256, 256, 32, Default, 32, 32, 4, 4, " diff --git a/experimental/builder/test/test_conv_grp_fwd_2d.cpp b/experimental/builder/test/test_conv_grp_fwd_2d.cpp index a27e2c3dc6..591abba130 100644 --- a/experimental/builder/test/test_conv_grp_fwd_2d.cpp +++ b/experimental/builder/test/test_conv_grp_fwd_2d.cpp @@ -23,14 +23,11 @@ struct DefaultAlgorithm }; static_assert(ckb::ConvAlgorithmDescriptor); -constexpr char API_VERSION[] = "0.1.0"; -static_assert(ckb::SupportedVersion); - TEST(ConvBuilderGrpFwd2d, TestDefaultInstance) { static constexpr const ConvSignature SIGNATURE; static constexpr const DefaultAlgorithm ALGORITHM; - using Builder = ckb::ConvBuilder; + using Builder = ckb::ConvBuilder; EXPECT_EQ( Builder::Instance::TypeString(), "DeviceGroupedConvFwdMultipleABD_Xdl_CShuffle_V3<256, 256, 256, 32, Default, 32, 32, 4, 4, " @@ -41,7 +38,7 @@ TEST(ConvBuilderGrpFwd2d, TestDefaultFP32Instance) { static constexpr const ConvSignature SIGNATURE{.data_type = ckb::DataType::FP32}; static constexpr const DefaultAlgorithm ALGORITHM; - using Builder = ckb::ConvBuilder; + using Builder = ckb::ConvBuilder; EXPECT_NE(Builder::Instance::TypeString(), ""); // It's difficult to check the types direction on the kernel, so we instead // check that the builder has the correct data type aliases for FP32. @@ -56,7 +53,7 @@ TEST(ConvBuilderGrpFwd2d, TestDefaultFP16Instance) { static constexpr const ConvSignature SIGNATURE{.data_type = ckb::DataType::FP16}; static constexpr const DefaultAlgorithm ALGORITHM; - using Builder = ckb::ConvBuilder; + using Builder = ckb::ConvBuilder; // Check that the builder has the correct data type aliases for FP16. EXPECT_TRUE((std::is_same_v)); @@ -70,7 +67,7 @@ TEST(ConvBuilderGrpFwd2d, TestDefaultBF16Instance) { static constexpr const ConvSignature SIGNATURE{.data_type = ckb::DataType::BF16}; static constexpr const DefaultAlgorithm ALGORITHM; - using Builder = ckb::ConvBuilder; + using Builder = ckb::ConvBuilder; // Check that the builder has the correct data type aliases for BF16. EXPECT_TRUE((std::is_same_v)); diff --git a/experimental/builder/test/test_conv_grp_fwd_3d.cpp b/experimental/builder/test/test_conv_grp_fwd_3d.cpp index c29b3f46cf..322b3f0ee5 100644 --- a/experimental/builder/test/test_conv_grp_fwd_3d.cpp +++ b/experimental/builder/test/test_conv_grp_fwd_3d.cpp @@ -23,14 +23,11 @@ struct DefaultAlgorithm }; static_assert(ckb::ConvAlgorithmDescriptor); -constexpr char API_VERSION[] = "0.1.0"; -static_assert(ckb::SupportedVersion); - TEST(ConvBuilderGrpFwd3d, TestDefaultInstance) { static constexpr const ConvSignature SIGNATURE; static constexpr const DefaultAlgorithm ALGORITHM; - using Builder = ckb::ConvBuilder; + using Builder = ckb::ConvBuilder; EXPECT_EQ( Builder::Instance::TypeString(), "DeviceGroupedConvFwdMultipleABD_Xdl_CShuffle_V3<256, 256, 256, 32, Default, 32, 32, 4, 4, " diff --git a/experimental/builder/test/test_conv_instances.cpp b/experimental/builder/test/test_conv_instances.cpp index e0aea59116..b86e33fbae 100644 --- a/experimental/builder/test/test_conv_instances.cpp +++ b/experimental/builder/test/test_conv_instances.cpp @@ -24,9 +24,6 @@ struct ConvSignature }; static_assert(ckb::ConvSignatureDescriptor); -constexpr char API_VERSION[] = "0.1.0"; -static_assert(ckb::SupportedVersion); - // Defines the tunable algorithmic parameters for the convolution kernel. // This includes thread block configuration, tuning parameters, data transfer // settings, and the GEMM pipeline version. @@ -259,7 +256,7 @@ TYPED_TEST(ConvBuilderInstancesTest, KernelParamsConfigured) static constexpr const FwdConvAlgorithm& ALGORITHM = ConvBuilderInstancesTest::ALGORITHM; static constexpr const ConvSignature SIGNATURE; - using Builder = ckb::ConvBuilder; + using Builder = ckb::ConvBuilder; EXPECT_EQ(Builder::Instance::TypeString(), ConvBuilderInstancesTest::EXPECTED_TYPE); const auto& tp = ALGORITHM.tuning_params; EXPECT_EQ(Builder::Factory::TUNING.ak1, tp.ak1);