From 7b1ec41e5bd0a644a74f275149b2d625b3f13982 Mon Sep 17 00:00:00 2001 From: Chao Liu Date: Fri, 6 Aug 2021 20:50:01 +0000 Subject: [PATCH] refactor --- host/driver_offline/CMakeLists.txt | 1 + host/driver_online/CMakeLists.txt | 1 + ...tion_forward_implicit_gemm_v6r1_dlops_nchw_kcyx_nkhw.hpp | 6 +++--- .../include/conv_igemm_fwd_v6r1_dlops_nchw_kcyx_nkhw.hpp | 0 .../include/conv_tunable_fwd_v4r4_dlops_nchw_kcyx_nkhw.hpp | 0 .../include/conv_tunable_fwd_v4r4_xdlops_nchw_kcyx_nkhw.hpp | 0 .../include/conv_tunable_fwd_v4r4_xdlops_nhwc_kyxc_nhwk.hpp | 0 .../include/convolution_problem_descriptor.hpp | 0 8 files changed, 5 insertions(+), 3 deletions(-) rename host/{driver_online => solver}/include/conv_igemm_fwd_v6r1_dlops_nchw_kcyx_nkhw.hpp (100%) rename host/{driver_online => solver}/include/conv_tunable_fwd_v4r4_dlops_nchw_kcyx_nkhw.hpp (100%) rename host/{driver_online => solver}/include/conv_tunable_fwd_v4r4_xdlops_nchw_kcyx_nkhw.hpp (100%) rename host/{driver_online => solver}/include/conv_tunable_fwd_v4r4_xdlops_nhwc_kyxc_nhwk.hpp (100%) rename host/{driver_online => solver}/include/convolution_problem_descriptor.hpp (100%) diff --git a/host/driver_offline/CMakeLists.txt b/host/driver_offline/CMakeLists.txt index 85bd31fbca..927975d449 100644 --- a/host/driver_offline/CMakeLists.txt +++ b/host/driver_offline/CMakeLists.txt @@ -1,6 +1,7 @@ include_directories(BEFORE include ${PROJECT_SOURCE_DIR}/host/host_tensor/include + ${PROJECT_SOURCE_DIR}/host/solver/include ${PROJECT_SOURCE_DIR}/composable_kernel/include ${PROJECT_SOURCE_DIR}/composable_kernel/include/utility ${PROJECT_SOURCE_DIR}/composable_kernel/include/tensor_description diff --git a/host/driver_online/CMakeLists.txt b/host/driver_online/CMakeLists.txt index 152eb270e3..077e3218a0 100644 --- a/host/driver_online/CMakeLists.txt +++ b/host/driver_online/CMakeLists.txt @@ -3,6 +3,7 @@ include_directories(BEFORE ${PROJECT_BINARY_DIR}/host/online_compile/include ${PROJECT_SOURCE_DIR}/host/online_compile/include ${PROJECT_SOURCE_DIR}/host/host_tensor/include + ${PROJECT_SOURCE_DIR}/host/solver/include ${PROJECT_SOURCE_DIR}/composable_kernel/include ${PROJECT_SOURCE_DIR}/composable_kernel/include/utility ${PROJECT_SOURCE_DIR}/composable_kernel/include/tensor_description diff --git a/host/driver_online/include/online_device_dynamic_convolution_forward_implicit_gemm_v6r1_dlops_nchw_kcyx_nkhw.hpp b/host/driver_online/include/online_device_dynamic_convolution_forward_implicit_gemm_v6r1_dlops_nchw_kcyx_nkhw.hpp index 46a9f61acc..92467a7668 100644 --- a/host/driver_online/include/online_device_dynamic_convolution_forward_implicit_gemm_v6r1_dlops_nchw_kcyx_nkhw.hpp +++ b/host/driver_online/include/online_device_dynamic_convolution_forward_implicit_gemm_v6r1_dlops_nchw_kcyx_nkhw.hpp @@ -106,7 +106,7 @@ void online_device_dynamic_convolution_forward_implicit_gemm_v6r1_dlops_nchw_kcy std::vector kernel1_times; std::vector kernel2_times; - for(index_t i = 0; i < nrepeat; ++i) + for(index_t i = 0; i < nrepeat + 1; ++i) { KernelTimer timer1, timer2; std::string kernel_name; @@ -164,11 +164,11 @@ void online_device_dynamic_convolution_forward_implicit_gemm_v6r1_dlops_nchw_kcy auto ave_time1 = std::accumulate( std::next(kernel1_times.begin()), kernel1_times.end(), 0., std::plus{}) / - (nrepeat - 1); + nrepeat; auto ave_time2 = std::accumulate( std::next(kernel2_times.begin()), kernel2_times.end(), 0., std::plus{}) / - (nrepeat - 1); + nrepeat; float perf = (float)(conv_problem_desc.CalculateFlop()) / (std::size_t(1000) * 1000 * 1000) / (ave_time1 + ave_time2); diff --git a/host/driver_online/include/conv_igemm_fwd_v6r1_dlops_nchw_kcyx_nkhw.hpp b/host/solver/include/conv_igemm_fwd_v6r1_dlops_nchw_kcyx_nkhw.hpp similarity index 100% rename from host/driver_online/include/conv_igemm_fwd_v6r1_dlops_nchw_kcyx_nkhw.hpp rename to host/solver/include/conv_igemm_fwd_v6r1_dlops_nchw_kcyx_nkhw.hpp diff --git a/host/driver_online/include/conv_tunable_fwd_v4r4_dlops_nchw_kcyx_nkhw.hpp b/host/solver/include/conv_tunable_fwd_v4r4_dlops_nchw_kcyx_nkhw.hpp similarity index 100% rename from host/driver_online/include/conv_tunable_fwd_v4r4_dlops_nchw_kcyx_nkhw.hpp rename to host/solver/include/conv_tunable_fwd_v4r4_dlops_nchw_kcyx_nkhw.hpp diff --git a/host/driver_online/include/conv_tunable_fwd_v4r4_xdlops_nchw_kcyx_nkhw.hpp b/host/solver/include/conv_tunable_fwd_v4r4_xdlops_nchw_kcyx_nkhw.hpp similarity index 100% rename from host/driver_online/include/conv_tunable_fwd_v4r4_xdlops_nchw_kcyx_nkhw.hpp rename to host/solver/include/conv_tunable_fwd_v4r4_xdlops_nchw_kcyx_nkhw.hpp diff --git a/host/driver_online/include/conv_tunable_fwd_v4r4_xdlops_nhwc_kyxc_nhwk.hpp b/host/solver/include/conv_tunable_fwd_v4r4_xdlops_nhwc_kyxc_nhwk.hpp similarity index 100% rename from host/driver_online/include/conv_tunable_fwd_v4r4_xdlops_nhwc_kyxc_nhwk.hpp rename to host/solver/include/conv_tunable_fwd_v4r4_xdlops_nhwc_kyxc_nhwk.hpp diff --git a/host/driver_online/include/convolution_problem_descriptor.hpp b/host/solver/include/convolution_problem_descriptor.hpp similarity index 100% rename from host/driver_online/include/convolution_problem_descriptor.hpp rename to host/solver/include/convolution_problem_descriptor.hpp