From acbbc784307330ebb155850ee9935d8dcfdded20 Mon Sep 17 00:00:00 2001
From: dangqingqing <dangqingqing@baidu.com>
Date: Tue, 7 Nov 2017 19:33:13 +0800
Subject: [PATCH] Check errors for the cuda kernel calls.

---
 paddle/framework/operator.cc                   | 3 +++
 paddle/operators/math/detail/lstm_gpu_kernel.h | 5 -----
 paddle/platform/device_context.cc              | 5 +++++
 paddle/platform/device_context.h               | 5 +++++
 4 files changed, 13 insertions(+), 5 deletions(-)

diff --git a/paddle/framework/operator.cc b/paddle/framework/operator.cc
index 22a7d9728a059..8150bf923926e 100644
--- a/paddle/framework/operator.cc
+++ b/paddle/framework/operator.cc
@@ -440,6 +440,9 @@ void OperatorWithKernel::Run(const Scope& scope,
   }
 
   kernel_iter->second->Compute(ctx);
+
+  // throws errors if have.
+  dev_ctx.Finish();
 }
 
 }  // namespace framework
diff --git a/paddle/operators/math/detail/lstm_gpu_kernel.h b/paddle/operators/math/detail/lstm_gpu_kernel.h
index 41a54a359daa1..8b46510db05fb 100644
--- a/paddle/operators/math/detail/lstm_gpu_kernel.h
+++ b/paddle/operators/math/detail/lstm_gpu_kernel.h
@@ -244,11 +244,6 @@ void gpu_lstm_backward(const platform::DeviceContext& context, Op op,
         op, value, grad, frameSize, batchSize, active_node, active_gate,
         active_state);
   }
-
-  cudaStreamSynchronize(stream);
-  // TODO(qingqing): Add cuda error check for each kernel.
-  cudaError_t err = cudaGetLastError();
-  PADDLE_ENFORCE(err, cudaGetErrorString(err));
 }
 
 }  // namespace detail
diff --git a/paddle/platform/device_context.cc b/paddle/platform/device_context.cc
index 36450e9268913..7afcdfce9371e 100644
--- a/paddle/platform/device_context.cc
+++ b/paddle/platform/device_context.cc
@@ -124,6 +124,11 @@ void CUDADeviceContext::Wait() const {
   PADDLE_ENFORCE(cudaStreamSynchronize(stream_));
 }
 
+void CUDADeviceContext::Finish() const {
+  Wait();
+  PADDLE_ENFORCE(cudaGetLastError());
+}
+
 Eigen::GpuDevice* CUDADeviceContext::eigen_device() const {
   return eigen_device_.get();
 }
diff --git a/paddle/platform/device_context.h b/paddle/platform/device_context.h
index ef5f19214d9cc..526d089e35da9 100644
--- a/paddle/platform/device_context.h
+++ b/paddle/platform/device_context.h
@@ -46,6 +46,8 @@ class DeviceContext {
   DeviceType* GetEigenDevice() const;
 
   virtual void Wait() const {}
+
+  virtual void Finish() const {}
 };
 
 class CPUDeviceContext : public DeviceContext {
@@ -77,6 +79,9 @@ class CUDADeviceContext : public DeviceContext {
   /*! \brief  Wait for all operations completion in the stream. */
   void Wait() const override;
 
+  /*! \brief  Check potential errors for the cuda kernel calls. */
+  void Finish() const override;
+
   /*! \brief  Return place in the device context. */
   Place GetPlace() const override;