Skip to content

GitLab

"...composable_kernel-1.git" did not exist on "4a1e97cf865fbf9fee3f02164e54c1d227299334"
Commit f5b25095 authored by Bartlomiej Kocot's avatar Bartlomiej Kocot
Browse files

Revert "Disable add instances functions for disabled data types" 

This reverts commit 728b8695.
parent 98f349c6
Hide whitespace changes
Inline Side-by-side
Showing with 112 additions and 64 deletions
library/src/tensor_operation_instance/gpu/column_to_image/device_column_to_image_nhwc_1d_instance.cpp
View file @ f5b25095
...@@ -11,41 +11,49 @@ namespace instance { ...@@ -11,41 +11,49 @@ namespace instance {
using namespace ck::conv_tensor_rearrange_op; using namespace ck::conv_tensor_rearrange_op;
#ifdef CK_ENABLE_BF16 void add_device_column_to_image_nhwc_1d_bf16_instances(
void add_device_column_to_image_nwc_1d_bf16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, BF16, BF16, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, BF16, BF16, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_BF16
add_device_operation_instances(instances, device_column_to_image_bf16_instances<1, GNWC>{}); add_device_operation_instances(instances, device_column_to_image_bf16_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP16 void add_device_column_to_image_nhwc_1d_f16_instances(
void add_device_column_to_image_nwc_1d_f16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F16, F16, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F16, F16, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP16
add_device_operation_instances(instances, device_column_to_image_f16_instances<1, GNWC>{}); add_device_operation_instances(instances, device_column_to_image_f16_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP32 void add_device_column_to_image_nhwc_1d_f32_instances(
void add_device_column_to_image_nwc_1d_f32_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F32, F32, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F32, F32, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP32
add_device_operation_instances(instances, device_column_to_image_f32_instances<1, GNWC>{}); add_device_operation_instances(instances, device_column_to_image_f32_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_INT8 void add_device_column_to_image_nhwc_1d_i8_instances(
void add_device_column_to_image_nwc_1d_i8_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, int8_t, int8_t, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, int8_t, int8_t, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_INT8
add_device_operation_instances(instances, device_column_to_image_i8_instances<1, GNWC>{}); add_device_operation_instances(instances, device_column_to_image_i8_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
} // namespace instance } // namespace instance
} // namespace device } // namespace device
......
library/src/tensor_operation_instance/gpu/column_to_image/device_column_to_image_nhwc_2d_instance.cpp
View file @ f5b25095
...@@ -11,42 +11,50 @@ namespace instance { ...@@ -11,42 +11,50 @@ namespace instance {
using namespace ck::conv_tensor_rearrange_op; using namespace ck::conv_tensor_rearrange_op;
#ifdef CK_ENABLE_BF16
void add_device_column_to_image_nhwc_2d_bf16_instances( void add_device_column_to_image_nhwc_2d_bf16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, BF16, BF16, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, BF16, BF16, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_BF16
add_device_operation_instances(instances, device_column_to_image_bf16_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_column_to_image_bf16_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP16
void add_device_column_to_image_nhwc_2d_f16_instances( void add_device_column_to_image_nhwc_2d_f16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F16, F16, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F16, F16, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP16
add_device_operation_instances(instances, device_column_to_image_f16_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_column_to_image_f16_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP32
void add_device_column_to_image_nhwc_2d_f32_instances( void add_device_column_to_image_nhwc_2d_f32_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F32, F32, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F32, F32, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP32
add_device_operation_instances(instances, device_column_to_image_f32_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_column_to_image_f32_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_INT8
void add_device_column_to_image_nhwc_2d_i8_instances( void add_device_column_to_image_nhwc_2d_i8_instances(
std::vector< std::vector<
std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, int8_t, int8_t, ColumnToImage>>>& std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, int8_t, int8_t, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_INT8
add_device_operation_instances(instances, device_column_to_image_i8_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_column_to_image_i8_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
} // namespace instance } // namespace instance
} // namespace device } // namespace device
......
library/src/tensor_operation_instance/gpu/column_to_image/device_column_to_image_nhwc_3d_instance.cpp
View file @ f5b25095
...@@ -11,42 +11,50 @@ namespace instance { ...@@ -11,42 +11,50 @@ namespace instance {
using namespace ck::conv_tensor_rearrange_op; using namespace ck::conv_tensor_rearrange_op;
#ifdef CK_ENABLE_BF16 void add_device_column_to_image_nhwc_3d_bf16_instances(
void add_device_column_to_image_ndhwc_3d_bf16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, BF16, BF16, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, BF16, BF16, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_BF16
add_device_operation_instances(instances, device_column_to_image_bf16_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_column_to_image_bf16_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP16 void add_device_column_to_image_nhwc_3d_f16_instances(
void add_device_column_to_image_ndhwc_3d_f16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F16, F16, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F16, F16, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP16
add_device_operation_instances(instances, device_column_to_image_f16_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_column_to_image_f16_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP32 void add_device_column_to_image_nhwc_3d_f32_instances(
void add_device_column_to_image_ndhwc_3d_f32_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F32, F32, ColumnToImage>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F32, F32, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP32
add_device_operation_instances(instances, device_column_to_image_f32_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_column_to_image_f32_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_INT8 void add_device_column_to_image_nhwc_3d_i8_instances(
void add_device_column_to_image_ndhwc_3d_i8_instances(
std::vector< std::vector<
std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, int8_t, int8_t, ColumnToImage>>>& std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, int8_t, int8_t, ColumnToImage>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_INT8
add_device_operation_instances(instances, device_column_to_image_i8_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_column_to_image_i8_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
} // namespace instance } // namespace instance
} // namespace device } // namespace device
......
library/src/tensor_operation_instance/gpu/image_to_column/device_image_to_column_nhwc_1d_instance.cpp
View file @ f5b25095
...@@ -11,41 +11,49 @@ namespace instance { ...@@ -11,41 +11,49 @@ namespace instance {
using namespace ck::conv_tensor_rearrange_op; using namespace ck::conv_tensor_rearrange_op;
#ifdef CK_ENABLE_BF16 void add_device_image_to_column_nhwc_1d_bf16_instances(
void add_device_image_to_column_nwc_1d_bf16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, BF16, BF16, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, BF16, BF16, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_BF16
add_device_operation_instances(instances, device_image_to_column_bf16_instances<1, GNWC>{}); add_device_operation_instances(instances, device_image_to_column_bf16_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP16 void add_device_image_to_column_nhwc_1d_f16_instances(
void add_device_image_to_column_nwc_1d_f16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F16, F16, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F16, F16, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP16
add_device_operation_instances(instances, device_image_to_column_f16_instances<1, GNWC>{}); add_device_operation_instances(instances, device_image_to_column_f16_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP32 void add_device_image_to_column_nhwc_1d_f32_instances(
void add_device_image_to_column_nwc_1d_f32_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F32, F32, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, F32, F32, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP32
add_device_operation_instances(instances, device_image_to_column_f32_instances<1, GNWC>{}); add_device_operation_instances(instances, device_image_to_column_f32_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_INT8 void add_device_image_to_column_nhwc_1d_i8_instances(
void add_device_image_to_column_nwc_1d_i8_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, int8_t, int8_t, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<1, GNWC, int8_t, int8_t, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_INT8
add_device_operation_instances(instances, device_image_to_column_i8_instances<1, GNWC>{}); add_device_operation_instances(instances, device_image_to_column_i8_instances<1, GNWC>{});
} #else
ignore = instances;
#endif #endif
}
} // namespace instance } // namespace instance
} // namespace device } // namespace device
......
library/src/tensor_operation_instance/gpu/image_to_column/device_image_to_column_nhwc_2d_instance.cpp
View file @ f5b25095
...@@ -11,42 +11,50 @@ namespace instance { ...@@ -11,42 +11,50 @@ namespace instance {
using namespace ck::conv_tensor_rearrange_op; using namespace ck::conv_tensor_rearrange_op;
#ifdef CK_ENABLE_BF16
void add_device_image_to_column_nhwc_2d_bf16_instances( void add_device_image_to_column_nhwc_2d_bf16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, BF16, BF16, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, BF16, BF16, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_BF16
add_device_operation_instances(instances, device_image_to_column_bf16_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_image_to_column_bf16_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP16
void add_device_image_to_column_nhwc_2d_f16_instances( void add_device_image_to_column_nhwc_2d_f16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F16, F16, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F16, F16, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP16
add_device_operation_instances(instances, device_image_to_column_f16_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_image_to_column_f16_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP32
void add_device_image_to_column_nhwc_2d_f32_instances( void add_device_image_to_column_nhwc_2d_f32_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F32, F32, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, F32, F32, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP32
add_device_operation_instances(instances, device_image_to_column_f32_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_image_to_column_f32_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_INT8
void add_device_image_to_column_nhwc_2d_i8_instances( void add_device_image_to_column_nhwc_2d_i8_instances(
std::vector< std::vector<
std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, int8_t, int8_t, ImageToColumn>>>& std::unique_ptr<DeviceConvTensorRearrange<2, GNHWC, int8_t, int8_t, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_INT8
add_device_operation_instances(instances, device_image_to_column_i8_instances<2, GNHWC>{}); add_device_operation_instances(instances, device_image_to_column_i8_instances<2, GNHWC>{});
} #else
ignore = instances;
#endif #endif
}
} // namespace instance } // namespace instance
} // namespace device } // namespace device
......
library/src/tensor_operation_instance/gpu/image_to_column/device_image_to_column_nhwc_3d_instance.cpp
View file @ f5b25095
...@@ -11,42 +11,50 @@ namespace instance { ...@@ -11,42 +11,50 @@ namespace instance {
using namespace ck::conv_tensor_rearrange_op; using namespace ck::conv_tensor_rearrange_op;
#ifdef CK_ENABLE_BF16 void add_device_image_to_column_nhwc_3d_bf16_instances(
void add_device_image_to_column_ndhwc_3d_bf16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, BF16, BF16, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, BF16, BF16, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_BF16
add_device_operation_instances(instances, device_image_to_column_bf16_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_image_to_column_bf16_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP16 void add_device_image_to_column_nhwc_3d_f16_instances(
void add_device_image_to_column_ndhwc_3d_f16_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F16, F16, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F16, F16, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP16
add_device_operation_instances(instances, device_image_to_column_f16_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_image_to_column_f16_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_FP32 void add_device_image_to_column_nhwc_3d_f32_instances(
void add_device_image_to_column_ndhwc_3d_f32_instances(
std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F32, F32, ImageToColumn>>>& std::vector<std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, F32, F32, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_FP32
add_device_operation_instances(instances, device_image_to_column_f32_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_image_to_column_f32_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
#ifdef CK_ENABLE_INT8 void add_device_image_to_column_nhwc_3d_i8_instances(
void add_device_image_to_column_ndhwc_3d_i8_instances(
std::vector< std::vector<
std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, int8_t, int8_t, ImageToColumn>>>& std::unique_ptr<DeviceConvTensorRearrange<3, GNDHWC, int8_t, int8_t, ImageToColumn>>>&
instances) instances)
{ {
#ifdef CK_ENABLE_INT8
add_device_operation_instances(instances, device_image_to_column_i8_instances<3, GNDHWC>{}); add_device_operation_instances(instances, device_image_to_column_i8_instances<3, GNDHWC>{});
} #else
ignore = instances;
#endif #endif
}
} // namespace instance } // namespace instance
} // namespace device } // namespace device
......
Markdown is supported
0% or .
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment