From cd14c410ee2f577eddc05af72e5049943ec68b06 Mon Sep 17 00:00:00 2001 From: Koichi Yoshigoe Date: Thu, 12 Mar 2020 11:19:08 +0900 Subject: [PATCH 01/18] Update dependency packages (#910) --- setup.cfg | 20 ++++++++++---------- 1 file changed, 10 insertions(+), 10 deletions(-) diff --git a/setup.cfg b/setup.cfg index 969bc73c4..5f02bd158 100644 --- a/setup.cfg +++ b/setup.cfg @@ -14,15 +14,15 @@ include_package_data = True packages = find: install_requires = jinja2 - click==6.7 - easydict==1.6 - matplotlib==2.2.2 - numpy==1.17.4 + click==7.0 + easydict==1.9 + matplotlib==3.1.1 + numpy==1.18.1 pandas==1.0.1 Pillow==6.2.1 prompt_toolkit==3.0.3 pytablewriter==0.47.0 - PyYAML==4.2b4 + PyYAML==5.3 tensorflow-datasets==1.0.2 inquirer==2.6.3 cython==0.27.3 @@ -36,15 +36,15 @@ dist = horovod==0.18.1 mpi4py==3.0.2 docs = - better-apidoc==0.1.2 + better-apidoc==0.3.1 sphinx-rtd-theme==0.2.4 - Sphinx==1.7.2 + Sphinx==2.4.4 sphinx-autobuild==0.7.1 test = flake8==3.5.0 - pytest==5.2.0 - pytest-xdist==1.30.0 - parameterized==0.7.0 + pytest==5.3.5 + pytest-xdist==1.31.0 + parameterized==0.7.1 validate = pycodestyle autopep8 From f77b0daa09990f722a16d1cb7e5fbb30cc557e7c Mon Sep 17 00:00:00 2001 From: prime number Date: Thu, 12 Mar 2020 13:40:55 +0900 Subject: [PATCH 02/18] Include types.h instead of global.h (#891) --- blueoil/converter/templates/include/dlk_test.h | 2 +- blueoil/converter/templates/include/func/add.h | 1 - .../converter/templates/include/func/average_pool.h | 2 +- .../templates/include/func/batch_normalization.h | 2 +- .../templates/include/func/concat_on_depth.h | 2 +- .../templates/include/func/depth_to_space.h | 12 +++++++----- .../converter/templates/include/func/leaky_relu.h | 1 - blueoil/converter/templates/include/func/matmul.h | 2 +- blueoil/converter/templates/include/func/max.h | 1 - blueoil/converter/templates/include/func/max_pool.h | 2 +- blueoil/converter/templates/include/func/minimum.h | 1 - blueoil/converter/templates/include/func/mul.h | 1 - blueoil/converter/templates/include/func/pad.h | 2 +- blueoil/converter/templates/include/func/real_div.h | 1 - blueoil/converter/templates/include/func/relu.h | 1 - .../templates/include/func/resize_nearest_neighbor.h | 12 +++++++----- blueoil/converter/templates/include/func/softmax.h | 2 +- blueoil/converter/templates/include/func/sub.h | 1 - blueoil/converter/templates/include/types.h | 1 + .../src/func/arm_neon/batch_normalization.cpp | 2 +- .../src/func/generic/batch_normalization.cpp | 2 +- blueoil/converter/templates/src/func/matmul.cpp | 2 +- blueoil/converter/templates/src/func/max_pool.cpp | 2 +- blueoil/converter/templates/src/func/pad.cpp | 2 +- blueoil/converter/templates/src/func/softmax.cpp | 2 +- .../src/func/x86_avx/batch_normalization.cpp | 2 +- blueoil/converter/templates/src/write_to_file.cpp | 2 +- 27 files changed, 31 insertions(+), 34 deletions(-) diff --git a/blueoil/converter/templates/include/dlk_test.h b/blueoil/converter/templates/include/dlk_test.h index d09895b90..79f8d5751 100644 --- a/blueoil/converter/templates/include/dlk_test.h +++ b/blueoil/converter/templates/include/dlk_test.h @@ -23,7 +23,7 @@ limitations under the License. #include #include #include -#include "global.h" +#include "types.h" template struct Diff { diff --git a/blueoil/converter/templates/include/func/add.h b/blueoil/converter/templates/include/func/add.h index a64a0fc2e..615a54256 100644 --- a/blueoil/converter/templates/include/func/add.h +++ b/blueoil/converter/templates/include/func/add.h @@ -17,7 +17,6 @@ limitations under the License. #include -#include "global.h" #include "tensor_view.h" #include "func/impl/binary_op.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/include/func/average_pool.h b/blueoil/converter/templates/include/func/average_pool.h index 0b78d7900..e47ed98ad 100644 --- a/blueoil/converter/templates/include/func/average_pool.h +++ b/blueoil/converter/templates/include/func/average_pool.h @@ -16,7 +16,7 @@ limitations under the License. #ifndef DLK_FUNC_AVERAGE_POOL_H_INCLUDED #define DLK_FUNC_AVERAGE_POOL_H_INCLUDED -#include "global.h" +#include "types.h" #include "tensor_view.h" struct avg_pooling_parameters { diff --git a/blueoil/converter/templates/include/func/batch_normalization.h b/blueoil/converter/templates/include/func/batch_normalization.h index 4ad7c23eb..01922047f 100644 --- a/blueoil/converter/templates/include/func/batch_normalization.h +++ b/blueoil/converter/templates/include/func/batch_normalization.h @@ -16,7 +16,7 @@ limitations under the License. #ifndef DLK_FUNC_BATCH_NORMALIZATION_H_INCLUDED #define DLK_FUNC_BATCH_NORMALIZATION_H_INCLUDED -#include "global.h" +#include "types.h" #include "tensor_view.h" void func_BatchNormalizationOptimized(const TensorView& input, diff --git a/blueoil/converter/templates/include/func/concat_on_depth.h b/blueoil/converter/templates/include/func/concat_on_depth.h index fe595ee4f..c75ce58fb 100644 --- a/blueoil/converter/templates/include/func/concat_on_depth.h +++ b/blueoil/converter/templates/include/func/concat_on_depth.h @@ -19,7 +19,7 @@ limitations under the License. #include #include -#include "global.h" +#include "types.h" #include "tensor_view.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/include/func/depth_to_space.h b/blueoil/converter/templates/include/func/depth_to_space.h index ee2da3efc..930a2c3c9 100644 --- a/blueoil/converter/templates/include/func/depth_to_space.h +++ b/blueoil/converter/templates/include/func/depth_to_space.h @@ -16,7 +16,7 @@ limitations under the License. #ifndef DLK_FUNC_DEPTH_TO_SPACE_H_INCLUDED #define DLK_FUNC_DEPTH_TO_SPACE_H_INCLUDED -#include "global.h" +#include "types.h" #include "time_measurement.h" #include "tensor_view.h" @@ -48,8 +48,9 @@ inline void func_DepthToSpace(const TensorView& input Measurement::Stop(); } -inline void func_DepthToSpace(const TensorView& input, - const TensorView& output, +template +void func_DepthToSpace(const TensorView, MemoryLayout::HWChBCl>& input, + const TensorView, MemoryLayout::HWChBCl>& output, T_UINT a, T_UINT b, T_UINT kernel_size, T_UINT stride) { Measurement::Start("DepthToSpace"); @@ -79,8 +80,9 @@ inline void func_DepthToSpace(const TensorView& input, - const TensorView& output, +template +void func_DepthToSpace(const TensorView, MemoryLayout::ChHWBCl>& input, + const TensorView, MemoryLayout::ChHWBCl>& output, T_UINT a, T_UINT b, T_UINT kernel_size, T_UINT stride) { Measurement::Start("DepthToSpace"); diff --git a/blueoil/converter/templates/include/func/leaky_relu.h b/blueoil/converter/templates/include/func/leaky_relu.h index d14431919..769c2171a 100644 --- a/blueoil/converter/templates/include/func/leaky_relu.h +++ b/blueoil/converter/templates/include/func/leaky_relu.h @@ -16,7 +16,6 @@ limitations under the License. #ifndef DLK_FUNC_LEAKY_RELU_H_INCLUDED #define DLK_FUNC_LEAKY_RELU_H_INCLUDED -#include "global.h" #include "tensor_view.h" #include "func/impl/unary_op.h" diff --git a/blueoil/converter/templates/include/func/matmul.h b/blueoil/converter/templates/include/func/matmul.h index a4907099d..fbbeb5435 100644 --- a/blueoil/converter/templates/include/func/matmul.h +++ b/blueoil/converter/templates/include/func/matmul.h @@ -16,7 +16,7 @@ limitations under the License. #ifndef DLK_FUNC_MATMUL_H_INCLUDED #define DLK_FUNC_MATMUL_H_INCLUDED -#include "global.h" +#include "types.h" #include "tensor_view.h" void func_Matmul(const TensorView& input, diff --git a/blueoil/converter/templates/include/func/max.h b/blueoil/converter/templates/include/func/max.h index 6d75d67cc..91db00998 100644 --- a/blueoil/converter/templates/include/func/max.h +++ b/blueoil/converter/templates/include/func/max.h @@ -16,7 +16,6 @@ limitations under the License. #ifndef DLK_FUNC_MAX_H_INCLUDED #define DLK_FUNC_MAX_H_INCLUDED -#include "global.h" #include "tensor_view.h" #include "func/impl/binary_op.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/include/func/max_pool.h b/blueoil/converter/templates/include/func/max_pool.h index c555df6d9..06660b686 100644 --- a/blueoil/converter/templates/include/func/max_pool.h +++ b/blueoil/converter/templates/include/func/max_pool.h @@ -16,7 +16,7 @@ limitations under the License. #ifndef DLK_FUNC_MAX_POOLING_H_INCLUDED #define DLK_FUNC_MAX_POOLING_H_INCLUDED -#include "global.h" +#include "types.h" #include "tensor_view.h" struct max_pooling_parameters { diff --git a/blueoil/converter/templates/include/func/minimum.h b/blueoil/converter/templates/include/func/minimum.h index f32ceebfe..62fe7f51b 100644 --- a/blueoil/converter/templates/include/func/minimum.h +++ b/blueoil/converter/templates/include/func/minimum.h @@ -16,7 +16,6 @@ limitations under the License. #ifndef DLK_FUNC_MINIMUM_H_INCLUDED #define DLK_FUNC_MINIMUM_H_INCLUDED -#include "global.h" #include "tensor_view.h" #include "func/impl/binary_op.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/include/func/mul.h b/blueoil/converter/templates/include/func/mul.h index 178fdc53f..67542d5ef 100644 --- a/blueoil/converter/templates/include/func/mul.h +++ b/blueoil/converter/templates/include/func/mul.h @@ -16,7 +16,6 @@ limitations under the License. #ifndef DLK_FUNC_MUL_H_INCLUDED #define DLK_FUNC_MUL_H_INCLUDED -#include "global.h" #include "tensor_view.h" #include "func/impl/binary_op.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/include/func/pad.h b/blueoil/converter/templates/include/func/pad.h index faeade871..b6f39ff72 100644 --- a/blueoil/converter/templates/include/func/pad.h +++ b/blueoil/converter/templates/include/func/pad.h @@ -1,7 +1,7 @@ #ifndef DLK_FUNC_PAD_H_INCLUDED #define DLK_FUNC_PAD_H_INCLUDED -#include "global.h" +#include "types.h" #include "tensor_view.h" void func_Pad(const TensorView& input, diff --git a/blueoil/converter/templates/include/func/real_div.h b/blueoil/converter/templates/include/func/real_div.h index 5b53385ca..c6dd23233 100644 --- a/blueoil/converter/templates/include/func/real_div.h +++ b/blueoil/converter/templates/include/func/real_div.h @@ -16,7 +16,6 @@ limitations under the License. #ifndef DLK_FUNC_REAL_DIV_H #define DLK_FUNC_REAL_DIV_H -#include "global.h" #include "tensor_view.h" #include "func/impl/binary_op.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/include/func/relu.h b/blueoil/converter/templates/include/func/relu.h index aa03861e4..3d6fddd67 100644 --- a/blueoil/converter/templates/include/func/relu.h +++ b/blueoil/converter/templates/include/func/relu.h @@ -16,7 +16,6 @@ limitations under the License. #ifndef DLK_FUNC_RELU_H_INCLUDED #define DLK_FUNC_RELU_H_INCLUDED -#include "global.h" #include "tensor_view.h" #include "func/impl/unary_op.h" diff --git a/blueoil/converter/templates/include/func/resize_nearest_neighbor.h b/blueoil/converter/templates/include/func/resize_nearest_neighbor.h index 63a549aa7..a9ff1483d 100644 --- a/blueoil/converter/templates/include/func/resize_nearest_neighbor.h +++ b/blueoil/converter/templates/include/func/resize_nearest_neighbor.h @@ -16,7 +16,7 @@ limitations under the License. #ifndef DLK_FUNC_RESIZE_NEAREST_NEIGHBOR_H_INCLUDED #define DLK_FUNC_RESIZE_NEAREST_NEIGHBOR_H_INCLUDED -#include "global.h" +#include "types.h" #include "time_measurement.h" #include "tensor_view.h" @@ -51,8 +51,9 @@ inline void func_ResizeNearestNeighbor(const TensorView& input, - const TensorView& output) { +template +void func_ResizeNearestNeighbor(const TensorView, MemoryLayout::HWChBCl>& input, + const TensorView, MemoryLayout::HWChBCl>& output) { Measurement::Start("ResizeNearestNeighbor"); const auto in_shape = input.get_shape(); @@ -83,8 +84,9 @@ inline void func_ResizeNearestNeighbor(const TensorView& input, - const TensorView& output) { +template +void func_ResizeNearestNeighbor(const TensorView, MemoryLayout::ChHWBCl>& input, + const TensorView, MemoryLayout::ChHWBCl>& output) { Measurement::Start("ResizeNearestNeighbor"); const auto in_shape = input.get_shape(); diff --git a/blueoil/converter/templates/include/func/softmax.h b/blueoil/converter/templates/include/func/softmax.h index a3043e354..47f58bb86 100644 --- a/blueoil/converter/templates/include/func/softmax.h +++ b/blueoil/converter/templates/include/func/softmax.h @@ -16,7 +16,7 @@ limitations under the License. #ifndef DLK_FUNC_SOFTMAX_H_INCLUDED #define DLK_FUNC_SOFTMAX_H_INCLUDED -#include "global.h" +#include "types.h" #include "tensor_view.h" void func_Softmax(const TensorView& input, diff --git a/blueoil/converter/templates/include/func/sub.h b/blueoil/converter/templates/include/func/sub.h index f0c480d46..25ae662a1 100644 --- a/blueoil/converter/templates/include/func/sub.h +++ b/blueoil/converter/templates/include/func/sub.h @@ -16,7 +16,6 @@ limitations under the License. #ifndef DLK_FUNC_SUB_H_INCLUDED #define DLK_FUNC_SUB_H_INCLUDED -#include "global.h" #include "tensor_view.h" #include "func/impl/binary_op.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/include/types.h b/blueoil/converter/templates/include/types.h index 035a05256..8ada4151c 100644 --- a/blueoil/converter/templates/include/types.h +++ b/blueoil/converter/templates/include/types.h @@ -16,6 +16,7 @@ limitations under the License. #ifndef TYPES_H #define TYPES_H +#include #include #include "func/impl/pop_count.h" diff --git a/blueoil/converter/templates/src/func/arm_neon/batch_normalization.cpp b/blueoil/converter/templates/src/func/arm_neon/batch_normalization.cpp index 79d49c7cc..db64292db 100644 --- a/blueoil/converter/templates/src/func/arm_neon/batch_normalization.cpp +++ b/blueoil/converter/templates/src/func/arm_neon/batch_normalization.cpp @@ -16,7 +16,7 @@ limitations under the License. #include #include -#include "global.h" +#include "types.h" #include "func/batch_normalization.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/src/func/generic/batch_normalization.cpp b/blueoil/converter/templates/src/func/generic/batch_normalization.cpp index 8bc5aa2e9..82a62e01f 100644 --- a/blueoil/converter/templates/src/func/generic/batch_normalization.cpp +++ b/blueoil/converter/templates/src/func/generic/batch_normalization.cpp @@ -16,7 +16,7 @@ limitations under the License. #include #include -#include "global.h" +#include "types.h" #include "func/batch_normalization.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/src/func/matmul.cpp b/blueoil/converter/templates/src/func/matmul.cpp index a1ff65b3d..75071030d 100644 --- a/blueoil/converter/templates/src/func/matmul.cpp +++ b/blueoil/converter/templates/src/func/matmul.cpp @@ -13,7 +13,7 @@ See the License for the specific language governing permissions and limitations under the License. ==============================================================================*/ -#include "global.h" +#include "types.h" #include "func/matmul.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/src/func/max_pool.cpp b/blueoil/converter/templates/src/func/max_pool.cpp index dc6f22e01..3cdcf542a 100644 --- a/blueoil/converter/templates/src/func/max_pool.cpp +++ b/blueoil/converter/templates/src/func/max_pool.cpp @@ -14,7 +14,7 @@ limitations under the License. ==============================================================================*/ #include #include -#include "global.h" +#include "types.h" #include "func/max_pool.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/src/func/pad.cpp b/blueoil/converter/templates/src/func/pad.cpp index f1747afba..71c755db8 100644 --- a/blueoil/converter/templates/src/func/pad.cpp +++ b/blueoil/converter/templates/src/func/pad.cpp @@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. See the License for the specific language governing permissions and limitations under the License. ==============================================================================*/ -#include "global.h" +#include "types.h" #include "func/pad.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/src/func/softmax.cpp b/blueoil/converter/templates/src/func/softmax.cpp index 2a7c48ba9..a641a0763 100644 --- a/blueoil/converter/templates/src/func/softmax.cpp +++ b/blueoil/converter/templates/src/func/softmax.cpp @@ -15,7 +15,7 @@ limitations under the License. #include -#include "global.h" +#include "types.h" #include "func/softmax.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/src/func/x86_avx/batch_normalization.cpp b/blueoil/converter/templates/src/func/x86_avx/batch_normalization.cpp index 6a47eabcd..a0a3174f6 100644 --- a/blueoil/converter/templates/src/func/x86_avx/batch_normalization.cpp +++ b/blueoil/converter/templates/src/func/x86_avx/batch_normalization.cpp @@ -16,7 +16,7 @@ limitations under the License. #include #include -#include "global.h" +#include "types.h" #include "func/batch_normalization.h" #include "time_measurement.h" diff --git a/blueoil/converter/templates/src/write_to_file.cpp b/blueoil/converter/templates/src/write_to_file.cpp index 376eca065..63cd91f03 100644 --- a/blueoil/converter/templates/src/write_to_file.cpp +++ b/blueoil/converter/templates/src/write_to_file.cpp @@ -1,4 +1,4 @@ -#include "global.h" +#include "types.h" #include #include From d428ef01e4ca0c44baacee54d57f819398f2aae6 Mon Sep 17 00:00:00 2001 From: Yuta Fukasawa <30371920+ytfksw@users.noreply.github.com> Date: Thu, 12 Mar 2020 20:14:24 +0900 Subject: [PATCH 03/18] convert tf2 for blueoil/generate_lmnet_config.py (#902) --- blueoil/generate_lmnet_config.py | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/blueoil/generate_lmnet_config.py b/blueoil/generate_lmnet_config.py index 99cb90789..b1e131731 100644 --- a/blueoil/generate_lmnet_config.py +++ b/blueoil/generate_lmnet_config.py @@ -157,9 +157,9 @@ def _blueoil_to_lmnet(blueoil_config): keep_checkpoint_max = default_keep_checkpoint_max if optimizer == 'Adam': - optimizer_class = "tf.train.AdamOptimizer" + optimizer_class = "tf.compat.v1.train.AdamOptimizer" elif optimizer == 'Momentum': - optimizer_class = "tf.train.MomentumOptimizer" + optimizer_class = "tf.compat.v1.train.MomentumOptimizer" else: raise ValueError("not supported optimizer.") @@ -173,9 +173,9 @@ def _blueoil_to_lmnet(blueoil_config): if learning_rate_schedule == "constant": learning_rate_func = None elif learning_rate_schedule == "cosine": - learning_rate_func = "tf.train.cosine_decay" + learning_rate_func = "tf.compat.v1.train.cosine_decay" else: - learning_rate_func = "tf.train.piecewise_constant" + learning_rate_func = "tf.compat.v1.train.piecewise_constant" if learning_rate_schedule == "constant": if optimizer == 'Momentum': From 65e60638fd87c2e2da864eda0d3ec14e06f7142e Mon Sep 17 00:00:00 2001 From: Kentaro Iizuka Date: Fri, 13 Mar 2020 09:06:15 +0900 Subject: [PATCH 04/18] Update setup.cfg (#911) --- setup.cfg | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/setup.cfg b/setup.cfg index 5f02bd158..e69a5cc74 100644 --- a/setup.cfg +++ b/setup.cfg @@ -1,6 +1,6 @@ [metadata] name = blueoil -version = 0.20.0 +version = 0.21.0 url = https://blueoil.org author = Blueoil development team author_email = test@example.com From d4da2b9204c0b1877dbb015d86f42eab89467c23 Mon Sep 17 00:00:00 2001 From: prime number Date: Mon, 16 Mar 2020 09:16:02 +0900 Subject: [PATCH 05/18] Fix debug (#918) --- blueoil/converter/templates/src/network.tpl.cpp | 12 ++++++------ 1 file changed, 6 insertions(+), 6 deletions(-) diff --git a/blueoil/converter/templates/src/network.tpl.cpp b/blueoil/converter/templates/src/network.tpl.cpp index 6f3392278..f7f6026ce 100644 --- a/blueoil/converter/templates/src/network.tpl.cpp +++ b/blueoil/converter/templates/src/network.tpl.cpp @@ -361,18 +361,18 @@ bool Network::run(float *network_input, float *network_output) {% if config.debug -%} {# Temporary: better access to the quantizer #} + {% for out_k in node.output_ops.keys() -%} {% if node.dtype.cpptype() in ['int', 'int32_t'] -%} - save_int32_data("debug/{{ node.name }}", {{ node.view.size_in_words_as_cpp }}, 0, {{ node.name }}.data(), 3.0 / 2.0 ); + save_int32_data("debug/{{ node.name }}_{{ out_k }}", {{ node.view.size_in_words_as_cpp }}, 0, {{ node.name }}_{{ out_k }}.data(), 3.0 / 2.0 ); {% elif node.dtype.cpptype() in ['unsigned', 'uint32_t'] -%} - save_uint32_data("debug/{{ node.name }}", {{ node.view.size_in_words_as_cpp }}, 0, {{ node.name }}.data(), 1.0); + save_uint32_data("debug/{{ node.name }}_{{ out_k }}", {{ node.view.size_in_words_as_cpp }}, 0, {{ node.name }}_{{ out_k }}.data(), 1.0); {% elif node.dtype.cpptype() == 'QUANTIZED_PACKED' -%} - save_uint32_data("debug/{{ node.name }}", {{ node.view.size_in_words_as_cpp }}, 0, reinterpret_cast({{ node.name }}.data()), 1.0); + save_uint32_data("debug/{{ node.name }}_{{ out_k }}", {{ node.view.size_in_words_as_cpp }}, 0, reinterpret_cast({{ node.name }}_{{ out_k }}.data()), 1.0); {% elif node.dtype.cpptype() == 'float' -%} - {% for out_k in node.output_ops.keys() -%} - save_float32_data("debug/{{ node.name }}", {{ node.view.size_in_words_as_cpp }}, {{ loop.index0 }}, {{ node.name + '_' + out_k }}.data(), 1.0); + save_float32_data("debug/{{ node.name }}_{{ out_k }}", {{ node.view.size_in_words_as_cpp }}, {{ loop.index0 }}, {{ node.name }}_{{ out_k }}.data(), 1.0); {{ '\n' -}} - {%- endfor %} {% endif %} + {%- endfor %} {% endif %} {% endfor -%} From f353c132b3efdd4cf17228445e437666620f9b0b Mon Sep 17 00:00:00 2001 From: Yuta Fukasawa <30371920+ytfksw@users.noreply.github.com> Date: Mon, 16 Mar 2020 11:02:42 +0900 Subject: [PATCH 06/18] Apply tf_upgrade_v2 (#914) --- blueoil/utils/executor.py | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/blueoil/utils/executor.py b/blueoil/utils/executor.py index a7cfe6ddc..e63b03023 100644 --- a/blueoil/utils/executor.py +++ b/blueoil/utils/executor.py @@ -88,7 +88,7 @@ def search_restore_filename(checkpoints_dir): def convert_variables_to_constants(sess, output_node_names=["output"]): - minimal_graph_def = tf.graph_util.convert_variables_to_constants( + minimal_graph_def = tf.compat.v1.graph_util.convert_variables_to_constants( sess, sess.graph.as_graph_def(add_shapes=True), output_node_names, @@ -114,7 +114,7 @@ def prepare_metrics(metrics_ops_dict): metrics_placeholders: list of metrics placeholder. """ - with tf.name_scope("metrics"): + with tf.compat.v1.name_scope("metrics"): metrics_placeholders = [] metrics_summaries = [] for (metrics_key, metrics_op) in metrics_ops_dict.items(): @@ -125,6 +125,6 @@ def prepare_metrics(metrics_ops_dict): metrics_placeholders.append(metrics_placeholder) metrics_summaries.append(summary) - metrics_summary_op = tf.summary.merge(metrics_summaries) + metrics_summary_op = tf.compat.v1.summary.merge(metrics_summaries) return metrics_summary_op, metrics_placeholders From 68463c2f9ff86d862b7d84aea906b5726d83bfdc Mon Sep 17 00:00:00 2001 From: Yuta Fukasawa <30371920+ytfksw@users.noreply.github.com> Date: Mon, 16 Mar 2020 11:33:14 +0900 Subject: [PATCH 07/18] Apply tf_upgrade_v2 (#916) --- output_template/python/lmnet/tensorflow_graph_runner.py | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/output_template/python/lmnet/tensorflow_graph_runner.py b/output_template/python/lmnet/tensorflow_graph_runner.py index 51e8a8378..6a258be52 100644 --- a/output_template/python/lmnet/tensorflow_graph_runner.py +++ b/output_template/python/lmnet/tensorflow_graph_runner.py @@ -37,16 +37,16 @@ def init(self): with graph.as_default(): with open(self.model_path, 'rb') as f: - graph_def = tf.GraphDef() + graph_def = tf.compat.v1.GraphDef() graph_def.ParseFromString(f.read()) tf.import_graph_def(graph_def, name="") - init_op = tf.global_variables_initializer() + init_op = tf.compat.v1.global_variables_initializer() self.images_placeholder = graph.get_tensor_by_name('images_placeholder:0') self.output_op = graph.get_tensor_by_name('output:0') - session_config = tf.ConfigProto() + session_config = tf.compat.v1.ConfigProto() session_config.gpu_options.allow_growth = True - self.sess = tf.Session(graph=graph, config=session_config) + self.sess = tf.compat.v1.Session(graph=graph, config=session_config) self.sess.run(init_op) From d537bdc7218408ae8c7db0314a5909a0877be55e Mon Sep 17 00:00:00 2001 From: Wasin Kalintha <6939637+oatawa1@users.noreply.github.com> Date: Tue, 17 Mar 2020 10:40:08 +0900 Subject: [PATCH 08/18] Fix very small comment typo (#926) --- blueoil/converter/core/operators.py | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/blueoil/converter/core/operators.py b/blueoil/converter/core/operators.py index 369a2716e..693bba974 100644 --- a/blueoil/converter/core/operators.py +++ b/blueoil/converter/core/operators.py @@ -2562,7 +2562,7 @@ def __init__(self, def _check_consistency(self) -> None: """ This check the following constraints: - 1. qunatized-packed data requires depth of input must be multiple of kernel_size^2 * 32 + 1. quantized-packed data requires depth of input must be multiple of kernel_size^2 * 32 """ super()._check_consistency() if self.input_ops['input'].op_type == 'QTZ_linear_mid_tread_half' and \ From a0541f7fe422b256ca468b2f8e8a3abc29c1f414 Mon Sep 17 00:00:00 2001 From: prime number Date: Tue, 17 Mar 2020 13:36:33 +0900 Subject: [PATCH 09/18] Add Comment (#925) --- blueoil/converter/templates/src/network.tpl.cpp | 1 + 1 file changed, 1 insertion(+) diff --git a/blueoil/converter/templates/src/network.tpl.cpp b/blueoil/converter/templates/src/network.tpl.cpp index f7f6026ce..0bea8b34a 100644 --- a/blueoil/converter/templates/src/network.tpl.cpp +++ b/blueoil/converter/templates/src/network.tpl.cpp @@ -280,6 +280,7 @@ bool Network::init() {% endfor -%} #endif // RUN_ON_FPGA +// Initialize OpenMP Thread-pool #pragma omp parallel std::cout << std::flush; From 4c5014ad69063eab0246fcade8aceb00694dc723 Mon Sep 17 00:00:00 2001 From: Koichi Yoshigoe Date: Wed, 18 Mar 2020 00:03:11 +0900 Subject: [PATCH 10/18] Add make commands (#922) --- Makefile | 14 ++++++++++++++ 1 file changed, 14 insertions(+) diff --git a/Makefile b/Makefile index 00bb58ad8..716fd548f 100644 --- a/Makefile +++ b/Makefile @@ -10,6 +10,20 @@ deps: # Update dependencies git submodule update --init --recursive +.PHONY: install +install: deps + pip install -e .[cpu,tests,docs] + pip install pycocotools==2.0.0 + +.PHONY: install-gpu +install-gpu: install + pip install -e .[gpu] + pip install -e .[dist] + +.PHONY: lint +lint: + flake8 ./blueoil --exclude=templates,converter + .PHONY: build build: deps # Build docker image From 620ba3b404dea142ff53461206c31e987b26cb6e Mon Sep 17 00:00:00 2001 From: Taketoshi Fujiwara Date: Wed, 18 Mar 2020 10:32:44 +0900 Subject: [PATCH 11/18] Include the tests directory to flake8 targets (#932) --- Makefile | 4 ++-- tests/e2e/test_object_detection.py | 1 + .../classification_test/test_lm_resnet_quantize.py | 2 +- 3 files changed, 4 insertions(+), 3 deletions(-) diff --git a/Makefile b/Makefile index 716fd548f..8cce9fb8b 100644 --- a/Makefile +++ b/Makefile @@ -22,7 +22,7 @@ install-gpu: install .PHONY: lint lint: - flake8 ./blueoil --exclude=templates,converter + flake8 ./blueoil ./tests --exclude=templates,converter .PHONY: build build: deps @@ -60,7 +60,7 @@ test-lmnet: test-blueoil-pep8 test-unit-main test-blueoil-pep8: build # Check blueoil pep8 # FIXME: blueoil/templates and blueoil/converter have a lot of errors with flake8 - docker run ${DOCKER_OPT} $(IMAGE_NAME):$(BUILD_VERSION) /bin/bash -c "flake8 ./blueoil --exclude=templates,converter" + docker run ${DOCKER_OPT} $(IMAGE_NAME):$(BUILD_VERSION) /bin/bash -c "flake8 ./blueoil ./tests --exclude=templates,converter" .PHONY: test-unit-main test-unit-main: build diff --git a/tests/e2e/test_object_detection.py b/tests/e2e/test_object_detection.py index d1c9b4d17..4e59dca38 100644 --- a/tests/e2e/test_object_detection.py +++ b/tests/e2e/test_object_detection.py @@ -123,6 +123,7 @@ } } + @pytest.mark.parametrize( "config", [ openimagesv4_object_detection, diff --git a/tests/unit/networks_tests/classification_test/test_lm_resnet_quantize.py b/tests/unit/networks_tests/classification_test/test_lm_resnet_quantize.py index acab6aaad..234be9cbf 100644 --- a/tests/unit/networks_tests/classification_test/test_lm_resnet_quantize.py +++ b/tests/unit/networks_tests/classification_test/test_lm_resnet_quantize.py @@ -24,7 +24,7 @@ from blueoil.datasets.image_folder import ImageFolderBase from blueoil.pre_processor import Resize from blueoil.quantizations import ( - binary_mean_scaling_quantizer, + binary_mean_scaling_quantizer, linear_mid_tread_half_quantizer, ) from blueoil.utils.executor import prepare_dirs From ae823056dfb2a4a588dda8fc1013e3f79db99ff2 Mon Sep 17 00:00:00 2001 From: Koichi Yoshigoe Date: Wed, 18 Mar 2020 15:21:07 +0900 Subject: [PATCH 12/18] Fix run.py time measurement log output (#930) --- output_template/python/run.py | 3 +-- 1 file changed, 1 insertion(+), 2 deletions(-) diff --git a/output_template/python/run.py b/output_template/python/run.py index 8e079a97f..0c7102cb4 100644 --- a/output_template/python/run.py +++ b/output_template/python/run.py @@ -177,8 +177,7 @@ def run_prediction(input_image, model, config_file, trial=1): filename_images = image_from_json(json_obj, raw_images, image_files) _save_images(output_dir, filename_images) logger.info("Benchmark avg result(sec) for {} trials: pre_process: {} inference: {} post_process: {} Total: {}" - .format(trial, bench_pre / trial, bench_inference / trial, bench_post / trial, - (bench_pre + bench_post + bench_inference) / trial,)) + .format(trial, bench_pre, bench_inference, bench_post, bench_pre + bench_post + bench_inference,)) @click.command(context_settings=dict(help_option_names=['-h', '--help'])) From 4f7bba656229a8a5a83fe834a6d2a8a34d97be4e Mon Sep 17 00:00:00 2001 From: Koichi Yoshigoe Date: Wed, 18 Mar 2020 16:30:59 +0900 Subject: [PATCH 13/18] Update flake8 and fixed lint (#923) --- blueoil/cmd/output_event.py | 31 ++++++++++++++++-------------- blueoil/datasets/base.py | 4 ++-- blueoil/datasets/mscoco.py | 8 ++++---- blueoil/datasets/open_images_v4.py | 2 +- blueoil/datasets/pascalvoc_base.py | 4 ++-- blueoil/networks/base.py | 10 +++++----- setup.cfg | 2 +- 7 files changed, 32 insertions(+), 29 deletions(-) diff --git a/blueoil/cmd/output_event.py b/blueoil/cmd/output_event.py index 1d7109827..1ca6cd697 100644 --- a/blueoil/cmd/output_event.py +++ b/blueoil/cmd/output_event.py @@ -34,7 +34,7 @@ def _value_step_list(event_accumulator, metrics_key): events = event_accumulator.Scalars(metrics_key) return [(event.value, event.step) for event in events] except KeyError as e: - print("Key {} was not found in {}".format(metrics_key, event_accumulator.path)) + print("Key {} was not found in {}\n{}".format(metrics_key, event_accumulator.path, e)) return [] @@ -55,14 +55,16 @@ def output(tensorboard_dir, output_dir, metrics_keys, steps, output_file_base="m event_accumulators.append(event_accumulator) if not metrics_keys: - metrics_keys = {metrics_key - for event_accumulator in event_accumulators - for metrics_key in _get_metrics_keys(event_accumulator)} + metrics_keys = { + metrics_key + for event_accumulator in event_accumulators + for metrics_key in _get_metrics_keys(event_accumulator) + } columns = [_column_name(event_accumulator, metrics_key) for event_accumulator, metrics_key in itertools.product(event_accumulators, metrics_keys)] columns.sort() - df = pd.DataFrame([], columns=columns) + df = pd.DataFrame([], columns=columns) for event_accumulator in event_accumulators: for metrics_key in metrics_keys: @@ -102,13 +104,8 @@ def output(tensorboard_dir, output_dir, metrics_keys, steps, output_file_base="m print(message) -@click.command(context_settings=dict(help_option_names=['-h', '--help'])) -@click.option( - "-i", - "--experiment_id", - help="id of target experiment", - required=True, -) +@click.command(context_settings=dict(help_option_names=["-h", "--help"])) +@click.option("-i", "--experiment_id", help="id of target experiment", required=True) @click.option( "-k", "--metrics_keys", @@ -135,8 +132,14 @@ def output(tensorboard_dir, output_dir, metrics_keys, steps, output_file_base="m def main(output_file_base, metrics_keys, steps, experiment_id): environment.init(experiment_id) - output(environment.TENSORBOARD_DIR, environment.EXPERIMENT_DIR, metrics_keys, steps, output_file_base="metrics",) + output( + environment.TENSORBOARD_DIR, + environment.EXPERIMENT_DIR, + metrics_keys, + steps, + output_file_base="metrics", + ) -if __name__ == '__main__': +if __name__ == "__main__": main() diff --git a/blueoil/datasets/base.py b/blueoil/datasets/base.py index 255e133ee..96b65417c 100644 --- a/blueoil/datasets/base.py +++ b/blueoil/datasets/base.py @@ -259,8 +259,8 @@ def _validation_data_dir(self): @property def data_dir(self): - if self.subset is "train": + if self.subset == "train": return self._train_data_dir - if self.subset is "validation": + if self.subset == "validation": return self._validation_data_dir diff --git a/blueoil/datasets/mscoco.py b/blueoil/datasets/mscoco.py index 36a03669e..fd34fe3ef 100644 --- a/blueoil/datasets/mscoco.py +++ b/blueoil/datasets/mscoco.py @@ -79,7 +79,7 @@ def coco(self): def _image_ids(self): """Return all files and gt_boxes list.""" - classes = [class_name for class_name in self.classes if class_name is not "__background__"] + classes = [class_name for class_name in self.classes if class_name != "__background__"] target_class_ids = self.coco.getCatIds(catNms=classes) image_ids = [] for target_class_id in target_class_ids: @@ -96,7 +96,7 @@ def _label_from_image_id(self, image_id): width = coco_image["width"] label = np.zeros((height, width), dtype='uint8') - classes = [class_name for class_name in self.classes if class_name is not "__background__"] + classes = [class_name for class_name in self.classes if class_name != "__background__"] for target_class in classes: target_class_id = self.coco.getCatIds(catNms=[target_class])[0] annotation_ids = self.coco.getAnnIds(imgIds=[image_id], catIds=[target_class_id], iscrowd=None) @@ -197,7 +197,7 @@ def coco(self): @functools.lru_cache(maxsize=None) def _image_ids(self): """Return all files and gt_boxes list.""" - classes = [class_name for class_name in self.classes if class_name is not "__background__"] + classes = [class_name for class_name in self.classes if class_name != "__background__"] target_class_ids = self.coco.getCatIds(catNms=classes) image_ids = [] for target_class_id in target_class_ids: @@ -225,7 +225,7 @@ def coco_category_id_to_lmnet_class_id(self, cat_id): @functools.lru_cache(maxsize=None) def _gt_boxes_from_image_id(self, image_id): """Return gt boxes list ([[x, y, w, h, class_id]]) of a image.""" - classes = [class_name for class_name in self.classes if class_name is not "__background__"] + classes = [class_name for class_name in self.classes if class_name != "__background__"] class_ids = set(self.coco.getCatIds(catNms=classes)) boxes = [] diff --git a/blueoil/datasets/open_images_v4.py b/blueoil/datasets/open_images_v4.py index f2c8350b7..af102e42f 100644 --- a/blueoil/datasets/open_images_v4.py +++ b/blueoil/datasets/open_images_v4.py @@ -56,7 +56,7 @@ def _classes_meta(self): with open(self.class_descriptions_csv, 'r') as f: reader = csv.reader(f) for row in reader: - class_name = re.sub('\W', '', row[1]) + class_name = re.sub(r'\W', '', row[1]) classes_meta[row[0]] = class_name return classes_meta diff --git a/blueoil/datasets/pascalvoc_base.py b/blueoil/datasets/pascalvoc_base.py index 335917735..8e09ba757 100644 --- a/blueoil/datasets/pascalvoc_base.py +++ b/blueoil/datasets/pascalvoc_base.py @@ -171,7 +171,7 @@ def _image_ids(self, data_type=None): return all_image_ids image_ids = [ - image_id for image_id, gt_boxes in zip(all_image_ids, all_gt_boxes_list) if len(gt_boxes) is not 0 + image_id for image_id, gt_boxes in zip(all_image_ids, all_gt_boxes_list) if len(gt_boxes) != 0 ] return image_ids @@ -196,7 +196,7 @@ def _all_image_ids(self, data_type=None, is_debug=False): return df.image_id.tolist() def _files_and_annotations(self): - raise NotImplemented() + raise NotImplementedError() def _init_files_and_annotations(self): """Init files and gt_boxes list, Cache these.""" diff --git a/blueoil/networks/base.py b/blueoil/networks/base.py index c74ac0c2c..e93d4ee8a 100644 --- a/blueoil/networks/base.py +++ b/blueoil/networks/base.py @@ -75,7 +75,7 @@ def base(self, images, is_training, *args, **kwargs): tf.Tensor: Inference result. """ - raise NotImplemented() + raise NotImplementedError() def placeholders(self): """Placeholders. @@ -86,7 +86,7 @@ def placeholders(self): tf.compat.v1.placeholder: Placeholders. """ - raise NotImplemented() + raise NotImplementedError() def metrics(self, output, labels): """Metrics. @@ -96,7 +96,7 @@ def metrics(self, output, labels): labels: labels tensor. """ - raise NotImplemented() + raise NotImplementedError() # TODO(wakisaka): Deal with many networks. def summary(self, output, labels=None): @@ -119,7 +119,7 @@ def inference(self, images, is_training): images: images tensor. shape is (batch_num, height, width, channel) """ - raise NotImplemented() + raise NotImplementedError() def loss(self, output, labels): """Loss. @@ -129,7 +129,7 @@ def loss(self, output, labels): labels: labels tensor. """ - raise NotImplemented() + raise NotImplementedError() def optimizer(self, global_step): assert ("learning_rate" in self.optimizer_kwargs.keys()) or \ diff --git a/setup.cfg b/setup.cfg index e69a5cc74..ea4556fe4 100644 --- a/setup.cfg +++ b/setup.cfg @@ -41,7 +41,7 @@ docs = Sphinx==2.4.4 sphinx-autobuild==0.7.1 test = - flake8==3.5.0 + flake8==3.7.9 pytest==5.3.5 pytest-xdist==1.31.0 parameterized==0.7.1 From f0c4b32f63f44b5abcb483dedb241b7cdf59df89 Mon Sep 17 00:00:00 2001 From: tk26eng <35796618+tk26eng@users.noreply.github.com> Date: Wed, 18 Mar 2020 17:20:21 +0900 Subject: [PATCH 14/18] Add time measurement flag to all elf binaries in Makefile (#928) --- blueoil/cmd/convert.py | 7 ------- blueoil/converter/templates/Makefile | 8 ++++---- 2 files changed, 4 insertions(+), 11 deletions(-) diff --git a/blueoil/cmd/convert.py b/blueoil/cmd/convert.py index 80778e721..0fcd1aa5b 100644 --- a/blueoil/cmd/convert.py +++ b/blueoil/cmd/convert.py @@ -121,15 +121,8 @@ def make_all(project_dir, output_dir): # Change current directory to project directory os.chdir(project_dir) - cxxflags_cache = os.getenv("CXXFLAGS", "") - # Make each target and move output files for target, output in make_list: - if target in {"lm_x86", "lm_x86_avx", "lm_arm", "lm_fpga", "lm_aarch64"}: - os.environ["CXXFLAGS"] = cxxflags_cache + " -DFUNC_TIME_MEASUREMENT" - else: - os.environ["CXXFLAGS"] = cxxflags_cache - subprocess.run(("make", "clean", "--quiet")) subprocess.run(("make", target, "-j4", "--quiet")) strip_binary(output) diff --git a/blueoil/converter/templates/Makefile b/blueoil/converter/templates/Makefile index 307df13d5..2f08c1cdd 100644 --- a/blueoil/converter/templates/Makefile +++ b/blueoil/converter/templates/Makefile @@ -116,19 +116,19 @@ clean: -$(RM) $(OBJ) lm_x86: CXX = g++ -lm_x86: FLAGS += $(INCLUDES) -O3 -std=c++14 -DUSE_PNG -pthread -g +lm_x86: FLAGS += $(INCLUDES) -O3 -std=c++14 -DUSE_PNG -pthread -g -DFUNC_TIME_MEASUREMENT lm_x86: CXXFLAGS += lm_x86_avx: CXX = g++ -lm_x86_avx: FLAGS += $(INCLUDES) -O3 -std=c++14 -mavx2 -mfma -DUSE_AVX -DUSE_PNG -pthread -g -fopenmp +lm_x86_avx: FLAGS += $(INCLUDES) -O3 -std=c++14 -mavx2 -mfma -DUSE_AVX -DUSE_PNG -pthread -g -fopenmp -DFUNC_TIME_MEASUREMENT lm_x86_avx: CXXFLAGS += lm_aarch64: CXX = aarch64-linux-gnu-g++ -lm_aarch64: FLAGS += $(INCLUDES) -std=c++14 -O3 -DUSE_NEON -DUSE_PNG -pthread -g -fopenmp +lm_aarch64: FLAGS += $(INCLUDES) -std=c++14 -O3 -DUSE_NEON -DUSE_PNG -pthread -g -fopenmp -DFUNC_TIME_MEASUREMENT lm_aarch64: CXXFLAGS += lm_arm: CXX = arm-linux-gnueabihf-g++ -lm_arm: FLAGS += $(INCLUDES) -std=c++14 -O3 -DUSE_NEON -DUSE_PNG -DAARCH32 -mcpu=cortex-a9 -mfpu=neon -mthumb -s -pthread -g -fopenmp +lm_arm: FLAGS += $(INCLUDES) -std=c++14 -O3 -DUSE_NEON -DUSE_PNG -DAARCH32 -mcpu=cortex-a9 -mfpu=neon -mthumb -s -pthread -g -fopenmp -DFUNC_TIME_MEASUREMENT lm_arm: CXXFLAGS += lm_fpga: CXX = arm-linux-gnueabihf-g++ From b8edc89608a51ed16f50c77d2a588af75210610c Mon Sep 17 00:00:00 2001 From: prime number Date: Wed, 18 Mar 2020 17:57:04 +0900 Subject: [PATCH 15/18] Fix threshold skipping if flg = -1 (#917) --- .../impl/arm_neon/quantized_conv2d_tiling.cpp | 24 +++++++++---------- 1 file changed, 12 insertions(+), 12 deletions(-) diff --git a/blueoil/converter/templates/src/func/impl/arm_neon/quantized_conv2d_tiling.cpp b/blueoil/converter/templates/src/func/impl/arm_neon/quantized_conv2d_tiling.cpp index 4a7e2608f..b651a50df 100644 --- a/blueoil/converter/templates/src/func/impl/arm_neon/quantized_conv2d_tiling.cpp +++ b/blueoil/converter/templates/src/func/impl/arm_neon/quantized_conv2d_tiling.cpp @@ -281,7 +281,7 @@ void QuantizedConv2DTiling(const tiling_input_t& input, if (p.thresholds != nullptr) { #define LOAD_TH(k) \ const auto ts##k = vld4q_s16(p.thresholds + NUM_OF_A2W1_THRESHOLD * (out_ch_high * OutChUnroll2 + Om + 8 * k)); \ - const auto is_neg##k = vreinterpretq_s16_u16(vcltq_s16(ts##k.val[3], vdupq_n_s16(0))); \ + const auto mask##k = vreinterpretq_s16_u16(0x0003 & vcltq_s16(ts##k.val[3], vdupq_n_s16(0))); \ const auto m2_##k = vsubq_s16(ts##k.val[3], vdupq_n_s16(2)); \ const auto is_const##k = vcgeq_s16(m2_##k, vdupq_n_s16(0)); LOAD_TH(0) @@ -292,11 +292,11 @@ void QuantizedConv2DTiling(const tiling_input_t& input, if (col_high + col >= out_width) break; #define APPLY(k) \ const auto d##k = vld1q_s16(out_tile + buf_index + 8 * k); \ - const auto f##k##0 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[0])) & ts##k.val[3]; \ - const auto f##k##1 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[1])) & ts##k.val[3]; \ - const auto f##k##2 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[2])) & ts##k.val[3]; \ - const auto tmp##k = f##k##0 + f##k##1 + f##k##2 + is_neg##k; \ - const auto res##k = vreinterpretq_u8_s16(vbslq_s16(is_const##k, m2_##k, tmp##k)); + const auto f##k##0 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[0]) & 0x0001u); \ + const auto f##k##1 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[1]) & 0x0001u); \ + const auto f##k##2 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[2]) & 0x0001u); \ + const auto tmp##k = f##k##0 + f##k##1 + f##k##2; \ + const auto res##k = vreinterpretq_u8_s16(vbslq_s16(is_const##k, m2_##k, mask##k ^ tmp##k)); const auto buf_index = row * TileWidth * OutChUnroll + col * OutChUnroll; APPLY(0) @@ -546,7 +546,7 @@ void QuantizedConv2DTiling(const tiling_input_t& input, if (p.thresholds != nullptr) { #define LOAD_TH(k) \ const auto ts##k = vld4q_s16(p.thresholds + NUM_OF_A2W1_THRESHOLD * (out_ch_high * OutChUnroll2 + Om + 8 * k)); \ - const auto is_neg##k = vreinterpretq_s16_u16(vcltq_s16(ts##k.val[3], vdupq_n_s16(0))); \ + const auto mask##k = vreinterpretq_s16_u16(0x0003 & vcltq_s16(ts##k.val[3], vdupq_n_s16(0))); \ const auto m2_##k = vsubq_s16(ts##k.val[3], vdupq_n_s16(2)); \ const auto is_const##k = vcgeq_s16(m2_##k, vdupq_n_s16(0)); LOAD_TH(0) @@ -557,11 +557,11 @@ void QuantizedConv2DTiling(const tiling_input_t& input, if (col_high + col >= out_width) break; #define APPLY(k) \ const auto d##k = vld1q_s16(out_tile + buf_index + 8 * k); \ - const auto f##k##0 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[0])) & ts##k.val[3]; \ - const auto f##k##1 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[1])) & ts##k.val[3]; \ - const auto f##k##2 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[2])) & ts##k.val[3]; \ - const auto tmp##k = f##k##0 + f##k##1 + f##k##2 + is_neg##k; \ - const auto res##k = vreinterpretq_u8_s16(vbslq_s16(is_const##k, m2_##k, tmp##k)); + const auto f##k##0 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[0]) & 0x0001u); \ + const auto f##k##1 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[1]) & 0x0001u); \ + const auto f##k##2 = vreinterpretq_s16_u16(vcgeq_s16(d##k, ts##k.val[2]) & 0x0001u); \ + const auto tmp##k = f##k##0 + f##k##1 + f##k##2; \ + const auto res##k = vreinterpretq_u8_s16(vbslq_s16(is_const##k, m2_##k, mask##k ^ tmp##k)); const auto buf_index = row * TileWidth * OutChUnroll + col * OutChUnroll; APPLY(0) From 940bc55721f4f1aa09f332982cd0022d405bcc98 Mon Sep 17 00:00:00 2001 From: Wasin Kalintha <6939637+oatawa1@users.noreply.github.com> Date: Wed, 18 Mar 2020 18:25:14 +0900 Subject: [PATCH 16/18] Add Converter (DLK) limitation to blueoil (#920) --- docs/converter/index.rst | 1 + docs/converter/supported_ops.md | 60 +++++++++++++++++++++++++++++++++ 2 files changed, 61 insertions(+) create mode 100644 docs/converter/supported_ops.md diff --git a/docs/converter/index.rst b/docs/converter/index.rst index 1441064a2..7e4655a76 100644 --- a/docs/converter/index.rst +++ b/docs/converter/index.rst @@ -13,6 +13,7 @@ Converter is a converter part in Blueoil for edge devices. :maxdepth: 1 overview + supported_ops modules auto_script shared_library diff --git a/docs/converter/supported_ops.md b/docs/converter/supported_ops.md new file mode 100644 index 000000000..b34fa7082 --- /dev/null +++ b/docs/converter/supported_ops.md @@ -0,0 +1,60 @@ +# Supported Ops +## Ops with Limitations +### Base Limitations +- **Output** + - Requires each output channel size <= `1024`. + +### Blueoil Customized Ops +- **[QTZ_binary_channel_wise_mean_scaling](https://github.com/blue-oil/blueoil/blob/620ba3b404dea142ff53461206c31e987b26cb6e/blueoil/converter/core/operators.py#L2352)**: Quantization operator using binary channel wise scaling. +- **[QTZ_binary_mean_scaling](https://github.com/blue-oil/blueoil/blob/620ba3b404dea142ff53461206c31e987b26cb6e/blueoil/converter/core/operators.py#L709)**: Quantization operator using binary scaling. + - Input tensor must have float values. +- **[QTZ_linear_mid_tread_half](https://github.com/blue-oil/blueoil/blob/620ba3b404dea142ff53461206c31e987b26cb6e/blueoil/converter/core/operators.py#L1373)**: Quantization operator with 'linear mid tread half'. + +### Tensorflow Ops with Limitations +- **[tf.layers.AveragePooling2D](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/layers/AveragePooling2D)** + - Currently, support only `2D`. + - Do ***not*** support `kernel depth = 1`. +- **[tf.concat](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/concat)** + - Do ***not*** support concat of mixed data types (e.g., quantized values and float values). + - All tensor channels must be equal. + - If inputs are quantized, requires `Each input channel size = multiple of 32`. +- **[tf.layers.Conv2D](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/layers/Conv2D)** + - Support only convolution `2D`. + - Do ***not*** support transpose. + - Requires `kernel size = 1x1` or `3x3` or `5x5`. + - Accelerator is not supported `kernel size = 5x5` (CPU supported only). + - Requires `Input channel size = multiple of 32`, otherwise zero padding is used. + - If output is quantized by later operations, `Output channel size = multiple of 32`, otherwise output channel size is free from limitation (but performance will be worse). +- **[tf.nn.depth_to_space](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/nn/depth_to_space)** + - Requires `depth of input = multiple of block_size^2 * 32`. +- **[tf.nn.fused_batch_norm](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/nn/fused_batch_norm)** + - `scale`, `offset`, `mean`, `variance`, and `epsilon` must be constants or computable from constants. +- **[tf.linalg.matmul](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/linalg/matmul)** + - Do ***not*** support `scalar`. +- **[tf.layers.max_pooling2d](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/layers/max_pooling2d)** + - Currently, support only `2D`. +- **[tf.pad](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/pad)** + - Supports only `channel-wise paddings`. +- **[tf.nn.space_to_depth](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/nn/space_to_depth)** + - Requires `output depth = (multiple of block_size^2 * 32)` or `(block_size^2 * {8, 16})`. +- **[tf.split](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/split)** + - Currently, all of output tensors must have `same` shape. + - For quantized tensor, requires `number of channel of each output tensor = multiple of 32`. + +### Tensorflow Ops without Limitations +- **[tf.math.add](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/math/add)** +- **[tf.identity](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/identity)** +- **[tf.nn.leaky_relu](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/nn/leaky_relu)** +- **[tf.math.maximum](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/math/maximum)** +- **[tf.math.minimum](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/math/minimum)** +- **[tf.math.multiply](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/math/multiply)** +- **[tf.nn.relu](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/nn/relu)** +- **[tf.reshape](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/reshape)** +- **[tf.image.resize_nearest_neighbor](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/image/resize_nearest_neighbor)** +- **[tf.nn.softmax](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/nn/softmax)** +- **[tf.transpose](https://www.tensorflow.org/versions/r1.15/api_docs/python/tf/transpose)** + + +## Data Types +- **Floating point** + - [tf.float32](https://www.tensorflow.org/api_docs/python/tf#float32): 32-bit single-precision floating-point. From 4e2ddc811d0c554bda0c6dec2bdc28ae00eb3410 Mon Sep 17 00:00:00 2001 From: prime number Date: Thu, 19 Mar 2020 09:09:17 +0900 Subject: [PATCH 17/18] Fix comparison order in generationg thresholds (#912) --- blueoil/converter/core/optimizer.py | 41 ++++++++++++++++------------- 1 file changed, 22 insertions(+), 19 deletions(-) diff --git a/blueoil/converter/core/optimizer.py b/blueoil/converter/core/optimizer.py index b68cdee01..7f63b3fc2 100644 --- a/blueoil/converter/core/optimizer.py +++ b/blueoil/converter/core/optimizer.py @@ -277,19 +277,35 @@ def pass_compute_thresholds(graph: Graph) -> None: # The threshold_table is numpy array that holds the threshold values for all channels threshold_table = np.empty([ch, n + 1], dtype=np.int32) + # Compute increasing order or decreasing order + is_inc_order = [True for i in range(ch)] + if quantizer_conv_weights.op_type == 'BinaryChannelWiseMeanScalingQuantizer': + for c in range(ch): + is_inc_order[c] = scaling_factor[c] > 0 + else: + for c in range(ch): + is_inc_order[c] = scaling_factor > 0 + + for op in p[:-1]: + if op.op_type == 'BatchNormalization': + bn_scale = op.input_ops['scale'].data + for c in range(ch): + if bn_scale[c] < 0: + is_inc_order[c] = not is_inc_order[c] + + for c in range(ch): + threshold_table[c, -1] = 1 \ + if is_inc_order[c] else -1 + # Compute threshold (t0, t1, t2) th_val = [0.5 + i for i in range(n)] for th_id, th_v in enumerate(th_val): init_threshold = np.full(ch, th_v, dtype=np.float64) # run calculation in reverse order, for example, q -> bn -> scaling - bn_nega_idx = [] trans_th = {'data': init_threshold} for op in p[:-1]: trans_th = op.de_run(**trans_th) - if op.op_type == 'BatchNormalization': - bn_scale = op.input_ops['scale'].data - bn_nega_idx = [v for v in range(len(bn_scale)) if bn_scale[v] < 0] threshold = (trans_th['data'] * np.float64(n)) / (np.float64(max_v) * scaling_factor) # take care of threshold values that are larger than 13-bit signed integer @@ -297,21 +313,8 @@ def pass_compute_thresholds(graph: Graph) -> None: threshold[threshold < -max_th_value] = -max_th_value for ch_id, th_per_ch in enumerate(threshold): - if quantizer_conv_weights.op_type == 'BinaryChannelWiseMeanScalingQuantizer': - threshold_table[ch_id, th_id] = int(math.floor(th_per_ch)) \ - if (scaling_factor[ch_id] < 0) ^ (ch_id in bn_nega_idx) \ - else int(math.ceil(th_per_ch)) - else: - threshold_table[ch_id, th_id] = int(math.floor(th_per_ch)) \ - if (scaling_factor < 0) ^ (ch_id in bn_nega_idx) \ - else int(math.ceil(th_per_ch)) - - for c in range(ch): - threshold_table[c, -1] = 1 \ - if np.all(threshold_table[c, 1:-1] > threshold_table[c, :-2], axis=0) else -1 - if np.all(threshold_table[c, 1:-1] == threshold_table[c, :-2], axis=0): - threshold_table[c, -1] = 1 - threshold_table[c, 0:-1] = max_th_value + threshold_table[ch_id, th_id] = int(math.ceil(th_per_ch)) \ + if is_inc_order[ch_id] else int(math.floor(th_per_ch)) bits_per_word = 32 rem = (bits_per_word - ch % bits_per_word) % bits_per_word From 969633c20aeea7ee964b2742efd5dfae331e6e80 Mon Sep 17 00:00:00 2001 From: Koichi Yoshigoe Date: Mon, 23 Mar 2020 16:02:58 +0900 Subject: [PATCH 18/18] Fix checkpoint loader to use GFile (#938) --- blueoil/cmd/train.py | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/blueoil/cmd/train.py b/blueoil/cmd/train.py index 8ec7abe96..43fbfc3e7 100644 --- a/blueoil/cmd/train.py +++ b/blueoil/cmd/train.py @@ -395,7 +395,7 @@ def train(config_file, experiment_id=None, recreate=False): if not tf.io.gfile.exists(checkpoint): raise Exception('Checkpoints are not created in {}'.format(experiment_dir)) - with open(checkpoint) as stream: + with tf.io.gfile.GFile(checkpoint) as stream: data = yaml.load(stream) checkpoint_name = os.path.basename(data['model_checkpoint_path'])