mirror of
https://github.com/ROCm/composable_kernel.git
synced 2026-04-19 22:39:03 +00:00
[rocm-libraries] ROCm/rocm-libraries#5101 (commit d4754fa)
[CK] Remove log spam for deprecated convolutions ## Motivation The `CK_BUILD_DEPRECATED` flag guards legacy non-grouped convolution instances, but both branches of every guard emit a `#pragma` message on every build, adding noise without actionable information. According to some recent testing, these non-grouped instances can outperform their grouped replacements in certain configurations, so their continued availability behind the flag remains valuable. This change removes only the warning directives while preserving all guards and guarded code paths. ## Technical Details Removed all `#pragma` message lines referencing deprecated instances from 25 convolution instance source files spanning conv1d_bwd_data, conv2d_fwd, conv2d_bwd_data, conv2d_fwd_bias_relu, conv2d_fwd_bias_relu_add, conv3d_bwd_data, grouped_conv3d_fwd, grouped_conv3d_bwd_data, and grouped_conv3d_bwd_weight. The `#if CK_BUILD_DEPRECATED` / `#else` / `#endif` preprocessor guards and all guarded code remain unchanged. ## Test Plan No functional change. The CK_BUILD_DEPRECATED conditional logic is unmodified; only #pragma message directives were removed. ## Submission Checklist - [x] Look over the contributing guidelines at https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests. Co-authored-by: Claude Opus 4.6 <noreply@anthropic.com>
This commit is contained in:
committed by
assistant-librarian[bot]
parent
03ce21ddcb
commit
7ee6c44387
@@ -91,13 +91,11 @@ void add_device_conv1d_bwd_data_xdl_nwc_kxc_nwk_bf16_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are deprecated. They may be removed in a future release."
|
||||
add_device_operation_instances(instances,
|
||||
device_conv1d_bwd_data_xdl_nwc_kxc_nwk_bf16_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv1d_bwd_data_xdl_nwc_kxc_nwk_1x1_s1_p0_bf16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -84,13 +84,11 @@ void add_device_conv1d_bwd_data_xdl_nwc_kxc_nwk_f16_instances(
|
||||
instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv1d_bwd_data_xdl_nwc_kxc_nwk_f16_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv1d_bwd_data_xdl_nwc_kxc_nwk_1x1_s1_p0_f16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -83,13 +83,11 @@ void add_device_conv1d_bwd_data_xdl_nwc_kxc_nwk_f32_instances(
|
||||
instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv1d_bwd_data_xdl_nwc_kxc_nwk_f32_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv1d_bwd_data_xdl_nwc_kxc_nwk_1x1_s1_p0_f32_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -88,13 +88,11 @@ void add_device_conv1d_bwd_data_xdl_nwc_kxc_nwk_int8_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv1d_bwd_data_xdl_nwc_kxc_nwk_int8_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv1d_bwd_data_xdl_nwc_kxc_nwk_1x1_s1_p0_int8_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -72,13 +72,11 @@ void add_device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_f16_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_f16_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_1x1_s1_p0_f16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -72,13 +72,11 @@ void add_device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_f32_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_f32_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_1x1_s1_p0_f32_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -72,13 +72,11 @@ void add_device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_int8_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_int8_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv2d_bwd_data_dl_nhwc_kyxc_nhwk_1x1_s1_p0_int8_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -141,7 +141,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_bf16_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_bf16_instances{});
|
||||
add_device_operation_instances(
|
||||
@@ -152,7 +151,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_bf16_instances(
|
||||
instances,
|
||||
device_conv_dedidecate_2d_bwd_data_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_bf16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -143,7 +143,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_f16_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_f16_instances{});
|
||||
add_device_operation_instances(
|
||||
@@ -153,7 +152,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_f16_instances(
|
||||
add_device_operation_instances(
|
||||
instances, device_conv_dedicated_2d_bwd_data_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_f16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -140,7 +140,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_f32_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_f32_instances{});
|
||||
add_device_operation_instances(
|
||||
@@ -150,7 +149,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_f32_instances(
|
||||
add_device_operation_instances(
|
||||
instances, device_conv_dedicated_2d_bwd_data_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_f32_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -137,7 +137,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_int8_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_int8_instances{});
|
||||
add_device_operation_instances(
|
||||
@@ -147,7 +146,6 @@ void add_device_conv2d_bwd_data_xdl_nhwc_kyxc_nhwk_int8_instances(
|
||||
add_device_operation_instances(
|
||||
instances, device_conv_dedicated_2d_bwd_data_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_int8_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -181,7 +181,6 @@ void add_device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk_f16_instances(
|
||||
instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk_f16_instances{});
|
||||
add_device_operation_instances(
|
||||
@@ -203,7 +202,6 @@ void add_device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk_f16_instances(
|
||||
instances, device_conv2d_fwd_xdl_c_shuffle_nhwc_kyxc_nhwk_odd_c_f16_instances_2x{});
|
||||
}
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -115,7 +115,6 @@ void add_device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_bf16_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_bf16_instances{});
|
||||
add_device_operation_instances(instances,
|
||||
@@ -123,7 +122,6 @@ void add_device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_bf16_instances(
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_bf16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -108,14 +108,12 @@ void add_device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_f16_instances(
|
||||
instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances, device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_f16_instances{});
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_1x1_p0_f16_instances{});
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_f16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -107,14 +107,12 @@ void add_device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_f32_instances(
|
||||
instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances, device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_f32_instances{});
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_1x1_p0_f32_instances{});
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_f32_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -112,7 +112,6 @@ void add_device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_int8_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_int8_instances{});
|
||||
add_device_operation_instances(instances,
|
||||
@@ -120,7 +119,6 @@ void add_device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_int8_instances(
|
||||
add_device_operation_instances(instances,
|
||||
device_conv2d_fwd_xdl_nhwc_kyxc_nhwk_1x1_s1_p0_int8_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -180,7 +180,6 @@ void add_device_conv2d_fwd_xdl_c_shuffle_bias_relu_nhwc_kyxc_nhwk_f16_instances(
|
||||
std::vector<DeviceConvFwdBiasActivationPtr<PassThrough, PassThrough, AddRelu>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(
|
||||
instances, device_conv2d_fwd_xdl_c_shuffle_bias_relu_nhwc_kyxc_nhwk_f16_instances{});
|
||||
add_device_operation_instances(
|
||||
@@ -206,7 +205,6 @@ void add_device_conv2d_fwd_xdl_c_shuffle_bias_relu_nhwc_kyxc_nhwk_f16_instances(
|
||||
device_conv2d_fwd_xdl_c_shuffle_bias_relu_nhwc_kyxc_nhwk_odd_c_f16_instances_2x{});
|
||||
}
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -178,7 +178,6 @@ void add_device_conv2d_fwd_xdl_c_shuffle_bias_relu_add_nhwc_kyxc_nhwk_f16_instan
|
||||
std::vector<DeviceConvFwdBiasActivationAddPtr<PassThrough, PassThrough, AddReluAdd>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(
|
||||
instances, device_conv2d_fwd_xdl_c_shuffle_bias_relu_add_nhwc_kyxc_nhwk_f16_instances{});
|
||||
add_device_operation_instances(
|
||||
@@ -207,7 +206,6 @@ void add_device_conv2d_fwd_xdl_c_shuffle_bias_relu_add_nhwc_kyxc_nhwk_f16_instan
|
||||
device_conv2d_fwd_xdl_c_shuffle_bias_relu_add_nhwc_kyxc_nhwk_odd_c_f16_instances_2x{});
|
||||
}
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -91,13 +91,11 @@ void add_device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_bf16_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_bf16_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_1x1_s1_p0_bf16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -91,13 +91,11 @@ void add_device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_f16_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_f16_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_1x1_s1_p0_f16_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -90,13 +90,11 @@ void add_device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_f32_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_f32_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_1x1_s1_p0_f32_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -88,13 +88,11 @@ void add_device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_int8_instances(
|
||||
PassThrough>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(instances,
|
||||
device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_int8_instances{});
|
||||
add_device_operation_instances(
|
||||
instances, device_conv3d_bwd_data_xdl_ndhwc_kzyxc_ndhwk_1x1_s1_p0_int8_instances{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -26,7 +26,6 @@ void add_device_grouped_conv3d_bwd_data_xdl_ndhwgk_gkzyxc_ndhwgc_input_f16_comp_
|
||||
F8>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
// 1. Default
|
||||
add_device_operation_instances(
|
||||
instances,
|
||||
@@ -46,7 +45,6 @@ void add_device_grouped_conv3d_bwd_data_xdl_ndhwgk_gkzyxc_ndhwgc_input_f16_comp_
|
||||
NDHWGC,
|
||||
ConvBwdDataFilter1x1Stride1Pad0>{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -24,7 +24,6 @@ void add_device_grouped_conv3d_bwd_weight_xdl_ndhwgc_gkzyxc_ndhwgk_f16_comp_bf8_
|
||||
F8>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
// 1. Default
|
||||
add_device_operation_instances(
|
||||
instances,
|
||||
@@ -44,7 +43,6 @@ void add_device_grouped_conv3d_bwd_weight_xdl_ndhwgc_gkzyxc_ndhwgk_f16_comp_bf8_
|
||||
NDHWGK,
|
||||
ConvBwdWeightFilter1x1Stride1Pad0>{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -25,7 +25,6 @@ void add_device_grouped_conv3d_fwd_xdl_ndhwgc_gkzyxc_ndhwgk_f16_comp_f8_instance
|
||||
F8>>>& instances)
|
||||
{
|
||||
#if CK_BUILD_DEPRECATED
|
||||
#pragma message "These instances are getting deprecated"
|
||||
add_device_operation_instances(
|
||||
instances,
|
||||
device_grouped_conv_fwd_xdl_f16_comp_f8_instances<3,
|
||||
@@ -51,7 +50,6 @@ void add_device_grouped_conv3d_fwd_xdl_ndhwgc_gkzyxc_ndhwgk_f16_comp_f8_instance
|
||||
NDHWGK,
|
||||
ConvFwd1x1S1P0>{});
|
||||
#else
|
||||
#pragma message "These instances were deprecated"
|
||||
std::ignore = instances;
|
||||
#endif
|
||||
}
|
||||
|
||||
Reference in New Issue
Block a user