From 849a6874ad6d3b2a0a25237728ffcd0a15de06de Mon Sep 17 00:00:00 2001 From: nhzlx Date: Tue, 16 Oct 2018 16:22:05 +0000 Subject: [PATCH 01/22] fix googlenet bug with relu --- .../inference/tensorrt/convert/conv2d_op.cc | 21 ++++++++++++++++++- paddle/fluid/inference/tensorrt/engine.h | 10 +++++++++ 2 files changed, 30 insertions(+), 1 deletion(-) diff --git a/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc b/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc index 0a37d3968c..c8fc0bedfd 100644 --- a/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc +++ b/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc @@ -18,6 +18,21 @@ namespace paddle { namespace inference { namespace tensorrt { +bool if_skip_merging_optimize(TensorRTEngine* engine_, + const std::vector& filters, + const std::vector& strides, + const std::vector& paddings, + std::string input_name) { + if (engine_->itensor_quote_num[input_name] > 0) { + return true; + } + if (filters[0] == 1 && filters[1] == 1 && strides[0] == 1 && + strides[1] == 1 && paddings[0] == 0 && paddings[1] == 0) + engine_->itensor_quote_num[input_name] += 1; + + return false; +} + class Conv2dOpConverter : public OpConverter { public: void operator()(const framework::proto::OpDesc& op, @@ -31,6 +46,7 @@ class Conv2dOpConverter : public OpConverter { PADDLE_ENFORCE_EQ(op_desc.Output("Output").size(), 1); auto* X = engine_->GetITensor(op_desc.Input("Input").front()); + // Declare weights auto* Y_v = scope.FindVar(op_desc.Input("Filter").front()); PADDLE_ENFORCE_NOT_NULL(Y_v); @@ -83,7 +99,10 @@ class Conv2dOpConverter : public OpConverter { std::move(weight_tensor); layer->getOutput(0)->setName(output_name.c_str()); engine_->SetITensor(output_name, layer->getOutput(0)); - if (test_mode) { + + if (test_mode || + if_skip_merging_optimize(engine_, {filter_h, filter_w}, strides, + paddings, op_desc.Input("Input").front())) { engine_->DeclareOutput(output_name); } } diff --git a/paddle/fluid/inference/tensorrt/engine.h b/paddle/fluid/inference/tensorrt/engine.h index bd3ba4cea6..e828d2077d 100644 --- a/paddle/fluid/inference/tensorrt/engine.h +++ b/paddle/fluid/inference/tensorrt/engine.h @@ -132,6 +132,16 @@ class TensorRTEngine : public EngineBase { std::unordered_map> weight_map; + // TODO: (NHZLX) + // In the normal case, the paddle-trt exists bug when runing the googlenet. + // When there are more than two convolutions of 1 * 1 with the same input, the + // paddle-tensorrt will do the merging optimization, which fuse those conv + // into + // one conv, and then trigger bug. So, We should use strategy to avoid this + // optimization for the time being. This bug will be fixed in the future. + std::unordered_map + itensor_quote_num; + private: // the max batch size int max_batch_; From 9e4e9e9b6e21bbfdfa9b441badde28908ed36a0d Mon Sep 17 00:00:00 2001 From: Qiao Longfei Date: Tue, 6 Nov 2018 10:17:08 +0800 Subject: [PATCH 02/22] clean rpc server profiler --- .../distributed/grpc_variable_response.cc | 6 +++- .../distributed/request_handler_impl.cc | 1 - .../fluid/operators/distributed/rpc_server.cc | 32 ------------------- .../fluid/operators/distributed/rpc_server.h | 16 ---------- paddle/fluid/operators/listen_and_serv_op.cc | 1 - paddle/fluid/platform/profiler.cc | 2 +- python/paddle/fluid/__init__.py | 1 - 7 files changed, 6 insertions(+), 53 deletions(-) diff --git a/paddle/fluid/operators/distributed/grpc_variable_response.cc b/paddle/fluid/operators/distributed/grpc_variable_response.cc index 34d47f3ec0..eda4c45d3b 100644 --- a/paddle/fluid/operators/distributed/grpc_variable_response.cc +++ b/paddle/fluid/operators/distributed/grpc_variable_response.cc @@ -22,6 +22,9 @@ #include "paddle/fluid/operators/distributed/grpc_variable_response.h" #include "paddle/fluid/platform/profiler.h" +DEFINE_string(rpc_server_profile_path, "/tmp/profile_ps", + "the profile log file path"); + namespace paddle { namespace operators { namespace distributed { @@ -289,7 +292,8 @@ int GRPCVariableResponse::Parse(Source* source) { // TODO(panyx0718): Should we allow to customize file dir. platform::DisableProfiler( platform::EventSortingKey::kDefault, - string::Sprintf("/tmp/profile_ps_%lld", listener_id)); + string::Sprintf("%s_%lld", FLAGS_rpc_server_profile_path, + listener_id)); } break; } diff --git a/paddle/fluid/operators/distributed/request_handler_impl.cc b/paddle/fluid/operators/distributed/request_handler_impl.cc index 849e412504..a89ae59666 100644 --- a/paddle/fluid/operators/distributed/request_handler_impl.cc +++ b/paddle/fluid/operators/distributed/request_handler_impl.cc @@ -50,7 +50,6 @@ bool RequestSendHandler::Handle(const std::string& varname, // Async if (!sync_mode_) { VLOG(3) << "async process var: " << varname; - rpc_server_->Profiler().OneStep(); try { executor_->RunPreparedContext((*grad_to_prepared_ctx_)[varname].get(), scope); diff --git a/paddle/fluid/operators/distributed/rpc_server.cc b/paddle/fluid/operators/distributed/rpc_server.cc index 084480ae48..3e30ed4ac8 100644 --- a/paddle/fluid/operators/distributed/rpc_server.cc +++ b/paddle/fluid/operators/distributed/rpc_server.cc @@ -20,42 +20,10 @@ #include "paddle/fluid/operators/distributed/rpc_server.h" #include "paddle/fluid/platform/profiler.h" -DEFINE_int32(rpc_server_profile_period, 0, - "the period of listen_and_serv to do profile"); -DEFINE_string(rpc_server_profile_path, "/dev/null", - "the profile log file path"); - namespace paddle { namespace operators { namespace distributed { -RPCServerProfiler::RPCServerProfiler(int profile_period, - const std::string& profile_log_path) - : profile_period_(profile_period), profile_log_path_(profile_log_path) { - step_ = 0; -} - -void RPCServerProfiler::OneStep() { - PADDLE_ENFORCE_LE(step_, profile_period_, - "step_ should not be larger then " - "profile_period_"); - if (profile_period_ <= 0) { - return; - } - - if (step_ == 0) { - auto pf_state = paddle::platform::ProfilerState::kCPU; - paddle::platform::EnableProfiler(pf_state); - } - if (step_ == profile_period_) { - paddle::platform::DisableProfiler(paddle::platform::EventSortingKey::kTotal, - profile_log_path_); - step_ = 0; - } else { - step_++; - } -} - void RPCServer::ShutDown() { LOG(INFO) << "RPCServer ShutDown "; ShutDownImpl(); diff --git a/paddle/fluid/operators/distributed/rpc_server.h b/paddle/fluid/operators/distributed/rpc_server.h index f3e61e1575..c6934f8ace 100644 --- a/paddle/fluid/operators/distributed/rpc_server.h +++ b/paddle/fluid/operators/distributed/rpc_server.h @@ -23,30 +23,16 @@ #include "paddle/fluid/operators/distributed/request_handler.h" -DECLARE_int32(rpc_server_profile_period); DECLARE_string(rpc_server_profile_path); namespace paddle { namespace operators { namespace distributed { -class RPCServerProfiler { - public: - RPCServerProfiler(int profile_period, const std::string& profile_log_path); - void OneStep(); - - private: - const int profile_period_; - std::string profile_log_path_; - int step_; -}; - class RPCServer { public: explicit RPCServer(const std::string& address, int client_num) : cur_cond_(0), - profiler_(FLAGS_rpc_server_profile_period, - FLAGS_rpc_server_profile_path), bind_address_(address), exit_flag_(false), selected_port_(0), @@ -86,7 +72,6 @@ class RPCServer { void Complete(); void ResetBarrierCounter(); - RPCServerProfiler& Profiler() { return profiler_; } bool NeedResetAllVars(); @@ -101,7 +86,6 @@ class RPCServer { std::unordered_map rpc_cond_map_; std::atomic cur_cond_; std::condition_variable rpc_cond_; - RPCServerProfiler profiler_; protected: std::string bind_address_; diff --git a/paddle/fluid/operators/listen_and_serv_op.cc b/paddle/fluid/operators/listen_and_serv_op.cc index a038bad701..7e8a0225c6 100644 --- a/paddle/fluid/operators/listen_and_serv_op.cc +++ b/paddle/fluid/operators/listen_and_serv_op.cc @@ -134,7 +134,6 @@ void ListenAndServOp::RunSyncLoop( rpc_service_->ResetBarrierCounter(); while (true) { - rpc_service_->Profiler().OneStep(); // Get from multiple trainers, we don't care about the order in which // the gradients arrives, just add suffix 0~n and merge the gradient. rpc_service_->SetCond(distributed::kRequestSend); diff --git a/paddle/fluid/platform/profiler.cc b/paddle/fluid/platform/profiler.cc index da46a1abe1..56bf9e31a3 100644 --- a/paddle/fluid/platform/profiler.cc +++ b/paddle/fluid/platform/profiler.cc @@ -226,7 +226,7 @@ RecordBlock::~RecordBlock() { void EnableProfiler(ProfilerState state) { PADDLE_ENFORCE(state != ProfilerState::kDisabled, - "Can't enbale profling, since the input state is ", + "Can't enable profiling, since the input state is ", "ProfilerState::kDisabled"); std::lock_guard l(profiler_mu); diff --git a/python/paddle/fluid/__init__.py b/python/paddle/fluid/__init__.py index 737c8be814..c4cfd8e468 100644 --- a/python/paddle/fluid/__init__.py +++ b/python/paddle/fluid/__init__.py @@ -118,7 +118,6 @@ def __bootstrap__(): ] if core.is_compiled_with_dist(): read_env_flags.append('rpc_deadline') - read_env_flags.append('rpc_server_profile_period') read_env_flags.append('rpc_server_profile_path') read_env_flags.append('enable_rpc_profiler') read_env_flags.append('rpc_send_thread_num') From bb09e310204b4cd5016da96f33c017aeb052c8c5 Mon Sep 17 00:00:00 2001 From: tensor-tang Date: Tue, 6 Nov 2018 05:29:21 +0000 Subject: [PATCH 03/22] add vadd jitcode test=develop --- paddle/fluid/operators/math/jit_code.cc | 36 ++++++ paddle/fluid/operators/math/jit_code.h | 24 ++++ paddle/fluid/operators/math/jit_kernel.h | 2 +- .../fluid/operators/math/jit_kernel_blas.cc | 118 ++++++++++-------- paddle/fluid/operators/math/jit_kernel_rnn.cc | 10 +- .../fluid/operators/math/jit_kernel_test.cc | 10 +- 6 files changed, 135 insertions(+), 65 deletions(-) diff --git a/paddle/fluid/operators/math/jit_code.cc b/paddle/fluid/operators/math/jit_code.cc index 9e2cc18c7a..9375ca2067 100644 --- a/paddle/fluid/operators/math/jit_code.cc +++ b/paddle/fluid/operators/math/jit_code.cc @@ -66,6 +66,42 @@ void VMulJitCode::generate() { ret(); } +bool VAddJitCode::init(int d) { return MayIUse(avx); } + +void VAddJitCode::generate() { + int offset = 0; + for (int i = 0; i < num_ / AVX_FLOAT_BLOCK; ++i) { + vmovups(ymm_src1, ptr[param1 + offset]); + vmovups(ymm_src2, ptr[param2 + offset]); + vaddps(ymm_dst, ymm_src1, ymm_src2); + vmovups(ptr[param3 + offset], ymm_dst); + offset += sizeof(float) * AVX_FLOAT_BLOCK; + } + int rest = num_ % AVX_FLOAT_BLOCK; + if (rest >= 4) { + vmovups(xmm_src1, ptr[param1 + offset]); + vmovups(xmm_src2, ptr[param2 + offset]); + vaddps(xmm_dst, xmm_src1, xmm_src2); + vmovups(ptr[param3 + offset], xmm_dst); + offset += sizeof(float) * 4; + rest -= 4; + } + if (rest >= 2) { + vmovq(xmm_src1, ptr[param1 + offset]); + vmovq(xmm_src2, ptr[param2 + offset]); + vaddps(xmm_dst, xmm_src1, xmm_src2); + vmovq(ptr[param3 + offset], xmm_dst); + offset += sizeof(float) * 2; + rest -= 2; + } + if (rest > 0) { + vmovss(xmm_src1, ptr[param1 + offset]); + vmovss(xmm_src2, ptr[param2 + offset]); + vaddss(xmm_dst, xmm_src1, xmm_src2); + vmovss(ptr[param3 + offset], xmm_dst); + } + ret(); +} } // namespace gen } // namespace jitkernel } // namespace math diff --git a/paddle/fluid/operators/math/jit_code.h b/paddle/fluid/operators/math/jit_code.h index 6007b29081..0c4b75d030 100644 --- a/paddle/fluid/operators/math/jit_code.h +++ b/paddle/fluid/operators/math/jit_code.h @@ -53,6 +53,30 @@ class VMulJitCode : public JitCode { ymm_t ymm_dst = ymm_t(2); }; +class VAddJitCode : public JitCode { + public: + DECLARE_JIT_CODE(VAddJitCode); + explicit VAddJitCode(int d, size_t code_size = 256 * 1024, + void* code_ptr = nullptr) + : JitCode(code_size, code_ptr), num_(d) {} + static bool init(int d); + void generate() override; + + private: + int num_; + reg64_t param1{abi_param1}; + reg64_t param2{abi_param2}; + reg64_t param3{abi_param3}; + + xmm_t xmm_src1 = xmm_t(0); + xmm_t xmm_src2 = xmm_t(1); + xmm_t xmm_dst = xmm_t(2); + + ymm_t ymm_src1 = ymm_t(0); + ymm_t ymm_src2 = ymm_t(1); + ymm_t ymm_dst = ymm_t(2); +}; + } // namespace gen } // namespace jitkernel } // namespace math diff --git a/paddle/fluid/operators/math/jit_kernel.h b/paddle/fluid/operators/math/jit_kernel.h index 7b6027aa26..7c3fb5de9b 100644 --- a/paddle/fluid/operators/math/jit_kernel.h +++ b/paddle/fluid/operators/math/jit_kernel.h @@ -71,7 +71,7 @@ class VMulKernel : public Kernel { template class VAddKernel : public Kernel { public: - virtual void Compute(const T *x, const T *y, T *z) const = 0; + void (*Compute)(const T *, const T *, T *, int); }; template diff --git a/paddle/fluid/operators/math/jit_kernel_blas.cc b/paddle/fluid/operators/math/jit_kernel_blas.cc index 7d38d51172..16eab62dda 100644 --- a/paddle/fluid/operators/math/jit_kernel_blas.cc +++ b/paddle/fluid/operators/math/jit_kernel_blas.cc @@ -39,6 +39,13 @@ void VMulRefer(const T* x, const T* y, T* z, int n) { } } +template +void VAddRefer(const T* x, const T* y, T* z, int n) { + for (int i = 0; i < n; ++i) { + z[i] = x[i] + y[i]; + } +} + #ifdef PADDLE_WITH_MKLML template void VMulMKL(const T* x, const T* y, T* z, int n); @@ -47,22 +54,38 @@ template <> void VMulMKL(const float* x, const float* y, float* z, int n) { platform::dynload::vsMul(n, x, y, z); } + template <> void VMulMKL(const double* x, const double* y, double* z, int n) { platform::dynload::vdMul(n, x, y, z); } + +template +void VAddMKL(const T* x, const T* y, T* z, int n); + +template <> +void VAddMKL(const float* x, const float* y, float* z, int n) { + platform::dynload::vsAdd(n, x, y, z); +} + +template <> +void VAddMKL(const double* x, const double* y, double* z, int n) { + platform::dynload::vdAdd(n, x, y, z); +} #endif +#define DECLARE_STATIC_FUNC \ + static inline std::string name(int d) { \ + PADDLE_THROW("DType should be either float or double"); \ + } \ + static inline bool useJIT(int d) { return false; } \ + static inline bool useMKL(int d) { return false; } + /* VMUL JitKernel */ template class VMulKernelImpl : public VMulKernel { public: - static inline std::string name(int d) { - PADDLE_THROW("DType should be either float or double"); - } - static inline bool useJIT(int d) { return false; } - static inline bool useMKL(int d) { return false; } - + DECLARE_STATIC_FUNC; explicit VMulKernelImpl(int d) : VMulKernel() { if (useJIT(d)) { // roughly estimate the size of code @@ -100,63 +123,51 @@ bool VMulKernelImpl::useMKL(int d) { return true; } -REGISTER_JITKERNEL(vmul, VMulKernel); - -/* VADD JitKernel */ -template +/* VAdd JitKernel */ +template class VAddKernelImpl : public VAddKernel { public: - explicit VAddKernelImpl(int d) : VAddKernel() { this->num_ = d; } - void Compute(const T* x, const T* y, T* z) const override { - for (int i = 0; i < this->num_; ++i) { - z[i] = x[i] + y[i]; + DECLARE_STATIC_FUNC; + explicit VAddKernelImpl(int d) : VAddKernel() { + if (useJIT(d)) { + size_t sz = 96 + d / AVX_FLOAT_BLOCK * 4 * 8; + jitcode_.reset(new gen::VAddJitCode(d, sz > 4096 ? sz : 4096)); + this->Compute = + jitcode_->getCode(); + return; } +#ifdef PADDLE_WITH_MKLML + if (useMKL(d)) { + this->Compute = VAddMKL; + return; + } +#endif + this->Compute = VAddRefer; } + + private: + std::unique_ptr jitcode_{nullptr}; }; -#ifdef PADDLE_WITH_MKLML -#define MKL_FLOAT(isa, block) \ - template <> \ - void VAddKernelImpl::Compute( \ - const float* x, const float* y, float* z) const { \ - platform::dynload::vsAdd(this->num_, x, y, z); \ - } +template <> +bool VAddKernelImpl::useJIT(int d) { + return gen::VAddJitCode::init(d); +} -#define MKL_DOUBLE(isa, block) \ - template <> \ - void VAddKernelImpl::Compute( \ - const double* x, const double* y, double* z) const { \ - platform::dynload::vdAdd(this->num_, x, y, z); \ - } +template <> +bool VAddKernelImpl::useMKL(int d) { + return d > 512; +} -FOR_EACH_ISA(MKL_FLOAT, kGT16); -FOR_EACH_ISA_BLOCK(MKL_DOUBLE); -#endif +template <> +bool VAddKernelImpl::useMKL(int d) { + return true; +} -#define INTRI8_FLOAT(isa) \ - template <> \ - void VAddKernelImpl::Compute( \ - const float* x, const float* y, float* z) const { \ - __m256 tmpx, tmpy; \ - tmpx = _mm256_loadu_ps(x); \ - tmpy = _mm256_loadu_ps(y); \ - tmpx = _mm256_add_ps(tmpx, tmpy); \ - _mm256_storeu_ps(z, tmpx); \ - } -#ifdef __AVX__ -INTRI8_FLOAT(jit::avx); -#endif -#ifdef __AVX2__ -INTRI8_FLOAT(jit::avx2); -#endif -#ifdef __AVX512F__ -INTRI8_FLOAT(jit::avx512f); -#endif -// TODO(TJ): eq16 test and complete avx512 +#undef DECLARE_STATIC_FUNC -#undef INTRI8_FLOAT -#undef MKL_FLOAT -#undef MKL_DOUBLE +REGISTER_JITKERNEL(vmul, VMulKernel); +REGISTER_JITKERNEL(vadd, VAddKernel); /* VSCAL JitKernel */ template @@ -480,7 +491,6 @@ INTRI_COMMON_FLOAT(jit::avx512f, kGT16); #undef INTRI16_FLOAT #undef INTRI_COMMON_FLOAT -REGISTER_JITKERNEL_DEPRECATED(vadd, VAddKernel); REGISTER_JITKERNEL_DEPRECATED(vscal, VScalKernel); REGISTER_JITKERNEL_DEPRECATED(vaddb, VAddBiasKernel); REGISTER_JITKERNEL_DEPRECATED(vrelu, VReluKernel); diff --git a/paddle/fluid/operators/math/jit_kernel_rnn.cc b/paddle/fluid/operators/math/jit_kernel_rnn.cc index d0932a37bb..ba3e917377 100644 --- a/paddle/fluid/operators/math/jit_kernel_rnn.cc +++ b/paddle/fluid/operators/math/jit_kernel_rnn.cc @@ -181,7 +181,7 @@ class LSTMKernelImpl : public LSTMKernel { act_cand_d_->Compute(gates, gates); vmul_d_->Compute(gates, gates + d_, gates + d_, d_); vmul_d_->Compute(ct_1, gates + d2_, gates + d2_, d_); - vadd_d_->Compute(gates + d_, gates + d2_, ct); + vadd_d_->Compute(gates + d_, gates + d2_, ct, d_); /* H_t = act_cell(C_t) * ogated */ act_cell_d_->Compute(ct, gates + d2_); @@ -291,16 +291,16 @@ class PeepholeKernelImpl : public LSTMKernel { /* get fgated and igated*/ vmul_d_->Compute(wp_data, ct_1, checked, d_); vmul_d_->Compute(wp_data + d_, ct_1, checked + d_, d_); - vadd_d2_->Compute(checked, gates + d_, gates + d_); + vadd_d2_->Compute(checked, gates + d_, gates + d_, d2_); act_gate_d2_->Compute(gates + d_, gates + d_); /* C_t = C_t-1 * fgated + cand_gated * igated*/ act_cand_d_->Compute(gates, gates); vmul_d_->Compute(gates, gates + d_, gates + d_, d_); vmul_d_->Compute(ct_1, gates + d2_, gates + d2_, d_); - vadd_d_->Compute(gates + d_, gates + d2_, ct); + vadd_d_->Compute(gates + d_, gates + d2_, ct, d_); /* get ogated*/ vmul_d_->Compute(wp_data + d2_, ct, gates + d_, d_); - vadd_d_->Compute(gates + d_, gates + d3_, gates + d3_); + vadd_d_->Compute(gates + d_, gates + d3_, gates + d3_, d_); act_gate_d_->Compute(gates + d3_, gates + d3_); /* H_t = act_cell(C_t) * ogated */ act_cell_d_->Compute(ct, gates + d2_); @@ -314,7 +314,7 @@ class PeepholeKernelImpl : public LSTMKernel { vmul_d_->Compute(gates, gates + d_, ct, d_); /* get outgated, put W_oc * C_t on igated */ vmul_d_->Compute(wp_data + d2_, ct, gates + d_, d_); - vadd_d_->Compute(gates + d_, gates + d3_, gates + d3_); + vadd_d_->Compute(gates + d_, gates + d3_, gates + d3_, d_); /* H_t = act_cell(C_t) * ogated */ act_gate_d_->Compute(gates + d3_, gates + d3_); act_cell_d_->Compute(ct, gates + d2_); diff --git a/paddle/fluid/operators/math/jit_kernel_test.cc b/paddle/fluid/operators/math/jit_kernel_test.cc index 667a95fe1a..f9064d8b2f 100644 --- a/paddle/fluid/operators/math/jit_kernel_test.cc +++ b/paddle/fluid/operators/math/jit_kernel_test.cc @@ -371,7 +371,7 @@ void lstm_ctht_better( vtanh_d->Compute(gates, gates); vmul_d->Compute(gates, gates + d, gates + d, d); vmul_d->Compute(ct_1, gates + d2, gates + d2, d); - vadd_d->Compute(gates + d, gates + d2, ct); + vadd_d->Compute(gates + d, gates + d2, ct, d); /* H_t = act_cell(C_t) * ogated */ vtanh_d->Compute(ct, gates + d2); vmul_d->Compute(gates + d2, gates + d * 3, ht, d); @@ -695,7 +695,7 @@ TEST(JitKernel, vadd) { auto ttgts = GetCurrentUS(); for (int i = 0; i < repeat; ++i) { - ker->Compute(x_data, y_data, ztgt_data); + ker->Compute(x_data, y_data, ztgt_data, d); } auto ttgte = GetCurrentUS(); @@ -723,8 +723,8 @@ void vaddrelu_better( const paddle::operators::math::jitkernel::VAddKernel>& vadd, const std::shared_ptr< const paddle::operators::math::jitkernel::VReluKernel>& vrelu, - const float* x, const float* y, float* z) { - vadd->Compute(x, y, z); + const float* x, const float* y, float* z, int d) { + vadd->Compute(x, y, z, d); vrelu->Compute(z, z); } @@ -752,7 +752,7 @@ TEST(JitKernel, vaddrelu) { auto trefe = GetCurrentUS(); auto tmkls = GetCurrentUS(); for (int i = 0; i < repeat; ++i) { - vaddrelu_better(vadd, vrelu, x_data, y_data, zref_data); + vaddrelu_better(vadd, vrelu, x_data, y_data, zref_data, d); } auto tmkle = GetCurrentUS(); auto ttgts = GetCurrentUS(); From 1fb1a0bc6b32b8d97a5d5f95e0f38cbdd6c67ca1 Mon Sep 17 00:00:00 2001 From: Shan Yi <35982308+shanyi15@users.noreply.github.com> Date: Tue, 6 Nov 2018 14:11:57 +0800 Subject: [PATCH 04/22] fix_recordio_internal_link test=develop --- python/paddle/fluid/recordio_writer.py | 3 --- 1 file changed, 3 deletions(-) diff --git a/python/paddle/fluid/recordio_writer.py b/python/paddle/fluid/recordio_writer.py index a69c0c29d4..076a942cdd 100644 --- a/python/paddle/fluid/recordio_writer.py +++ b/python/paddle/fluid/recordio_writer.py @@ -41,9 +41,6 @@ def convert_reader_to_recordio_file( """ Convert a Python Reader to a recordio file. - Please see :ref:`api_guide_python_reader` and :ref:`api_guide_reader_op` for - details. - Examples: >>> import paddle.fluid as fluid From b68ececb7327aab332b5a07346d73286bf4a8d74 Mon Sep 17 00:00:00 2001 From: tensor-tang Date: Tue, 6 Nov 2018 07:03:06 +0000 Subject: [PATCH 05/22] add vaddrelu jitcode test=develop --- paddle/fluid/operators/math/jit_code.cc | 15 +++ paddle/fluid/operators/math/jit_code.h | 15 ++- paddle/fluid/operators/math/jit_kernel.h | 12 +- .../fluid/operators/math/jit_kernel_blas.cc | 124 +++++------------- .../fluid/operators/math/jit_kernel_test.cc | 2 +- 5 files changed, 66 insertions(+), 102 deletions(-) diff --git a/paddle/fluid/operators/math/jit_code.cc b/paddle/fluid/operators/math/jit_code.cc index 9375ca2067..35f0bdb9b3 100644 --- a/paddle/fluid/operators/math/jit_code.cc +++ b/paddle/fluid/operators/math/jit_code.cc @@ -70,10 +70,16 @@ bool VAddJitCode::init(int d) { return MayIUse(avx); } void VAddJitCode::generate() { int offset = 0; + if (with_relu_) { + vxorps(ymm_zero, ymm_zero, ymm_zero); + } for (int i = 0; i < num_ / AVX_FLOAT_BLOCK; ++i) { vmovups(ymm_src1, ptr[param1 + offset]); vmovups(ymm_src2, ptr[param2 + offset]); vaddps(ymm_dst, ymm_src1, ymm_src2); + if (with_relu_) { + vmaxps(ymm_dst, ymm_zero, ymm_dst); + } vmovups(ptr[param3 + offset], ymm_dst); offset += sizeof(float) * AVX_FLOAT_BLOCK; } @@ -82,6 +88,9 @@ void VAddJitCode::generate() { vmovups(xmm_src1, ptr[param1 + offset]); vmovups(xmm_src2, ptr[param2 + offset]); vaddps(xmm_dst, xmm_src1, xmm_src2); + if (with_relu_) { + vmaxps(xmm_dst, xmm_zero, xmm_dst); + } vmovups(ptr[param3 + offset], xmm_dst); offset += sizeof(float) * 4; rest -= 4; @@ -90,6 +99,9 @@ void VAddJitCode::generate() { vmovq(xmm_src1, ptr[param1 + offset]); vmovq(xmm_src2, ptr[param2 + offset]); vaddps(xmm_dst, xmm_src1, xmm_src2); + if (with_relu_) { + vmaxps(xmm_dst, xmm_zero, xmm_dst); + } vmovq(ptr[param3 + offset], xmm_dst); offset += sizeof(float) * 2; rest -= 2; @@ -98,6 +110,9 @@ void VAddJitCode::generate() { vmovss(xmm_src1, ptr[param1 + offset]); vmovss(xmm_src2, ptr[param2 + offset]); vaddss(xmm_dst, xmm_src1, xmm_src2); + if (with_relu_) { + vmaxps(xmm_dst, xmm_zero, xmm_dst); + } vmovss(ptr[param3 + offset], xmm_dst); } ret(); diff --git a/paddle/fluid/operators/math/jit_code.h b/paddle/fluid/operators/math/jit_code.h index 0c4b75d030..6bfed4b22d 100644 --- a/paddle/fluid/operators/math/jit_code.h +++ b/paddle/fluid/operators/math/jit_code.h @@ -46,35 +46,38 @@ class VMulJitCode : public JitCode { xmm_t xmm_src1 = xmm_t(0); xmm_t xmm_src2 = xmm_t(1); - xmm_t xmm_dst = xmm_t(2); + xmm_t xmm_dst = xmm_t(1); ymm_t ymm_src1 = ymm_t(0); ymm_t ymm_src2 = ymm_t(1); - ymm_t ymm_dst = ymm_t(2); + ymm_t ymm_dst = ymm_t(1); }; class VAddJitCode : public JitCode { public: DECLARE_JIT_CODE(VAddJitCode); - explicit VAddJitCode(int d, size_t code_size = 256 * 1024, + explicit VAddJitCode(int d, bool with_relu, size_t code_size = 256 * 1024, void* code_ptr = nullptr) - : JitCode(code_size, code_ptr), num_(d) {} + : JitCode(code_size, code_ptr), num_(d), with_relu_(with_relu) {} static bool init(int d); void generate() override; private: int num_; + bool with_relu_; reg64_t param1{abi_param1}; reg64_t param2{abi_param2}; reg64_t param3{abi_param3}; xmm_t xmm_src1 = xmm_t(0); xmm_t xmm_src2 = xmm_t(1); - xmm_t xmm_dst = xmm_t(2); + xmm_t xmm_dst = xmm_t(1); + xmm_t xmm_zero = xmm_t(2); ymm_t ymm_src1 = ymm_t(0); ymm_t ymm_src2 = ymm_t(1); - ymm_t ymm_dst = ymm_t(2); + ymm_t ymm_dst = ymm_t(1); + ymm_t ymm_zero = ymm_t(2); }; } // namespace gen diff --git a/paddle/fluid/operators/math/jit_kernel.h b/paddle/fluid/operators/math/jit_kernel.h index 7c3fb5de9b..04e0b81d3e 100644 --- a/paddle/fluid/operators/math/jit_kernel.h +++ b/paddle/fluid/operators/math/jit_kernel.h @@ -75,22 +75,22 @@ class VAddKernel : public Kernel { }; template -class VScalKernel : public Kernel { +class VAddReluKernel : public Kernel { public: - virtual void Compute(const T a, const T *x, T *y) const = 0; - virtual void Compute(const T a, T *x) const = 0; + void (*Compute)(const T *, const T *, T *, int); }; template -class VAddBiasKernel : public Kernel { +class VScalKernel : public Kernel { public: virtual void Compute(const T a, const T *x, T *y) const = 0; + virtual void Compute(const T a, T *x) const = 0; }; template -class VAddReluKernel : public Kernel { +class VAddBiasKernel : public Kernel { public: - virtual void Compute(const T *x, const T *y, T *z) const = 0; + virtual void Compute(const T a, const T *x, T *y) const = 0; }; template diff --git a/paddle/fluid/operators/math/jit_kernel_blas.cc b/paddle/fluid/operators/math/jit_kernel_blas.cc index 16eab62dda..b3ac33043b 100644 --- a/paddle/fluid/operators/math/jit_kernel_blas.cc +++ b/paddle/fluid/operators/math/jit_kernel_blas.cc @@ -46,6 +46,14 @@ void VAddRefer(const T* x, const T* y, T* z, int n) { } } +template +void VAddReluRefer(const T* x, const T* y, T* z, int n) { + for (int i = 0; i < n; ++i) { + z[i] = x[i] + y[i]; + z[i] = z[i] > 0 ? z[i] : 0; + } +} + #ifdef PADDLE_WITH_MKLML template void VMulMKL(const T* x, const T* y, T* z, int n); @@ -131,7 +139,7 @@ class VAddKernelImpl : public VAddKernel { explicit VAddKernelImpl(int d) : VAddKernel() { if (useJIT(d)) { size_t sz = 96 + d / AVX_FLOAT_BLOCK * 4 * 8; - jitcode_.reset(new gen::VAddJitCode(d, sz > 4096 ? sz : 4096)); + jitcode_.reset(new gen::VAddJitCode(d, false, sz > 4096 ? sz : 4096)); this->Compute = jitcode_->getCode(); return; @@ -164,10 +172,36 @@ bool VAddKernelImpl::useMKL(int d) { return true; } +/* VAddRelu JitKernel */ +template +class VAddReluKernelImpl : public VAddReluKernel { + public: + DECLARE_STATIC_FUNC; + explicit VAddReluKernelImpl(int d) : VAddReluKernel() { + if (useJIT(d)) { + size_t sz = 96 + d / AVX_FLOAT_BLOCK * 4 * 8; + jitcode_.reset(new gen::VAddJitCode(d, true, sz > 4096 ? sz : 4096)); + this->Compute = + jitcode_->getCode(); + return; + } + this->Compute = VAddReluRefer; + } + + private: + std::unique_ptr jitcode_{nullptr}; +}; + +template <> +bool VAddReluKernelImpl::useJIT(int d) { + return gen::VAddJitCode::init(d); +} + #undef DECLARE_STATIC_FUNC REGISTER_JITKERNEL(vmul, VMulKernel); REGISTER_JITKERNEL(vadd, VAddKernel); +REGISTER_JITKERNEL(vaddrelu, VAddReluKernel); /* VSCAL JitKernel */ template @@ -404,97 +438,9 @@ class VIdentityKernelImpl : public VIdentityKernel { void Compute(const T* x, T* y) const override {} }; -/* VAddRelu JitKernel */ -template -class VAddReluKernelImpl : public VAddReluKernel { - public: - explicit VAddReluKernelImpl(int d) : VAddReluKernel() { this->num_ = d; } - void Compute(const T* x, const T* y, T* z) const override { - for (int i = 0; i < this->num_; ++i) { - z[i] = x[i] + y[i]; - z[i] = z[i] > 0 ? z[i] : 0; - } - } -}; - -#define INTRI8_FLOAT(isa) \ - template <> \ - void VAddReluKernelImpl::Compute( \ - const float* x, const float* y, float* z) const { \ - __m256 tmpx = _mm256_loadu_ps(x); \ - __m256 tmpy = _mm256_loadu_ps(y); \ - tmpy = _mm256_add_ps(tmpx, tmpy); \ - tmpy = _mm256_max_ps(tmpy, _mm256_setzero_ps()); \ - _mm256_storeu_ps(z, tmpy); \ - } - -#define INTRI16_FLOAT(isa) \ - template <> \ - void VAddReluKernelImpl::Compute( \ - const float* x, const float* y, float* z) const { \ - __m256 zeros = _mm256_setzero_ps(); \ - __m256 tmp0 = _mm256_loadu_ps(x); \ - __m256 tmp1 = _mm256_loadu_ps(y); \ - tmp0 = _mm256_add_ps(tmp0, tmp1); \ - tmp0 = _mm256_max_ps(tmp0, zeros); \ - tmp1 = _mm256_loadu_ps(x + 8); \ - __m256 tmp2 = _mm256_loadu_ps(y + 8); \ - tmp1 = _mm256_add_ps(tmp1, tmp2); \ - tmp1 = _mm256_max_ps(tmp1, zeros); \ - _mm256_storeu_ps(z, tmp0); \ - _mm256_storeu_ps(z + 8, tmp1); \ - } - -#define INTRI_COMMON_FLOAT(isa, block) \ - template <> \ - VAddReluKernelImpl::VAddReluKernelImpl(int d) \ - : VAddReluKernel() { \ - this->num_ = d; \ - this->end_ = d - d % AVX_FLOAT_BLOCK; \ - this->rest_ = d - this->end_; \ - } \ - template <> \ - void VAddReluKernelImpl::Compute( \ - const float* x, const float* y, float* z) const { \ - __m256 zeros = _mm256_setzero_ps(); \ - for (int i = 0; i < this->end_; i += AVX_FLOAT_BLOCK) { \ - __m256 tmpx = _mm256_loadu_ps(x + i); \ - __m256 tmpy = _mm256_loadu_ps(y + i); \ - tmpy = _mm256_add_ps(tmpx, tmpy); \ - tmpy = _mm256_max_ps(tmpy, zeros); \ - _mm256_storeu_ps(z + i, tmpy); \ - } \ - for (int i = this->end_; i < this->num_; ++i) { \ - z[i] = x[i] + y[i]; \ - z[i] = z[i] > 0 ? z[i] : 0; \ - } \ - } - -#ifdef __AVX__ -INTRI8_FLOAT(jit::avx); -INTRI16_FLOAT(jit::avx); -INTRI_COMMON_FLOAT(jit::avx, kGT16); -#endif -#ifdef __AVX2__ -INTRI8_FLOAT(jit::avx2); -INTRI16_FLOAT(jit::avx2); -INTRI_COMMON_FLOAT(jit::avx2, kGT16); -#endif -#ifdef __AVX512F__ -// TODO(TJ): refine avx512 -INTRI8_FLOAT(jit::avx512f); -INTRI16_FLOAT(jit::avx512f); -INTRI_COMMON_FLOAT(jit::avx512f, kGT16); -#endif - -#undef INTRI8_FLOAT -#undef INTRI16_FLOAT -#undef INTRI_COMMON_FLOAT - REGISTER_JITKERNEL_DEPRECATED(vscal, VScalKernel); REGISTER_JITKERNEL_DEPRECATED(vaddb, VAddBiasKernel); REGISTER_JITKERNEL_DEPRECATED(vrelu, VReluKernel); -REGISTER_JITKERNEL_DEPRECATED(vaddrelu, VAddReluKernel); REGISTER_JITKERNEL_DEPRECATED(videntity, VIdentityKernel); } // namespace jitkernel diff --git a/paddle/fluid/operators/math/jit_kernel_test.cc b/paddle/fluid/operators/math/jit_kernel_test.cc index f9064d8b2f..d990a0a982 100644 --- a/paddle/fluid/operators/math/jit_kernel_test.cc +++ b/paddle/fluid/operators/math/jit_kernel_test.cc @@ -757,7 +757,7 @@ TEST(JitKernel, vaddrelu) { auto tmkle = GetCurrentUS(); auto ttgts = GetCurrentUS(); for (int i = 0; i < repeat; ++i) { - ker->Compute(x_data, y_data, ztgt_data); + ker->Compute(x_data, y_data, ztgt_data, d); } auto ttgte = GetCurrentUS(); VLOG(3) << "Vec size " << d << ": refer takes: " << (trefe - trefs) / repeat From 86b99ac95339226d75b615e549eb41ffa2e10cca Mon Sep 17 00:00:00 2001 From: nhzlx Date: Tue, 6 Nov 2018 09:43:43 +0000 Subject: [PATCH 06/22] fix comments and fix bug --- .../inference/tensorrt/convert/conv2d_op.cc | 4 ++-- paddle/fluid/inference/tensorrt/engine.cc | 4 ++++ paddle/fluid/inference/tensorrt/engine.h | 2 ++ .../inference/tests/api/trt_models_tester.cc | 17 +++++++++++------ paddle/fluid/operators/tensorrt_engine_op.h | 4 +++- 5 files changed, 22 insertions(+), 9 deletions(-) diff --git a/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc b/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc index c8fc0bedfd..7bcf2dd1ee 100644 --- a/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc +++ b/paddle/fluid/inference/tensorrt/convert/conv2d_op.cc @@ -18,7 +18,7 @@ namespace paddle { namespace inference { namespace tensorrt { -bool if_skip_merging_optimize(TensorRTEngine* engine_, +bool to_skip_merging_optimize(TensorRTEngine* engine_, const std::vector& filters, const std::vector& strides, const std::vector& paddings, @@ -101,7 +101,7 @@ class Conv2dOpConverter : public OpConverter { engine_->SetITensor(output_name, layer->getOutput(0)); if (test_mode || - if_skip_merging_optimize(engine_, {filter_h, filter_w}, strides, + to_skip_merging_optimize(engine_, {filter_h, filter_w}, strides, paddings, op_desc.Input("Input").front())) { engine_->DeclareOutput(output_name); } diff --git a/paddle/fluid/inference/tensorrt/engine.cc b/paddle/fluid/inference/tensorrt/engine.cc index 14e9e14d33..9e0f958447 100644 --- a/paddle/fluid/inference/tensorrt/engine.cc +++ b/paddle/fluid/inference/tensorrt/engine.cc @@ -133,6 +133,10 @@ void TensorRTEngine::DeclareOutput(const nvinfer1::ILayer *layer, int offset, buffer_sizes_[name] = 0; } +bool TensorRTEngine::HasDeclared(const std::string &name) { + return buffer_sizes_.count(name) > 0; +} + void TensorRTEngine::DeclareOutput(const std::string &name) { PADDLE_ENFORCE_EQ(0, buffer_sizes_.count(name), "duplicate output name %s", name); diff --git a/paddle/fluid/inference/tensorrt/engine.h b/paddle/fluid/inference/tensorrt/engine.h index e828d2077d..d9d3827321 100644 --- a/paddle/fluid/inference/tensorrt/engine.h +++ b/paddle/fluid/inference/tensorrt/engine.h @@ -91,6 +91,8 @@ class TensorRTEngine : public EngineBase { const std::string& name); // Set the itensor_map_[name] as the network's output, and set its name. void DeclareOutput(const std::string& name); + // Check if the ITensor has been declared + bool HasDeclared(const std::string& name); // GPU memory address for an ITensor with specific name. One can operate on // these memory directly for acceleration, for example, output the converted diff --git a/paddle/fluid/inference/tests/api/trt_models_tester.cc b/paddle/fluid/inference/tests/api/trt_models_tester.cc index bf320a0cbc..a5635f911a 100644 --- a/paddle/fluid/inference/tests/api/trt_models_tester.cc +++ b/paddle/fluid/inference/tests/api/trt_models_tester.cc @@ -96,11 +96,16 @@ void CompareTensorRTWithFluid(int batch_size, std::string model_dirname) { } } -TEST(trt_models_test, main) { - std::vector infer_models = {"mobilenet", "resnet50", - "resnext50"}; - for (auto &model_dir : infer_models) { - CompareTensorRTWithFluid(1, FLAGS_dirname + "/" + model_dir); - } +TEST(trt_models_test, mobilenet) { + CompareTensorRTWithFluid(1, FLAGS_dirname + "/mobilenet"); +} + +TEST(trt_models_test, resnet50) { + CompareTensorRTWithFluid(1, FLAGS_dirname + "/resnet50"); } + +TEST(trt_models_test, resnext50) { + CompareTensorRTWithFluid(1, FLAGS_dirname + "/resnext50"); +} + } // namespace paddle diff --git a/paddle/fluid/operators/tensorrt_engine_op.h b/paddle/fluid/operators/tensorrt_engine_op.h index d4ba0f9c33..673f86da76 100644 --- a/paddle/fluid/operators/tensorrt_engine_op.h +++ b/paddle/fluid/operators/tensorrt_engine_op.h @@ -223,7 +223,9 @@ class TensorRTEngineKernel : public framework::OpKernel { // Add outputs for (auto& output : output_maps) { - engine->DeclareOutput(output); + if (!engine->HasDeclared(output)) { + engine->DeclareOutput(output); + } } engine->FreezeNetwork(); From 2f9a5a2e0a1de26095e7d28298974389d9268360 Mon Sep 17 00:00:00 2001 From: Tao Luo Date: Tue, 6 Nov 2018 19:37:26 +0800 Subject: [PATCH 07/22] add analyzer_face_tester --- .../tests/api/analyzer_resnet50_tester.cc | 20 +------------ .../fluid/inference/tests/api/tester_helper.h | 30 +++++++++++++++++++ paddle/fluid/inference/tests/test_helper.h | 6 ++-- 3 files changed, 33 insertions(+), 23 deletions(-) diff --git a/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc b/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc index c2151eea08..cd04d888a5 100644 --- a/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc +++ b/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc @@ -30,25 +30,7 @@ void SetConfig(AnalysisConfig *cfg) { } void SetInput(std::vector> *inputs) { - PADDLE_ENFORCE_EQ(FLAGS_test_all_data, 0, "Only have single batch of data."); - - PaddleTensor input; - // channel=3, height/width=318 - std::vector shape({FLAGS_batch_size, 3, 318, 318}); - input.shape = shape; - input.dtype = PaddleDType::FLOAT32; - - // fill input data, for profile easily, do not use random data here. - size_t size = FLAGS_batch_size * 3 * 318 * 318; - input.data.Resize(size * sizeof(float)); - float *input_data = static_cast(input.data.data()); - for (size_t i = 0; i < size; i++) { - *(input_data + i) = static_cast(i) / size; - } - - std::vector input_slots; - input_slots.assign({input}); - (*inputs).emplace_back(input_slots); + SetFakeImageInput(inputs, FLAGS_infer_model); } // Easy for profiling independently. diff --git a/paddle/fluid/inference/tests/api/tester_helper.h b/paddle/fluid/inference/tests/api/tester_helper.h index 19c3f532d5..79468da03a 100644 --- a/paddle/fluid/inference/tests/api/tester_helper.h +++ b/paddle/fluid/inference/tests/api/tester_helper.h @@ -25,6 +25,7 @@ #include "paddle/fluid/inference/api/analysis_predictor.h" #include "paddle/fluid/inference/api/helper.h" #include "paddle/fluid/inference/api/paddle_inference_pass.h" +#include "paddle/fluid/inference/tests/test_helper.h" #include "paddle/fluid/platform/profiler.h" DEFINE_string(infer_model, "", "model path"); @@ -105,6 +106,35 @@ std::unordered_map GetFuseStatis(PaddlePredictor *predictor, return fuse_statis; } +void SetFakeImageInput(std::vector> *inputs, + const std::string &dirname, + const bool is_combined = true) { + // Set fake_image_data + PADDLE_ENFORCE_EQ(FLAGS_test_all_data, 0, "Only have single batch of data."); + std::vector> feed_target_shapes = + GetFeedTargetShapes(dirname, is_combined); + int dim1 = feed_target_shapes[0][1]; + int dim2 = feed_target_shapes[0][2]; + int dim3 = feed_target_shapes[0][3]; + + PaddleTensor input; + std::vector shape({FLAGS_batch_size, dim1, dim2, dim3}); + input.shape = shape; + input.dtype = PaddleDType::FLOAT32; + + // fill input data, for profile easily, do not use random data here. + size_t size = FLAGS_batch_size * dim1 * dim2 * dim3; + input.data.Resize(size * sizeof(float)); + float *input_data = static_cast(input.data.data()); + for (size_t i = 0; i < size; i++) { + *(input_data + i) = static_cast(i) / size; + } + + std::vector input_slots; + input_slots.assign({input}); + (*inputs).emplace_back(input_slots); +} + void TestOneThreadPrediction( const AnalysisConfig &config, const std::vector> &inputs, diff --git a/paddle/fluid/inference/tests/test_helper.h b/paddle/fluid/inference/tests/test_helper.h index 94f0550df5..e26094c0db 100644 --- a/paddle/fluid/inference/tests/test_helper.h +++ b/paddle/fluid/inference/tests/test_helper.h @@ -101,8 +101,8 @@ std::unique_ptr InitProgram( // Hard-coding the file names of program and parameters in unittest. // The file names should be consistent with that used in Python API // `fluid.io.save_inference_model`. - std::string prog_filename = "__model_combined__"; - std::string param_filename = "__params_combined__"; + std::string prog_filename = "model"; + std::string param_filename = "params"; inference_program = paddle::inference::Load(executor, scope, dirname + "/" + prog_filename, dirname + "/" + param_filename); @@ -261,5 +261,3 @@ void TestInference(const std::string& dirname, delete scope; } - -USE_PASS(graph_to_program_pass); From 2ec65ae0db4390addc9d0820947ca806682a5429 Mon Sep 17 00:00:00 2001 From: Tao Luo Date: Tue, 6 Nov 2018 21:00:32 +0800 Subject: [PATCH 08/22] download face_model in CMakeLists.txt test=develop --- paddle/fluid/inference/CMakeLists.txt | 2 +- .../fluid/inference/tests/api/CMakeLists.txt | 43 ++++++++++++++++--- paddle/fluid/inference/{ => tests}/test.cmake | 0 3 files changed, 37 insertions(+), 8 deletions(-) rename paddle/fluid/inference/{ => tests}/test.cmake (100%) diff --git a/paddle/fluid/inference/CMakeLists.txt b/paddle/fluid/inference/CMakeLists.txt index d31c8e3b7d..e5678cf607 100644 --- a/paddle/fluid/inference/CMakeLists.txt +++ b/paddle/fluid/inference/CMakeLists.txt @@ -1,5 +1,5 @@ if(WITH_TESTING) - include(test.cmake) # some generic cmake funtion for inference + include(tests/test.cmake) # some generic cmake funtion for inference endif() # analysis and tensorrt must be added before creating static library, # otherwise, there would be undefined reference to them in static library. diff --git a/paddle/fluid/inference/tests/api/CMakeLists.txt b/paddle/fluid/inference/tests/api/CMakeLists.txt index b57a26b470..88e632bf9d 100644 --- a/paddle/fluid/inference/tests/api/CMakeLists.txt +++ b/paddle/fluid/inference/tests/api/CMakeLists.txt @@ -1,5 +1,11 @@ set(INFERENCE_EXTRA_DEPS paddle_inference_api paddle_fluid_api ir_pass_manager analysis_predictor) +function(download_model install_dir model_name) + if (NOT EXISTS ${install_dir}) + inference_download_and_uncompress(${install_dir} ${INFERENCE_URL} ${model_name}) + endif() +endfunction() + function(download_model_and_data install_dir model_name data_name) if (NOT EXISTS ${install_dir}) inference_download_and_uncompress(${install_dir} ${INFERENCE_URL} ${model_name}) @@ -13,6 +19,13 @@ function(inference_analysis_api_test target install_dir filename) ARGS --infer_model=${install_dir}/model --infer_data=${install_dir}/data.txt) endfunction() +function(inference_analysis_api_test_with_fake_data target install_dir filename model_name) + download_model(${install_dir} ${model_name}) + inference_analysis_test(${target} SRCS ${filename} + EXTRA_DEPS ${INFERENCE_EXTRA_DEPS} + ARGS --infer_model=${install_dir}/model) +endfunction() + # RNN1 if(NOT APPLE) set(RNN1_INSTALL_DIR "${INFERENCE_DEMO_INSTALL_DIR}/rnn1") @@ -61,17 +74,33 @@ inference_analysis_api_test(test_analyzer_seq_conv1 ${SEQ_CONV1_INSTALL_DIR} ana # ocr set(OCR_INSTALL_DIR "${INFERENCE_DEMO_INSTALL_DIR}/ocr") if (NOT EXISTS ${OCR_INSTALL_DIR}) - inference_download_and_uncompress(${OCR_INSTALL_DIR} "http://paddlemodels.cdn.bcebos.com/" "inference-vis-demos%2Focr.tar.gz") + inference_download_and_uncompress(${OCR_INSTALL_DIR} "http://paddlemodels.cdn.bcebos.com/" "inference-vis-demos%2Focr.tar.gz") endif() inference_analysis_api_test(test_analyzer_ocr ${OCR_INSTALL_DIR} analyzer_vis_tester.cc) # resnet50 -set(RESNET50_INSTALL_DIR "${INFERENCE_DEMO_INSTALL_DIR}/resnet50") -if (NOT EXISTS ${RESNET50_INSTALL_DIR}) - inference_download_and_uncompress(${RESNET50_INSTALL_DIR} ${INFERENCE_URL} "resnet50_model.tar.gz") -endif() -inference_analysis_test(test_analyzer_resnet50 SRCS analyzer_resnet50_tester.cc - EXTRA_DEPS ${INFERENCE_EXTRA_DEPS} ARGS --infer_model=${RESNET50_INSTALL_DIR}/model) +inference_analysis_api_test_with_fake_data(test_analyzer_resnet50 + "${INFERENCE_DEMO_INSTALL_DIR}/resnet50" analyzer_resnet50_tester.cc "resnet50_model.tar.gz") + +# face +set(FACE_INSTALL_DIR "${INFERENCE_DEMO_INSTALL_DIR}/face") +inference_analysis_api_test_with_fake_data(test_analyzer_face_align1 + "${FACE_INSTALL_DIR}/align1" analyzer_face_tester.cc "face%2Falign1_model.tar.gz") +inference_analysis_api_test_with_fake_data(test_analyzer_face_align2 + "${FACE_INSTALL_DIR}/align2" analyzer_face_tester.cc "face%2Falign2_model.tar.gz") +inference_analysis_api_test_with_fake_data(test_analyzer_face_feature1 + "${FACE_INSTALL_DIR}/feature1" analyzer_face_tester.cc "face%2Ffeature_id_model.tar.gz") +# TODO(luotao): Disable this test due to analysis is timeout 10 minutes. +# inference_analysis_api_test_with_fake_data(test_analyzer_face_feature2 +# "${FACE_INSTALL_DIR}/feature2" analyzer_face_tester.cc "face%2Ffeature_life_model.tar.gz") +inference_analysis_api_test_with_fake_data(test_analyzer_face_detect + "${FACE_INSTALL_DIR}/detect" analyzer_face_tester.cc "face%2Fdetect_model.tar.gz") +inference_analysis_api_test_with_fake_data(test_analyzer_face_demark + "${FACE_INSTALL_DIR}/demark" analyzer_face_tester.cc "face%2Fdemark_model.tar.gz") +inference_analysis_api_test_with_fake_data(test_analyzer_face_score + "${FACE_INSTALL_DIR}/score" analyzer_face_tester.cc "face%2Fscore_model.tar.gz") +inference_analysis_api_test_with_fake_data(test_analyzer_face_super_res + "${FACE_INSTALL_DIR}/super_res" analyzer_face_tester.cc "face%2Fsuper_res_model.tar.gz") # anakin if (WITH_ANAKIN AND WITH_MKL) # only needed in CI diff --git a/paddle/fluid/inference/test.cmake b/paddle/fluid/inference/tests/test.cmake similarity index 100% rename from paddle/fluid/inference/test.cmake rename to paddle/fluid/inference/tests/test.cmake From 7a2887d212ed9a6d9f1f7e59bb38b1dec0d64279 Mon Sep 17 00:00:00 2001 From: Tao Luo Date: Tue, 6 Nov 2018 22:29:03 +0800 Subject: [PATCH 09/22] add analyzer_face_tester test=develop --- .../tests/api/analyzer_face_tester.cc | 69 +++++++++++++++++++ 1 file changed, 69 insertions(+) create mode 100644 paddle/fluid/inference/tests/api/analyzer_face_tester.cc diff --git a/paddle/fluid/inference/tests/api/analyzer_face_tester.cc b/paddle/fluid/inference/tests/api/analyzer_face_tester.cc new file mode 100644 index 0000000000..b7db8887d5 --- /dev/null +++ b/paddle/fluid/inference/tests/api/analyzer_face_tester.cc @@ -0,0 +1,69 @@ +/* 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 "paddle/fluid/inference/tests/api/tester_helper.h" + +namespace paddle { +namespace inference { +namespace analysis { + +void SetConfig(AnalysisConfig *cfg) { + cfg->param_file = FLAGS_infer_model + "/params"; + cfg->prog_file = FLAGS_infer_model + "/model"; + cfg->use_gpu = false; + cfg->device = 0; + cfg->enable_ir_optim = true; + cfg->specify_input_name = true; +} + +void SetInput(std::vector> *inputs) { + SetFakeImageInput(inputs, FLAGS_infer_model); +} + +// Easy for profiling independently. +TEST(Analyzer_face, profile) { + AnalysisConfig cfg; + SetConfig(&cfg); + std::vector outputs; + + std::vector> input_slots_all; + SetInput(&input_slots_all); + TestPrediction(cfg, input_slots_all, &outputs, FLAGS_num_threads); +} + +// Check the fuse status +TEST(Analyzer_face, fuse_statis) { + AnalysisConfig cfg; + SetConfig(&cfg); + int num_ops; + auto predictor = CreatePaddlePredictor(cfg); + auto fuse_statis = GetFuseStatis( + static_cast(predictor.get()), &num_ops); +} + +// Compare result of NativeConfig and AnalysisConfig +TEST(Analyzer_face, compare) { + AnalysisConfig cfg; + SetConfig(&cfg); + + std::vector> input_slots_all; + SetInput(&input_slots_all); + CompareNativeAndAnalysis(cfg, input_slots_all); +} + +} // namespace analysis +} // namespace inference +} // namespace paddle From cb4083b9fa1f61f2453f744f6b823e4a72ac0089 Mon Sep 17 00:00:00 2001 From: tensor-tang Date: Tue, 6 Nov 2018 16:37:19 +0000 Subject: [PATCH 10/22] fix compile error test=develop --- paddle/fluid/operators/math/fc_compute.h | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/paddle/fluid/operators/math/fc_compute.h b/paddle/fluid/operators/math/fc_compute.h index 87220d4019..b072b4c20a 100644 --- a/paddle/fluid/operators/math/fc_compute.h +++ b/paddle/fluid/operators/math/fc_compute.h @@ -36,7 +36,7 @@ inline void FCCompute(const BlasT& blas, const int M, .template Get>(N); for (int i = 0; i < M; i++) { T* dst = Y + i * N; - vaddrelu->Compute(B, dst, dst); + vaddrelu->Compute(B, dst, dst, N); } } else { const auto& vadd = jitkernel::KernelPool::Instance() @@ -47,7 +47,7 @@ inline void FCCompute(const BlasT& blas, const int M, #endif for (int i = 0; i < M; i++) { T* dst = Y + i * N; - vadd->Compute(B, dst, dst); + vadd->Compute(B, dst, dst, N); } } } From 1ead9318d5ffae709fb4842c41248e0e6c530011 Mon Sep 17 00:00:00 2001 From: Tao Luo Date: Wed, 7 Nov 2018 10:58:22 +0800 Subject: [PATCH 11/22] remove unused code in test_helper.h to pass ci test=develop --- paddle/fluid/inference/tests/test_helper.h | 11 ----------- 1 file changed, 11 deletions(-) diff --git a/paddle/fluid/inference/tests/test_helper.h b/paddle/fluid/inference/tests/test_helper.h index e26094c0db..00976a3992 100644 --- a/paddle/fluid/inference/tests/test_helper.h +++ b/paddle/fluid/inference/tests/test_helper.h @@ -18,7 +18,6 @@ limitations under the License. */ #include #include -#include "paddle/fluid/framework/ir/graph_to_program_pass.h" #include "paddle/fluid/framework/lod_tensor.h" #include "paddle/fluid/inference/io.h" #include "paddle/fluid/platform/profiler.h" @@ -136,15 +135,6 @@ std::vector> GetFeedTargetShapes( return feed_target_shapes; } -void Compile(paddle::framework::ProgramDesc* program) { - std::unique_ptr g( - new paddle::framework::ir::Graph(*program)); - auto pass = paddle::framework::ir::PassRegistry::Instance().Get( - "graph_to_program_pass"); - pass->SetNotOwned("program", program); - pass->Apply(std::move(g)); -} - template void TestInference(const std::string& dirname, const std::vector& cpu_feeds, @@ -182,7 +172,6 @@ void TestInference(const std::string& dirname, paddle::platform::DeviceContextPool::Instance().Get(place)); inference_program = InitProgram(&executor, scope, dirname, is_combined); } - Compile(inference_program.get()); // Disable the profiler and print the timing information paddle::platform::DisableProfiler(paddle::platform::EventSortingKey::kDefault, From 2b791f1f639e3e4706a56dd2ba7b686a081112ba Mon Sep 17 00:00:00 2001 From: Tao Luo Date: Wed, 7 Nov 2018 12:16:34 +0800 Subject: [PATCH 12/22] unify analyzer_face_tester to analyzer_resnet50_tester test=develop --- .../fluid/inference/tests/api/CMakeLists.txt | 20 ------ .../tests/api/analyzer_face_tester.cc | 69 ------------------- .../tests/api/analyzer_resnet50_tester.cc | 10 +-- 3 files changed, 1 insertion(+), 98 deletions(-) delete mode 100644 paddle/fluid/inference/tests/api/analyzer_face_tester.cc diff --git a/paddle/fluid/inference/tests/api/CMakeLists.txt b/paddle/fluid/inference/tests/api/CMakeLists.txt index 88e632bf9d..2ca84c8005 100644 --- a/paddle/fluid/inference/tests/api/CMakeLists.txt +++ b/paddle/fluid/inference/tests/api/CMakeLists.txt @@ -82,26 +82,6 @@ inference_analysis_api_test(test_analyzer_ocr ${OCR_INSTALL_DIR} analyzer_vis_te inference_analysis_api_test_with_fake_data(test_analyzer_resnet50 "${INFERENCE_DEMO_INSTALL_DIR}/resnet50" analyzer_resnet50_tester.cc "resnet50_model.tar.gz") -# face -set(FACE_INSTALL_DIR "${INFERENCE_DEMO_INSTALL_DIR}/face") -inference_analysis_api_test_with_fake_data(test_analyzer_face_align1 - "${FACE_INSTALL_DIR}/align1" analyzer_face_tester.cc "face%2Falign1_model.tar.gz") -inference_analysis_api_test_with_fake_data(test_analyzer_face_align2 - "${FACE_INSTALL_DIR}/align2" analyzer_face_tester.cc "face%2Falign2_model.tar.gz") -inference_analysis_api_test_with_fake_data(test_analyzer_face_feature1 - "${FACE_INSTALL_DIR}/feature1" analyzer_face_tester.cc "face%2Ffeature_id_model.tar.gz") -# TODO(luotao): Disable this test due to analysis is timeout 10 minutes. -# inference_analysis_api_test_with_fake_data(test_analyzer_face_feature2 -# "${FACE_INSTALL_DIR}/feature2" analyzer_face_tester.cc "face%2Ffeature_life_model.tar.gz") -inference_analysis_api_test_with_fake_data(test_analyzer_face_detect - "${FACE_INSTALL_DIR}/detect" analyzer_face_tester.cc "face%2Fdetect_model.tar.gz") -inference_analysis_api_test_with_fake_data(test_analyzer_face_demark - "${FACE_INSTALL_DIR}/demark" analyzer_face_tester.cc "face%2Fdemark_model.tar.gz") -inference_analysis_api_test_with_fake_data(test_analyzer_face_score - "${FACE_INSTALL_DIR}/score" analyzer_face_tester.cc "face%2Fscore_model.tar.gz") -inference_analysis_api_test_with_fake_data(test_analyzer_face_super_res - "${FACE_INSTALL_DIR}/super_res" analyzer_face_tester.cc "face%2Fsuper_res_model.tar.gz") - # anakin if (WITH_ANAKIN AND WITH_MKL) # only needed in CI # anakin rnn1 diff --git a/paddle/fluid/inference/tests/api/analyzer_face_tester.cc b/paddle/fluid/inference/tests/api/analyzer_face_tester.cc deleted file mode 100644 index b7db8887d5..0000000000 --- a/paddle/fluid/inference/tests/api/analyzer_face_tester.cc +++ /dev/null @@ -1,69 +0,0 @@ -/* 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 "paddle/fluid/inference/tests/api/tester_helper.h" - -namespace paddle { -namespace inference { -namespace analysis { - -void SetConfig(AnalysisConfig *cfg) { - cfg->param_file = FLAGS_infer_model + "/params"; - cfg->prog_file = FLAGS_infer_model + "/model"; - cfg->use_gpu = false; - cfg->device = 0; - cfg->enable_ir_optim = true; - cfg->specify_input_name = true; -} - -void SetInput(std::vector> *inputs) { - SetFakeImageInput(inputs, FLAGS_infer_model); -} - -// Easy for profiling independently. -TEST(Analyzer_face, profile) { - AnalysisConfig cfg; - SetConfig(&cfg); - std::vector outputs; - - std::vector> input_slots_all; - SetInput(&input_slots_all); - TestPrediction(cfg, input_slots_all, &outputs, FLAGS_num_threads); -} - -// Check the fuse status -TEST(Analyzer_face, fuse_statis) { - AnalysisConfig cfg; - SetConfig(&cfg); - int num_ops; - auto predictor = CreatePaddlePredictor(cfg); - auto fuse_statis = GetFuseStatis( - static_cast(predictor.get()), &num_ops); -} - -// Compare result of NativeConfig and AnalysisConfig -TEST(Analyzer_face, compare) { - AnalysisConfig cfg; - SetConfig(&cfg); - - std::vector> input_slots_all; - SetInput(&input_slots_all); - CompareNativeAndAnalysis(cfg, input_slots_all); -} - -} // namespace analysis -} // namespace inference -} // namespace paddle diff --git a/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc b/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc index cd04d888a5..e5c8dfd22a 100644 --- a/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc +++ b/paddle/fluid/inference/tests/api/analyzer_resnet50_tester.cc @@ -43,13 +43,6 @@ void profile(bool use_mkldnn = false) { std::vector> input_slots_all; SetInput(&input_slots_all); TestPrediction(cfg, input_slots_all, &outputs, FLAGS_num_threads); - - if (FLAGS_num_threads == 1 && !FLAGS_test_all_data) { - PADDLE_ENFORCE_EQ(outputs.size(), 1UL); - size_t size = GetSize(outputs[0]); - // output is a 512-dimension feature - EXPECT_EQ(size, 512 * FLAGS_batch_size); - } } TEST(Analyzer_resnet50, profile) { profile(); } @@ -65,8 +58,7 @@ TEST(Analyzer_resnet50, fuse_statis) { auto predictor = CreatePaddlePredictor(cfg); auto fuse_statis = GetFuseStatis( static_cast(predictor.get()), &num_ops); - ASSERT_TRUE(fuse_statis.count("fc_fuse")); - EXPECT_EQ(fuse_statis.at("fc_fuse"), 1); + LOG(INFO) << "num_ops: " << num_ops; } // Compare result of NativeConfig and AnalysisConfig From 3b8dd9ebbd5eee808d8fa891c2a91f4bd1680910 Mon Sep 17 00:00:00 2001 From: Qiao Longfei Date: Wed, 7 Nov 2018 12:41:50 +0800 Subject: [PATCH 13/22] optimize code test=develop --- paddle/fluid/operators/distributed/grpc_variable_response.cc | 4 ---- paddle/fluid/operators/distributed/rpc_server.h | 2 -- paddle/fluid/operators/distributed/variable_response.cc | 3 +++ paddle/fluid/operators/distributed/variable_response.h | 2 ++ 4 files changed, 5 insertions(+), 6 deletions(-) diff --git a/paddle/fluid/operators/distributed/grpc_variable_response.cc b/paddle/fluid/operators/distributed/grpc_variable_response.cc index 7076bae205..d6d219d436 100644 --- a/paddle/fluid/operators/distributed/grpc_variable_response.cc +++ b/paddle/fluid/operators/distributed/grpc_variable_response.cc @@ -22,9 +22,6 @@ #include "paddle/fluid/operators/distributed/grpc_variable_response.h" #include "paddle/fluid/platform/profiler.h" -DEFINE_string(rpc_server_profile_path, "/tmp/profile_ps", - "the profile log file path"); - namespace paddle { namespace operators { namespace distributed { @@ -289,7 +286,6 @@ int GRPCVariableResponse::Parse(Source* source) { platform::EnableProfiler(platform::ProfilerState::kCPU); } else if (profiling == platform::kDisableProfiler && platform::IsProfileEnabled()) { - // TODO(panyx0718): Should we allow to customize file dir. platform::DisableProfiler( platform::EventSortingKey::kDefault, string::Sprintf("%s_%lld", FLAGS_rpc_server_profile_path, diff --git a/paddle/fluid/operators/distributed/rpc_server.h b/paddle/fluid/operators/distributed/rpc_server.h index c6934f8ace..c78c5007a7 100644 --- a/paddle/fluid/operators/distributed/rpc_server.h +++ b/paddle/fluid/operators/distributed/rpc_server.h @@ -23,8 +23,6 @@ #include "paddle/fluid/operators/distributed/request_handler.h" -DECLARE_string(rpc_server_profile_path); - namespace paddle { namespace operators { namespace distributed { diff --git a/paddle/fluid/operators/distributed/variable_response.cc b/paddle/fluid/operators/distributed/variable_response.cc index c4854d50b6..b2f73b67dc 100644 --- a/paddle/fluid/operators/distributed/variable_response.cc +++ b/paddle/fluid/operators/distributed/variable_response.cc @@ -16,6 +16,9 @@ #include #include "paddle/fluid/operators/distributed/sendrecvop_utils.h" +DEFINE_string(rpc_server_profile_path, "./profile_ps", + "the profile log file path"); + namespace paddle { namespace operators { namespace distributed { diff --git a/paddle/fluid/operators/distributed/variable_response.h b/paddle/fluid/operators/distributed/variable_response.h index f20a6038ce..4c7fcbbdfb 100644 --- a/paddle/fluid/operators/distributed/variable_response.h +++ b/paddle/fluid/operators/distributed/variable_response.h @@ -27,6 +27,8 @@ #include "paddle/fluid/framework/tensor.h" #include "paddle/fluid/operators/distributed/send_recv.pb.h" +DECLARE_string(rpc_server_profile_path); + namespace paddle { namespace operators { namespace distributed { From e564eb341ff0b79d8ffeb89f6380538113ba2387 Mon Sep 17 00:00:00 2001 From: tangwei12 Date: Wed, 7 Nov 2018 13:28:13 +0800 Subject: [PATCH 14/22] Fix mkdir conflict in save_inference_model (#14285) * fix mkdir conflict test=develop --- python/paddle/fluid/io.py | 9 +++++++-- 1 file changed, 7 insertions(+), 2 deletions(-) diff --git a/python/paddle/fluid/io.py b/python/paddle/fluid/io.py index 22c60c1cbe..8936d884dd 100644 --- a/python/paddle/fluid/io.py +++ b/python/paddle/fluid/io.py @@ -65,7 +65,7 @@ def is_persistable(var): Examples: .. code-block:: python - param = fluid.default_main_program().global_block().var('fc.w') + param = fluid.default_main_program().global_block().var('fc.b') res = fluid.io.is_persistable(param) """ if var.desc.type() == core.VarDesc.VarType.FEED_MINIBATCH or \ @@ -625,8 +625,13 @@ def save_inference_model(dirname, main_program._distributed_lookup_table, main_program._endpoints) - if not os.path.isdir(dirname): + # when a pserver and a trainer running on the same machine, mkdir may conflict + try: os.makedirs(dirname) + except OSError as e: + if e.errno != errno.EEXIST: + raise + if model_filename is not None: model_basename = os.path.basename(model_filename) else: From eea36739ccc2f5cde74a13ee8dd46da4de1d2223 Mon Sep 17 00:00:00 2001 From: Tao Luo Date: Wed, 7 Nov 2018 13:40:54 +0800 Subject: [PATCH 15/22] refine test_helper.h test=develop --- paddle/fluid/inference/tests/api/tester_helper.h | 5 ++--- paddle/fluid/inference/tests/test_helper.h | 13 ++++++++----- 2 files changed, 10 insertions(+), 8 deletions(-) diff --git a/paddle/fluid/inference/tests/api/tester_helper.h b/paddle/fluid/inference/tests/api/tester_helper.h index 79468da03a..8c5888d8da 100644 --- a/paddle/fluid/inference/tests/api/tester_helper.h +++ b/paddle/fluid/inference/tests/api/tester_helper.h @@ -107,12 +107,11 @@ std::unordered_map GetFuseStatis(PaddlePredictor *predictor, } void SetFakeImageInput(std::vector> *inputs, - const std::string &dirname, - const bool is_combined = true) { + const std::string &dirname) { // Set fake_image_data PADDLE_ENFORCE_EQ(FLAGS_test_all_data, 0, "Only have single batch of data."); std::vector> feed_target_shapes = - GetFeedTargetShapes(dirname, is_combined); + GetFeedTargetShapes(dirname, true, "model", "params"); int dim1 = feed_target_shapes[0][1]; int dim2 = feed_target_shapes[0][2]; int dim3 = feed_target_shapes[0][3]; diff --git a/paddle/fluid/inference/tests/test_helper.h b/paddle/fluid/inference/tests/test_helper.h index 00976a3992..2118fcfd4b 100644 --- a/paddle/fluid/inference/tests/test_helper.h +++ b/paddle/fluid/inference/tests/test_helper.h @@ -93,15 +93,15 @@ void CheckError(const paddle::framework::LoDTensor& output1, std::unique_ptr InitProgram( paddle::framework::Executor* executor, paddle::framework::Scope* scope, - const std::string& dirname, const bool is_combined = false) { + const std::string& dirname, const bool is_combined = false, + const std::string& prog_filename = "__model_combined__", + const std::string& param_filename = "__params_combined__") { std::unique_ptr inference_program; if (is_combined) { // All parameters are saved in a single file. // Hard-coding the file names of program and parameters in unittest. // The file names should be consistent with that used in Python API // `fluid.io.save_inference_model`. - std::string prog_filename = "model"; - std::string param_filename = "params"; inference_program = paddle::inference::Load(executor, scope, dirname + "/" + prog_filename, dirname + "/" + param_filename); @@ -114,12 +114,15 @@ std::unique_ptr InitProgram( } std::vector> GetFeedTargetShapes( - const std::string& dirname, const bool is_combined = false) { + const std::string& dirname, const bool is_combined = false, + const std::string& prog_filename = "__model_combined__", + const std::string& param_filename = "__params_combined__") { auto place = paddle::platform::CPUPlace(); auto executor = paddle::framework::Executor(place); auto* scope = new paddle::framework::Scope(); - auto inference_program = InitProgram(&executor, scope, dirname, is_combined); + auto inference_program = InitProgram(&executor, scope, dirname, is_combined, + prog_filename, param_filename); auto& global_block = inference_program->Block(0); const std::vector& feed_target_names = From ffc866159fcdf23bc38ce00e9af84cd80fed26e9 Mon Sep 17 00:00:00 2001 From: chengduo Date: Wed, 7 Nov 2018 14:50:10 +0800 Subject: [PATCH 16/22] hot fix log (#14293) test=develop --- paddle/fluid/operators/math/cross_entropy.cu | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/paddle/fluid/operators/math/cross_entropy.cu b/paddle/fluid/operators/math/cross_entropy.cu index a651e0265a..cb200ec8d6 100644 --- a/paddle/fluid/operators/math/cross_entropy.cu +++ b/paddle/fluid/operators/math/cross_entropy.cu @@ -28,7 +28,7 @@ __device__ __forceinline__ double real_log(double x) { return log(x); } __device__ __forceinline__ platform::float16 real_log( const platform::float16& val) { - return static_cast(hlog(static_cast(val))); + return static_cast(logf(static_cast(val))); } template From c9730d33d914b4032da1d8a6b411237fa7e6236d Mon Sep 17 00:00:00 2001 From: tensor-tang Date: Wed, 7 Nov 2018 07:24:28 +0000 Subject: [PATCH 17/22] fix run error on mac test=develop --- paddle/fluid/platform/init.cc | 4 +++- 1 file changed, 3 insertions(+), 1 deletion(-) diff --git a/paddle/fluid/platform/init.cc b/paddle/fluid/platform/init.cc index 07abe1dd5c..2211e55043 100644 --- a/paddle/fluid/platform/init.cc +++ b/paddle/fluid/platform/init.cc @@ -116,6 +116,7 @@ void InitDevices(bool init_p2p, const std::vector devices) { platform::SetNumThreads(FLAGS_paddle_num_threads); #endif +#if !defined(_WIN32) && !defined(__APPLE__) && !defined(__OSX__) if (platform::jit::MayIUse(platform::jit::avx)) { #ifndef __AVX__ LOG(WARNING) << "AVX is available, Please re-compile on local machine"; @@ -157,8 +158,9 @@ void InitDevices(bool init_p2p, const std::vector devices) { AVX_GUIDE(AVX, NonAVX); } #endif - #undef AVX_GUIDE + +#endif } void InitGLOG(const std::string &prog_name) { From c28beb8a3ce3471cd19cd96e69f6e0d2eb13f008 Mon Sep 17 00:00:00 2001 From: Yu Yang Date: Wed, 7 Nov 2018 15:50:28 +0800 Subject: [PATCH 18/22] test(Pe): add dry run tests for pe (#14254) Dry run tests will skip `Op.Run` and just perform job scheduling. It helps to analysis dead lock in PE. test=develop --- .../framework/details/execution_strategy.h | 2 + .../fast_threaded_ssa_graph_executor.cc | 4 +- .../details/threaded_ssa_graph_executor.cc | 4 +- .../details/threaded_ssa_graph_executor.h | 2 +- paddle/fluid/framework/parallel_executor.cc | 23 +++--- paddle/fluid/pybind/pybind.cc | 7 +- python/paddle/fluid/layers/io.py | 2 +- .../test_parallel_executor_dry_run.py | 80 +++++++++++++++++++ 8 files changed, 108 insertions(+), 16 deletions(-) create mode 100644 python/paddle/fluid/tests/unittests/test_parallel_executor_dry_run.py diff --git a/paddle/fluid/framework/details/execution_strategy.h b/paddle/fluid/framework/details/execution_strategy.h index 5183be878e..15c496130c 100644 --- a/paddle/fluid/framework/details/execution_strategy.h +++ b/paddle/fluid/framework/details/execution_strategy.h @@ -13,6 +13,7 @@ // limitations under the License. #pragma once +#include // for size_t namespace paddle { namespace framework { @@ -26,6 +27,7 @@ struct ExecutionStrategy { bool allow_op_delay_{false}; size_t num_iteration_per_drop_scope_{100}; ExecutorType type_{kDefault}; + bool dry_run_{false}; }; } // namespace details diff --git a/paddle/fluid/framework/details/fast_threaded_ssa_graph_executor.cc b/paddle/fluid/framework/details/fast_threaded_ssa_graph_executor.cc index 98fc390e72..2b2329b969 100644 --- a/paddle/fluid/framework/details/fast_threaded_ssa_graph_executor.cc +++ b/paddle/fluid/framework/details/fast_threaded_ssa_graph_executor.cc @@ -128,7 +128,9 @@ void FastThreadedSSAGraphExecutor::RunOpAsync( size_t complete = 0; while (op_to_run != nullptr) { try { - op_to_run->Run(strategy_.use_cuda_); + if (LIKELY(!strategy_.dry_run_)) { + op_to_run->Run(strategy_.use_cuda_); + } ++complete; } catch (...) { exception_.Catch(std::current_exception()); diff --git a/paddle/fluid/framework/details/threaded_ssa_graph_executor.cc b/paddle/fluid/framework/details/threaded_ssa_graph_executor.cc index dc63effd1b..2d2bdb604f 100644 --- a/paddle/fluid/framework/details/threaded_ssa_graph_executor.cc +++ b/paddle/fluid/framework/details/threaded_ssa_graph_executor.cc @@ -211,7 +211,9 @@ void ThreadedSSAGraphExecutor::RunOp( if (VLOG_IS_ON(10)) { VLOG(10) << op << " " << op->Name() << " : " << op->DebugString(); } - op->Run(strategy_.use_cuda_); + if (LIKELY(!strategy_.dry_run_)) { + op->Run(strategy_.use_cuda_); + } VLOG(10) << op << " " << op->Name() << " Done "; running_ops_--; ready_var_q->Extend(op->Outputs()); diff --git a/paddle/fluid/framework/details/threaded_ssa_graph_executor.h b/paddle/fluid/framework/details/threaded_ssa_graph_executor.h index dbb0b498d9..5c0bc169ea 100644 --- a/paddle/fluid/framework/details/threaded_ssa_graph_executor.h +++ b/paddle/fluid/framework/details/threaded_ssa_graph_executor.h @@ -48,7 +48,7 @@ class ThreadedSSAGraphExecutor : public SSAGraphExecutor { // Use topological sort algorithm FeedFetchList Run(const std::vector &fetch_tensors) override; - ~ThreadedSSAGraphExecutor() {} + ~ThreadedSSAGraphExecutor() final = default; private: void RunOp(const std::shared_ptr> &ready_var_q, diff --git a/paddle/fluid/framework/parallel_executor.cc b/paddle/fluid/framework/parallel_executor.cc index a45b9ec7a2..dfb107688a 100644 --- a/paddle/fluid/framework/parallel_executor.cc +++ b/paddle/fluid/framework/parallel_executor.cc @@ -38,9 +38,20 @@ class ParallelExecutorPrivate { explicit ParallelExecutorPrivate(const std::vector &places) : places_(places) {} + ~ParallelExecutorPrivate() { + if (own_local_scope_) { + for (size_t i = 1; i < local_scopes_.size(); ++i) { + // Skip the first scope, since it is the global scope. + Scope *local_scope = local_scopes_[i]; + if (global_scope_->HasKid(local_scope)) { + global_scope_->DeleteScope(local_scope); + } + } + } + } std::vector places_; std::vector local_scopes_; - Scope *global_scope_; + Scope *global_scope_; // not owned std::unique_ptr executor_; #ifdef PADDLE_WITH_CUDA @@ -306,16 +317,6 @@ ParallelExecutor::~ParallelExecutor() { for (auto &p : member_->places_) { platform::DeviceContextPool::Instance().Get(p)->Wait(); } - - if (member_->own_local_scope_) { - for (size_t i = 1; i < member_->local_scopes_.size(); ++i) { - Scope *local_scope = member_->local_scopes_[i]; - if (member_->global_scope_->HasKid(local_scope)) { - member_->global_scope_->DeleteScope(local_scope); - } - } - } - // member_ must be destructed before gcs_ since the destructor of // ReferenceCountOpHandle use raw pointers of gcs_ inside. member_.reset(); diff --git a/paddle/fluid/pybind/pybind.cc b/paddle/fluid/pybind/pybind.cc index fc821e04a0..238cc19189 100644 --- a/paddle/fluid/pybind/pybind.cc +++ b/paddle/fluid/pybind/pybind.cc @@ -742,7 +742,12 @@ All parameter, weight, gradient are variables in Paddle. will clean up the temp variables at the end of the current iteration. 2. In some NLP model, it may cause the GPU memory is insufficient, in this case, you should reduce `num_iteration_per_drop_scope`. - )DOC"); + )DOC") + .def_property("_dry_run", + [](const ExecutionStrategy &self) { return self.dry_run_; }, + [](ExecutionStrategy &self, bool dry_run) { + self.dry_run_ = dry_run; + }); exec_strategy.def_property( "use_experimental_executor", diff --git a/python/paddle/fluid/layers/io.py b/python/paddle/fluid/layers/io.py index 80b50022dd..d1c926c4e4 100644 --- a/python/paddle/fluid/layers/io.py +++ b/python/paddle/fluid/layers/io.py @@ -60,7 +60,7 @@ def data(name, For example if shape=[1], the resulting shape is [-1, 1]. 2. If shape contains -1, such as shape=[1, -1], append_batch_size will be enforced to be be False (ineffective). - dtype(int|float): The type of data : float32, float_16, int etc + dtype(basestring): The type of data : float32, float_16, int etc type(VarType): The output type. By default it is LOD_TENSOR. lod_level(int): The LoD Level. 0 means the input data is not a sequence. stop_gradient(bool): A boolean that mentions whether gradient should flow. diff --git a/python/paddle/fluid/tests/unittests/test_parallel_executor_dry_run.py b/python/paddle/fluid/tests/unittests/test_parallel_executor_dry_run.py new file mode 100644 index 0000000000..c93740669f --- /dev/null +++ b/python/paddle/fluid/tests/unittests/test_parallel_executor_dry_run.py @@ -0,0 +1,80 @@ +# 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. + +import paddle.fluid as fluid +import unittest +import logging +import six + + +class TestBase(unittest.TestCase): + def main(self, + network_func, + iter=100, + iter_per_pe=100, + use_gpu=True, + use_experimental_executor=False): + if use_gpu and not fluid.core.is_compiled_with_cuda(): + logging.warning( + "Paddle is not compiled with CUDA, skip GPU unittests") + return + + main_prog = fluid.Program() + startup_prog = fluid.Program() + scope = fluid.Scope() + with fluid.program_guard(main_prog, startup_prog): + with fluid.scope_guard(scope): + loss = network_func() + fluid.Executor( + fluid.CUDAPlace(0) + if use_gpu else fluid.CPUPlace()).run(startup_prog) + + for _ in six.moves.xrange(iter): + exe_strategy = fluid.ExecutionStrategy() + exe_strategy._dry_run = True + exe_strategy.use_experimental_executor = use_experimental_executor + pe = fluid.ParallelExecutor( + use_cuda=True, + loss_name=loss.name, + main_program=main_prog, + exec_strategy=exe_strategy) + for _ in six.moves.xrange(iter_per_pe): + pe.run([]) + + +class TestMNISTDryRun(TestBase): + def test_mnist_dry_run(self): + for use_gpu in (False, True): + for use_experimental_executor in (False, True): + self.main( + network_func=TestMNISTDryRun.network_func, + use_gpu=use_gpu, + use_experimental_executor=use_experimental_executor) + + @staticmethod + def network_func(): + img = fluid.layers.data(name='img', shape=[784], dtype='float32') + label = fluid.layers.data(name='label', shape=[1], dtype='int64') + hidden = img + for _ in six.moves.xrange(10): + hidden = fluid.layers.fc(input=img, size=200, act='tanh') + prediction = fluid.layers.fc(input=hidden, size=10, act='softmax') + loss = fluid.layers.cross_entropy(input=prediction, label=label) + avg_loss = fluid.layers.mean(loss) + fluid.optimizer.Adam().minimize(avg_loss) + return avg_loss + + +if __name__ == '__main__': + unittest.main() From 03992630b5f4d0ce44735ce689af3f6f70dfecec Mon Sep 17 00:00:00 2001 From: Yu Yang Date: Wed, 7 Nov 2018 17:25:54 +0800 Subject: [PATCH 19/22] fix(py): set `cwd` when get commit sha in setup.py (#14299) `cwd` was not set before when get commit SHA. The default `cwd` is the current build directory. However, the build directory might not be the subdirectory of source. The `git` command will fail when that happened. test=develop --- python/setup.py.in | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/python/setup.py.in b/python/setup.py.in index ee19294ad5..b1ff9f3a5c 100644 --- a/python/setup.py.in +++ b/python/setup.py.in @@ -14,7 +14,8 @@ RC = 0 def git_commit(): try: cmd = ['git', 'rev-parse', 'HEAD'] - git_commit = subprocess.Popen(cmd, stdout = subprocess.PIPE).communicate()[0].strip() + git_commit = subprocess.Popen(cmd, stdout = subprocess.PIPE, + cwd="@PADDLE_SOURCE_DIR@").communicate()[0].strip() except: git_commit = 'Unknown' git_commit = git_commit.decode() @@ -44,7 +45,7 @@ def get_patch(): def is_taged(): try: cmd = ['git', 'describe', '--exact-match', '--tags', 'HEAD', '2>/dev/null'] - git_tag = subprocess.Popen(cmd, stdout = subprocess.PIPE).communicate()[0].strip() + git_tag = subprocess.Popen(cmd, stdout = subprocess.PIPE, cwd="@PADDLE_SOURCE_DIR@").communicate()[0].strip() git_tag = git_tag.decode() except: return False @@ -55,8 +56,7 @@ def is_taged(): return False def write_version_py(filename='paddle/version.py'): - cnt = ''' -# THIS FILE IS GENERATED FROM PADDLEPADDLE SETUP.PY + cnt = '''# THIS FILE IS GENERATED FROM PADDLEPADDLE SETUP.PY # full_version = '%(major)d.%(minor)d.%(patch)s' major = '%(major)d' From 2466ca13ec80c6181b4ad1b3e6bf66fe95d7f904 Mon Sep 17 00:00:00 2001 From: Yu Yang Date: Wed, 7 Nov 2018 17:27:30 +0800 Subject: [PATCH 20/22] test(Pe): remove unittests for recordio in test_pe_mnist (#14262) recordio is not the official API in Fluid 1.0. Remove unittests for it. test=develop --- .../unittests/test_parallel_executor_mnist.py | 61 +++---------------- 1 file changed, 9 insertions(+), 52 deletions(-) diff --git a/python/paddle/fluid/tests/unittests/test_parallel_executor_mnist.py b/python/paddle/fluid/tests/unittests/test_parallel_executor_mnist.py index af3745987a..3eecc46701 100644 --- a/python/paddle/fluid/tests/unittests/test_parallel_executor_mnist.py +++ b/python/paddle/fluid/tests/unittests/test_parallel_executor_mnist.py @@ -14,30 +14,18 @@ from __future__ import print_function -from parallel_executor_test_base import TestParallelExecutorBase -import paddle.fluid as fluid -import paddle.fluid.core as core -import numpy as np -import paddle -import paddle.dataset.mnist as mnist import unittest -import os -MNIST_RECORDIO_FILE = "./mnist_test_pe.recordio" +import numpy as np +import paddle.fluid.core as core +import os +import paddle.fluid as fluid +from parallel_executor_test_base import TestParallelExecutorBase def simple_fc_net(use_feed): - if use_feed: - img = fluid.layers.data(name='image', shape=[784], dtype='float32') - label = fluid.layers.data(name='label', shape=[1], dtype='int64') - else: - reader = fluid.layers.open_files( - filenames=[MNIST_RECORDIO_FILE], - shapes=[[-1, 784], [-1, 1]], - lod_levels=[0, 0], - dtypes=['float32', 'int64']) - reader = fluid.layers.io.double_buffer(reader) - img, label = fluid.layers.read_file(reader) + img = fluid.layers.data(name='image', shape=[784], dtype='float32') + label = fluid.layers.data(name='label', shape=[1], dtype='int64') hidden = img for _ in range(4): hidden = fluid.layers.fc( @@ -53,17 +41,8 @@ def simple_fc_net(use_feed): def fc_with_batchnorm(use_feed): - if use_feed: - img = fluid.layers.data(name='image', shape=[784], dtype='float32') - label = fluid.layers.data(name='label', shape=[1], dtype='int64') - else: - reader = fluid.layers.open_files( - filenames=[MNIST_RECORDIO_FILE], - shapes=[[-1, 784], [-1, 1]], - lod_levels=[0, 0], - dtypes=['float32', 'int64']) - reader = fluid.layers.io.double_buffer(reader) - img, label = fluid.layers.read_file(reader) + img = fluid.layers.data(name='image', shape=[784], dtype='float32') + label = fluid.layers.data(name='label', shape=[1], dtype='int64') hidden = img for _ in range(1): @@ -88,19 +67,6 @@ class TestMNIST(TestParallelExecutorBase): @classmethod def setUpClass(cls): os.environ['CPU_NUM'] = str(4) - # Convert mnist to recordio file - with fluid.program_guard(fluid.Program(), fluid.Program()): - reader = paddle.batch(mnist.train(), batch_size=4) - feeder = fluid.DataFeeder( - feed_list=[ # order is image and label - fluid.layers.data( - name='image', shape=[784]), - fluid.layers.data( - name='label', shape=[1], dtype='int64'), - ], - place=fluid.CPUPlace()) - fluid.recordio_writer.convert_reader_to_recordio_file( - MNIST_RECORDIO_FILE, reader, feeder) def _init_data(self): np.random.seed(5) @@ -111,10 +77,6 @@ class TestMNIST(TestParallelExecutorBase): def _compare_reduce_and_allreduce(self, model, use_cuda): if use_cuda and not core.is_compiled_with_cuda(): return - self.check_network_convergence( - model, use_cuda=use_cuda, use_reduce=True) - self.check_network_convergence( - model, use_cuda=use_cuda, allow_op_delay=True, use_reduce=True) img, label = self._init_data() @@ -140,9 +102,6 @@ class TestMNIST(TestParallelExecutorBase): def check_simple_fc_convergence(self, use_cuda, use_reduce=False): if use_cuda and not core.is_compiled_with_cuda(): return - self.check_network_convergence(simple_fc_net, use_cuda=use_cuda) - self.check_network_convergence( - simple_fc_net, use_cuda=use_cuda, allow_op_delay=True) img, label = self._init_data() @@ -199,8 +158,6 @@ class TestMNIST(TestParallelExecutorBase): if use_cuda and not core.is_compiled_with_cuda(): return - self.check_network_convergence(fc_with_batchnorm, use_cuda=use_cuda) - img, label = self._init_data() self.check_network_convergence( From 3319072858fe051035bc8f5c986db8d6c4bb32de Mon Sep 17 00:00:00 2001 From: tensor-tang Date: Wed, 7 Nov 2018 09:29:59 +0000 Subject: [PATCH 21/22] fix jit kernel test on mac test=develop --- paddle/fluid/operators/math/jit_kernel_test.cc | 6 +++++- 1 file changed, 5 insertions(+), 1 deletion(-) diff --git a/paddle/fluid/operators/math/jit_kernel_test.cc b/paddle/fluid/operators/math/jit_kernel_test.cc index 667a95fe1a..34fa2b9a78 100644 --- a/paddle/fluid/operators/math/jit_kernel_test.cc +++ b/paddle/fluid/operators/math/jit_kernel_test.cc @@ -801,7 +801,11 @@ TEST(JitKernel, pool) { std::dynamic_pointer_cast(pvmul_d)); const auto& pvmul_from_key = jit::KernelPool::Instance().Get("vmulfjit4"); - EXPECT_EQ(pvmul_f, pvmul_from_key); +#if defined(__APPLE__) || defined(__OSX__) || defined(_WIN32) + EXPECT_EQ(pvmul_from_key, nullptr); +#else + EXPECT_EQ(pvmul_from_key, pvmul_f); +#endif const auto& pvmul_from_key2 = jit::KernelPool::Instance().Get("vmulfjit"); EXPECT_TRUE(pvmul_from_key2 == nullptr); } From 382307b94345916dd4094623e06c5ade7a87e32e Mon Sep 17 00:00:00 2001 From: tensor-tang Date: Wed, 7 Nov 2018 06:16:45 +0000 Subject: [PATCH 22/22] refine code test=develop --- paddle/fluid/operators/math/jit_code.cc | 65 +++++++------------ paddle/fluid/operators/math/jit_code.h | 50 ++++++-------- .../fluid/operators/math/jit_kernel_blas.cc | 31 ++++++--- 3 files changed, 65 insertions(+), 81 deletions(-) diff --git a/paddle/fluid/operators/math/jit_code.cc b/paddle/fluid/operators/math/jit_code.cc index 35f0bdb9b3..a92e5d351e 100644 --- a/paddle/fluid/operators/math/jit_code.cc +++ b/paddle/fluid/operators/math/jit_code.cc @@ -24,51 +24,14 @@ namespace gen { using namespace platform::jit; // NOLINT -bool VMulJitCode::init(int d) { +bool VVVJitCode::init(int d) { // It's not necessary to use avx512 since it would slow down the frequency // and this kernel is not compute bound. return MayIUse(avx); } -void VMulJitCode::generate() { +void VVVJitCode::generate() { // do not need push stack, and do not need save avx512reg if do not use avx512 - int offset = 0; - for (int i = 0; i < num_ / AVX_FLOAT_BLOCK; ++i) { - vmovups(ymm_src1, ptr[param1 + offset]); - vmovups(ymm_src2, ptr[param2 + offset]); - vmulps(ymm_dst, ymm_src1, ymm_src2); - vmovups(ptr[param3 + offset], ymm_dst); - offset += sizeof(float) * AVX_FLOAT_BLOCK; - } - int rest = num_ % AVX_FLOAT_BLOCK; - if (rest >= 4) { - vmovups(xmm_src1, ptr[param1 + offset]); - vmovups(xmm_src2, ptr[param2 + offset]); - vmulps(xmm_dst, xmm_src1, xmm_src2); - vmovups(ptr[param3 + offset], xmm_dst); - offset += sizeof(float) * 4; - rest -= 4; - } - if (rest >= 2) { - vmovq(xmm_src1, ptr[param1 + offset]); - vmovq(xmm_src2, ptr[param2 + offset]); - vmulps(xmm_dst, xmm_src1, xmm_src2); - vmovq(ptr[param3 + offset], xmm_dst); - offset += sizeof(float) * 2; - rest -= 2; - } - if (rest > 0) { - vmovss(xmm_src1, ptr[param1 + offset]); - vmovss(xmm_src2, ptr[param2 + offset]); - vmulss(xmm_dst, xmm_src1, xmm_src2); - vmovss(ptr[param3 + offset], xmm_dst); - } - ret(); -} - -bool VAddJitCode::init(int d) { return MayIUse(avx); } - -void VAddJitCode::generate() { int offset = 0; if (with_relu_) { vxorps(ymm_zero, ymm_zero, ymm_zero); @@ -76,7 +39,11 @@ void VAddJitCode::generate() { for (int i = 0; i < num_ / AVX_FLOAT_BLOCK; ++i) { vmovups(ymm_src1, ptr[param1 + offset]); vmovups(ymm_src2, ptr[param2 + offset]); - vaddps(ymm_dst, ymm_src1, ymm_src2); + if (type_ == operand_type::mul) { + vmulps(ymm_dst, ymm_src1, ymm_src2); + } else if (type_ == operand_type::add) { + vaddps(ymm_dst, ymm_src1, ymm_src2); + } if (with_relu_) { vmaxps(ymm_dst, ymm_zero, ymm_dst); } @@ -87,7 +54,11 @@ void VAddJitCode::generate() { if (rest >= 4) { vmovups(xmm_src1, ptr[param1 + offset]); vmovups(xmm_src2, ptr[param2 + offset]); - vaddps(xmm_dst, xmm_src1, xmm_src2); + if (type_ == operand_type::mul) { + vmulps(xmm_dst, xmm_src1, xmm_src2); + } else if (type_ == operand_type::add) { + vaddps(xmm_dst, xmm_src1, xmm_src2); + } if (with_relu_) { vmaxps(xmm_dst, xmm_zero, xmm_dst); } @@ -98,7 +69,11 @@ void VAddJitCode::generate() { if (rest >= 2) { vmovq(xmm_src1, ptr[param1 + offset]); vmovq(xmm_src2, ptr[param2 + offset]); - vaddps(xmm_dst, xmm_src1, xmm_src2); + if (type_ == operand_type::mul) { + vmulps(xmm_dst, xmm_src1, xmm_src2); + } else if (type_ == operand_type::add) { + vaddps(xmm_dst, xmm_src1, xmm_src2); + } if (with_relu_) { vmaxps(xmm_dst, xmm_zero, xmm_dst); } @@ -109,7 +84,11 @@ void VAddJitCode::generate() { if (rest > 0) { vmovss(xmm_src1, ptr[param1 + offset]); vmovss(xmm_src2, ptr[param2 + offset]); - vaddss(xmm_dst, xmm_src1, xmm_src2); + if (type_ == operand_type::mul) { + vmulss(xmm_dst, xmm_src1, xmm_src2); + } else if (type_ == operand_type::add) { + vaddss(xmm_dst, xmm_src1, xmm_src2); + } if (with_relu_) { vmaxps(xmm_dst, xmm_zero, xmm_dst); } diff --git a/paddle/fluid/operators/math/jit_code.h b/paddle/fluid/operators/math/jit_code.h index 6bfed4b22d..73692ebc67 100644 --- a/paddle/fluid/operators/math/jit_code.h +++ b/paddle/fluid/operators/math/jit_code.h @@ -14,8 +14,8 @@ limitations under the License. */ #pragma once +#include #include "paddle/fluid/operators/math/jit_gen.h" - namespace paddle { namespace operators { namespace math { @@ -29,41 +29,33 @@ using ymm_t = const Xbyak::Ymm; using zmm_t = const Xbyak::Zmm; using Label = Xbyak::Label; -class VMulJitCode : public JitCode { - public: - DECLARE_JIT_CODE(VMulJitCode); - explicit VMulJitCode(int d, size_t code_size = 256 * 1024, - void* code_ptr = nullptr) - : JitCode(code_size, code_ptr), num_(d) {} - static bool init(int d); - void generate() override; - - private: - int num_; - reg64_t param1{abi_param1}; - reg64_t param2{abi_param2}; - reg64_t param3{abi_param3}; - - xmm_t xmm_src1 = xmm_t(0); - xmm_t xmm_src2 = xmm_t(1); - xmm_t xmm_dst = xmm_t(1); - - ymm_t ymm_src1 = ymm_t(0); - ymm_t ymm_src2 = ymm_t(1); - ymm_t ymm_dst = ymm_t(1); -}; +// function: vec = Operand(vec, vec) (maybe with relu) +typedef enum { mul = 0, add } operand_type; -class VAddJitCode : public JitCode { +class VVVJitCode : public JitCode { public: - DECLARE_JIT_CODE(VAddJitCode); - explicit VAddJitCode(int d, bool with_relu, size_t code_size = 256 * 1024, - void* code_ptr = nullptr) - : JitCode(code_size, code_ptr), num_(d), with_relu_(with_relu) {} + const char* name() const override { + std::string base = "VVVJitCode"; + if (type_ == operand_type::mul) { + base += "_Mul"; + } else if (type_ == operand_type::add) { + base += "_Add"; + } + base += (with_relu_ ? "_relu" : ""); + return base.c_str(); + } + explicit VVVJitCode(int d, operand_type type, bool with_relu, + size_t code_size = 256 * 1024, void* code_ptr = nullptr) + : JitCode(code_size, code_ptr), + num_(d), + type_(type), + with_relu_(with_relu) {} static bool init(int d); void generate() override; private: int num_; + operand_type type_; bool with_relu_; reg64_t param1{abi_param1}; reg64_t param2{abi_param2}; diff --git a/paddle/fluid/operators/math/jit_kernel_blas.cc b/paddle/fluid/operators/math/jit_kernel_blas.cc index 27801f4c63..9acb349f66 100644 --- a/paddle/fluid/operators/math/jit_kernel_blas.cc +++ b/paddle/fluid/operators/math/jit_kernel_blas.cc @@ -102,7 +102,8 @@ class VMulKernelImpl : public VMulKernel { if (useJIT(d)) { // roughly estimate the size of code size_t sz = 96 + d / AVX_FLOAT_BLOCK * 4 * 8; - jitcode_.reset(new gen::VMulJitCode(d, sz > 4096 ? sz : 4096)); + jitcode_.reset(new gen::VVVJitCode(d, gen::operand_type::mul, false, + sz > 4096 ? sz : 4096)); this->Compute = jitcode_->getCode(); return; @@ -120,14 +121,14 @@ class VMulKernelImpl : public VMulKernel { #ifdef PADDLE_WITH_XBYAK private: - std::unique_ptr jitcode_{nullptr}; + std::unique_ptr jitcode_{nullptr}; #endif }; #ifdef PADDLE_WITH_XBYAK template <> bool VMulKernelImpl::useJIT(int d) { - return gen::VMulJitCode::init(d); + return gen::VVVJitCode::init(d); } #endif @@ -149,13 +150,16 @@ class VAddKernelImpl : public VAddKernel { public: DECLARE_STATIC_FUNC; explicit VAddKernelImpl(int d) : VAddKernel() { +#ifdef PADDLE_WITH_XBYAK if (useJIT(d)) { size_t sz = 96 + d / AVX_FLOAT_BLOCK * 4 * 8; - jitcode_.reset(new gen::VAddJitCode(d, false, sz > 4096 ? sz : 4096)); + jitcode_.reset(new gen::VVVJitCode(d, gen::operand_type::add, false, + sz > 4096 ? sz : 4096)); this->Compute = jitcode_->getCode(); return; } +#endif #ifdef PADDLE_WITH_MKLML if (useMKL(d)) { this->Compute = VAddMKL; @@ -166,14 +170,17 @@ class VAddKernelImpl : public VAddKernel { } private: - std::unique_ptr jitcode_{nullptr}; + std::unique_ptr jitcode_{nullptr}; }; +#ifdef PADDLE_WITH_XBYAK template <> bool VAddKernelImpl::useJIT(int d) { - return gen::VAddJitCode::init(d); + return gen::VVVJitCode::init(d); } +#endif +#ifdef PADDLE_WITH_MKLML template <> bool VAddKernelImpl::useMKL(int d) { return d > 512; @@ -183,6 +190,7 @@ template <> bool VAddKernelImpl::useMKL(int d) { return true; } +#endif /* VAddRelu JitKernel */ template @@ -190,24 +198,29 @@ class VAddReluKernelImpl : public VAddReluKernel { public: DECLARE_STATIC_FUNC; explicit VAddReluKernelImpl(int d) : VAddReluKernel() { +#ifdef PADDLE_WITH_XBYAK if (useJIT(d)) { size_t sz = 96 + d / AVX_FLOAT_BLOCK * 4 * 8; - jitcode_.reset(new gen::VAddJitCode(d, true, sz > 4096 ? sz : 4096)); + jitcode_.reset(new gen::VVVJitCode(d, gen::operand_type::add, true, + sz > 4096 ? sz : 4096)); this->Compute = jitcode_->getCode(); return; } +#endif this->Compute = VAddReluRefer; } private: - std::unique_ptr jitcode_{nullptr}; + std::unique_ptr jitcode_{nullptr}; }; +#ifdef PADDLE_WITH_XBYAK template <> bool VAddReluKernelImpl::useJIT(int d) { - return gen::VAddJitCode::init(d); + return gen::VVVJitCode::init(d); } +#endif #undef DECLARE_STATIC_FUNC