diff --git a/src/common/types.cpp b/src/common/types.cpp index 399c808e425c90e6d8003e2978dab11ed209cabd..cd5e66517f159f5f9db118313b78ccd2a8c216a8 100644 --- a/src/common/types.cpp +++ b/src/common/types.cpp @@ -75,6 +75,7 @@ const char *G_OP_TYPE_TOP_K = "top_k"; const char *G_OP_TYPE_CAST = "cast"; const char *G_OP_TYPE_LOG = "log"; const char *G_OP_TYPE_LOD_RESET = "lod_reset"; +const char *G_OP_TYPE_LESS_THAN = "less_than"; const char *G_OP_TYPE_QUANTIZE = "quantize"; const char *G_OP_TYPE_DEQUANTIZE = "dequantize"; @@ -175,5 +176,6 @@ std::unordered_map< {G_OP_TYPE_SEQUENCE_SOFTMAX, {{"X"}, {"Out"}}}, {G_OP_TYPE_NORM, {{"X"}, {"Out", "Norm"}}}, {G_OP_TYPE_LOG, {{"X"}, {"Out"}}}, - {G_OP_TYPE_LOD_RESET, {{"X", "Y"}, {"Out"}}}}; + {G_OP_TYPE_LOD_RESET, {{"X", "Y"}, {"Out"}}}, + {G_OP_TYPE_LESS_THAN, {{"X", "Y"}, {"Out"}}}}; } // namespace paddle_mobile diff --git a/src/common/types.h b/src/common/types.h index a5ad00c5a13a50332d35784d436b9cb6f6d08b89..9719be623cf7d92147416de1b5ebf276c866edcd 100644 --- a/src/common/types.h +++ b/src/common/types.h @@ -158,6 +158,7 @@ extern const char *G_OP_TYPE_TOP_K; extern const char *G_OP_TYPE_CAST; extern const char *G_OP_TYPE_LOG; extern const char *G_OP_TYPE_LOD_RESET; +extern const char *G_OP_TYPE_LESS_THAN; extern const char *G_OP_TYPE_QUANTIZE; extern const char *G_OP_TYPE_DEQUANTIZE; diff --git a/src/framework/load_ops.h b/src/framework/load_ops.h index 264f72523e2f876f332db53c357a73f4de017ec0..1caefe5ae77c9f4328d4d99af9e0b5c3b408d921 100644 --- a/src/framework/load_ops.h +++ b/src/framework/load_ops.h @@ -273,3 +273,9 @@ LOAD_OP1(sequence_pool, CPU); #ifdef LOG_OP LOAD_OP1(log, CPU); #endif +#ifdef LOD_RESET_OP +LOAD_OP1(lod_reset, CPU); +#endif +#ifdef LESS_THAN_OP +LOAD_OP1(less_than, CPU); +#endif diff --git a/src/operators/activation_op.cpp b/src/operators/activation_op.cpp index 24962e5f066696c0099d191ef040c16bcfce81c8..bcff87c9276721c19a970eb328fc0a183ed6c003 100644 --- a/src/operators/activation_op.cpp +++ b/src/operators/activation_op.cpp @@ -17,7 +17,7 @@ limitations under the License. */ namespace paddle_mobile { namespace operators { -#define DEFINE_INFERSHAPE(OpName) \ +#define DEFINE_ACTIVATION_INFERSHAPE(OpName) \ template \ void OpName##Op::InferShape() const { \ const auto &input_dims = this->param_.InputX()->dims(); \ @@ -25,20 +25,20 @@ namespace operators { } #ifdef RELU_OP -DEFINE_INFERSHAPE(Relu); -DEFINE_INFERSHAPE(Relu6); +DEFINE_ACTIVATION_INFERSHAPE(Relu); +DEFINE_ACTIVATION_INFERSHAPE(Relu6); #endif // RELU_OP #ifdef SIGMOID_OP -DEFINE_INFERSHAPE(Sigmoid); +DEFINE_ACTIVATION_INFERSHAPE(Sigmoid); #endif // SIGMOID_OP #ifdef TANH_OP -DEFINE_INFERSHAPE(Tanh); +DEFINE_ACTIVATION_INFERSHAPE(Tanh); #endif // TANH_OP #ifdef LOG_OP -DEFINE_INFERSHAPE(Log); +DEFINE_ACTIVATION_INFERSHAPE(Log); #endif // LOG_OP } // namespace operators diff --git a/src/operators/compare_op.cpp b/src/operators/compare_op.cpp new file mode 100644 index 0000000000000000000000000000000000000000..312173a30bad7c5b3762de6f6abc3a735895704c --- /dev/null +++ b/src/operators/compare_op.cpp @@ -0,0 +1,34 @@ +/* Copyright (c) 2018 PaddlePaddle Authors. All Rights Reserved. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +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 "operators/compare_op.h" + +namespace paddle_mobile { +namespace operators { + +#ifdef LESS_THAN_OP +template +void LessThanOp::InferShape() const { + const auto &input_dims = this->param_.input_x_->dims(); + this->param_.output_->Resize(input_dims); +} +#endif // LESS_THAN_OP + +} // namespace operators +} // namespace paddle_mobile + +namespace ops = paddle_mobile::operators; +#ifdef LESS_THAN_OP +REGISTER_OPERATOR_CPU(less_than, ops::LessThanOp); +#endif // LESS_THAN_OP diff --git a/src/operators/compare_op.h b/src/operators/compare_op.h new file mode 100644 index 0000000000000000000000000000000000000000..83ae88eae948694a8eef0845da699f1af8e2b8a9 --- /dev/null +++ b/src/operators/compare_op.h @@ -0,0 +1,30 @@ +/* Copyright (c) 2018 PaddlePaddle Authors. All Rights Reserved. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +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. */ + +#pragma once + +#include +#include "framework/operator.h" +#include "operators/kernel/compare_kernel.h" +#include "operators/op_param.h" + +namespace paddle_mobile { +namespace operators { + +#ifdef LESS_THAN_OP +DECLARE_OPERATOR(LessThan, CompareParam, LessThanKernel); +#endif // LESS_THAN_OP + +} // namespace operators +} // namespace paddle_mobile diff --git a/src/operators/kernel/arm/compare_kernel.cpp b/src/operators/kernel/arm/compare_kernel.cpp new file mode 100644 index 0000000000000000000000000000000000000000..c2ba6b583a1f80c3545f565464ff614ba4aaf52e --- /dev/null +++ b/src/operators/kernel/arm/compare_kernel.cpp @@ -0,0 +1,209 @@ +/* Copyright (c) 2018 PaddlePaddle Authors. All Rights Reserved. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +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 "operators/kernel/compare_kernel.h" +#if defined(__ARM_NEON__) || defined(__ARM_NEON) +#include +#endif + +namespace paddle_mobile { +namespace operators { + +typedef enum { + LESS_THAN = 0, + LESS_EQUAL = 1, + GREATER_THAN = 2, + GREATER_EQUAL = 3, + EQUAL = 4, + NOT_EQUAL = 5, +} CompareType; + +#if defined(__ARM_NEON__) || defined(__ARM_NEON) +template +inline uint32x4_t vcmpq_f32(const float32x4_t x, const float32x4_t y) { + return vcleq_f32(x, y); +} +#endif + +template +inline uint8_t Compare(const float x, const float y) { + return static_cast(x < y); +} + +template +inline uint8_t Compare(const int64_t x, const int64_t y) { + return static_cast(x < y); +} + +template +struct CompareCompute { + void operator()(const Tensor *X, const Tensor *Y, const int Axis, + Tensor *Out) {} +}; + +template +struct CompareCompute { + void operator()(const Tensor *X, const Tensor *Y, const int Axis, + Tensor *Out) { + const float *x = X->data(); + const float *y = Y->data(); + uint8_t *output = reinterpret_cast(Out->mutable_data()); + const auto &x_dims = X->dims(); + const auto &y_dims = Y->dims(); + /// axis = -1 represent the last dimensions. + int axis = (Axis == -1 ? x_dims.size() - y_dims.size() : Axis); + int batch = 1; + int channels = 1; + int elementwise_num = 1; + for (int i = 0; i < axis; ++i) { + batch *= x_dims[i]; + } + for (int i = 0; i < y_dims.size(); ++i) { + channels *= y_dims[i]; + } + for (int i = y_dims.size() + axis; i < x_dims.size(); ++i) { + elementwise_num *= x_dims[i]; + } + // if elementwise_num == 1, compare rowwise + if (elementwise_num == 1) { + int remain_start = 0; +#if defined(__ARM_NEON__) || defined(__ARM_NEON) + remain_start = channels & 0xfff8; + uint8x8_t __mask = vdup_n_u8(0x1); + for (int i = 0; i < batch; ++i) { + for (int j = 0; j < channels - 7; j += 8) { + int x_offset = i * channels + j; + float32x4_t __x0 = vld1q_f32(x + x_offset); + float32x4_t __x1 = vld1q_f32(x + x_offset + 4); + float32x4_t __y0 = vld1q_f32(y + j); + float32x4_t __y1 = vld1q_f32(y + j + 4); + uint32x4_t __cmp0 = vcmpq_f32(__x0, __y0); + uint32x4_t __cmp1 = vcmpq_f32(__x1, __y1); + uint16x4_t __ncmp0 = vmovn_u32(__cmp0); + uint16x4_t __ncmp1 = vmovn_u32(__cmp1); + uint16x8_t __ncmp = vcombine_u16(__ncmp0, __ncmp1); + uint8x8_t __nncmp = vmovn_u16(__ncmp); + __nncmp = vand_u8(__nncmp, __mask); + vst1_u8(output + x_offset, __nncmp); + } + } +#endif // __ARM_NEON__ + for (int i = 0; i < batch; ++i) { + for (int j = remain_start; j < channels; ++j) { + int x_offset = i * channels + j; + output[x_offset] = Compare(x[x_offset], y[j]); + } + } + } else { + for (int i = 0; i < batch; ++i) { + for (int j = 0; j < channels; ++j) { + int x_offset = (i * channels + j) * elementwise_num; + int y_offset = j * elementwise_num; + int remain_start = 0; +#if defined(__ARM_NEON__) || defined(__ARM_NEON) + remain_start = elementwise_num & 0xfff8; + uint8x8_t __mask = vdup_n_u8(0x1); + for (int k = 0; k < elementwise_num - 7; k += 8) { + float32x4_t __x0 = vld1q_f32(x + x_offset); + float32x4_t __x1 = vld1q_f32(x + x_offset + 4); + float32x4_t __y0 = vld1q_f32(y + y_offset); + uint32x4_t __cmp0 = vcmpq_f32(__x0, __y0); + uint32x4_t __cmp1 = vcmpq_f32(__x1, __y0); + uint16x4_t __ncmp0 = vmovn_u32(__cmp0); + uint16x4_t __ncmp1 = vmovn_u32(__cmp1); + uint16x8_t __ncmp = vcombine_u16(__ncmp0, __ncmp1); + uint8x8_t __nncmp = vmovn_u16(__ncmp); + __nncmp = vand_u8(__nncmp, __mask); + vst1_u8(output + x_offset, __nncmp); + x_offset += 8; + y_offset += 8; + } +#endif // __ARM_NEON__ + for (int k = remain_start; k < elementwise_num; ++k) { + output[x_offset + k] = Compare(x[x_offset + k], y[y_offset]); + } + } + } + } + } +}; + +template +struct CompareCompute { + void operator()(const Tensor *X, const Tensor *Y, const int Axis, + Tensor *Out) { + const int64_t *x = X->data(); + const int64_t *y = Y->data(); + uint8_t *output = reinterpret_cast(Out->mutable_data()); + const auto &x_dims = X->dims(); + const auto &y_dims = Y->dims(); + /// axis = -1 represent the last dimensions. + int axis = (Axis == -1 ? x_dims.size() - y_dims.size() : Axis); + int batch = 1; + int channels = 1; + int elementwise_num = 1; + for (int i = 0; i < axis; ++i) { + batch *= x_dims[i]; + } + for (int i = 0; i < y_dims.size(); ++i) { + channels *= y_dims[i]; + } + for (int i = y_dims.size() + axis; i < x_dims.size(); ++i) { + elementwise_num *= x_dims[i]; + } + // if elementwise_num == 1, compare rowwise + if (elementwise_num == 1) { + for (int i = 0; i < batch; ++i) { + for (int j = 0; j < channels; ++j) { + int x_offset = i * channels + j; + output[x_offset] = Compare(x[x_offset], y[j]); + } + } + } else { + for (int i = 0; i < batch; ++i) { + for (int j = 0; j < channels; ++j) { + int x_offset = (i * channels + j) * elementwise_num; + int y_offset = j * elementwise_num; + for (int k = 0; k < elementwise_num; ++k) { + output[x_offset + k] = Compare(x[x_offset + k], y[y_offset]); + } + } + } + } + } +}; + +#ifdef LESS_THAN_OP +template <> +bool LessThanKernel::Init(CompareParam *param) { + return true; +} + +template <> +void LessThanKernel::Compute(const CompareParam ¶m) { + if (param.input_x_->type() == typeid(int64_t)) { + CompareCompute()(param.input_x_, param.input_y_, + param.axis_, param.output_); + } else if (param.input_x_->type() == typeid(float)) { + CompareCompute()(param.input_x_, param.input_y_, + param.axis_, param.output_); + } else { + PADDLE_MOBILE_THROW_EXCEPTION( + "LessThan only support int64_t and float data type."); + } +} +#endif // LESS_THAN_OP + +} // namespace operators +} // namespace paddle_mobile diff --git a/src/operators/kernel/compare_kernel.h b/src/operators/kernel/compare_kernel.h new file mode 100644 index 0000000000000000000000000000000000000000..0f8a8ac74131054ebca53566dc38c374e9dafe4c --- /dev/null +++ b/src/operators/kernel/compare_kernel.h @@ -0,0 +1,28 @@ +/* Copyright (c) 2018 PaddlePaddle Authors. All Rights Reserved. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +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. */ + +#pragma once + +#include "framework/operator.h" +#include "operators/op_param.h" + +namespace paddle_mobile { +namespace operators { + +#ifdef LESS_THAN_OP +DECLARE_KERNEL(LessThan, CompareParam); +#endif // LESS_THAN_OP + +} // namespace operators +} // namespace paddle_mobile diff --git a/src/operators/kernel/kernels.h b/src/operators/kernel/kernels.h index 7b57d50668861416f341248b40715432a501bb35..668344674c91827775b78b71c4c50ada6d7108b7 100644 --- a/src/operators/kernel/kernels.h +++ b/src/operators/kernel/kernels.h @@ -21,15 +21,15 @@ namespace paddle_mobile { namespace operators { #ifdef TOP_K_OP -DECLARE_KERNEL(TopK, TopKParam) +DECLARE_KERNEL(TopK, TopKParam); #endif // TOP_K_OP #ifdef CAST_OP -DECLARE_KERNEL(Cast, CastParam) +DECLARE_KERNEL(Cast, CastParam); #endif // CAST_OP #ifdef LOD_RESET_OP -DECLARE_KERNEL(LodReset, LodResetParam) +DECLARE_KERNEL(LodReset, LodResetParam); #endif // LOD_RESET_OP } // namespace operators diff --git a/src/operators/lod_reset_op.cpp b/src/operators/lod_reset_op.cpp index 871129002da3b38d9f471beccc1c687c5849a21e..1eca733fd95e0893fe811da15290ee5878d1ac77 100644 --- a/src/operators/lod_reset_op.cpp +++ b/src/operators/lod_reset_op.cpp @@ -21,7 +21,7 @@ namespace operators { template void LodResetOp::InferShape() const { - const auto &input_dims = this->param_.input_->dims(); + const auto &input_dims = this->param_.input_x_->dims(); this->param_.output_->Resize(input_dims); } diff --git a/src/operators/op_param.h b/src/operators/op_param.h index e19856fdeb3140aefea169926fc03670219327b7..d20075f89195bbbd3a35577b19f84b9bb91b2a1c 100644 --- a/src/operators/op_param.h +++ b/src/operators/op_param.h @@ -2856,5 +2856,28 @@ class LodResetParam : public OpParam { }; #endif // LOD_RESET_OP +#ifdef LESS_THAN_OP +template +class CompareParam : public OpParam { + typedef typename DtypeTensorTrait::gtype GType; + typedef typename DtypeTensorTrait::rtype RType; + + public: + CompareParam(const VariableNameMap &inputs, const VariableNameMap &outputs, + const AttributeMap &attrs, const Scope &scope) { + input_x_ = InputXFrom(inputs, scope); + input_y_ = InputYFrom(inputs, scope); + output_ = OutFrom(outputs, scope); + axis_ = OpParam::GetAttr("axis", attrs); + } + + public: + GType *input_x_; + GType *input_y_; + GType *output_; + int axis_; +}; +#endif // LESS_THAN_OP + } // namespace operators } // namespace paddle_mobile diff --git a/test/CMakeLists.txt b/test/CMakeLists.txt index c1166db57587b587ca048e676ab5355eedc0dc99..75ad5348aa0743c081e2df3a59c190fa4c9a64f9 100644 --- a/test/CMakeLists.txt +++ b/test/CMakeLists.txt @@ -250,6 +250,9 @@ if (NOT FOUND_MATCH) ADD_EXECUTABLE(test-cast-op operators/test_cast_op.cpp test_helper.h test_include.h) target_link_libraries(test-cast-op paddle-mobile) + ADD_EXECUTABLE(test-less-than-op operators/test_less_than_op.cpp test_helper.h test_include.h) + target_link_libraries(test-less-than-op paddle-mobile) + # gen test ADD_EXECUTABLE(test-fc-op operators/test_fusion_fc_op.cpp test_helper.h test_include.h) target_link_libraries(test-fc-op paddle-mobile) diff --git a/test/operators/test_less_than_op.cpp b/test/operators/test_less_than_op.cpp new file mode 100644 index 0000000000000000000000000000000000000000..5c8fa8910d9abfb6ac0a834d9d00274b35fc790b --- /dev/null +++ b/test/operators/test_less_than_op.cpp @@ -0,0 +1,122 @@ +/* Copyright (c) 2018 PaddlePaddle Authors. All Rights Reserved. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +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 +#include +#include "../test_include.h" +#include "operators/compare_op.h" + +namespace paddle_mobile { + +template +void LessThan(const framework::Tensor *X, const framework::Tensor *Y, + const int Axis, framework::Tensor *Out) { + const T *x = X->data(); + const T *y = Y->data(); + bool *output = Out->mutable_data(); + const auto &x_dims = X->dims(); + const auto &y_dims = Y->dims(); + /// axis = -1 represent the last dimensions. + int axis = (Axis == -1 ? x_dims.size() - y_dims.size() : Axis); + int batch = 1; + int channels = 1; + int elementwise_num = 1; + for (int i = 0; i < axis; ++i) { + batch *= x_dims[i]; + } + for (int i = 0; i < y_dims.size(); ++i) { + channels *= y_dims[i]; + } + for (int i = y_dims.size() + axis; i < x_dims.size(); ++i) { + elementwise_num *= x_dims[i]; + } + // less than + for (int i = 0; i < batch; ++i) { + for (int j = 0; j < channels; ++j) { + int x_offset = (i * channels + j) * elementwise_num; + int y_offset = j * elementwise_num; + for (int k = 0; k < elementwise_num; ++k) { + output[x_offset + k] = (x[x_offset + k] < y[y_offset]); + } + } + } +} + +template +int TestLessThanOp(const std::vector &x_shape, + const std::vector &y_shape, const int axis) { + framework::DDim xdims = framework::make_ddim(x_shape); + framework::DDim ydims = framework::make_ddim(y_shape); + VariableNameMap inputs; + VariableNameMap outputs; + auto scope = std::make_shared(); + inputs["X"] = std::vector({"inputx"}); + inputs["Y"] = std::vector({"inputy"}); + outputs["Out"] = std::vector({"output"}); + + auto inputx_var = scope.get()->Var("inputx"); + auto inputx = inputx_var->template GetMutable(); + SetupTensor(inputx, xdims, static_cast(-100), static_cast(100)); + auto inputy_var = scope.get()->Var("inputy"); + auto inputy = inputy_var->template GetMutable(); + SetupTensor(inputy, ydims, static_cast(-100), static_cast(100)); + + auto output_var = scope.get()->Var("output"); + + framework::AttributeMap attrs; + attrs["axis"].Set(axis); + auto *op = new operators::LessThanOp("less_than", inputs, outputs, + attrs, scope); + op->InferShape(); + op->Init(); + op->Run(); + + auto output = output_var->template Get(); + + framework::Tensor output_cmp; + bool *output_cmp_data = output_cmp.mutable_data(output->dims()); + LessThan(inputx, inputy, axis, &output_cmp); + + const bool *output_data = output->data(); + for (int i = 0; i < output->numel(); ++i) { + if (output_data[i] != output_cmp_data[i]) { + LOG(kLOG_INFO) << "output_data[" << i << "] = " << output_data[i] + << ", output_cmp_data[" << i + << "] = " << output_cmp_data[i]; + delete op; + exit(1); + } + } + delete op; + return 0; +} + +} // namespace paddle_mobile + +int main() { + paddle_mobile::TestLessThanOp({1, 2, 3}, {1, 2, 3}, 0); + paddle_mobile::TestLessThanOp({10, 2, 1}, {10, 2, 1}, 0); + + paddle_mobile::TestLessThanOp({2, 10, 1}, {1, 10, 1}, 1); + paddle_mobile::TestLessThanOp({10, 2, 1}, {1, 2, 1}, 1); + + paddle_mobile::TestLessThanOp({1, 2, 3}, {1, 2, 3}, 0); + paddle_mobile::TestLessThanOp({10, 2, 1}, {10, 2, 1}, 0); + + paddle_mobile::TestLessThanOp({2, 10, 1}, {1, 10, 1}, 1); + paddle_mobile::TestLessThanOp({10, 2, 1}, {1, 2, 1}, 1); + + std::cout << "test less_than op pass." << std::endl; + return 0; +} diff --git a/tools/build.sh b/tools/build.sh index 5cb0887e3aa43038f1cc58c915f1e6809dd4ac89..117e74adeafae185693c751462454ecbefbd48ca 100755 --- a/tools/build.sh +++ b/tools/build.sh @@ -173,7 +173,7 @@ build_error() { } if [ $# -lt 1 ]; then - echo "error: target missing!" + echo "error: target missing!" echo "available targets: ios|android" echo "sample usage: ./build.sh android" else diff --git a/tools/op.cmake b/tools/op.cmake index d4c5d7b61df4c130879b639c38e23c7093d55a2b..c29d6eb0f4d1324be75a69b0c29aac56b76ed421 100644 --- a/tools/op.cmake +++ b/tools/op.cmake @@ -278,6 +278,8 @@ if(NOT FOUND_MATCH) set(SEQUENCE_SOFTMAX_OP ON) set(LOG_OP ON) set(TANH_OP ON) + set(LOD_RESET_OP ON) + set(LESS_THAN_OP ON) endif() # option(BATCHNORM_OP "" ON) @@ -517,6 +519,12 @@ endif() if (LOG_OP) add_definitions(-DLOG_OP) endif() +if (LOD_RESET_OP) + add_definitions(-DLOD_RESET_OP) +endif() +if (LESS_THAN_OP) + add_definitions(-DLESS_THAN_OP) +endif() if (TANH_OP) add_definitions(-DTANH_OP)