diff --git a/cmake/operators.cmake b/cmake/operators.cmake
new file mode 100644
index 0000000000000000000000000000000000000000..c9d0f80da2962ee1578af0f6f06f07438898f399
--- /dev/null
+++ b/cmake/operators.cmake
@@ -0,0 +1,214 @@
+set(PART_CUDA_KERNEL_FILES)
+function(op_library TARGET)
+    # op_library is a function to create op library. The interface is same as
+    # cc_library. But it handle split GPU/CPU code and link some common library
+    # for ops.
+    set(cc_srcs)
+    set(cu_srcs)
+    set(hip_cu_srcs)
+    set(miopen_hip_cc_srcs)
+    set(cu_cc_srcs)
+    set(cudnn_cu_cc_srcs)
+    set(CUDNN_FILE)
+    set(mkldnn_cc_srcs)
+    set(MKLDNN_FILE)
+    set(op_common_deps operator op_registry math_function)
+    set(options "")
+    set(oneValueArgs "")
+    set(multiValueArgs SRCS DEPS)
+    set(pybind_flag 0)
+    cmake_parse_arguments(op_library "${options}" "${oneValueArgs}"
+            "${multiValueArgs}" ${ARGN})
+
+    list(LENGTH op_library_SRCS op_library_SRCS_len)
+    if (${op_library_SRCS_len} EQUAL 0)
+        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.cc)
+            list(APPEND cc_srcs ${TARGET}.cc)
+        endif()
+        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.cu.cc)
+            list(APPEND cu_cc_srcs ${TARGET}.cu.cc)
+        endif()
+        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.cu)
+            list(APPEND cu_srcs ${TARGET}.cu)
+        endif()
+        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.part.cu)
+            set(PART_CUDA_KERNEL_FILES ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.part.cu
+                    ${PART_CUDA_KERNEL_FILES} PARENT_SCOPE)
+            list(APPEND cu_srcs ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.part.cu)
+        endif()
+
+        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.hip.cu)
+            list(APPEND hip_cu_srcs ${TARGET}.hip.cu)
+        endif()
+        string(REPLACE "_op" "_cudnn_op" CUDNN_FILE "${TARGET}")
+        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${CUDNN_FILE}.cu.cc)
+            list(APPEND cudnn_cu_cc_srcs ${CUDNN_FILE}.cu.cc)
+        endif()
+        if(WITH_AMD_GPU)
+            string(REPLACE "_op" "_miopen_op" MIOPEN_FILE "${TARGET}")
+            if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${MIOPEN_FILE}.hip.cc)
+                list(APPEND miopen_hip_cc_srcs ${MIOPEN_FILE}.hip.cc)
+            endif()
+        endif()
+        if(WITH_MKLDNN)
+            string(REPLACE "_op" "_mkldnn_op" MKLDNN_FILE "${TARGET}")
+            if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${MKLDNN_FILE}.cc)
+                list(APPEND mkldnn_cc_srcs ${MKLDNN_FILE}.cc)
+            endif()
+        endif()
+    else()
+        foreach(src ${op_library_SRCS})
+            if (${src} MATCHES ".*\\.hip.cu$")
+                list(APPEND hip_cu_srcs ${src})
+            elseif (${src} MATCHES ".*\\.cu$")
+                list(APPEND cu_srcs ${src})
+            elseif(${src} MATCHES ".*_cudnn_op.cu.cc$")
+                list(APPEND cudnn_cu_cc_srcs ${src})
+            elseif(WITH_AMD_GPU AND ${src} MATCHES ".*_miopen_op.hip.cc$")
+                list(APPEND miopen_hip_cc_srcs ${src})
+            elseif(WITH_MKLDNN AND ${src} MATCHES ".*_mkldnn_op.cc$")
+                list(APPEND mkldnn_cc_srcs ${src})
+            elseif(${src} MATCHES ".*\\.cu.cc$")
+                list(APPEND cu_cc_srcs ${src})
+            elseif(${src} MATCHES ".*\\.cc$")
+                list(APPEND cc_srcs ${src})
+            else()
+                message(FATAL_ERROR "${TARGET} Source file ${src} should only be .cc or .cu")
+            endif()
+        endforeach()
+    endif()
+
+    list(LENGTH cc_srcs cc_srcs_len)
+    if (${cc_srcs_len} EQUAL 0)
+        message(FATAL_ERROR "The op library ${TARGET} should contains at least one .cc file")
+    endif()
+    if (WIN32)
+    # remove windows unsupported op, because windows has no nccl, no warpctc such ops.
+    foreach(windows_unsupport_op "nccl_op" "gen_nccl_id_op" "warpctc_op" "hierarchical_sigmoid_op"
+     "crf_decoding_op" "select_op" "lstmp_op" "gru_op" "fusion_gru_op" "lstm_op" "fusion_lstm_op" "cumsum_op"
+      "fusion_seqconv_eltadd_relu_op" "channel_send_op" "channel_create_op" "channel_close_op" "channel_recv_op")
+        if ("${TARGET}" STREQUAL "${windows_unsupport_op}")
+          return()
+        endif()
+    endforeach()
+    endif(WIN32)
+    set(OP_LIBRARY ${TARGET} ${OP_LIBRARY} CACHE INTERNAL "op libs")
+
+    list(LENGTH op_library_DEPS op_library_DEPS_len)
+    if (${op_library_DEPS_len} GREATER 0)
+        set(DEPS_OPS ${TARGET} ${DEPS_OPS} PARENT_SCOPE)
+    endif()
+    if (WITH_GPU)
+        nv_library(${TARGET} SRCS ${cc_srcs} ${cu_cc_srcs} ${cudnn_cu_cc_srcs} ${mkldnn_cc_srcs} ${cu_srcs} DEPS ${op_library_DEPS}
+                ${op_common_deps})
+    elseif (WITH_AMD_GPU)
+        hip_library(${TARGET} SRCS ${cc_srcs} ${hip_cu_srcs} ${miopen_hip_cc_srcs} ${mkldnn_cc_srcs} DEPS ${op_library_DEPS}
+                ${op_common_deps})
+    else()
+        cc_library(${TARGET} SRCS ${cc_srcs} ${mkldnn_cc_srcs} DEPS ${op_library_DEPS}
+            ${op_common_deps})
+    endif()
+
+    # Define operators that don't need pybind here.
+    foreach(manual_pybind_op "compare_op" "logical_op" "nccl_op"
+"tensor_array_read_write_op" "tensorrt_engine_op")
+        if ("${TARGET}" STREQUAL "${manual_pybind_op}")
+            set(pybind_flag 1)
+        endif()
+    endforeach()
+
+    # The registration of USE_OP, please refer to paddle/fluid/framework/op_registry.h.
+    # Note that it's enough to just adding one operator to pybind in a *_op.cc file.
+    # And for detail pybind information, please see generated paddle/pybind/pybind.h.
+    file(READ ${TARGET}.cc TARGET_CONTENT)
+    string(REGEX MATCH "REGISTER_OPERATOR\\(.*REGISTER_OPERATOR\\(" multi_register "${TARGET_CONTENT}")
+    string(REGEX MATCH "REGISTER_OPERATOR\\([a-z0-9_]*," one_register "${multi_register}")
+    if (one_register STREQUAL "")
+        string(REPLACE "_op" "" TARGET "${TARGET}")
+    else ()
+        string(REPLACE "REGISTER_OPERATOR(" "" TARGET "${one_register}")
+        string(REPLACE "," "" TARGET "${TARGET}")
+    endif()
+
+    # pybind USE_NO_KERNEL_OP
+    # HACK: if REGISTER_OP_CPU_KERNEL presents the operator must have kernel
+    string(REGEX MATCH "REGISTER_OP_CPU_KERNEL" regex_result "${TARGET_CONTENT}")
+    string(REPLACE "_op" "" TARGET "${TARGET}")
+    if (${pybind_flag} EQUAL 0 AND regex_result STREQUAL "")
+        file(APPEND ${pybind_file} "USE_NO_KERNEL_OP(${TARGET});\n")
+        set(pybind_flag 1)
+    endif()
+
+    # pybind USE_CPU_ONLY_OP
+    list(LENGTH cu_srcs cu_srcs_len)
+    list(LENGTH cu_cc_srcs cu_cc_srcs_len)
+    list(LENGTH mkldnn_cc_srcs mkldnn_cc_srcs_len)
+    list(LENGTH hip_cu_srcs hip_cu_srcs_len)
+    list(LENGTH miopen_hip_cc_srcs miopen_hip_cc_srcs_len)
+    if (${pybind_flag} EQUAL 0 AND ${mkldnn_cc_srcs_len} EQUAL 0 AND ${cu_srcs_len} EQUAL 0 AND ${cu_cc_srcs_len} EQUAL 0 AND
+        ${hip_cu_srcs_len} EQUAL 0 AND ${miopen_hip_cc_srcs_len} EQUAL 0)
+        file(APPEND ${pybind_file} "USE_CPU_ONLY_OP(${TARGET});\n")
+        set(pybind_flag 1)
+    endif()
+
+    # pybind USE_OP_DEVICE_KERNEL for CUDNN
+    list(LENGTH cudnn_cu_cc_srcs cudnn_cu_cc_srcs_len)
+    if (WITH_GPU AND ${cudnn_cu_cc_srcs_len} GREATER 0)
+        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${TARGET}, CUDNN);\n")
+    endif()
+
+    # pybind USE_OP_DEVICE_KERNEL for MIOPEN
+    if (WITH_AMD_GPU AND ${miopen_hip_cc_srcs_len} GREATER 0)
+        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${TARGET}, MIOPEN);\n")
+    endif()
+
+    # pybind USE_OP_DEVICE_KERNEL for MKLDNN
+    if (WITH_MKLDNN AND ${mkldnn_cc_srcs_len} GREATER 0)
+      # Append first implemented MKLDNN activation operator
+      if (${MKLDNN_FILE} STREQUAL "activation_mkldnn_op")
+        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(relu, MKLDNN);\n")
+      else()
+        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${TARGET}, MKLDNN);\n")
+      endif()
+    endif()
+
+    # pybind USE_OP
+    if (${pybind_flag} EQUAL 0)
+      # NOTE(*): activation use macro to regist the kernels, set use_op manually.
+      if(${TARGET} STREQUAL "activation")
+        file(APPEND ${pybind_file} "USE_OP(relu);\n")
+      elseif(${TARGET} STREQUAL "fake_dequantize")
+        file(APPEND ${pybind_file} "USE_OP(fake_dequantize_max_abs);\n")
+      elseif(${TARGET} STREQUAL "fake_quantize")
+        file(APPEND ${pybind_file} "USE_OP(fake_quantize_abs_max);\n")
+      elseif(${TARGET} STREQUAL "tensorrt_engine_op")
+          message(STATUS "Pybind skips [tensorrt_engine_op], for this OP is only used in inference")
+      elseif(${TARGET} STREQUAL "fc")
+        # HACK: fc only have mkldnn and cpu, which would mismatch the cpu only condition
+        file(APPEND ${pybind_file} "USE_CPU_ONLY_OP(${TARGET});\n")
+      else()
+        file(APPEND ${pybind_file} "USE_OP(${TARGET});\n")
+      endif()
+    endif()
+endfunction()
+
+
+function(register_operators)
+    set(options "")
+    set(oneValueArgs "")
+    set(multiValueArgs EXCLUDES)
+    cmake_parse_arguments(register_operators "${options}" "${oneValueArgs}"
+            "${multiValueArgs}" ${ARGN})
+
+    file(GLOB OPS RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*_op.cc")
+    string(REPLACE "_mkldnn" "" OPS "${OPS}")
+    string(REPLACE ".cc" "" OPS "${OPS}")
+    list(REMOVE_DUPLICATES OPS)
+
+    foreach(src ${OPS})
+        list(FIND register_operators_EXCLUDES ${src} _index)
+        if (${_index} EQUAL -1)
+            op_library(${src})
+        endif()
+    endforeach()
+endfunction()
diff --git a/paddle/fluid/framework/data_device_transform_test.cu b/paddle/fluid/framework/data_device_transform_test.cu
index 21e0cb3f91cc0ae05513c3bbd470650ca71194d7..2d2323edc3a6636bec72ea2ae7329ebd4e619348 100644
--- a/paddle/fluid/framework/data_device_transform_test.cu
+++ b/paddle/fluid/framework/data_device_transform_test.cu
@@ -17,7 +17,7 @@ limitations under the License. */
 #include "paddle/fluid/framework/lod_tensor.h"
 #include "paddle/fluid/framework/op_info.h"
 #include "paddle/fluid/framework/op_registry.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 #include "paddle/fluid/operators/math/math_function.h"
 #include "paddle/fluid/platform/device_context.h"
 #include "paddle/fluid/platform/init.h"
diff --git a/paddle/fluid/inference/CMakeLists.txt b/paddle/fluid/inference/CMakeLists.txt
index fc656613010d18608c2780c96212f2d7bb674704..2c5364b72402befd2c34e5f542ce5c6b2add621d 100644
--- a/paddle/fluid/inference/CMakeLists.txt
+++ b/paddle/fluid/inference/CMakeLists.txt
@@ -13,7 +13,7 @@ set(FLUID_CORE_MODULES proto_desc memory lod_tensor executor)
 # TODO(panyx0718): Should this be called paddle_fluid_inference_api_internal?
 cc_library(paddle_fluid_api
     SRCS io.cc
-    DEPS ${FLUID_CORE_MODULES} ${GLOB_OP_LIB}) 
+    DEPS ${FLUID_CORE_MODULES} ${GLOB_OP_LIB} ${GLOB_OPERATOR_DEPS})
 
 get_property(fluid_modules GLOBAL PROPERTY FLUID_MODULES)
 get_property(cuda_modules GLOBAL PROPERTY CUDA_MODULES)
diff --git a/paddle/fluid/inference/tensorrt/CMakeLists.txt b/paddle/fluid/inference/tensorrt/CMakeLists.txt
index e09705e3c69eb2b2370bd1ad2d9cf178ef041ee6..17f6c6d9f10abf99fd93364d1356e2b3ef1b3934 100644
--- a/paddle/fluid/inference/tensorrt/CMakeLists.txt
+++ b/paddle/fluid/inference/tensorrt/CMakeLists.txt
@@ -1,4 +1,4 @@
-nv_library(tensorrt_engine SRCS engine.cc DEPS framework_proto device_context)
+nv_library(tensorrt_engine SRCS engine.cc DEPS ${GLOB_OPERATOR_DEPS} framework_proto device_context)
 nv_test(test_tensorrt SRCS test_tensorrt.cc DEPS dynload_cuda device_context dynamic_loader)
 nv_test(test_tensorrt_engine SRCS test_engine.cc DEPS dynload_cuda tensorrt_engine)
 add_subdirectory(plugin)
diff --git a/paddle/fluid/inference/tensorrt/convert/CMakeLists.txt b/paddle/fluid/inference/tensorrt/convert/CMakeLists.txt
index aa4126392bf1f4b16a123403b68f506d6f76b5a2..85ad5ffe7875cdc205b5bdff28cc90ef01b236a4 100644
--- a/paddle/fluid/inference/tensorrt/convert/CMakeLists.txt
+++ b/paddle/fluid/inference/tensorrt/convert/CMakeLists.txt
@@ -6,34 +6,34 @@ pad_op.cc split_op.cc prelu_op.cc
   DEPS tensorrt_engine tensorrt_plugin operator scope framework_proto op_registry)
 
 nv_test(test_op_converter SRCS test_op_converter.cc DEPS
-  ${FLUID_CORE_MODULES} tensorrt_engine tensorrt_converter)
+  ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine tensorrt_converter)
 
 nv_test(test_io_converter SRCS test_io_converter.cc io_converter.cc DEPS dynload_cuda dynamic_loader lod_tensor)
 nv_test(test_trt_mul_op SRCS test_mul_op.cc mul_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine mul_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine mul_op SERIAL)
 nv_test(test_trt_fc_op SRCS test_fc_op.cc fc_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine mul_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine mul_op SERIAL)
 nv_test(test_trt_activation_op SRCS test_activation_op.cc activation_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine activation_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine activation_op SERIAL)
 nv_test(test_trt_conv_op SRCS test_conv2d_op.cc conv2d_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine conv_op conv_transpose_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine conv_op conv_transpose_op SERIAL)
 nv_test(test_trt_pool2d_op SRCS test_pool2d_op.cc pool2d_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine pool_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine pool_op SERIAL)
 nv_test(test_trt_elementwise_op SRCS test_elementwise_op.cc elementwise_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine elementwise_add_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine elementwise_add_op SERIAL)
 nv_test(test_trt_softmax_op SRCS test_softmax_op.cc softmax_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine softmax_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine softmax_op SERIAL)
 nv_test(test_trt_batch_norm_op SRCS test_batch_norm_op.cc batch_norm_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine batch_norm_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine batch_norm_op SERIAL)
 nv_test(test_trt_concat_op SRCS test_concat_op.cc concat_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine concat_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine concat_op SERIAL)
 nv_test(test_trt_dropout_op SRCS test_dropout_op.cc dropout_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine dropout_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine dropout_op SERIAL)
 nv_test(test_trt_pad_op SRCS test_pad_op.cc pad_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine pad_op SERIAL)
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine pad_op SERIAL)
 nv_test(test_trt_split_op SRCS test_split_op.cc split_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine tensorrt_plugin
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine tensorrt_plugin
         split_op concat_op SERIAL)
 nv_test(test_trt_prelu_op SRCS test_prelu_op.cc prelu_op.cc
-        DEPS ${FLUID_CORE_MODULES} tensorrt_engine tensorrt_plugin
+        DEPS ${FLUID_CORE_MODULES} ${GLOB_OPERATOR_DEPS} tensorrt_engine tensorrt_plugin
         prelu_op SERIAL)
diff --git a/paddle/fluid/operators/CMakeLists.txt b/paddle/fluid/operators/CMakeLists.txt
index 0117a24c1b308e528e19b9e948cc1decd2b5eab5..df2a3e7aa635c9ba41dad85ccb8316823f875639 100644
--- a/paddle/fluid/operators/CMakeLists.txt
+++ b/paddle/fluid/operators/CMakeLists.txt
@@ -1,367 +1,73 @@
-file(GLOB GENERAL_OPS RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*_op.cc")
-string(REPLACE "_mkldnn" "" GENERAL_OPS "${GENERAL_OPS}")
-string(REPLACE ".cc" "" GENERAL_OPS "${GENERAL_OPS}")
-list(REMOVE_DUPLICATES GENERAL_OPS)
-set(DEPS_OPS "")
-set(pybind_file ${PADDLE_BINARY_DIR}/paddle/fluid/pybind/pybind.h)
-file(WRITE ${pybind_file} "// Generated by the paddle/fluid/operator/CMakeLists.txt.  DO NOT EDIT!\n\n")
-
-set(PART_CUDA_KERNEL_FILES)
-function(op_library TARGET)
-    # op_library is a function to create op library. The interface is same as
-    # cc_library. But it handle split GPU/CPU code and link some common library
-    # for ops.
-    set(cc_srcs)
-    set(cu_srcs)
-    set(hip_cu_srcs)
-    set(miopen_hip_cc_srcs)
-    set(cu_cc_srcs)
-    set(cudnn_cu_cc_srcs)
-    set(CUDNN_FILE)
-    set(mkldnn_cc_srcs)
-    set(MKLDNN_FILE)
-    set(op_common_deps operator op_registry math_function)
-    set(options "")
-    set(oneValueArgs "")
-    set(multiValueArgs SRCS DEPS)
-    set(pybind_flag 0)
-    cmake_parse_arguments(op_library "${options}" "${oneValueArgs}"
-            "${multiValueArgs}" ${ARGN})
-
-    list(LENGTH op_library_SRCS op_library_SRCS_len)
-    if (${op_library_SRCS_len} EQUAL 0)
-        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.cc)
-            list(APPEND cc_srcs ${TARGET}.cc)
-        endif()
-        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.cu.cc)
-            list(APPEND cu_cc_srcs ${TARGET}.cu.cc)
-        endif()
-        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.cu)
-            list(APPEND cu_srcs ${TARGET}.cu)
-        endif()
-        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.part.cu)
-            set(PART_CUDA_KERNEL_FILES ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.part.cu
-                    ${PART_CUDA_KERNEL_FILES} PARENT_SCOPE)
-            list(APPEND cu_srcs ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.part.cu)
-        endif()
-
-        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${TARGET}.hip.cu)
-            list(APPEND hip_cu_srcs ${TARGET}.hip.cu)
-        endif()
-        string(REPLACE "_op" "_cudnn_op" CUDNN_FILE "${TARGET}")
-        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${CUDNN_FILE}.cu.cc)
-            list(APPEND cudnn_cu_cc_srcs ${CUDNN_FILE}.cu.cc)
-        endif()
-        if(WITH_AMD_GPU)
-            string(REPLACE "_op" "_miopen_op" MIOPEN_FILE "${TARGET}")
-            if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${MIOPEN_FILE}.hip.cc)
-                list(APPEND miopen_hip_cc_srcs ${MIOPEN_FILE}.hip.cc)
-            endif()
-        endif()
-        if(WITH_MKLDNN)
-            string(REPLACE "_op" "_mkldnn_op" MKLDNN_FILE "${TARGET}")
-            if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${MKLDNN_FILE}.cc)
-                list(APPEND mkldnn_cc_srcs ${MKLDNN_FILE}.cc)
-            endif()
-        endif()
-    else()
-        foreach(src ${op_library_SRCS})
-            if (${src} MATCHES ".*\\.hip.cu$")
-                list(APPEND hip_cu_srcs ${src})
-            elseif (${src} MATCHES ".*\\.cu$")
-                list(APPEND cu_srcs ${src})
-            elseif(${src} MATCHES ".*_cudnn_op.cu.cc$")
-                list(APPEND cudnn_cu_cc_srcs ${src})
-            elseif(WITH_AMD_GPU AND ${src} MATCHES ".*_miopen_op.hip.cc$")
-                list(APPEND miopen_hip_cc_srcs ${src})
-            elseif(WITH_MKLDNN AND ${src} MATCHES ".*_mkldnn_op.cc$")
-                list(APPEND mkldnn_cc_srcs ${src})
-            elseif(${src} MATCHES ".*\\.cu.cc$")
-                list(APPEND cu_cc_srcs ${src})
-            elseif(${src} MATCHES ".*\\.cc$")
-                list(APPEND cc_srcs ${src})
-            else()
-                message(FATAL_ERROR "${TARGET} Source file ${src} should only be .cc or .cu")
-            endif()
-        endforeach()
-    endif()
-
-    list(LENGTH cc_srcs cc_srcs_len)
-    if (${cc_srcs_len} EQUAL 0)
-        message(FATAL_ERROR "The op library ${TARGET} should contains at least one .cc file")
-    endif()
-    if (WIN32)
-    # remove windows unsupported op, because windows has no nccl, no warpctc such ops.
-    foreach(windows_unsupport_op "nccl_op" "gen_nccl_id_op" "warpctc_op" "hierarchical_sigmoid_op"
-     "crf_decoding_op" "select_op" "lstmp_op" "gru_op" "fusion_gru_op" "lstm_op" "fusion_lstm_op" "cumsum_op"
-      "fusion_seqconv_eltadd_relu_op" "channel_send_op" "channel_create_op" "channel_close_op" "channel_recv_op"
-            "fusion_seqexpand_concat_fc_op" "attention_lstm_op" "fused_embedding_fc_lstm_op" "fc_op")
-        if ("${TARGET}" STREQUAL "${windows_unsupport_op}")
-          return()
-        endif()
-    endforeach()
-    endif(WIN32)
-    set(OP_LIBRARY ${TARGET} ${OP_LIBRARY} PARENT_SCOPE)
+include(operators)
 
-    list(LENGTH op_library_DEPS op_library_DEPS_len)
-    if (${op_library_DEPS_len} GREATER 0)
-        set(DEPS_OPS ${TARGET} ${DEPS_OPS} PARENT_SCOPE)
-    endif()
-    if (WITH_GPU)
-        nv_library(${TARGET} SRCS ${cc_srcs} ${cu_cc_srcs} ${cudnn_cu_cc_srcs} ${mkldnn_cc_srcs} ${cu_srcs} DEPS ${op_library_DEPS}
-                ${op_common_deps})
-    elseif (WITH_AMD_GPU)
-        hip_library(${TARGET} SRCS ${cc_srcs} ${hip_cu_srcs} ${miopen_hip_cc_srcs} ${mkldnn_cc_srcs} DEPS ${op_library_DEPS}
-                ${op_common_deps})
-    else()
-        cc_library(${TARGET} SRCS ${cc_srcs} ${mkldnn_cc_srcs} DEPS ${op_library_DEPS}
-            ${op_common_deps})
-    endif()
-
-    # Define operators that don't need pybind here.
-    foreach(manual_pybind_op "compare_op" "logical_op" "nccl_op"
-"tensor_array_read_write_op" "tensorrt_engine_op")
-        if ("${TARGET}" STREQUAL "${manual_pybind_op}")
-            set(pybind_flag 1)
-        endif()
-    endforeach()
-
-    # The registration of USE_OP, please refer to paddle/fluid/framework/op_registry.h.
-    # Note that it's enough to just adding one operator to pybind in a *_op.cc file.
-    # And for detail pybind information, please see generated paddle/pybind/pybind.h.
-    file(READ ${TARGET}.cc TARGET_CONTENT)
-    string(REGEX MATCH "REGISTER_OPERATOR\\(.*REGISTER_OPERATOR\\(" multi_register "${TARGET_CONTENT}")
-    string(REGEX MATCH "REGISTER_OPERATOR\\([a-z0-9_]*," one_register "${multi_register}")
-    if (one_register STREQUAL "")
-        string(REPLACE "_op" "" TARGET "${TARGET}")
-    else ()
-        string(REPLACE "REGISTER_OPERATOR(" "" TARGET "${one_register}")
-        string(REPLACE "," "" TARGET "${TARGET}")
-    endif()
-
-    # pybind USE_NO_KERNEL_OP
-    # HACK: if REGISTER_OP_CPU_KERNEL presents the operator must have kernel
-    string(REGEX MATCH "REGISTER_OP_CPU_KERNEL" regex_result "${TARGET_CONTENT}")
-    string(REPLACE "_op" "" TARGET "${TARGET}")
-    if (${pybind_flag} EQUAL 0 AND regex_result STREQUAL "")
-        file(APPEND ${pybind_file} "USE_NO_KERNEL_OP(${TARGET});\n")
-        set(pybind_flag 1)
-    endif()
-
-    # pybind USE_CPU_ONLY_OP
-    list(LENGTH cu_srcs cu_srcs_len)
-    list(LENGTH cu_cc_srcs cu_cc_srcs_len)
-    list(LENGTH mkldnn_cc_srcs mkldnn_cc_srcs_len)
-    list(LENGTH hip_cu_srcs hip_cu_srcs_len)
-    list(LENGTH miopen_hip_cc_srcs miopen_hip_cc_srcs_len)
-    if (${pybind_flag} EQUAL 0 AND ${mkldnn_cc_srcs_len} EQUAL 0 AND ${cu_srcs_len} EQUAL 0 AND ${cu_cc_srcs_len} EQUAL 0 AND
-        ${hip_cu_srcs_len} EQUAL 0 AND ${miopen_hip_cc_srcs_len} EQUAL 0)
-        file(APPEND ${pybind_file} "USE_CPU_ONLY_OP(${TARGET});\n")
-        set(pybind_flag 1)
-    endif()
-
-    # pybind USE_OP_DEVICE_KERNEL for CUDNN
-    list(LENGTH cudnn_cu_cc_srcs cudnn_cu_cc_srcs_len)
-    if (WITH_GPU AND ${cudnn_cu_cc_srcs_len} GREATER 0)
-        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${TARGET}, CUDNN);\n")
-    endif()
-
-    # pybind USE_OP_DEVICE_KERNEL for MIOPEN
-    if (WITH_AMD_GPU AND ${miopen_hip_cc_srcs_len} GREATER 0)
-        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${TARGET}, MIOPEN);\n")
-    endif()
-
-    # pybind USE_OP_DEVICE_KERNEL for MKLDNN
-    if (WITH_MKLDNN AND ${mkldnn_cc_srcs_len} GREATER 0)
-      # Append first implemented MKLDNN activation operator
-      if (${MKLDNN_FILE} STREQUAL "activation_mkldnn_op")
-        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(relu, MKLDNN);\n")
-      else()
-        file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${TARGET}, MKLDNN);\n")
-      endif()
-    endif()
-
-    # pybind USE_OP
-    if (${pybind_flag} EQUAL 0)
-      # NOTE(*): activation use macro to regist the kernels, set use_op manually.
-      if(${TARGET} STREQUAL "activation")
-        file(APPEND ${pybind_file} "USE_OP(relu);\n")
-      elseif(${TARGET} STREQUAL "fake_dequantize")
-        file(APPEND ${pybind_file} "USE_OP(fake_dequantize_max_abs);\n")
-      elseif(${TARGET} STREQUAL "fake_quantize")
-        file(APPEND ${pybind_file} "USE_OP(fake_quantize_abs_max);\n")
-      elseif(${TARGET} STREQUAL "tensorrt_engine_op")
-          message(STATUS "Pybind skips [tensorrt_engine_op], for this OP is only used in inference")
-      elseif(${TARGET} STREQUAL "fc")
-        # HACK: fc only have mkldnn and cpu, which would mismatch the cpu only condition
-        file(APPEND ${pybind_file} "USE_CPU_ONLY_OP(${TARGET});\n")
-      else()
-        file(APPEND ${pybind_file} "USE_OP(${TARGET});\n")
-      endif()
-    endif()
-endfunction()
+# clean cache and pybind_file content first when rebuild
+unset(GLOB_OP_LIB CACHE)
+unset(OP_LIBRARY CACHE)
+set(pybind_file ${PADDLE_BINARY_DIR}/paddle/fluid/pybind/pybind.h CACHE INTERNAL "pybind.h file")
+file(WRITE ${pybind_file} "// Generated by the paddle/fluid/operator/CMakeLists.txt.  DO NOT EDIT!\n\n")
 
 add_subdirectory(math)
-if (NOT WIN32)
-add_subdirectory(nccl)
-if(WITH_GPU)
-    op_library(nccl_op DEPS nccl_common)
-    file(APPEND ${pybind_file} "USE_CUDA_ONLY_OP(ncclAllReduce);\n")
-else()
-    set(DEPS_OPS ${DEPS_OPS} nccl_op)
-endif()
-endif() # NOT WIN32
+add_subdirectory(controlflow)
+add_subdirectory(csp)
+add_subdirectory(detection)
+add_subdirectory(elementwise)
+add_subdirectory(fused)
+add_subdirectory(metrics)
+add_subdirectory(optimizers)
+add_subdirectory(reduce_ops)
+add_subdirectory(sequence_ops)
 
-set(DISTRIBUTE_DEPS "")
 if(WITH_DISTRIBUTE)
     add_subdirectory(distributed)
-    set(DISTRIBUTE_DEPS "")
-    if(WITH_GRPC)
-        set(DISTRIBUTE_DEPS sendrecvop_grpc grpc++_unsecure grpc_unsecure gpr cares zlib protobuf node)
-    else()
-        set(DISTRIBUTE_DEPS sendrecvop_brpc brpc leveldb snappystream snappy protobuf ssl crypto zlib node)
-        if(WITH_BRPC_RDMA)
-            find_library(IBVERBS_LIBRARY NAMES ibverbs)
-            ADD_LIBRARY(ibverbs SHARED IMPORTED GLOBAL)
-            SET_PROPERTY(TARGET ibverbs PROPERTY IMPORTED_LOCATION ${IBVERBS_LIBRARY})
-
-
-            find_library(RDMACM_LIBRARY NAMES rdmacm)
-            ADD_LIBRARY(rdmacm SHARED IMPORTED GLOBAL)
-            SET_PROPERTY(TARGET rdmacm PROPERTY IMPORTED_LOCATION ${RDMACM_LIBRARY})
-
-            set(DISTRIBUTE_DEPS ${DISTRIBUTE_DEPS} ibverbs rdmacm)
-        endif()
-    endif()
-
-    set(DISTRIBUTE_COMPILE_FLAGS "-Wno-non-virtual-dtor -Wno-error=non-virtual-dtor -Wno-error=delete-non-virtual-dtor")
-    foreach(dist_op "prefetch_op" "checkpoint_notify_op" "listen_and_serv_op" "send_op" "recv_op" "send_barrier_op" "fetch_barrier_op")
-        op_library(${dist_op} DEPS ${DISTRIBUTE_DEPS})
-        set_source_files_properties(${dist_op}.cc PROPERTIES COMPILE_FLAGS ${DISTRIBUTE_COMPILE_FLAGS})
-    endforeach()
+    add_subdirectory(distributed_ops)
+endif()
 
-    #set_source_files_properties(send_recv_op_test.cc PROPERTIES COMPILE_FLAGS ${DISTRIBUTE_COMPILE_FLAGS})
-    #cc_test(test_send_recv SRCS send_recv_op_test.cc DEPS prefetch_op send_op
-    #        listen_and_serv_op sum_op executor SERIAL)
-    if(WITH_GPU AND NOT WIN32)
-        set_source_files_properties(test_send_nccl_id.cc PROPERTIES COMPILE_FLAGS ${DISTRIBUTE_COMPILE_FLAGS})
-        cc_test(test_send_nccl_id SRCS test_send_nccl_id.cc DEPS listen_and_serv_op ${DISTRIBUTE_DEPS} executor SERIAL)
-        if(WITH_GRPC)
-            op_library(gen_nccl_id_op DEPS nccl_common sendrecvop_grpc)
-        else()
-            op_library(gen_nccl_id_op DEPS nccl_common sendrecvop_brpc)
-        endif()
-        set_source_files_properties(gen_nccl_id_op.cc PROPERTIES COMPILE_FLAGS ${DISTRIBUTE_COMPILE_FLAGS})
-    else()
-        set(DEPS_OPS ${DEPS_OPS} gen_nccl_id_op)
-    endif() # WITH_GPU AND NOT WIN32
-else()
-    set(DEPS_OPS ${DEPS_OPS}  checkpoint_notify_op prefetch_op recv_op listen_and_serv_op send_op send_barrier_op fetch_barrier_op gen_nccl_id_op)
+if (NOT WIN32)
+    add_subdirectory(reader)
 endif()
 
-op_library(cross_entropy_op DEPS cross_entropy)
-if(WITH_GPU)
-  op_library(softmax_with_cross_entropy_op DEPS cross_entropy softmax cub)
-  op_library(sequence_softmax_op DEPS cub)
-else()
-  op_library(softmax_with_cross_entropy_op DEPS cross_entropy softmax)
+if (NOT WIN32)
+    add_subdirectory(nccl)
 endif()
 
-op_library(softmax_op DEPS softmax)
 if (WITH_GPU AND TENSORRT_FOUND)
-    op_library(tensorrt_engine_op DEPS tensorrt_engine tensorrt_converter)
-    file(APPEND ${pybind_file} "USE_CUDA_ONLY_OP(tensorrt_engine);\n")
-    nv_test(test_tensorrt_engine_op SRCS tensorrt_engine_op_test.cc
-      DEPS tensorrt_engine_op
-      analysis)
-else()
-    set(DEPS_OPS ${DEPS_OPS} tensorrt_engine_op)
+    add_subdirectory(tensorrt)
 endif()
-op_library(hash_op DEPS xxhash)
-op_library(clip_by_norm_op DEPS selected_rows_functor selected_rows)
-op_library(sum_op DEPS selected_rows_functor)
-op_library(sgd_op DEPS selected_rows_functor)
-op_library(print_op DEPS lod_tensor)
-op_library(adagrad_op DEPS selected_rows_functor)
-op_library(maxout_op DEPS maxouting)
-op_library(unpool_op DEPS unpooling)
-op_library(pool_op DEPS pooling)
-op_library(pool_with_index_op DEPS pooling)
-op_library(lod_rank_table_op DEPS lod_rank_table)
-op_library(lod_tensor_to_array_op DEPS lod_rank_table_op)
-op_library(array_to_lod_tensor_op DEPS lod_rank_table_op)
-op_library(max_sequence_len_op DEPS lod_rank_table)
-op_library(sequence_conv_op DEPS context_project)
-op_library(sequence_pool_op DEPS sequence_pooling)
-if (NOT WIN32)
-    op_library(lstm_op DEPS sequence2batch lstm_compute)
-    op_library(hierarchical_sigmoid_op DEPS matrix_bit_code)
-    op_library(lstmp_op DEPS sequence2batch lstm_compute)
-    op_library(gru_op DEPS sequence2batch gru_compute)
-endif(NOT WIN32)
-op_library(recurrent_op DEPS executor)
-op_library(cos_sim_op DEPS cos_sim_functor)
-op_library(parallel_do_op DEPS executor)
-op_library(unsqueeze_op DEPS reshape_op)
-op_library(squeeze_op DEPS reshape_op)
-op_library(flatten_op DEPS reshape_op)
-op_library(sequence_pad_op DEPS sequence_padding)
-op_library(unstack_op DEPS stack_op)
-op_library(fake_quantize_op DEPS memory)
-op_library(nce_op DEPS sampler)
-if (NOT WIN32)
-op_library(crf_decoding_op DEPS jit_kernel)
-op_library(fusion_lstm_op DEPS jit_kernel)
-endif(NOT WIN32)
-if (WITH_GPU)
-    op_library(conv_op DEPS vol2col depthwise_conv im2col)
-    op_library(layer_norm_op DEPS cub)
-    op_library(reduce_mean_op DEPS cub)
-    op_library(affine_channel_op DEPS cub)
-else()
-    op_library(conv_op DEPS vol2col im2col)
-endif()
-op_library(conv_transpose_op DEPS vol2col im2col)
 
-# FIXME(typhoonzero): save/load depends lodtensor serialization functions
-op_library(save_op DEPS lod_tensor)
-op_library(load_op DEPS lod_tensor)
-op_library(save_combine_op DEPS lod_tensor)
-op_library(load_combine_op DEPS lod_tensor)
-op_library(concat_op DEPS concat_and_split)
-op_library(tensor_array_to_tensor_op DEPS concat_op)
+register_operators(EXCLUDES warpctc_op)
 
-set(DEPS_OPS ${DEPS_OPS} warpctc_op)
+# warpctc_cudnn need cudnn 7 above
 if (WITH_GPU)
     if (${CUDNN_MAJOR_VERSION} VERSION_LESS 7)
         op_library(warpctc_op DEPS dynload_warpctc sequence_padding sequence_scale SRCS warpctc_op.cc warpctc_op.cu.cc)
+    else()
+        op_library(warpctc_op DEPS dynload_warpctc sequence_padding sequence_scale)
     endif()
+else()
+    op_library(warpctc_op DEPS dynload_warpctc sequence_padding sequence_scale)
 endif()
-op_library(warpctc_op DEPS dynload_warpctc sequence_padding sequence_scale)
-
-list(REMOVE_ITEM GENERAL_OPS ${DEPS_OPS})
-
-foreach(src ${GENERAL_OPS})
-    op_library(${src})
-endforeach()
-
-file(APPEND ${pybind_file} "USE_OP(less_than);\nUSE_OP(logical_and);\nUSE_NO_KERNEL_OP(read_from_array);\n")
 
+set(COMMON_OP_DEPS "")
 
+set(COMMON_OP_DEPS ${COMMON_OP_DEPS} xxhash selected_rows_functor selected_rows lod_tensor maxouting unpooling pooling lod_rank_table context_project sequence_pooling executor dynload_warpctc sequence_padding sequence_scale cos_sim_functor memory jit_kernel concat_and_split cross_entropy softmax vol2col im2col sampler)
 if (NOT WIN32)
-add_subdirectory(reader)
-endif(NOT WIN32)
-foreach(src ${READER_LIBRARY})
-    set(OP_LIBRARY ${src} ${OP_LIBRARY})
-endforeach()
+  set(COMMON_OP_DEPS ${COMMON_OP_DEPS} sequence2batch lstm_compute matrix_bit_code gru_compute activation_functions)
+endif()
+if (WITH_GPU)
+  set(COMMON_OP_DEPS ${COMMON_OP_DEPS} depthwise_conv cub)
+endif()
 
-add_subdirectory(detection)
-foreach(src ${DETECTION_LIBRARY})
-    set(OP_LIBRARY ${src} ${OP_LIBRARY})
-endforeach()
+# FIXME(typhoonzero): operator deps may not needed.
+# op_library(lod_tensor_to_array_op DEPS lod_rank_table_op)
+# op_library(array_to_lod_tensor_op DEPS lod_rank_table_op)
+# op_library(unsqueeze_op DEPS reshape_op)
+# op_library(squeeze_op DEPS reshape_op)
+# op_library(flatten_op DEPS reshape_op)
+# op_library(unstack_op DEPS stack_op)
+# op_library(tensor_array_to_tensor_op DEPS concat_op)
 
-set(GLOB_OP_LIB ${OP_LIBRARY} CACHE INTERNAL "Global OP library")
-set(GLOB_DISTRIBUTE_DEPS ${DISTRIBUTE_DEPS} CACHE INTERNAL "distributed dependency")
+set(OPERATOR_DEPS ${OPERATOR_DEPS} ${COMMON_OP_DEPS})
+set(GLOB_OPERATOR_DEPS ${OPERATOR_DEPS} CACHE INTERNAL "Global Op dependencies")
 
 cc_test(gather_test SRCS gather_test.cc DEPS tensor)
 cc_test(scatter_test SRCS scatter_test.cc DEPS tensor)
@@ -370,18 +76,6 @@ cc_test(beam_search_op_test SRCS beam_search_op_test.cc DEPS lod_tensor beam_sea
 cc_test(strided_memcpy_test SRCS strided_memcpy_test.cc DEPS tensor memory)
 cc_test(save_load_op_test SRCS save_load_op_test.cc DEPS save_op load_op)
 cc_test(save_load_combine_op_test SRCS save_load_combine_op_test.cc DEPS save_combine_op load_combine_op)
-if(NOT WIN32)
-    nv_test(nccl_op_test SRCS nccl_op_test.cu.cc DEPS nccl_op gpu_info device_context)
-endif()
 nv_test(dropout_op_test SRCS dropout_op_test.cc DEPS dropout_op tensor)
 
-if(WITH_GPU)
-    foreach(CUDA_KERNEL_FILE ${PART_CUDA_KERNEL_FILES})
-        file(READ ${CUDA_KERNEL_FILE} TARGET_CONTENT)
-        string(REGEX MATCH "REGISTER_OP_CUDA_KERNEL\\(\\n?([^,]+),.*" MATCHED ${TARGET_CONTENT})
-        if (MATCHED)
-            string(STRIP ${CMAKE_MATCH_1} MATCHED)
-            file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${MATCHED}, CUDA);\n")
-        endif()
-    endforeach()
-endif()
+set(GLOB_OP_LIB ${OP_LIBRARY} CACHE INTERNAL "Global OP library")
diff --git a/paddle/fluid/operators/controlflow/CMakeLists.txt b/paddle/fluid/operators/controlflow/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..b1c2ee22951a3881b4ce5b82f9ff7eb01fde6e9e
--- /dev/null
+++ b/paddle/fluid/operators/controlflow/CMakeLists.txt
@@ -0,0 +1,4 @@
+include(operators)
+register_operators()
+
+file(APPEND ${pybind_file} "USE_OP(less_than);\nUSE_OP(logical_and);\nUSE_NO_KERNEL_OP(read_from_array);\n")
diff --git a/paddle/fluid/operators/compare_op.cc b/paddle/fluid/operators/controlflow/compare_op.cc
similarity index 98%
rename from paddle/fluid/operators/compare_op.cc
rename to paddle/fluid/operators/controlflow/compare_op.cc
index f40b1ba338d429c248103eeb930ac7e1bb690218..488ca7fe95f5119c59b861011993a379d08008ba 100644
--- a/paddle/fluid/operators/compare_op.cc
+++ b/paddle/fluid/operators/controlflow/compare_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/compare_op.h"
+#include "paddle/fluid/operators/controlflow/compare_op.h"
 #include <string>
 #include "paddle/fluid/framework/op_registry.h"
 
diff --git a/paddle/fluid/operators/compare_op.cu b/paddle/fluid/operators/controlflow/compare_op.cu
similarity index 94%
rename from paddle/fluid/operators/compare_op.cu
rename to paddle/fluid/operators/controlflow/compare_op.cu
index 1bf85c64fb5b4d79c62118959fd72b13ed1c63ed..b1f306358359764b919f9e570cf44f9733a7d178 100644
--- a/paddle/fluid/operators/compare_op.cu
+++ b/paddle/fluid/operators/controlflow/compare_op.cu
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/compare_op.h"
+#include "paddle/fluid/operators/controlflow/compare_op.h"
 
 REGISTER_COMPARE_KERNEL(less_than, CUDA, paddle::operators::LessThanFunctor);
 REGISTER_COMPARE_KERNEL(less_equal, CUDA, paddle::operators::LessEqualFunctor);
diff --git a/paddle/fluid/operators/compare_op.h b/paddle/fluid/operators/controlflow/compare_op.h
similarity index 97%
rename from paddle/fluid/operators/compare_op.h
rename to paddle/fluid/operators/controlflow/compare_op.h
index 1cbabdaf6767815c1fedba0eabec9b5de678e047..b7529e4ae632d31524846d9d5aa4b1883f4509a1 100644
--- a/paddle/fluid/operators/compare_op.h
+++ b/paddle/fluid/operators/controlflow/compare_op.h
@@ -16,7 +16,7 @@ limitations under the License. */
 #include <math.h>
 #include <type_traits>
 #include "paddle/fluid/framework/op_registry.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 #include "paddle/fluid/platform/transform.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/conditional_block_op.cc b/paddle/fluid/operators/controlflow/conditional_block_op.cc
similarity index 100%
rename from paddle/fluid/operators/conditional_block_op.cc
rename to paddle/fluid/operators/controlflow/conditional_block_op.cc
diff --git a/paddle/fluid/operators/feed_op.cc b/paddle/fluid/operators/controlflow/feed_op.cc
similarity index 100%
rename from paddle/fluid/operators/feed_op.cc
rename to paddle/fluid/operators/controlflow/feed_op.cc
diff --git a/paddle/fluid/operators/fetch_op.cc b/paddle/fluid/operators/controlflow/fetch_op.cc
similarity index 100%
rename from paddle/fluid/operators/fetch_op.cc
rename to paddle/fluid/operators/controlflow/fetch_op.cc
diff --git a/paddle/fluid/operators/get_places_op.cc b/paddle/fluid/operators/controlflow/get_places_op.cc
similarity index 100%
rename from paddle/fluid/operators/get_places_op.cc
rename to paddle/fluid/operators/controlflow/get_places_op.cc
diff --git a/paddle/fluid/operators/logical_op.cc b/paddle/fluid/operators/controlflow/logical_op.cc
similarity index 99%
rename from paddle/fluid/operators/logical_op.cc
rename to paddle/fluid/operators/controlflow/logical_op.cc
index 26970db8d2af62bb06fce4eb1a1f21fd41617bd1..6446cab5ec5f889dccaef90484476e55c4852dee 100644
--- a/paddle/fluid/operators/logical_op.cc
+++ b/paddle/fluid/operators/controlflow/logical_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/logical_op.h"
+#include "paddle/fluid/operators/controlflow/logical_op.h"
 #include <string>
 #include "paddle/fluid/framework/op_registry.h"
 
diff --git a/paddle/fluid/operators/logical_op.cu b/paddle/fluid/operators/controlflow/logical_op.cu
similarity index 94%
rename from paddle/fluid/operators/logical_op.cu
rename to paddle/fluid/operators/controlflow/logical_op.cu
index 7ffe4dfc268b1ad3894dd54cb13c2f424818aa05..7ca54b488bfbb260c422941b82145f092a150be7 100644
--- a/paddle/fluid/operators/logical_op.cu
+++ b/paddle/fluid/operators/controlflow/logical_op.cu
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/logical_op.h"
+#include "paddle/fluid/operators/controlflow/logical_op.h"
 
 REGISTER_BINARY_LOGICAL_KERNEL(logical_and, CUDA,
                                paddle::operators::LogicalAndFunctor);
diff --git a/paddle/fluid/operators/logical_op.h b/paddle/fluid/operators/controlflow/logical_op.h
similarity index 100%
rename from paddle/fluid/operators/logical_op.h
rename to paddle/fluid/operators/controlflow/logical_op.h
diff --git a/paddle/fluid/operators/parallel_do_op.cc b/paddle/fluid/operators/controlflow/parallel_do_op.cc
similarity index 100%
rename from paddle/fluid/operators/parallel_do_op.cc
rename to paddle/fluid/operators/controlflow/parallel_do_op.cc
diff --git a/paddle/fluid/operators/tensor_array_read_write_op.cc b/paddle/fluid/operators/controlflow/tensor_array_read_write_op.cc
similarity index 100%
rename from paddle/fluid/operators/tensor_array_read_write_op.cc
rename to paddle/fluid/operators/controlflow/tensor_array_read_write_op.cc
diff --git a/paddle/fluid/operators/while_op.cc b/paddle/fluid/operators/controlflow/while_op.cc
similarity index 100%
rename from paddle/fluid/operators/while_op.cc
rename to paddle/fluid/operators/controlflow/while_op.cc
diff --git a/paddle/fluid/operators/csp/CMakeLists.txt b/paddle/fluid/operators/csp/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..5d468316e8eacb73c4a4ce81c784880bb5e46c2d
--- /dev/null
+++ b/paddle/fluid/operators/csp/CMakeLists.txt
@@ -0,0 +1,2 @@
+include(operators)
+register_operators()
diff --git a/paddle/fluid/operators/go_op.cc b/paddle/fluid/operators/csp/go_op.cc
similarity index 100%
rename from paddle/fluid/operators/go_op.cc
rename to paddle/fluid/operators/csp/go_op.cc
diff --git a/paddle/fluid/operators/detection/CMakeLists.txt b/paddle/fluid/operators/detection/CMakeLists.txt
index e5c3f0eeb385e1a15fdbb12a989603996420efe3..58f6f48467310ffb2429ad440f58fcd823edf079 100644
--- a/paddle/fluid/operators/detection/CMakeLists.txt
+++ b/paddle/fluid/operators/detection/CMakeLists.txt
@@ -40,4 +40,8 @@ endif()
 
 detection_library(roi_perspective_transform_op SRCS roi_perspective_transform_op.cc roi_perspective_transform_op.cu)
 #Export local libraries to parent
-set(DETECTION_LIBRARY ${LOCAL_DETECTION_LIBS} PARENT_SCOPE)
+# set(DETECTION_LIBRARY ${LOCAL_DETECTION_LIBS} PARENT_SCOPE)
+
+foreach(src ${LOCAL_DETECTION_LIBS})
+    set(OP_LIBRARY ${src} ${OP_LIBRARY} CACHE INTERNAL "op libs")
+endforeach()
diff --git a/paddle/fluid/operators/distributed_ops/CMakeLists.txt b/paddle/fluid/operators/distributed_ops/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..a071babc8227d76504b3404532146e3f0692358e
--- /dev/null
+++ b/paddle/fluid/operators/distributed_ops/CMakeLists.txt
@@ -0,0 +1,40 @@
+include(operators)
+
+set(DISTRIBUTE_DEPS "")
+if(WITH_GRPC)
+    set(DISTRIBUTE_DEPS sendrecvop_grpc grpc++_unsecure grpc_unsecure gpr cares zlib protobuf node)
+else()
+    set(DISTRIBUTE_DEPS sendrecvop_brpc brpc leveldb snappystream snappy protobuf ssl crypto zlib node)
+    if(WITH_BRPC_RDMA)
+        find_library(IBVERBS_LIBRARY NAMES ibverbs)
+        ADD_LIBRARY(ibverbs SHARED IMPORTED GLOBAL)
+        SET_PROPERTY(TARGET ibverbs PROPERTY IMPORTED_LOCATION ${IBVERBS_LIBRARY})
+
+
+        find_library(RDMACM_LIBRARY NAMES rdmacm)
+        ADD_LIBRARY(rdmacm SHARED IMPORTED GLOBAL)
+        SET_PROPERTY(TARGET rdmacm PROPERTY IMPORTED_LOCATION ${RDMACM_LIBRARY})
+
+        set(DISTRIBUTE_DEPS ${DISTRIBUTE_DEPS} ibverbs rdmacm)
+    endif()
+endif()
+
+set(DISTRIBUTE_COMPILE_FLAGS "-Wno-non-virtual-dtor -Wno-error=non-virtual-dtor -Wno-error=delete-non-virtual-dtor")
+
+
+file(GLOB OPS RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*_op.cc")
+list(REMOVE_DUPLICATES OPS)
+
+foreach(src ${OPS})
+    set_source_files_properties(${src} PROPERTIES COMPILE_FLAGS ${DISTRIBUTE_COMPILE_FLAGS})
+endforeach()
+
+register_operators(EXCLUDES gen_nccl_id_op)
+
+if(WITH_GPU AND NOT WIN32)
+    set(DISTRIBUTE_DEPS ${DISTRIBUTE_DEPS} nccl_common)
+    op_library(gen_nccl_id_op)
+endif()
+
+set(OPERATOR_DEPS ${OPERATOR_DEPS} ${DISTRIBUTE_DEPS} PARENT_SCOPE)
+set(GLOB_DISTRIBUTE_DEPS ${DISTRIBUTE_DEPS} CACHE INTERNAL "distributed dependency")
diff --git a/paddle/fluid/operators/checkpoint_notify_op.cc b/paddle/fluid/operators/distributed_ops/checkpoint_notify_op.cc
similarity index 98%
rename from paddle/fluid/operators/checkpoint_notify_op.cc
rename to paddle/fluid/operators/distributed_ops/checkpoint_notify_op.cc
index defa287bdb913e406aa7e2a383cefc3cca8c4d94..ed4dced51356515d5910e2962c9ee91a1997dbf0 100644
--- a/paddle/fluid/operators/checkpoint_notify_op.cc
+++ b/paddle/fluid/operators/distributed_ops/checkpoint_notify_op.cc
@@ -19,7 +19,7 @@ limitations under the License. */
 #include "paddle/fluid/framework/lod_tensor.h"
 #include "paddle/fluid/framework/op_registry.h"
 #include "paddle/fluid/operators/detail/macros.h"
-#include "paddle/fluid/operators/send_recv_util.h"
+#include "paddle/fluid/operators/distributed_ops/send_recv_util.h"
 #include "paddle/fluid/string/printf.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/fake_init_op.cc b/paddle/fluid/operators/distributed_ops/fake_init_op.cc
similarity index 100%
rename from paddle/fluid/operators/fake_init_op.cc
rename to paddle/fluid/operators/distributed_ops/fake_init_op.cc
diff --git a/paddle/fluid/operators/fetch_barrier_op.cc b/paddle/fluid/operators/distributed_ops/fetch_barrier_op.cc
similarity index 100%
rename from paddle/fluid/operators/fetch_barrier_op.cc
rename to paddle/fluid/operators/distributed_ops/fetch_barrier_op.cc
diff --git a/paddle/fluid/operators/gen_nccl_id_op.cc b/paddle/fluid/operators/distributed_ops/gen_nccl_id_op.cc
similarity index 100%
rename from paddle/fluid/operators/gen_nccl_id_op.cc
rename to paddle/fluid/operators/distributed_ops/gen_nccl_id_op.cc
diff --git a/paddle/fluid/operators/listen_and_serv_op.cc b/paddle/fluid/operators/distributed_ops/listen_and_serv_op.cc
similarity index 99%
rename from paddle/fluid/operators/listen_and_serv_op.cc
rename to paddle/fluid/operators/distributed_ops/listen_and_serv_op.cc
index e3d09e2d14817fe0f2ccda18ed90c9436b399ae3..9f0c7db0e1133f6d73e73a9d162a945ba4c17dc6 100644
--- a/paddle/fluid/operators/listen_and_serv_op.cc
+++ b/paddle/fluid/operators/distributed_ops/listen_and_serv_op.cc
@@ -25,7 +25,7 @@ limitations under the License. */
 #include "paddle/fluid/operators/math/math_function.h"
 
 #include "paddle/fluid/operators/distributed/request_handler_impl.h"
-#include "paddle/fluid/operators/listen_and_serv_op.h"
+#include "paddle/fluid/operators/distributed_ops/listen_and_serv_op.h"
 
 DEFINE_int32(rpc_send_thread_num, 5, "number of threads for rpc send");
 DEFINE_int32(rpc_get_thread_num, 5, "number of threads for rpc get");
diff --git a/paddle/fluid/operators/listen_and_serv_op.h b/paddle/fluid/operators/distributed_ops/listen_and_serv_op.h
similarity index 100%
rename from paddle/fluid/operators/listen_and_serv_op.h
rename to paddle/fluid/operators/distributed_ops/listen_and_serv_op.h
diff --git a/paddle/fluid/operators/merge_ids_op.cc b/paddle/fluid/operators/distributed_ops/merge_ids_op.cc
similarity index 98%
rename from paddle/fluid/operators/merge_ids_op.cc
rename to paddle/fluid/operators/distributed_ops/merge_ids_op.cc
index 6e0e13698097ade36449f2e8ff6ab981a1b24311..252a63cb605f65e8572281a05e884fb8b020a820 100644
--- a/paddle/fluid/operators/merge_ids_op.cc
+++ b/paddle/fluid/operators/distributed_ops/merge_ids_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/merge_ids_op.h"
+#include "paddle/fluid/operators/distributed_ops/merge_ids_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/merge_ids_op.h b/paddle/fluid/operators/distributed_ops/merge_ids_op.h
similarity index 100%
rename from paddle/fluid/operators/merge_ids_op.h
rename to paddle/fluid/operators/distributed_ops/merge_ids_op.h
diff --git a/paddle/fluid/operators/prefetch_op.cc b/paddle/fluid/operators/distributed_ops/prefetch_op.cc
similarity index 98%
rename from paddle/fluid/operators/prefetch_op.cc
rename to paddle/fluid/operators/distributed_ops/prefetch_op.cc
index 55853d25460bf6e3d07c829d686e71cc9367118c..faa67a28d86235625a87b8bd7b87685e09c75f0b 100644
--- a/paddle/fluid/operators/prefetch_op.cc
+++ b/paddle/fluid/operators/distributed_ops/prefetch_op.cc
@@ -19,7 +19,7 @@ limitations under the License. */
 #include "paddle/fluid/framework/lod_tensor.h"
 #include "paddle/fluid/framework/op_registry.h"
 #include "paddle/fluid/operators/detail/macros.h"
-#include "paddle/fluid/operators/send_recv_util.h"
+#include "paddle/fluid/operators/distributed_ops/send_recv_util.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/recv_op.cc b/paddle/fluid/operators/distributed_ops/recv_op.cc
similarity index 100%
rename from paddle/fluid/operators/recv_op.cc
rename to paddle/fluid/operators/distributed_ops/recv_op.cc
diff --git a/paddle/fluid/operators/ref_by_trainer_id_op.cc b/paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.cc
similarity index 97%
rename from paddle/fluid/operators/ref_by_trainer_id_op.cc
rename to paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.cc
index 6cb651af6dc3d8e301365968787c199acc4c60ee..98b0af7688b928f21573247b327bee1d22a73f17 100644
--- a/paddle/fluid/operators/ref_by_trainer_id_op.cc
+++ b/paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/ref_by_trainer_id_op.h"
+#include "paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.h"
 #include <string>
 
 namespace paddle {
diff --git a/paddle/fluid/operators/ref_by_trainer_id_op.cu.cc b/paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.cu.cc
similarity index 94%
rename from paddle/fluid/operators/ref_by_trainer_id_op.cu.cc
rename to paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.cu.cc
index b98e2b5c9c7341f2a424fb4b32ff1e8bc45a056c..168cd51355de56c2e2a83ba73d7eb14f6ba6e533 100644
--- a/paddle/fluid/operators/ref_by_trainer_id_op.cu.cc
+++ b/paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.cu.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/ref_by_trainer_id_op.h"
+#include "paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.h"
 
 REGISTER_OP_CUDA_KERNEL(
     ref_by_trainer_id,
diff --git a/paddle/fluid/operators/ref_by_trainer_id_op.h b/paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.h
similarity index 100%
rename from paddle/fluid/operators/ref_by_trainer_id_op.h
rename to paddle/fluid/operators/distributed_ops/ref_by_trainer_id_op.h
diff --git a/paddle/fluid/operators/send_barrier_op.cc b/paddle/fluid/operators/distributed_ops/send_barrier_op.cc
similarity index 100%
rename from paddle/fluid/operators/send_barrier_op.cc
rename to paddle/fluid/operators/distributed_ops/send_barrier_op.cc
diff --git a/paddle/fluid/operators/send_op.cc b/paddle/fluid/operators/distributed_ops/send_op.cc
similarity index 98%
rename from paddle/fluid/operators/send_op.cc
rename to paddle/fluid/operators/distributed_ops/send_op.cc
index 0ad43d56d3cd7500290dc1e386a2dbaf4453a191..be53a1a32b59d7c0235382f5db18d2203b4a035a 100644
--- a/paddle/fluid/operators/send_op.cc
+++ b/paddle/fluid/operators/distributed_ops/send_op.cc
@@ -20,7 +20,7 @@ limitations under the License. */
 #include "paddle/fluid/framework/lod_tensor.h"
 #include "paddle/fluid/framework/op_registry.h"
 #include "paddle/fluid/operators/detail/macros.h"
-#include "paddle/fluid/operators/send_recv_util.h"
+#include "paddle/fluid/operators/distributed_ops/send_recv_util.h"
 #include "paddle/fluid/platform/profiler.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/send_recv_op_test.cc b/paddle/fluid/operators/distributed_ops/send_recv_op_test.cc
similarity index 99%
rename from paddle/fluid/operators/send_recv_op_test.cc
rename to paddle/fluid/operators/distributed_ops/send_recv_op_test.cc
index d79b16e3cca714d44c88834082cea9367480da9a..bf798a8251fcb4148db486f26d32525b59299c81 100644
--- a/paddle/fluid/operators/send_recv_op_test.cc
+++ b/paddle/fluid/operators/distributed_ops/send_recv_op_test.cc
@@ -20,7 +20,7 @@ limitations under the License. */
 #include "paddle/fluid/framework/op_registry.h"
 #include "paddle/fluid/framework/operator.h"
 #include "paddle/fluid/framework/program_desc.h"
-#include "paddle/fluid/operators/listen_and_serv_op.h"
+#include "paddle/fluid/operators/distributed_ops/listen_and_serv_op.h"
 #include "paddle/fluid/operators/math/math_function.h"
 #include "paddle/fluid/operators/math/selected_rows_functor.h"
 #include "paddle/fluid/string/printf.h"
diff --git a/paddle/fluid/operators/send_recv_util.h b/paddle/fluid/operators/distributed_ops/send_recv_util.h
similarity index 100%
rename from paddle/fluid/operators/send_recv_util.h
rename to paddle/fluid/operators/distributed_ops/send_recv_util.h
diff --git a/paddle/fluid/operators/split_byref_op.cc b/paddle/fluid/operators/distributed_ops/split_byref_op.cc
similarity index 98%
rename from paddle/fluid/operators/split_byref_op.cc
rename to paddle/fluid/operators/distributed_ops/split_byref_op.cc
index bc998e1abbd7131a7497288cc9d66315a6fedc85..d65e7ffe5a492fe5df038bb6bd469e09de6f95ca 100644
--- a/paddle/fluid/operators/split_byref_op.cc
+++ b/paddle/fluid/operators/distributed_ops/split_byref_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/split_byref_op.h"
+#include "paddle/fluid/operators/distributed_ops/split_byref_op.h"
 #include "paddle/fluid/operators/split_op.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/split_byref_op.cu.cc b/paddle/fluid/operators/distributed_ops/split_byref_op.cu.cc
similarity index 91%
rename from paddle/fluid/operators/split_byref_op.cu.cc
rename to paddle/fluid/operators/distributed_ops/split_byref_op.cu.cc
index 5ee6186f3541b7dcb845ce0c6d28081685925da0..056659c3ea61f6233a6dda56ca1e272e72770d4a 100644
--- a/paddle/fluid/operators/split_byref_op.cu.cc
+++ b/paddle/fluid/operators/distributed_ops/split_byref_op.cu.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/split_byref_op.h"
+#include "paddle/fluid/operators/distributed_ops/split_byref_op.h"
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
     split_byref,
diff --git a/paddle/fluid/operators/split_byref_op.h b/paddle/fluid/operators/distributed_ops/split_byref_op.h
similarity index 100%
rename from paddle/fluid/operators/split_byref_op.h
rename to paddle/fluid/operators/distributed_ops/split_byref_op.h
diff --git a/paddle/fluid/operators/split_ids_op.cc b/paddle/fluid/operators/distributed_ops/split_ids_op.cc
similarity index 98%
rename from paddle/fluid/operators/split_ids_op.cc
rename to paddle/fluid/operators/distributed_ops/split_ids_op.cc
index 01d432e13068f7b718d08dc15d8cc99a7fbb0afe..f61d387fbef636298c412c227bf7a56a04f69c63 100644
--- a/paddle/fluid/operators/split_ids_op.cc
+++ b/paddle/fluid/operators/distributed_ops/split_ids_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/split_ids_op.h"
+#include "paddle/fluid/operators/distributed_ops/split_ids_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/split_ids_op.h b/paddle/fluid/operators/distributed_ops/split_ids_op.h
similarity index 100%
rename from paddle/fluid/operators/split_ids_op.h
rename to paddle/fluid/operators/distributed_ops/split_ids_op.h
diff --git a/paddle/fluid/operators/test_send_nccl_id.cc b/paddle/fluid/operators/distributed_ops/test_send_nccl_id.cc
similarity index 96%
rename from paddle/fluid/operators/test_send_nccl_id.cc
rename to paddle/fluid/operators/distributed_ops/test_send_nccl_id.cc
index b5426e17aac19dc07ee56545fac8472d9ef0d93c..a73cb08eca272b044501d48e7b8c5b7dc8553a50 100644
--- a/paddle/fluid/operators/test_send_nccl_id.cc
+++ b/paddle/fluid/operators/distributed_ops/test_send_nccl_id.cc
@@ -22,14 +22,14 @@ limitations under the License. */
 #include "paddle/fluid/framework/program_desc.h"
 #include "paddle/fluid/operators/detail/macros.h"
 #include "paddle/fluid/operators/distributed/request_handler_impl.h"
-#include "paddle/fluid/operators/listen_and_serv_op.h"
+#include "paddle/fluid/operators/distributed_ops/listen_and_serv_op.h"
 #include "paddle/fluid/operators/math/math_function.h"
 #include "paddle/fluid/operators/math/selected_rows_functor.h"
 #include "paddle/fluid/platform/nccl_helper.h"
 #include "paddle/fluid/string/printf.h"
 
 #ifdef PADDLE_WITH_GRPC
-#include "paddle/fluid/operators/send_recv_util.h"
+#include "paddle/fluid/operators/distributed_ops/send_recv_util.h"
 #endif
 
 USE_NO_KERNEL_OP(listen_and_serv);
diff --git a/paddle/fluid/operators/elementwise/CMakeLists.txt b/paddle/fluid/operators/elementwise/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..5d468316e8eacb73c4a4ce81c784880bb5e46c2d
--- /dev/null
+++ b/paddle/fluid/operators/elementwise/CMakeLists.txt
@@ -0,0 +1,2 @@
+include(operators)
+register_operators()
diff --git a/paddle/fluid/operators/elementwise_add_mkldnn_op.cc b/paddle/fluid/operators/elementwise/elementwise_add_mkldnn_op.cc
similarity index 97%
rename from paddle/fluid/operators/elementwise_add_mkldnn_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_add_mkldnn_op.cc
index 9ad82aec8182d6ba06b67391d71317a3d0df1833..6a6741d8fc54d22addca91b75dfabf5950c1a35a 100644
--- a/paddle/fluid/operators/elementwise_add_mkldnn_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_add_mkldnn_op.cc
@@ -13,8 +13,8 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #include "paddle/fluid/memory/memcpy.h"
-#include "paddle/fluid/operators/elementwise_add_op.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_add_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 
 #include "paddle/fluid/platform/mkldnn_helper.h"
 
diff --git a/paddle/fluid/operators/elementwise_add_op.cc b/paddle/fluid/operators/elementwise/elementwise_add_op.cc
similarity index 92%
rename from paddle/fluid/operators/elementwise_add_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_add_op.cc
index 3c97ac995c649ecd0d196a584240e1e7ac04f08e..7e789cd8d9143164c2346b067855eb904e00075f 100644
--- a/paddle/fluid/operators/elementwise_add_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_add_op.cc
@@ -12,8 +12,8 @@ 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 "paddle/fluid/operators/elementwise_add_op.h"
-#include "paddle/fluid/operators/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_add_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
 namespace ops = paddle::operators;
 REGISTER_ELEMWISE_GRAD_MAKER(elementwise_add, Add);
 REGISTER_ELEMWISE_EXPLICIT_OP(elementwise_add, "Add", "Out = X + Y", "Out",
diff --git a/paddle/fluid/operators/elementwise_add_op.cu b/paddle/fluid/operators/elementwise/elementwise_add_op.cu
similarity index 95%
rename from paddle/fluid/operators/elementwise_add_op.cu
rename to paddle/fluid/operators/elementwise/elementwise_add_op.cu
index f9f5c66d34fa1d73db00173e493f9953b8579518..2fb7eeb4b9e3119a6eea3e69a2a6002a80f6c0f3 100644
--- a/paddle/fluid/operators/elementwise_add_op.cu
+++ b/paddle/fluid/operators/elementwise/elementwise_add_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/elementwise_add_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_add_op.h"
 #include "paddle/fluid/platform/float16.h"
 
 namespace ops = paddle::operators;
diff --git a/paddle/fluid/operators/elementwise_add_op.h b/paddle/fluid/operators/elementwise/elementwise_add_op.h
similarity index 97%
rename from paddle/fluid/operators/elementwise_add_op.h
rename to paddle/fluid/operators/elementwise/elementwise_add_op.h
index 9edbdbefe76600dc4bf937d95e70d11450206cd4..69f640ab6649df673f07ac0cef81bf80d16eb98d 100644
--- a/paddle/fluid/operators/elementwise_add_op.h
+++ b/paddle/fluid/operators/elementwise/elementwise_add_op.h
@@ -15,8 +15,8 @@ limitations under the License. */
 #pragma once
 
 #include "paddle/fluid/framework/eigen.h"
-#include "paddle/fluid/operators/elementwise_op.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 #include "paddle/fluid/operators/math/blas.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/elementwise_div_op.cc b/paddle/fluid/operators/elementwise/elementwise_div_op.cc
similarity index 91%
rename from paddle/fluid/operators/elementwise_div_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_div_op.cc
index 84c8a65e5f859d276ae6d5f1a3f25c9d713a7a61..85612ba47448a7b0d712e9314e3980019c96e9c3 100644
--- a/paddle/fluid/operators/elementwise_div_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_div_op.cc
@@ -12,8 +12,8 @@ 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 "paddle/fluid/operators/elementwise_div_op.h"
-#include "paddle/fluid/operators/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_div_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
 namespace ops = paddle::operators;
 
 REGISTER_ELEMWISE_OP(elementwise_div, "Div", "Out = X / Y");
diff --git a/paddle/fluid/operators/elementwise_div_op.cu b/paddle/fluid/operators/elementwise/elementwise_div_op.cu
similarity index 95%
rename from paddle/fluid/operators/elementwise_div_op.cu
rename to paddle/fluid/operators/elementwise/elementwise_div_op.cu
index 588d1f7420241ba1697e5141e4e4a2870f2dc87c..c5a1a7e08d89f3ef205af4c37246f8fa288189f3 100644
--- a/paddle/fluid/operators/elementwise_div_op.cu
+++ b/paddle/fluid/operators/elementwise/elementwise_div_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/elementwise_div_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_div_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/elementwise_div_op.h b/paddle/fluid/operators/elementwise/elementwise_div_op.h
similarity index 94%
rename from paddle/fluid/operators/elementwise_div_op.h
rename to paddle/fluid/operators/elementwise/elementwise_div_op.h
index cdb1264d298ef48d6b3da39d63ff1d09e1561aa4..8a07339077aeaa4403ffd1e1e30e0d58a9cc30e7 100644
--- a/paddle/fluid/operators/elementwise_div_op.h
+++ b/paddle/fluid/operators/elementwise/elementwise_div_op.h
@@ -14,8 +14,8 @@ limitations under the License. */
 
 #pragma once
 
-#include "paddle/fluid/operators/elementwise_op.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 namespace paddle {
 namespace operators {
 
diff --git a/paddle/fluid/operators/elementwise_max_op.cc b/paddle/fluid/operators/elementwise/elementwise_max_op.cc
similarity index 91%
rename from paddle/fluid/operators/elementwise_max_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_max_op.cc
index 411671335a19ae2283ca9db8b8f6bcbb6a6b630a..ea0dcd736e5700fb0f341938ac3e3e3b178f29c1 100644
--- a/paddle/fluid/operators/elementwise_max_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_max_op.cc
@@ -12,8 +12,8 @@ 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 "paddle/fluid/operators/elementwise_max_op.h"
-#include "paddle/fluid/operators/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_max_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
 namespace ops = paddle::operators;
 REGISTER_ELEMWISE_OP(elementwise_max, "Max", "Out = max(X, Y)");
 REGISTER_OP_CPU_KERNEL(
diff --git a/paddle/fluid/operators/elementwise_max_op.cu b/paddle/fluid/operators/elementwise/elementwise_max_op.cu
similarity index 95%
rename from paddle/fluid/operators/elementwise_max_op.cu
rename to paddle/fluid/operators/elementwise/elementwise_max_op.cu
index 32c99835d66d8b11b72af162230aa383c7e4a57c..a90dcd3ecf0da114110db5946e111a8b3a925e42 100644
--- a/paddle/fluid/operators/elementwise_max_op.cu
+++ b/paddle/fluid/operators/elementwise/elementwise_max_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/elementwise_max_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_max_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/elementwise_max_op.h b/paddle/fluid/operators/elementwise/elementwise_max_op.h
similarity index 94%
rename from paddle/fluid/operators/elementwise_max_op.h
rename to paddle/fluid/operators/elementwise/elementwise_max_op.h
index 367489dd563f7d8bdf430517cadf49d4ef2a0105..3ee0c32e0d5d5df02d5d157416918fb4fb3aca92 100644
--- a/paddle/fluid/operators/elementwise_max_op.h
+++ b/paddle/fluid/operators/elementwise/elementwise_max_op.h
@@ -14,8 +14,8 @@ limitations under the License. */
 
 #pragma once
 
-#include "paddle/fluid/operators/elementwise_op.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/elementwise_min_op.cc b/paddle/fluid/operators/elementwise/elementwise_min_op.cc
similarity index 91%
rename from paddle/fluid/operators/elementwise_min_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_min_op.cc
index 816192083d2275b26e6dd9afc76f2c021a01cf73..b263b9addd40cfd329d2cc8588c278df2cb008e9 100644
--- a/paddle/fluid/operators/elementwise_min_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_min_op.cc
@@ -12,8 +12,8 @@ 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 "paddle/fluid/operators/elementwise_min_op.h"
-#include "paddle/fluid/operators/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_min_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
 namespace ops = paddle::operators;
 REGISTER_ELEMWISE_OP(elementwise_min, "Min", "Out = min(X, Y)");
 REGISTER_OP_CPU_KERNEL(
diff --git a/paddle/fluid/operators/elementwise_min_op.cu b/paddle/fluid/operators/elementwise/elementwise_min_op.cu
similarity index 95%
rename from paddle/fluid/operators/elementwise_min_op.cu
rename to paddle/fluid/operators/elementwise/elementwise_min_op.cu
index a237c9c503ec998fd74fec50a1d7949279bb38f0..ab77709c28c15a925bd3deac07c43e12b12cb781 100644
--- a/paddle/fluid/operators/elementwise_min_op.cu
+++ b/paddle/fluid/operators/elementwise/elementwise_min_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/elementwise_min_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_min_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/elementwise_min_op.h b/paddle/fluid/operators/elementwise/elementwise_min_op.h
similarity index 94%
rename from paddle/fluid/operators/elementwise_min_op.h
rename to paddle/fluid/operators/elementwise/elementwise_min_op.h
index 1bd0a6279766c8eba92d1e3a76191c59410286b2..d04e372faaa4e6296e982afe6155cdde2fec4f81 100644
--- a/paddle/fluid/operators/elementwise_min_op.h
+++ b/paddle/fluid/operators/elementwise/elementwise_min_op.h
@@ -14,8 +14,8 @@ limitations under the License. */
 
 #pragma once
 
-#include "paddle/fluid/operators/elementwise_op.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 namespace paddle {
 namespace operators {
 
diff --git a/paddle/fluid/operators/elementwise_mul_op.cc b/paddle/fluid/operators/elementwise/elementwise_mul_op.cc
similarity index 95%
rename from paddle/fluid/operators/elementwise_mul_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_mul_op.cc
index 86a8459a79135d1fbcba6886172acc5a2abdb88b..d5e3300ac954aebf34a9c65fbca8de8fa2685932 100644
--- a/paddle/fluid/operators/elementwise_mul_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_mul_op.cc
@@ -12,9 +12,9 @@ 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 "paddle/fluid/operators/elementwise_mul_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_mul_op.h"
 #include <string>
-#include "paddle/fluid/operators/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/elementwise_mul_op.cu b/paddle/fluid/operators/elementwise/elementwise_mul_op.cu
similarity index 95%
rename from paddle/fluid/operators/elementwise_mul_op.cu
rename to paddle/fluid/operators/elementwise/elementwise_mul_op.cu
index 2fb1b4bee689c059625e3dbd59f80c541ace83a0..4d16bc38e1d8e4cbbe3afbe08f233e14329e0f2e 100644
--- a/paddle/fluid/operators/elementwise_mul_op.cu
+++ b/paddle/fluid/operators/elementwise/elementwise_mul_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/elementwise_mul_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_mul_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/elementwise_mul_op.h b/paddle/fluid/operators/elementwise/elementwise_mul_op.h
similarity index 96%
rename from paddle/fluid/operators/elementwise_mul_op.h
rename to paddle/fluid/operators/elementwise/elementwise_mul_op.h
index 29e4ab7db1377b6aa80e94a26ab3cb8669f9154a..dc25bc57103286ce183a4649964fd96c62169b7f 100644
--- a/paddle/fluid/operators/elementwise_mul_op.h
+++ b/paddle/fluid/operators/elementwise/elementwise_mul_op.h
@@ -13,8 +13,8 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #pragma once
-#include "paddle/fluid/operators/elementwise_op.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 #include "paddle/fluid/operators/math/blas.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/elementwise_op.h b/paddle/fluid/operators/elementwise/elementwise_op.h
similarity index 100%
rename from paddle/fluid/operators/elementwise_op.h
rename to paddle/fluid/operators/elementwise/elementwise_op.h
diff --git a/paddle/fluid/operators/elementwise_op_function.h b/paddle/fluid/operators/elementwise/elementwise_op_function.h
similarity index 100%
rename from paddle/fluid/operators/elementwise_op_function.h
rename to paddle/fluid/operators/elementwise/elementwise_op_function.h
diff --git a/paddle/fluid/operators/elementwise_pow_op.cc b/paddle/fluid/operators/elementwise/elementwise_pow_op.cc
similarity index 90%
rename from paddle/fluid/operators/elementwise_pow_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_pow_op.cc
index 5fd6bde9ba0930e29f2161f1ff23ff9f5e7dc85d..6335e67a8a48c8702f0cb14ce947275d47e01d17 100644
--- a/paddle/fluid/operators/elementwise_pow_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_pow_op.cc
@@ -12,9 +12,9 @@ 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 "paddle/fluid/operators/elementwise_pow_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_pow_op.h"
 #include <string>
-#include "paddle/fluid/operators/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/elementwise_pow_op.cu b/paddle/fluid/operators/elementwise/elementwise_pow_op.cu
similarity index 92%
rename from paddle/fluid/operators/elementwise_pow_op.cu
rename to paddle/fluid/operators/elementwise/elementwise_pow_op.cu
index 1f19ebd470973137b465381e498ab07a36323c14..6ee0779f23bc2c734aa1d439abb12f366227e686 100644
--- a/paddle/fluid/operators/elementwise_pow_op.cu
+++ b/paddle/fluid/operators/elementwise/elementwise_pow_op.cu
@@ -10,7 +10,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/elementwise_pow_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_pow_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/elementwise_pow_op.h b/paddle/fluid/operators/elementwise/elementwise_pow_op.h
similarity index 95%
rename from paddle/fluid/operators/elementwise_pow_op.h
rename to paddle/fluid/operators/elementwise/elementwise_pow_op.h
index 8c1c5f9f98018d8d4368a9333e2004141615775d..dc584b4c32fc3063da0c6de50577d28afcb63b83 100644
--- a/paddle/fluid/operators/elementwise_pow_op.h
+++ b/paddle/fluid/operators/elementwise/elementwise_pow_op.h
@@ -15,7 +15,7 @@ limitations under the License. */
 #pragma once
 
 #include <cmath>
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/elementwise_sub_op.cc b/paddle/fluid/operators/elementwise/elementwise_sub_op.cc
similarity index 92%
rename from paddle/fluid/operators/elementwise_sub_op.cc
rename to paddle/fluid/operators/elementwise/elementwise_sub_op.cc
index b7224261e6a7ca82dff92a25f5fe8818c08e676d..efc66374c812cbd07adef6ac25c9616b880ec383 100644
--- a/paddle/fluid/operators/elementwise_sub_op.cc
+++ b/paddle/fluid/operators/elementwise/elementwise_sub_op.cc
@@ -12,8 +12,8 @@ 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 "paddle/fluid/operators/elementwise_sub_op.h"
-#include "paddle/fluid/operators/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_sub_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
 namespace ops = paddle::operators;
 REGISTER_ELEMWISE_GRAD_MAKER(elementwise_sub, Sub);
 REGISTER_ELEMWISE_EXPLICIT_OP(elementwise_sub, "Sub", "Out = X - Y", "Out",
diff --git a/paddle/fluid/operators/elementwise_sub_op.cu b/paddle/fluid/operators/elementwise/elementwise_sub_op.cu
similarity index 95%
rename from paddle/fluid/operators/elementwise_sub_op.cu
rename to paddle/fluid/operators/elementwise/elementwise_sub_op.cu
index 8709f686f9af1bf4dacbc2dfc3e2d5dcc1c59b9a..8d9bf7c4d81d49d83b5d1cf0369be5c9957242b4 100644
--- a/paddle/fluid/operators/elementwise_sub_op.cu
+++ b/paddle/fluid/operators/elementwise/elementwise_sub_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/elementwise_sub_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_sub_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/elementwise_sub_op.h b/paddle/fluid/operators/elementwise/elementwise_sub_op.h
similarity index 94%
rename from paddle/fluid/operators/elementwise_sub_op.h
rename to paddle/fluid/operators/elementwise/elementwise_sub_op.h
index 7204c43464e0b81126148b86f64a36b0e299368b..770323fe5a8fe7c1051b418b2541ab4c669635b4 100644
--- a/paddle/fluid/operators/elementwise_sub_op.h
+++ b/paddle/fluid/operators/elementwise/elementwise_sub_op.h
@@ -13,8 +13,8 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #pragma once
-#include "paddle/fluid/operators/elementwise_op.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/fused/CMakeLists.txt b/paddle/fluid/operators/fused/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..5d468316e8eacb73c4a4ce81c784880bb5e46c2d
--- /dev/null
+++ b/paddle/fluid/operators/fused/CMakeLists.txt
@@ -0,0 +1,2 @@
+include(operators)
+register_operators()
diff --git a/paddle/fluid/operators/fused_elemwise_activation_op.cc b/paddle/fluid/operators/fused/fused_elemwise_activation_op.cc
similarity index 99%
rename from paddle/fluid/operators/fused_elemwise_activation_op.cc
rename to paddle/fluid/operators/fused/fused_elemwise_activation_op.cc
index d88ef15949da3809bffe41e4bf303d1fee568675..3771aac0dfd98a52dcd8b789e5a6114e977e22f8 100644
--- a/paddle/fluid/operators/fused_elemwise_activation_op.cc
+++ b/paddle/fluid/operators/fused/fused_elemwise_activation_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/fused_elemwise_activation_op.h"
+#include "paddle/fluid/operators/fused/fused_elemwise_activation_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/fused_elemwise_activation_op.cu b/paddle/fluid/operators/fused/fused_elemwise_activation_op.cu
similarity index 94%
rename from paddle/fluid/operators/fused_elemwise_activation_op.cu
rename to paddle/fluid/operators/fused/fused_elemwise_activation_op.cu
index e1d2b16b4b5e3a480777f834c2cbeb6d00a755e4..e10693bae1859307c9cf266965d4ce20e6de1bf9 100644
--- a/paddle/fluid/operators/fused_elemwise_activation_op.cu
+++ b/paddle/fluid/operators/fused/fused_elemwise_activation_op.cu
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/fused_elemwise_activation_op.h"
+#include "paddle/fluid/operators/fused/fused_elemwise_activation_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/fused_elemwise_activation_op.h b/paddle/fluid/operators/fused/fused_elemwise_activation_op.h
similarity index 99%
rename from paddle/fluid/operators/fused_elemwise_activation_op.h
rename to paddle/fluid/operators/fused/fused_elemwise_activation_op.h
index 5ae9aea959c268985c17643f2f47199c852c2bcb..01dc2dbfd61cc88f72174233382aa49f61c9b60f 100644
--- a/paddle/fluid/operators/fused_elemwise_activation_op.h
+++ b/paddle/fluid/operators/fused/fused_elemwise_activation_op.h
@@ -19,7 +19,7 @@ limitations under the License. */
 #include "paddle/fluid/framework/op_desc.h"
 #include "paddle/fluid/framework/op_registry.h"
 #include "paddle/fluid/operators/detail/safe_ref.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 #include "paddle/fluid/operators/math/compound_functors.h"
 #include "paddle/fluid/operators/math/functors.h"
 
diff --git a/paddle/fluid/operators/fused_embedding_fc_lstm_op.cc b/paddle/fluid/operators/fused/fused_embedding_fc_lstm_op.cc
similarity index 99%
rename from paddle/fluid/operators/fused_embedding_fc_lstm_op.cc
rename to paddle/fluid/operators/fused/fused_embedding_fc_lstm_op.cc
index fdc9cb4888b3468b85abfa0c693ed8ac5b0d450b..6d463538d232e1a38f845e7abc3786568ca3bb21 100644
--- a/paddle/fluid/operators/fused_embedding_fc_lstm_op.cc
+++ b/paddle/fluid/operators/fused/fused_embedding_fc_lstm_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/fused_embedding_fc_lstm_op.h"
+#include "paddle/fluid/operators/fused/fused_embedding_fc_lstm_op.h"
 #include <string>
 #include "paddle/fluid/operators/math/blas.h"
 #include "paddle/fluid/operators/math/cpu_vec.h"
diff --git a/paddle/fluid/operators/fused_embedding_fc_lstm_op.h b/paddle/fluid/operators/fused/fused_embedding_fc_lstm_op.h
similarity index 100%
rename from paddle/fluid/operators/fused_embedding_fc_lstm_op.h
rename to paddle/fluid/operators/fused/fused_embedding_fc_lstm_op.h
diff --git a/paddle/fluid/operators/fusion_gru_op.cc b/paddle/fluid/operators/fused/fusion_gru_op.cc
similarity index 99%
rename from paddle/fluid/operators/fusion_gru_op.cc
rename to paddle/fluid/operators/fused/fusion_gru_op.cc
index 120b2ab440156f6020fd6005dd64a48e9a6918ec..7e34d1019c9e6577b50ff8c2fa3d767124b5ff3b 100644
--- a/paddle/fluid/operators/fusion_gru_op.cc
+++ b/paddle/fluid/operators/fused/fusion_gru_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/fusion_gru_op.h"
+#include "paddle/fluid/operators/fused/fusion_gru_op.h"
 #include <cstring>  // for memcpy
 #include <string>
 #include "paddle/fluid/operators/math/blas.h"
diff --git a/paddle/fluid/operators/fusion_gru_op.h b/paddle/fluid/operators/fused/fusion_gru_op.h
similarity index 100%
rename from paddle/fluid/operators/fusion_gru_op.h
rename to paddle/fluid/operators/fused/fusion_gru_op.h
diff --git a/paddle/fluid/operators/fusion_lstm_op.cc b/paddle/fluid/operators/fused/fusion_lstm_op.cc
similarity index 99%
rename from paddle/fluid/operators/fusion_lstm_op.cc
rename to paddle/fluid/operators/fused/fusion_lstm_op.cc
index 067e6a3e7cccc1f15ebdd984f3a2441339a989ab..0959539068eef5b550a8e3997d3f11ea67ae0707 100644
--- a/paddle/fluid/operators/fusion_lstm_op.cc
+++ b/paddle/fluid/operators/fused/fusion_lstm_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/fusion_lstm_op.h"
+#include "paddle/fluid/operators/fused/fusion_lstm_op.h"
 #include <string>
 #include "paddle/fluid/operators/math/blas.h"
 #include "paddle/fluid/operators/math/fc_compute.h"
diff --git a/paddle/fluid/operators/fusion_lstm_op.h b/paddle/fluid/operators/fused/fusion_lstm_op.h
similarity index 100%
rename from paddle/fluid/operators/fusion_lstm_op.h
rename to paddle/fluid/operators/fused/fusion_lstm_op.h
diff --git a/paddle/fluid/operators/fusion_seqconv_eltadd_relu_op.cc b/paddle/fluid/operators/fused/fusion_seqconv_eltadd_relu_op.cc
similarity index 99%
rename from paddle/fluid/operators/fusion_seqconv_eltadd_relu_op.cc
rename to paddle/fluid/operators/fused/fusion_seqconv_eltadd_relu_op.cc
index b0910dc19edb246d9acfe3bdb15071c64cbdaba7..40bba09f3ef71021b7daff83b9d63005f7580395 100644
--- a/paddle/fluid/operators/fusion_seqconv_eltadd_relu_op.cc
+++ b/paddle/fluid/operators/fused/fusion_seqconv_eltadd_relu_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/fusion_seqconv_eltadd_relu_op.h"
+#include "paddle/fluid/operators/fused/fusion_seqconv_eltadd_relu_op.h"
 #include <algorithm>  // for min, max
 #include <string>
 #include "paddle/fluid/operators/math/blas.h"
diff --git a/paddle/fluid/operators/fusion_seqconv_eltadd_relu_op.h b/paddle/fluid/operators/fused/fusion_seqconv_eltadd_relu_op.h
similarity index 100%
rename from paddle/fluid/operators/fusion_seqconv_eltadd_relu_op.h
rename to paddle/fluid/operators/fused/fusion_seqconv_eltadd_relu_op.h
diff --git a/paddle/fluid/operators/fusion_seqexpand_concat_fc_op.cc b/paddle/fluid/operators/fused/fusion_seqexpand_concat_fc_op.cc
similarity index 99%
rename from paddle/fluid/operators/fusion_seqexpand_concat_fc_op.cc
rename to paddle/fluid/operators/fused/fusion_seqexpand_concat_fc_op.cc
index 8d2f055d53a0c5bbef624ff3b01b01724d0b3a21..288b56fc2485138b20c5b53af3e950f1c1886ba5 100644
--- a/paddle/fluid/operators/fusion_seqexpand_concat_fc_op.cc
+++ b/paddle/fluid/operators/fused/fusion_seqexpand_concat_fc_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/fusion_seqexpand_concat_fc_op.h"
+#include "paddle/fluid/operators/fused/fusion_seqexpand_concat_fc_op.h"
 #include <string>
 #include "paddle/fluid/operators/math/blas.h"
 #include "paddle/fluid/operators/math/cpu_vec.h"
diff --git a/paddle/fluid/operators/fusion_seqexpand_concat_fc_op.h b/paddle/fluid/operators/fused/fusion_seqexpand_concat_fc_op.h
similarity index 100%
rename from paddle/fluid/operators/fusion_seqexpand_concat_fc_op.h
rename to paddle/fluid/operators/fused/fusion_seqexpand_concat_fc_op.h
diff --git a/paddle/fluid/operators/layer_norm_op.h b/paddle/fluid/operators/layer_norm_op.h
index 2e54bb497dec11eaeda03a1aa6acfd4cc261dbfe..7bf79b08956885259e5ac3801274a1a675e6d975 100644
--- a/paddle/fluid/operators/layer_norm_op.h
+++ b/paddle/fluid/operators/layer_norm_op.h
@@ -15,7 +15,7 @@ limitations under the License. */
 #pragma once
 #include "paddle/fluid/framework/eigen.h"
 #include "paddle/fluid/framework/op_registry.h"
-#include "paddle/fluid/operators/elementwise_op_function.h"
+#include "paddle/fluid/operators/elementwise/elementwise_op_function.h"
 #include "paddle/fluid/operators/math/blas.h"
 #include "paddle/fluid/operators/math/math_function.h"
 
diff --git a/paddle/fluid/operators/metrics/CMakeLists.txt b/paddle/fluid/operators/metrics/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..5d468316e8eacb73c4a4ce81c784880bb5e46c2d
--- /dev/null
+++ b/paddle/fluid/operators/metrics/CMakeLists.txt
@@ -0,0 +1,2 @@
+include(operators)
+register_operators()
diff --git a/paddle/fluid/operators/accuracy_op.cc b/paddle/fluid/operators/metrics/accuracy_op.cc
similarity index 98%
rename from paddle/fluid/operators/accuracy_op.cc
rename to paddle/fluid/operators/metrics/accuracy_op.cc
index 42fcace17926641b5caf677eb3c8ba5222e37190..95aa76bc6947c9c39e56d39031c5184dc262acd0 100644
--- a/paddle/fluid/operators/accuracy_op.cc
+++ b/paddle/fluid/operators/metrics/accuracy_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/accuracy_op.h"
+#include "paddle/fluid/operators/metrics/accuracy_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/accuracy_op.cu b/paddle/fluid/operators/metrics/accuracy_op.cu
similarity index 98%
rename from paddle/fluid/operators/accuracy_op.cu
rename to paddle/fluid/operators/metrics/accuracy_op.cu
index 23b48c6fdf427348879de07c671c65327d6436d7..b255d2a7c413b4f965f6b874d342dcb93c7b5e66 100644
--- a/paddle/fluid/operators/accuracy_op.cu
+++ b/paddle/fluid/operators/metrics/accuracy_op.cu
@@ -14,7 +14,7 @@ limitations under the License. */
 
 #include <thrust/execution_policy.h>
 #include <thrust/reduce.h>
-#include "paddle/fluid/operators/accuracy_op.h"
+#include "paddle/fluid/operators/metrics/accuracy_op.h"
 #include "paddle/fluid/platform/cuda_primitives.h"
 #include "paddle/fluid/platform/gpu_info.h"
 
diff --git a/paddle/fluid/operators/accuracy_op.h b/paddle/fluid/operators/metrics/accuracy_op.h
similarity index 100%
rename from paddle/fluid/operators/accuracy_op.h
rename to paddle/fluid/operators/metrics/accuracy_op.h
diff --git a/paddle/fluid/operators/auc_op.cc b/paddle/fluid/operators/metrics/auc_op.cc
similarity index 98%
rename from paddle/fluid/operators/auc_op.cc
rename to paddle/fluid/operators/metrics/auc_op.cc
index cb98bc514083ad113fdebfbac043a9516fd9435a..335d4fded4a9543dabf984f7ed9c342b46dd04f0 100644
--- a/paddle/fluid/operators/auc_op.cc
+++ b/paddle/fluid/operators/metrics/auc_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/auc_op.h"
+#include "paddle/fluid/operators/metrics/auc_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/auc_op.h b/paddle/fluid/operators/metrics/auc_op.h
similarity index 100%
rename from paddle/fluid/operators/auc_op.h
rename to paddle/fluid/operators/metrics/auc_op.h
diff --git a/paddle/fluid/operators/precision_recall_op.cc b/paddle/fluid/operators/metrics/precision_recall_op.cc
similarity index 99%
rename from paddle/fluid/operators/precision_recall_op.cc
rename to paddle/fluid/operators/metrics/precision_recall_op.cc
index e7ce16f33fb5052ffb41fc05bd1538e2f0dc35be..0d733c47dd2fcaad776d8d4e6467ecd1872bce05 100644
--- a/paddle/fluid/operators/precision_recall_op.cc
+++ b/paddle/fluid/operators/metrics/precision_recall_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/precision_recall_op.h"
+#include "paddle/fluid/operators/metrics/precision_recall_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/precision_recall_op.h b/paddle/fluid/operators/metrics/precision_recall_op.h
similarity index 100%
rename from paddle/fluid/operators/precision_recall_op.h
rename to paddle/fluid/operators/metrics/precision_recall_op.h
diff --git a/paddle/fluid/operators/nccl/CMakeLists.txt b/paddle/fluid/operators/nccl/CMakeLists.txt
index cdcba8035762d8f442eb8b8ed52a4e3e99ac31b6..9b26e19cc7ed05038e05308f9277b200a885dc10 100644
--- a/paddle/fluid/operators/nccl/CMakeLists.txt
+++ b/paddle/fluid/operators/nccl/CMakeLists.txt
@@ -1,3 +1,13 @@
 if(WITH_GPU AND NOT WIN32)
   nv_library(nccl_common SRCS nccl_gpu_common.cc DEPS device_context operator )
 endif()
+
+if(WITH_GPU)
+    op_library(nccl_op DEPS nccl_common)
+    file(APPEND ${pybind_file} "USE_CUDA_ONLY_OP(ncclAllReduce);\n")
+    set(OPERATOR_DEPS ${OPERATOR_DEPS} nccl_common PARENT_SCOPE)
+endif()
+
+if(NOT WIN32)
+    nv_test(nccl_op_test SRCS nccl_op_test.cu.cc DEPS nccl_op gpu_info device_context)
+endif()
diff --git a/paddle/fluid/operators/nccl_op.cc b/paddle/fluid/operators/nccl/nccl_op.cc
similarity index 100%
rename from paddle/fluid/operators/nccl_op.cc
rename to paddle/fluid/operators/nccl/nccl_op.cc
diff --git a/paddle/fluid/operators/nccl_op.cu.cc b/paddle/fluid/operators/nccl/nccl_op.cu.cc
similarity index 100%
rename from paddle/fluid/operators/nccl_op.cu.cc
rename to paddle/fluid/operators/nccl/nccl_op.cu.cc
diff --git a/paddle/fluid/operators/nccl_op_test.cu.cc b/paddle/fluid/operators/nccl/nccl_op_test.cu.cc
similarity index 100%
rename from paddle/fluid/operators/nccl_op_test.cu.cc
rename to paddle/fluid/operators/nccl/nccl_op_test.cu.cc
diff --git a/paddle/fluid/operators/optimizers/CMakeLists.txt b/paddle/fluid/operators/optimizers/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..5d468316e8eacb73c4a4ce81c784880bb5e46c2d
--- /dev/null
+++ b/paddle/fluid/operators/optimizers/CMakeLists.txt
@@ -0,0 +1,2 @@
+include(operators)
+register_operators()
diff --git a/paddle/fluid/operators/adadelta_op.cc b/paddle/fluid/operators/optimizers/adadelta_op.cc
similarity index 98%
rename from paddle/fluid/operators/adadelta_op.cc
rename to paddle/fluid/operators/optimizers/adadelta_op.cc
index 89a7a49e0fa8427826f5d91274912a68f2316b61..9039d02b673b3403c840492c088179b30e23da9c 100644
--- a/paddle/fluid/operators/adadelta_op.cc
+++ b/paddle/fluid/operators/optimizers/adadelta_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/adadelta_op.h"
+#include "paddle/fluid/operators/optimizers/adadelta_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/adadelta_op.cu b/paddle/fluid/operators/optimizers/adadelta_op.cu
similarity index 93%
rename from paddle/fluid/operators/adadelta_op.cu
rename to paddle/fluid/operators/optimizers/adadelta_op.cu
index fc10c6657476e7f87b2f703a1d0cb88eeebc35cf..3fbfee5df05770a1206ab3170d3baffdd20bc77b 100644
--- a/paddle/fluid/operators/adadelta_op.cu
+++ b/paddle/fluid/operators/optimizers/adadelta_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/adadelta_op.h"
+#include "paddle/fluid/operators/optimizers/adadelta_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/adadelta_op.h b/paddle/fluid/operators/optimizers/adadelta_op.h
similarity index 100%
rename from paddle/fluid/operators/adadelta_op.h
rename to paddle/fluid/operators/optimizers/adadelta_op.h
diff --git a/paddle/fluid/operators/adagrad_op.cc b/paddle/fluid/operators/optimizers/adagrad_op.cc
similarity index 99%
rename from paddle/fluid/operators/adagrad_op.cc
rename to paddle/fluid/operators/optimizers/adagrad_op.cc
index c88297ff544ddb0e5a97452a8ad2e8f9f77825ba..e8d5a9e2c875570a198629bd745c9d58036746cb 100644
--- a/paddle/fluid/operators/adagrad_op.cc
+++ b/paddle/fluid/operators/optimizers/adagrad_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/adagrad_op.h"
+#include "paddle/fluid/operators/optimizers/adagrad_op.h"
 #include <vector>
 
 #include <cmath>
diff --git a/paddle/fluid/operators/adagrad_op.cu b/paddle/fluid/operators/optimizers/adagrad_op.cu
similarity index 98%
rename from paddle/fluid/operators/adagrad_op.cu
rename to paddle/fluid/operators/optimizers/adagrad_op.cu
index b99b33343d36fbb7f6b1a2928e142ca615b238b3..4efe56855a4bdca41d24f02c29a618a8d4232887 100644
--- a/paddle/fluid/operators/adagrad_op.cu
+++ b/paddle/fluid/operators/optimizers/adagrad_op.cu
@@ -13,9 +13,9 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/adagrad_op.h"
 #include "paddle/fluid/operators/math/math_function.h"
 #include "paddle/fluid/operators/math/selected_rows_functor.h"
+#include "paddle/fluid/operators/optimizers/adagrad_op.h"
 #include "paddle/fluid/platform/cuda_primitives.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/adagrad_op.h b/paddle/fluid/operators/optimizers/adagrad_op.h
similarity index 100%
rename from paddle/fluid/operators/adagrad_op.h
rename to paddle/fluid/operators/optimizers/adagrad_op.h
diff --git a/paddle/fluid/operators/adam_op.cc b/paddle/fluid/operators/optimizers/adam_op.cc
similarity index 99%
rename from paddle/fluid/operators/adam_op.cc
rename to paddle/fluid/operators/optimizers/adam_op.cc
index f3717af630017eba18aa265f3dbb496e18280a57..5710cda39acce53e35dfceec675fcd4979a84e31 100644
--- a/paddle/fluid/operators/adam_op.cc
+++ b/paddle/fluid/operators/optimizers/adam_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/adam_op.h"
+#include "paddle/fluid/operators/optimizers/adam_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/adam_op.cu b/paddle/fluid/operators/optimizers/adam_op.cu
similarity index 93%
rename from paddle/fluid/operators/adam_op.cu
rename to paddle/fluid/operators/optimizers/adam_op.cu
index 77f1991002e6007e8b8dff4746739a90e836145d..e8090ebacfe85153aba9e275c9cd1c55fd7af15e 100644
--- a/paddle/fluid/operators/adam_op.cu
+++ b/paddle/fluid/operators/optimizers/adam_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/adam_op.h"
+#include "paddle/fluid/operators/optimizers/adam_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/adam_op.h b/paddle/fluid/operators/optimizers/adam_op.h
similarity index 100%
rename from paddle/fluid/operators/adam_op.h
rename to paddle/fluid/operators/optimizers/adam_op.h
diff --git a/paddle/fluid/operators/adamax_op.cc b/paddle/fluid/operators/optimizers/adamax_op.cc
similarity index 99%
rename from paddle/fluid/operators/adamax_op.cc
rename to paddle/fluid/operators/optimizers/adamax_op.cc
index d4aa4d338a2379adf985ba7f89b528bc402eda06..4b244a76dc0ebee65b7c95db2d2754ebae03bbac 100644
--- a/paddle/fluid/operators/adamax_op.cc
+++ b/paddle/fluid/operators/optimizers/adamax_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/adamax_op.h"
+#include "paddle/fluid/operators/optimizers/adamax_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/adamax_op.cu b/paddle/fluid/operators/optimizers/adamax_op.cu
similarity index 93%
rename from paddle/fluid/operators/adamax_op.cu
rename to paddle/fluid/operators/optimizers/adamax_op.cu
index 05cafd7a8eef79588d1d5724084586cb9b51d3d4..e54adcb142fe0d50dad23fe5df14bd6f28220d8a 100644
--- a/paddle/fluid/operators/adamax_op.cu
+++ b/paddle/fluid/operators/optimizers/adamax_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/adamax_op.h"
+#include "paddle/fluid/operators/optimizers/adamax_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/adamax_op.h b/paddle/fluid/operators/optimizers/adamax_op.h
similarity index 100%
rename from paddle/fluid/operators/adamax_op.h
rename to paddle/fluid/operators/optimizers/adamax_op.h
diff --git a/paddle/fluid/operators/decayed_adagrad_op.cc b/paddle/fluid/operators/optimizers/decayed_adagrad_op.cc
similarity index 98%
rename from paddle/fluid/operators/decayed_adagrad_op.cc
rename to paddle/fluid/operators/optimizers/decayed_adagrad_op.cc
index d73ae9e2721b388212cb6efa354eb4b480df9cad..80278441c07203b03dbcff157193ea5976eefbf1 100644
--- a/paddle/fluid/operators/decayed_adagrad_op.cc
+++ b/paddle/fluid/operators/optimizers/decayed_adagrad_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/decayed_adagrad_op.h"
+#include "paddle/fluid/operators/optimizers/decayed_adagrad_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/decayed_adagrad_op.cu b/paddle/fluid/operators/optimizers/decayed_adagrad_op.cu
similarity index 92%
rename from paddle/fluid/operators/decayed_adagrad_op.cu
rename to paddle/fluid/operators/optimizers/decayed_adagrad_op.cu
index 7da16acf05eefc21cbe3dd0540dcbf69022431de..84d65e39329659f82099011f9ec60468d5db6328 100644
--- a/paddle/fluid/operators/decayed_adagrad_op.cu
+++ b/paddle/fluid/operators/optimizers/decayed_adagrad_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/decayed_adagrad_op.h"
+#include "paddle/fluid/operators/optimizers/decayed_adagrad_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/decayed_adagrad_op.h b/paddle/fluid/operators/optimizers/decayed_adagrad_op.h
similarity index 100%
rename from paddle/fluid/operators/decayed_adagrad_op.h
rename to paddle/fluid/operators/optimizers/decayed_adagrad_op.h
diff --git a/paddle/fluid/operators/ftrl_op.cc b/paddle/fluid/operators/optimizers/ftrl_op.cc
similarity index 99%
rename from paddle/fluid/operators/ftrl_op.cc
rename to paddle/fluid/operators/optimizers/ftrl_op.cc
index b77e12d6508eb07ae137b313ca91eac951afbcbe..1c9e91d9b610669def6d6d52e4753714745d1c0f 100644
--- a/paddle/fluid/operators/ftrl_op.cc
+++ b/paddle/fluid/operators/optimizers/ftrl_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/ftrl_op.h"
+#include "paddle/fluid/operators/optimizers/ftrl_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/ftrl_op.cu b/paddle/fluid/operators/optimizers/ftrl_op.cu
similarity index 93%
rename from paddle/fluid/operators/ftrl_op.cu
rename to paddle/fluid/operators/optimizers/ftrl_op.cu
index e7371c80da1d1cbb39247b50d8c6537ee8e948f8..f836b75df93861a0fd670f2a0e786e6a797a4661 100644
--- a/paddle/fluid/operators/ftrl_op.cu
+++ b/paddle/fluid/operators/optimizers/ftrl_op.cu
@@ -12,7 +12,7 @@ CONDITIONS OF ANY KIND, either express or implied. See the License for the
 specific language governing permissions and limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/ftrl_op.h"
+#include "paddle/fluid/operators/optimizers/ftrl_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/ftrl_op.h b/paddle/fluid/operators/optimizers/ftrl_op.h
similarity index 100%
rename from paddle/fluid/operators/ftrl_op.h
rename to paddle/fluid/operators/optimizers/ftrl_op.h
diff --git a/paddle/fluid/operators/lars_momentum_op.cc b/paddle/fluid/operators/optimizers/lars_momentum_op.cc
similarity index 96%
rename from paddle/fluid/operators/lars_momentum_op.cc
rename to paddle/fluid/operators/optimizers/lars_momentum_op.cc
index a8dda93902448fa1bd21b719ffd9c9b500caf755..574a03680b66962ac2d6ba249d0fc491a36794cd 100644
--- a/paddle/fluid/operators/lars_momentum_op.cc
+++ b/paddle/fluid/operators/optimizers/lars_momentum_op.cc
@@ -12,8 +12,8 @@ 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 "paddle/fluid/operators/lars_momentum_op.h"
-#include "paddle/fluid/operators/momentum_op.h"
+#include "paddle/fluid/operators/optimizers/lars_momentum_op.h"
+#include "paddle/fluid/operators/optimizers/momentum_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/lars_momentum_op.cu b/paddle/fluid/operators/optimizers/lars_momentum_op.cu
similarity index 98%
rename from paddle/fluid/operators/lars_momentum_op.cu
rename to paddle/fluid/operators/optimizers/lars_momentum_op.cu
index eb346851a2f690fa05422c84ddcb08307539048f..a277d6ff2bea917addac8c6ea4b24b63dcbc8dba 100644
--- a/paddle/fluid/operators/lars_momentum_op.cu
+++ b/paddle/fluid/operators/optimizers/lars_momentum_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #include "paddle/fluid/framework/op_registry.h"
-#include "paddle/fluid/operators/lars_momentum_op.h"
+#include "paddle/fluid/operators/optimizers/lars_momentum_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/lars_momentum_op.h b/paddle/fluid/operators/optimizers/lars_momentum_op.h
similarity index 100%
rename from paddle/fluid/operators/lars_momentum_op.h
rename to paddle/fluid/operators/optimizers/lars_momentum_op.h
diff --git a/paddle/fluid/operators/momentum_op.cc b/paddle/fluid/operators/optimizers/momentum_op.cc
similarity index 98%
rename from paddle/fluid/operators/momentum_op.cc
rename to paddle/fluid/operators/optimizers/momentum_op.cc
index 7f0b51580aa2591ac7338ad7c29ee4756d909925..cde238c076b6991eb52dac328c3e30a045420c92 100644
--- a/paddle/fluid/operators/momentum_op.cc
+++ b/paddle/fluid/operators/optimizers/momentum_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/momentum_op.h"
+#include "paddle/fluid/operators/optimizers/momentum_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/momentum_op.cu b/paddle/fluid/operators/optimizers/momentum_op.cu
similarity index 93%
rename from paddle/fluid/operators/momentum_op.cu
rename to paddle/fluid/operators/optimizers/momentum_op.cu
index b68fec34d43f0dee834f1045f192d5c6089d9356..8ce739de8dfd74cb43f9521bf39e3127a8a21925 100644
--- a/paddle/fluid/operators/momentum_op.cu
+++ b/paddle/fluid/operators/optimizers/momentum_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #include "paddle/fluid/framework/op_registry.h"
-#include "paddle/fluid/operators/momentum_op.h"
+#include "paddle/fluid/operators/optimizers/momentum_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/momentum_op.h b/paddle/fluid/operators/optimizers/momentum_op.h
similarity index 100%
rename from paddle/fluid/operators/momentum_op.h
rename to paddle/fluid/operators/optimizers/momentum_op.h
diff --git a/paddle/fluid/operators/proximal_adagrad_op.cc b/paddle/fluid/operators/optimizers/proximal_adagrad_op.cc
similarity index 98%
rename from paddle/fluid/operators/proximal_adagrad_op.cc
rename to paddle/fluid/operators/optimizers/proximal_adagrad_op.cc
index 8d8075d76111928ec9855eb0b70fe6dbd90a979b..7b07b3b7071cb39e4e81cb4612372eec96efe489 100644
--- a/paddle/fluid/operators/proximal_adagrad_op.cc
+++ b/paddle/fluid/operators/optimizers/proximal_adagrad_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/proximal_adagrad_op.h"
+#include "paddle/fluid/operators/optimizers/proximal_adagrad_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/proximal_adagrad_op.cu b/paddle/fluid/operators/optimizers/proximal_adagrad_op.cu
similarity index 92%
rename from paddle/fluid/operators/proximal_adagrad_op.cu
rename to paddle/fluid/operators/optimizers/proximal_adagrad_op.cu
index 7e0226c62bfd5d4804cc70c00391237deec33ebb..d1c1f747b70c3ceb806da06e6786a70b62a32995 100644
--- a/paddle/fluid/operators/proximal_adagrad_op.cu
+++ b/paddle/fluid/operators/optimizers/proximal_adagrad_op.cu
@@ -12,7 +12,7 @@ CONDITIONS OF ANY KIND, either express or implied. See the License for the
 specific language governing permissions and limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/proximal_adagrad_op.h"
+#include "paddle/fluid/operators/optimizers/proximal_adagrad_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/proximal_adagrad_op.h b/paddle/fluid/operators/optimizers/proximal_adagrad_op.h
similarity index 100%
rename from paddle/fluid/operators/proximal_adagrad_op.h
rename to paddle/fluid/operators/optimizers/proximal_adagrad_op.h
diff --git a/paddle/fluid/operators/proximal_gd_op.cc b/paddle/fluid/operators/optimizers/proximal_gd_op.cc
similarity index 98%
rename from paddle/fluid/operators/proximal_gd_op.cc
rename to paddle/fluid/operators/optimizers/proximal_gd_op.cc
index baf9cbcba2ed89f62afc9816e0ab9e0f112e6008..dcef4f7be249e04306732213a7c6209d32602048 100644
--- a/paddle/fluid/operators/proximal_gd_op.cc
+++ b/paddle/fluid/operators/optimizers/proximal_gd_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/proximal_gd_op.h"
+#include "paddle/fluid/operators/optimizers/proximal_gd_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/proximal_gd_op.cu b/paddle/fluid/operators/optimizers/proximal_gd_op.cu
similarity index 92%
rename from paddle/fluid/operators/proximal_gd_op.cu
rename to paddle/fluid/operators/optimizers/proximal_gd_op.cu
index 32ee9ab74cd58fd6f48b6c34e108f31315adaf71..7aa0e1015008eba0c1cf63ba1278dc2b8049b20b 100644
--- a/paddle/fluid/operators/proximal_gd_op.cu
+++ b/paddle/fluid/operators/optimizers/proximal_gd_op.cu
@@ -12,7 +12,7 @@ CONDITIONS OF ANY KIND, either express or implied. See the License for the
 specific language governing permissions and limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/proximal_gd_op.h"
+#include "paddle/fluid/operators/optimizers/proximal_gd_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/proximal_gd_op.h b/paddle/fluid/operators/optimizers/proximal_gd_op.h
similarity index 100%
rename from paddle/fluid/operators/proximal_gd_op.h
rename to paddle/fluid/operators/optimizers/proximal_gd_op.h
diff --git a/paddle/fluid/operators/rmsprop_op.cc b/paddle/fluid/operators/optimizers/rmsprop_op.cc
similarity index 99%
rename from paddle/fluid/operators/rmsprop_op.cc
rename to paddle/fluid/operators/optimizers/rmsprop_op.cc
index f06f87e61d3a4d1fc8b864b9dd84e697fb12a006..99d1156ee6d5fc88161e25bfa581a265707e6f92 100644
--- a/paddle/fluid/operators/rmsprop_op.cc
+++ b/paddle/fluid/operators/optimizers/rmsprop_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/rmsprop_op.h"
+#include "paddle/fluid/operators/optimizers/rmsprop_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/rmsprop_op.cu b/paddle/fluid/operators/optimizers/rmsprop_op.cu
similarity index 92%
rename from paddle/fluid/operators/rmsprop_op.cu
rename to paddle/fluid/operators/optimizers/rmsprop_op.cu
index cdc473769598be5aac87a14613d9acdd5c1a1204..69e35a309e04f61068d9ff1b6d9f1450d2524253 100644
--- a/paddle/fluid/operators/rmsprop_op.cu
+++ b/paddle/fluid/operators/optimizers/rmsprop_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #define EIGEN_USE_GPU
-#include "paddle/fluid/operators/rmsprop_op.h"
+#include "paddle/fluid/operators/optimizers/rmsprop_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/rmsprop_op.h b/paddle/fluid/operators/optimizers/rmsprop_op.h
similarity index 100%
rename from paddle/fluid/operators/rmsprop_op.h
rename to paddle/fluid/operators/optimizers/rmsprop_op.h
diff --git a/paddle/fluid/operators/sgd_op.cc b/paddle/fluid/operators/optimizers/sgd_op.cc
similarity index 98%
rename from paddle/fluid/operators/sgd_op.cc
rename to paddle/fluid/operators/optimizers/sgd_op.cc
index ea62acd08c5009556abf05c91726111870d1a462..690381a67f89d18fe81c3b856b7ddce25d496ed0 100644
--- a/paddle/fluid/operators/sgd_op.cc
+++ b/paddle/fluid/operators/optimizers/sgd_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sgd_op.h"
+#include "paddle/fluid/operators/optimizers/sgd_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sgd_op.cu b/paddle/fluid/operators/optimizers/sgd_op.cu
similarity index 98%
rename from paddle/fluid/operators/sgd_op.cu
rename to paddle/fluid/operators/optimizers/sgd_op.cu
index d3f4eba3b24ec1ac0328ef270256cdf3abe499db..a9d303d55d8f681fe3a014db36ede5ef6b2742bd 100644
--- a/paddle/fluid/operators/sgd_op.cu
+++ b/paddle/fluid/operators/optimizers/sgd_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #include <algorithm>
-#include "paddle/fluid/operators/sgd_op.h"
+#include "paddle/fluid/operators/optimizers/sgd_op.h"
 #include "paddle/fluid/platform/cuda_primitives.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sgd_op.h b/paddle/fluid/operators/optimizers/sgd_op.h
similarity index 100%
rename from paddle/fluid/operators/sgd_op.h
rename to paddle/fluid/operators/optimizers/sgd_op.h
diff --git a/paddle/fluid/operators/reader/CMakeLists.txt b/paddle/fluid/operators/reader/CMakeLists.txt
index 728197377df04df8c993a48bc282431473fe9959..6c919ee1782ebce6d56f7530daa9b748dfb26c47 100644
--- a/paddle/fluid/operators/reader/CMakeLists.txt
+++ b/paddle/fluid/operators/reader/CMakeLists.txt
@@ -1,3 +1,5 @@
+include(operators)
+
 cc_library(reader_op_registry SRCS reader_op_registry.cc DEPS operator op_registry reader)
 set(LOCAL_READER_LIBS)
 
@@ -28,4 +30,10 @@ reader_library(create_py_reader_op SRCS create_py_reader_op.cc)
 
 cc_test(reader_blocking_queue_test SRCS reader_blocking_queue_test.cc)
 # Export local libraries to parent
-set(READER_LIBRARY ${LOCAL_READER_LIBS} PARENT_SCOPE)
+# set(READER_LIBRARY ${LOCAL_READER_LIBS} PARENT_SCOPE)
+
+op_library(read_op)
+
+foreach(src ${LOCAL_READER_LIBS})
+    set(OP_LIBRARY ${src} ${OP_LIBRARY} CACHE INTERNAL "op libs")
+endforeach()
diff --git a/paddle/fluid/operators/read_op.cc b/paddle/fluid/operators/reader/read_op.cc
similarity index 100%
rename from paddle/fluid/operators/read_op.cc
rename to paddle/fluid/operators/reader/read_op.cc
diff --git a/paddle/fluid/operators/reduce_ops/CMakeLists.txt b/paddle/fluid/operators/reduce_ops/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..5fe4d15ae2c6254a50318813c852b6c314880aba
--- /dev/null
+++ b/paddle/fluid/operators/reduce_ops/CMakeLists.txt
@@ -0,0 +1,20 @@
+include(operators)
+register_operators()
+
+if(WITH_GPU)
+    file(GLOB OPS RELATIVE "${CMAKE_CURRENT_SOURCE_DIR}" "*.part.cu")
+    string(REPLACE ".part.cu" "" OPS "${OPS}")
+
+    foreach(src ${OPS})
+        if (EXISTS ${CMAKE_CURRENT_SOURCE_DIR}/${src}.part.cu)
+            set(CUDA_KERNEL_FILE ${CMAKE_CURRENT_SOURCE_DIR}/${src}.part.cu)
+            file(READ ${CUDA_KERNEL_FILE} TARGET_CONTENT)
+            string(REGEX MATCH "REGISTER_OP_CUDA_KERNEL\\(\\n?([^,]+),.*" MATCHED ${TARGET_CONTENT})
+            if (MATCHED)
+                string(STRIP ${CMAKE_MATCH_1} MATCHED)
+                file(APPEND ${pybind_file} "USE_OP_DEVICE_KERNEL(${MATCHED}, CUDA);\n")
+            endif()
+
+        endif()
+    endforeach()
+endif()
diff --git a/paddle/fluid/operators/cub_reduce.h b/paddle/fluid/operators/reduce_ops/cub_reduce.h
similarity index 100%
rename from paddle/fluid/operators/cub_reduce.h
rename to paddle/fluid/operators/reduce_ops/cub_reduce.h
diff --git a/paddle/fluid/operators/reduce_max_op.cc b/paddle/fluid/operators/reduce_ops/reduce_max_op.cc
similarity index 96%
rename from paddle/fluid/operators/reduce_max_op.cc
rename to paddle/fluid/operators/reduce_ops/reduce_max_op.cc
index 95d3768e1fdf6947659c7b3a1c9d57fad741472a..cb438b4a8057267015c8b3c15dd8468fca5a4b44 100644
--- a/paddle/fluid/operators/reduce_max_op.cc
+++ b/paddle/fluid/operators/reduce_ops/reduce_max_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_min_max_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_min_max_op.h"
 
 REGISTER_REDUCE_OP(reduce_max);
 REGISTER_OP_CPU_KERNEL(
diff --git a/paddle/fluid/operators/reduce_max_op.cu b/paddle/fluid/operators/reduce_ops/reduce_max_op.cu
similarity index 95%
rename from paddle/fluid/operators/reduce_max_op.cu
rename to paddle/fluid/operators/reduce_ops/reduce_max_op.cu
index b21da178f3eeaafa41bde5f64cc4abcf7944b032..832112ede833a06e053dcff5139e82f054b127c4 100644
--- a/paddle/fluid/operators/reduce_max_op.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_max_op.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_min_max_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_min_max_op.h"
 
 REGISTER_OP_CUDA_KERNEL(reduce_max,
                         ops::ReduceKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/reduce_max_op.part.cu b/paddle/fluid/operators/reduce_ops/reduce_max_op.part.cu
similarity index 94%
rename from paddle/fluid/operators/reduce_max_op.part.cu
rename to paddle/fluid/operators/reduce_ops/reduce_max_op.part.cu
index 6954c8d744faee6f8f0b715d6e4c8e3bcda7fb83..5ee38b8fa46290c86cd44ef1bcc71bd2fcd9bcd4 100644
--- a/paddle/fluid/operators/reduce_max_op.part.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_max_op.part.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_min_max_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_min_max_op.h"
 
 REGISTER_OP_CUDA_KERNEL(
     reduce_max_grad, ops::ReduceGradKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/reduce_mean_op.cc b/paddle/fluid/operators/reduce_ops/reduce_mean_op.cc
similarity index 96%
rename from paddle/fluid/operators/reduce_mean_op.cc
rename to paddle/fluid/operators/reduce_ops/reduce_mean_op.cc
index fc258c2496340b47d24dc89f16f7419dbb4b0d95..072bc34d3e23a48c8d856a51b0d5a6facc7ececf 100644
--- a/paddle/fluid/operators/reduce_mean_op.cc
+++ b/paddle/fluid/operators/reduce_ops/reduce_mean_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_mean_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_mean_op.h"
 
 REGISTER_REDUCE_OP(reduce_mean);
 REGISTER_OP_CPU_KERNEL(reduce_mean,
diff --git a/paddle/fluid/operators/reduce_mean_op.cu b/paddle/fluid/operators/reduce_ops/reduce_mean_op.cu
similarity index 94%
rename from paddle/fluid/operators/reduce_mean_op.cu
rename to paddle/fluid/operators/reduce_ops/reduce_mean_op.cu
index 4408200d2d052c2f68c2dd35619de6ed67f07f6e..4d3bce8fdd05e536baa5fecb4fc5a117e2031224 100644
--- a/paddle/fluid/operators/reduce_mean_op.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_mean_op.cu
@@ -13,8 +13,8 @@
 // limitations under the License.
 
 #include <vector>
-#include "paddle/fluid/operators/cub_reduce.h"
-#include "paddle/fluid/operators/reduce_mean_op.h"
+#include "paddle/fluid/operators/reduce_ops/cub_reduce.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_mean_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/reduce_mean_op.h b/paddle/fluid/operators/reduce_ops/reduce_mean_op.h
similarity index 95%
rename from paddle/fluid/operators/reduce_mean_op.h
rename to paddle/fluid/operators/reduce_ops/reduce_mean_op.h
index 1359679c4767d2032bf3e3a90849ad2a2ef3e829..240c43bc6d0af266e3500c14f894fe30abab728e 100644
--- a/paddle/fluid/operators/reduce_mean_op.h
+++ b/paddle/fluid/operators/reduce_ops/reduce_mean_op.h
@@ -14,7 +14,7 @@
 
 #pragma once
 
-#include "paddle/fluid/operators/reduce_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/reduce_mean_op.part.cu b/paddle/fluid/operators/reduce_ops/reduce_mean_op.part.cu
similarity index 95%
rename from paddle/fluid/operators/reduce_mean_op.part.cu
rename to paddle/fluid/operators/reduce_ops/reduce_mean_op.part.cu
index 4b663bcdca7c20f8802d962a362f429d8eafe9af..9324ec1e1db6f40e463b415e5d2bdc5cfe664ef4 100644
--- a/paddle/fluid/operators/reduce_mean_op.part.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_mean_op.part.cu
@@ -13,7 +13,7 @@
 // limitations under the License.
 
 // .part used to speed up nvcc compile
-#include "paddle/fluid/operators/reduce_mean_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_mean_op.h"
 
 REGISTER_OP_CUDA_KERNEL(
     reduce_mean_grad, ops::ReduceGradKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/reduce_min_max_op.h b/paddle/fluid/operators/reduce_ops/reduce_min_max_op.h
similarity index 96%
rename from paddle/fluid/operators/reduce_min_max_op.h
rename to paddle/fluid/operators/reduce_ops/reduce_min_max_op.h
index ec59f3e71c1c702655a3feed10935b2f5a29d8a8..2557e8dd488618dd4998845b0e6e3ba823b96986 100644
--- a/paddle/fluid/operators/reduce_min_max_op.h
+++ b/paddle/fluid/operators/reduce_ops/reduce_min_max_op.h
@@ -13,7 +13,7 @@
 // limitations under the License.
 #pragma once
 
-#include "paddle/fluid/operators/reduce_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/reduce_min_op.cc b/paddle/fluid/operators/reduce_ops/reduce_min_op.cc
similarity index 96%
rename from paddle/fluid/operators/reduce_min_op.cc
rename to paddle/fluid/operators/reduce_ops/reduce_min_op.cc
index 330a86d2e4237a10d8cf6fd40025540edf08d897..11aa78382e319331dc65ec22927f0d5762adfb43 100644
--- a/paddle/fluid/operators/reduce_min_op.cc
+++ b/paddle/fluid/operators/reduce_ops/reduce_min_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_min_max_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_min_max_op.h"
 
 REGISTER_REDUCE_OP(reduce_min);
 REGISTER_OP_CPU_KERNEL(
diff --git a/paddle/fluid/operators/reduce_min_op.cu b/paddle/fluid/operators/reduce_ops/reduce_min_op.cu
similarity index 95%
rename from paddle/fluid/operators/reduce_min_op.cu
rename to paddle/fluid/operators/reduce_ops/reduce_min_op.cu
index 5a04a12b79444dcea30d3c1140d9708a98b55fe3..7b2706866f594228cbceb084e99d83aa8f345dfd 100644
--- a/paddle/fluid/operators/reduce_min_op.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_min_op.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_min_max_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_min_max_op.h"
 
 REGISTER_OP_CUDA_KERNEL(reduce_min,
                         ops::ReduceKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/reduce_min_op.part.cu b/paddle/fluid/operators/reduce_ops/reduce_min_op.part.cu
similarity index 94%
rename from paddle/fluid/operators/reduce_min_op.part.cu
rename to paddle/fluid/operators/reduce_ops/reduce_min_op.part.cu
index 5b8f061b2d03eb76863401905ac87044fd5ea778..bf886063786a8c36884ed20fef41c99468156c01 100644
--- a/paddle/fluid/operators/reduce_min_op.part.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_min_op.part.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_min_max_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_min_max_op.h"
 
 REGISTER_OP_CUDA_KERNEL(
     reduce_min_grad, ops::ReduceGradKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/reduce_op.h b/paddle/fluid/operators/reduce_ops/reduce_op.h
similarity index 99%
rename from paddle/fluid/operators/reduce_op.h
rename to paddle/fluid/operators/reduce_ops/reduce_op.h
index 72b6cf1773d5bcc42e40e72111179d454d2bb4a9..540742c4cd8b0efc4c6cf095d7a8b3516f551d4c 100644
--- a/paddle/fluid/operators/reduce_op.h
+++ b/paddle/fluid/operators/reduce_ops/reduce_op.h
@@ -18,7 +18,7 @@ limitations under the License. */
 #include <string>
 #include <vector>
 
-#include "paddle/fluid/operators/reduce_op_function.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_op_function.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/reduce_op_function.h b/paddle/fluid/operators/reduce_ops/reduce_op_function.h
similarity index 100%
rename from paddle/fluid/operators/reduce_op_function.h
rename to paddle/fluid/operators/reduce_ops/reduce_op_function.h
diff --git a/paddle/fluid/operators/reduce_prod_op.cc b/paddle/fluid/operators/reduce_ops/reduce_prod_op.cc
similarity index 96%
rename from paddle/fluid/operators/reduce_prod_op.cc
rename to paddle/fluid/operators/reduce_ops/reduce_prod_op.cc
index 713728b99757a6f3bb128f665d5576ac64eef8ec..88935107df187da731e5b77bb6c24cd692d2994f 100644
--- a/paddle/fluid/operators/reduce_prod_op.cc
+++ b/paddle/fluid/operators/reduce_ops/reduce_prod_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_prod_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_prod_op.h"
 
 REGISTER_REDUCE_OP(reduce_prod);
 REGISTER_OP_CPU_KERNEL(reduce_prod,
diff --git a/paddle/fluid/operators/reduce_prod_op.cu b/paddle/fluid/operators/reduce_ops/reduce_prod_op.cu
similarity index 95%
rename from paddle/fluid/operators/reduce_prod_op.cu
rename to paddle/fluid/operators/reduce_ops/reduce_prod_op.cu
index d8692afb96e4d5d3206210060684dd12fb4d79a7..4434937f75397d8d5340a94abbd41efa7e7a8d4b 100644
--- a/paddle/fluid/operators/reduce_prod_op.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_prod_op.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_prod_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_prod_op.h"
 
 REGISTER_OP_CUDA_KERNEL(reduce_prod,
                         ops::ReduceKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/reduce_prod_op.h b/paddle/fluid/operators/reduce_ops/reduce_prod_op.h
similarity index 95%
rename from paddle/fluid/operators/reduce_prod_op.h
rename to paddle/fluid/operators/reduce_ops/reduce_prod_op.h
index 97748113e092719aceed9d806ca6242077111532..103e108e4bda1c33434ec0c5d6c58f24fa725f57 100644
--- a/paddle/fluid/operators/reduce_prod_op.h
+++ b/paddle/fluid/operators/reduce_ops/reduce_prod_op.h
@@ -14,7 +14,7 @@
 
 #pragma once
 
-#include "paddle/fluid/operators/reduce_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/reduce_prod_op.part.cu b/paddle/fluid/operators/reduce_ops/reduce_prod_op.part.cu
similarity index 94%
rename from paddle/fluid/operators/reduce_prod_op.part.cu
rename to paddle/fluid/operators/reduce_ops/reduce_prod_op.part.cu
index 486c578c64b9a2d80abc940a7c4266ef5fd23c7f..0610cdd94f89c0371988fac7955d07fc5498a69f 100644
--- a/paddle/fluid/operators/reduce_prod_op.part.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_prod_op.part.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_prod_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_prod_op.h"
 
 REGISTER_OP_CUDA_KERNEL(
     reduce_prod_grad, ops::ReduceGradKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/reduce_sum_op.cc b/paddle/fluid/operators/reduce_ops/reduce_sum_op.cc
similarity index 96%
rename from paddle/fluid/operators/reduce_sum_op.cc
rename to paddle/fluid/operators/reduce_ops/reduce_sum_op.cc
index f0e5f6580fbc9e70562cb2fdd7e0c5d8729bc9a7..c7742f45dd147ea87413aa17680d671bede5dd6c 100644
--- a/paddle/fluid/operators/reduce_sum_op.cc
+++ b/paddle/fluid/operators/reduce_ops/reduce_sum_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/reduce_sum_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_sum_op.h"
 
 REGISTER_REDUCE_OP(reduce_sum);
 REGISTER_OP_CPU_KERNEL(
diff --git a/paddle/fluid/operators/reduce_sum_op.cu b/paddle/fluid/operators/reduce_ops/reduce_sum_op.cu
similarity index 94%
rename from paddle/fluid/operators/reduce_sum_op.cu
rename to paddle/fluid/operators/reduce_ops/reduce_sum_op.cu
index 2b031e8df99768c9208146640bddbe51149b2614..9051740e83aabd783750e8f415da09921608e470 100644
--- a/paddle/fluid/operators/reduce_sum_op.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_sum_op.cu
@@ -12,8 +12,8 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/cub_reduce.h"
-#include "paddle/fluid/operators/reduce_sum_op.h"
+#include "paddle/fluid/operators/reduce_ops/cub_reduce.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_sum_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/reduce_sum_op.h b/paddle/fluid/operators/reduce_ops/reduce_sum_op.h
similarity index 98%
rename from paddle/fluid/operators/reduce_sum_op.h
rename to paddle/fluid/operators/reduce_ops/reduce_sum_op.h
index 3e8d1bbdba504669bc06e0637094e3bee840adf2..26f59c72b4b99ff92a63c2fc2f00a31df0f5df61 100644
--- a/paddle/fluid/operators/reduce_sum_op.h
+++ b/paddle/fluid/operators/reduce_ops/reduce_sum_op.h
@@ -16,7 +16,7 @@
 
 #include <vector>
 
-#include "paddle/fluid/operators/reduce_op.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/reduce_sum_op.part.cu b/paddle/fluid/operators/reduce_ops/reduce_sum_op.part.cu
similarity index 90%
rename from paddle/fluid/operators/reduce_sum_op.part.cu
rename to paddle/fluid/operators/reduce_ops/reduce_sum_op.part.cu
index 525633f62a95b2d0d677fcbebe551b75cb2a180d..eb3295731b047391a244bfb598c9d802bca1fc0c 100644
--- a/paddle/fluid/operators/reduce_sum_op.part.cu
+++ b/paddle/fluid/operators/reduce_ops/reduce_sum_op.part.cu
@@ -12,8 +12,8 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/cub_reduce.h"
-#include "paddle/fluid/operators/reduce_sum_op.h"
+#include "paddle/fluid/operators/reduce_ops/cub_reduce.h"
+#include "paddle/fluid/operators/reduce_ops/reduce_sum_op.h"
 
 REGISTER_OP_CUDA_KERNEL(
     reduce_sum_grad, ops::ReduceGradKernel<paddle::platform::CUDADeviceContext,
diff --git a/paddle/fluid/operators/sequence_ops/CMakeLists.txt b/paddle/fluid/operators/sequence_ops/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..5d468316e8eacb73c4a4ce81c784880bb5e46c2d
--- /dev/null
+++ b/paddle/fluid/operators/sequence_ops/CMakeLists.txt
@@ -0,0 +1,2 @@
+include(operators)
+register_operators()
diff --git a/paddle/fluid/operators/sequence_concat_op.cc b/paddle/fluid/operators/sequence_ops/sequence_concat_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_concat_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_concat_op.cc
index 3234b60861da3d0c6a8434eb11fd0488a95e171f..37f1b9dda50ba4b62d7cf75765125e0ad663d9d8 100644
--- a/paddle/fluid/operators/sequence_concat_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_concat_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_concat_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_concat_op.h"
 #include <vector>
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_concat_op.cu.cc b/paddle/fluid/operators/sequence_ops/sequence_concat_op.cu.cc
similarity index 94%
rename from paddle/fluid/operators/sequence_concat_op.cu.cc
rename to paddle/fluid/operators/sequence_ops/sequence_concat_op.cu.cc
index eb6535235df80a9267b22403ae1f35c6cefb7fe7..7b8043bc4538b486bb73e005769e1585e5c4817e 100644
--- a/paddle/fluid/operators/sequence_concat_op.cu.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_concat_op.cu.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_concat_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_concat_op.h"
 
 template <typename T>
 using Kernel =
diff --git a/paddle/fluid/operators/sequence_concat_op.h b/paddle/fluid/operators/sequence_ops/sequence_concat_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_concat_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_concat_op.h
diff --git a/paddle/fluid/operators/sequence_conv_op.cc b/paddle/fluid/operators/sequence_ops/sequence_conv_op.cc
similarity index 99%
rename from paddle/fluid/operators/sequence_conv_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_conv_op.cc
index 95a21a5d3ee6d8037431083edc25d1cddf05dedb..65cd9edbc7125f605d6fb437a2e056054eb9a6d7 100644
--- a/paddle/fluid/operators/sequence_conv_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_conv_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_conv_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_conv_op.h"
 
 #include <algorithm>
 
diff --git a/paddle/fluid/operators/sequence_conv_op.cu.cc b/paddle/fluid/operators/sequence_ops/sequence_conv_op.cu.cc
similarity index 93%
rename from paddle/fluid/operators/sequence_conv_op.cu.cc
rename to paddle/fluid/operators/sequence_ops/sequence_conv_op.cu.cc
index de482b7f10bafc4ac6f3838670e2da9a86374c26..600981b5e96c279329a67b608a8dd94dee7d88ef 100644
--- a/paddle/fluid/operators/sequence_conv_op.cu.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_conv_op.cu.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_conv_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_conv_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/sequence_conv_op.h b/paddle/fluid/operators/sequence_ops/sequence_conv_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_conv_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_conv_op.h
diff --git a/paddle/fluid/operators/sequence_enumerate_op.cc b/paddle/fluid/operators/sequence_ops/sequence_enumerate_op.cc
similarity index 97%
rename from paddle/fluid/operators/sequence_enumerate_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_enumerate_op.cc
index 58e48c228bb34814700fd0f7a3d62ef4b1a435dd..1eebadc2c980ddf1cbaaefef1568dd401d0c77ed 100644
--- a/paddle/fluid/operators/sequence_enumerate_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_enumerate_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_enumerate_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_enumerate_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sequence_enumerate_op.cu b/paddle/fluid/operators/sequence_ops/sequence_enumerate_op.cu
similarity index 97%
rename from paddle/fluid/operators/sequence_enumerate_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_enumerate_op.cu
index bdc9a615aa9a1ecd99c1f6995361f8c5ff0aa383..28821e7129c1601f1214b0b56696fbf526a2123f 100644
--- a/paddle/fluid/operators/sequence_enumerate_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_enumerate_op.cu
@@ -14,7 +14,7 @@
 
 #include <thrust/device_vector.h>
 #include <thrust/host_vector.h>
-#include "paddle/fluid/operators/sequence_enumerate_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_enumerate_op.h"
 #include "paddle/fluid/platform/cuda_primitives.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_enumerate_op.h b/paddle/fluid/operators/sequence_ops/sequence_enumerate_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_enumerate_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_enumerate_op.h
diff --git a/paddle/fluid/operators/sequence_erase_op.cc b/paddle/fluid/operators/sequence_ops/sequence_erase_op.cc
similarity index 97%
rename from paddle/fluid/operators/sequence_erase_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_erase_op.cc
index 816ba123a6cbf84ec9b321d5d7cfef7fab9749b1..ddda80ee0824e261b0d737f86e03866d5fdfd77a 100644
--- a/paddle/fluid/operators/sequence_erase_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_erase_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_erase_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_erase_op.h"
 #include <vector>
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_erase_op.cu b/paddle/fluid/operators/sequence_ops/sequence_erase_op.cu
similarity index 98%
rename from paddle/fluid/operators/sequence_erase_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_erase_op.cu
index 3a58e47f1132cd1ac85584b2470e8c6cddcfb28a..619c40dbd10ad6b538f2d4e3567966b222fc5e2d 100644
--- a/paddle/fluid/operators/sequence_erase_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_erase_op.cu
@@ -14,7 +14,7 @@ limitations under the License. */
 
 #include <thrust/device_vector.h>
 #include <thrust/host_vector.h>
-#include "paddle/fluid/operators/sequence_erase_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_erase_op.h"
 #include "paddle/fluid/platform/cuda_primitives.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_erase_op.h b/paddle/fluid/operators/sequence_ops/sequence_erase_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_erase_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_erase_op.h
diff --git a/paddle/fluid/operators/sequence_expand_as_op.cc b/paddle/fluid/operators/sequence_ops/sequence_expand_as_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_expand_as_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_expand_as_op.cc
index 33c1e1c973c80ba3943924331380d35b225ac800..3b79d0c71975bb740b4085ce80f7d95b65f600c1 100644
--- a/paddle/fluid/operators/sequence_expand_as_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_expand_as_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_expand_as_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_expand_as_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sequence_expand_as_op.cu b/paddle/fluid/operators/sequence_ops/sequence_expand_as_op.cu
similarity index 98%
rename from paddle/fluid/operators/sequence_expand_as_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_expand_as_op.cu
index 7357f5ae6e732f28307af65d1f1b6b3cbed1f640..998bf82ab1ddcd815491de95a3f7cf987036ee65 100644
--- a/paddle/fluid/operators/sequence_expand_as_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_expand_as_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #include <algorithm>
-#include "paddle/fluid/operators/sequence_expand_as_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_expand_as_op.h"
 #include "paddle/fluid/platform/cuda_primitives.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_expand_as_op.h b/paddle/fluid/operators/sequence_ops/sequence_expand_as_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_expand_as_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_expand_as_op.h
diff --git a/paddle/fluid/operators/sequence_expand_op.cc b/paddle/fluid/operators/sequence_ops/sequence_expand_op.cc
similarity index 99%
rename from paddle/fluid/operators/sequence_expand_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_expand_op.cc
index 944c7f85e5f43679e1875fcce813382be2ba5526..c07e6962e673ceb274ef31cbf492f378ae696137 100644
--- a/paddle/fluid/operators/sequence_expand_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_expand_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_expand_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_expand_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sequence_expand_op.cu b/paddle/fluid/operators/sequence_ops/sequence_expand_op.cu
similarity index 98%
rename from paddle/fluid/operators/sequence_expand_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_expand_op.cu
index 550677b22694085059e914678a5361d914b455bc..afc08c7b3f6596efd3b6e0b74c17aa3c9268c47d 100644
--- a/paddle/fluid/operators/sequence_expand_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_expand_op.cu
@@ -13,7 +13,7 @@ See the License for the specific language governing permissions and
 limitations under the License. */
 
 #include <algorithm>
-#include "paddle/fluid/operators/sequence_expand_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_expand_op.h"
 #include "paddle/fluid/platform/cuda_primitives.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_expand_op.h b/paddle/fluid/operators/sequence_ops/sequence_expand_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_expand_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_expand_op.h
diff --git a/paddle/fluid/operators/sequence_mask_op.cc b/paddle/fluid/operators/sequence_ops/sequence_mask_op.cc
similarity index 95%
rename from paddle/fluid/operators/sequence_mask_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_mask_op.cc
index 798211f481659eb71248f7a6210e6522273d387f..7fc506aab4d3c6861282b68b09fdcb5fd8055f77 100644
--- a/paddle/fluid/operators/sequence_mask_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_mask_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_mask_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_mask_op.h"
 
 REGISTER_OPERATOR(sequence_mask, paddle::operators::SequenceMaskOp,
                   paddle::operators::SequenceMaskOpMaker,
diff --git a/paddle/fluid/operators/sequence_mask_op.cu b/paddle/fluid/operators/sequence_ops/sequence_mask_op.cu
similarity index 94%
rename from paddle/fluid/operators/sequence_mask_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_mask_op.cu
index 2ad23774579533b62b9189c1564ad7c7db5c298a..e963ce610e2c147d66087a1df59f67a04d899ccc 100644
--- a/paddle/fluid/operators/sequence_mask_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_mask_op.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_mask_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_mask_op.h"
 
 REGISTER_OP_CUDA_KERNEL(
     sequence_mask,
diff --git a/paddle/fluid/operators/sequence_mask_op.h b/paddle/fluid/operators/sequence_ops/sequence_mask_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_mask_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_mask_op.h
diff --git a/paddle/fluid/operators/sequence_pad_op.cc b/paddle/fluid/operators/sequence_ops/sequence_pad_op.cc
similarity index 99%
rename from paddle/fluid/operators/sequence_pad_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_pad_op.cc
index 4583b26256ba2e084bf7477c54d468df860d9b43..23c7bf7cea830bb0ccf5e81f99130043c2d5f80b 100644
--- a/paddle/fluid/operators/sequence_pad_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_pad_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_pad_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_pad_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sequence_pad_op.cu b/paddle/fluid/operators/sequence_ops/sequence_pad_op.cu
similarity index 95%
rename from paddle/fluid/operators/sequence_pad_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_pad_op.cu
index ff8f81a2f0ec4a72befc3be2a5fc48c3a586c824..7fc64a530ef5442ae927faac96ad92a4126febcd 100644
--- a/paddle/fluid/operators/sequence_pad_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_pad_op.cu
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_pad_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_pad_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/sequence_pad_op.h b/paddle/fluid/operators/sequence_ops/sequence_pad_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_pad_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_pad_op.h
diff --git a/paddle/fluid/operators/sequence_pool_op.cc b/paddle/fluid/operators/sequence_ops/sequence_pool_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_pool_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_pool_op.cc
index 7e80b8db5e90730e2df420466a33362620e15730..44b09bf7c2c776cdc455a8706cb2b2251f3be509 100644
--- a/paddle/fluid/operators/sequence_pool_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_pool_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_pool_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_pool_op.h"
 #include <string>
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_pool_op.cu b/paddle/fluid/operators/sequence_ops/sequence_pool_op.cu
similarity index 93%
rename from paddle/fluid/operators/sequence_pool_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_pool_op.cu
index 2bf0697af3c74ee922a832fecaa2cd2399a06849..63cd47a38a0ff6413c430c6be6284c5f4bfc2595 100644
--- a/paddle/fluid/operators/sequence_pool_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_pool_op.cu
@@ -14,7 +14,7 @@ limitations under the License. */
 
 #define EIGEN_USE_GPU
 
-#include "paddle/fluid/operators/sequence_pool_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_pool_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/sequence_pool_op.h b/paddle/fluid/operators/sequence_ops/sequence_pool_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_pool_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_pool_op.h
diff --git a/paddle/fluid/operators/sequence_reshape_op.cc b/paddle/fluid/operators/sequence_ops/sequence_reshape_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_reshape_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_reshape_op.cc
index 31d28d723498892f287246ba228df757d5b9f6c8..5421f35662b3b0a6a61748ac0b6b5f718d213b73 100644
--- a/paddle/fluid/operators/sequence_reshape_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_reshape_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_reshape_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_reshape_op.h"
 #include "paddle/fluid/framework/ddim.h"
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_reshape_op.cu b/paddle/fluid/operators/sequence_ops/sequence_reshape_op.cu
similarity index 95%
rename from paddle/fluid/operators/sequence_reshape_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_reshape_op.cu
index 232e031c0b022497d9e5141750dbf8fccffc7615..38bc599165d5f84f67e2fe08bf96ebef4b03d8a4 100644
--- a/paddle/fluid/operators/sequence_reshape_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_reshape_op.cu
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_reshape_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_reshape_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/sequence_reshape_op.h b/paddle/fluid/operators/sequence_ops/sequence_reshape_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_reshape_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_reshape_op.h
diff --git a/paddle/fluid/operators/sequence_reverse_op.cc b/paddle/fluid/operators/sequence_ops/sequence_reverse_op.cc
similarity index 94%
rename from paddle/fluid/operators/sequence_reverse_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_reverse_op.cc
index 1428cca1a6bf6150594f9cb72dbf00cd0eff7df5..dfbbf5f156983189ac1ab82fbff51d7eb4844f9a 100644
--- a/paddle/fluid/operators/sequence_reverse_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_reverse_op.cc
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_reverse_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_reverse_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/sequence_reverse_op.cu b/paddle/fluid/operators/sequence_ops/sequence_reverse_op.cu
similarity index 94%
rename from paddle/fluid/operators/sequence_reverse_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_reverse_op.cu
index ce65f4799e8661adca60d212eaa9c3f0f92c4c29..0a59ed7f9fee07bc3b12909973535f31ef049a4a 100644
--- a/paddle/fluid/operators/sequence_reverse_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_reverse_op.cu
@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "paddle/fluid/operators/sequence_reverse_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_reverse_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/sequence_reverse_op.h b/paddle/fluid/operators/sequence_ops/sequence_reverse_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_reverse_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_reverse_op.h
diff --git a/paddle/fluid/operators/sequence_scatter_op.cc b/paddle/fluid/operators/sequence_ops/sequence_scatter_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_scatter_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_scatter_op.cc
index adb81bffccb50069b3a2e5f391f3fdfde231b2be..c49d1ccb18427a1ec3c45f326b57bce32c60e1e2 100644
--- a/paddle/fluid/operators/sequence_scatter_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_scatter_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_scatter_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_scatter_op.h"
 #include "paddle/fluid/framework/eigen.h"
 #include "paddle/fluid/framework/op_registry.h"
 #include "paddle/fluid/operators/gather.h"
diff --git a/paddle/fluid/operators/sequence_scatter_op.h b/paddle/fluid/operators/sequence_ops/sequence_scatter_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_scatter_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_scatter_op.h
diff --git a/paddle/fluid/operators/sequence_slice_op.cc b/paddle/fluid/operators/sequence_ops/sequence_slice_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_slice_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_slice_op.cc
index df9243dc04c584d70dfa6ca78d5fac8423796466..6f84023e26dbf1280d9622946ab20184fb835be1 100644
--- a/paddle/fluid/operators/sequence_slice_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_slice_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_slice_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_slice_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sequence_slice_op.cu b/paddle/fluid/operators/sequence_ops/sequence_slice_op.cu
similarity index 92%
rename from paddle/fluid/operators/sequence_slice_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_slice_op.cu
index 059e802df0ebdba68f758decfb8b54a362996335..1e4a1b8323dbaacdf3f74c33e7aa4484d9be2478 100644
--- a/paddle/fluid/operators/sequence_slice_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_slice_op.cu
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_slice_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_slice_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/sequence_slice_op.h b/paddle/fluid/operators/sequence_ops/sequence_slice_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_slice_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_slice_op.h
diff --git a/paddle/fluid/operators/sequence_softmax_cudnn_op.cu.cc b/paddle/fluid/operators/sequence_ops/sequence_softmax_cudnn_op.cu.cc
similarity index 100%
rename from paddle/fluid/operators/sequence_softmax_cudnn_op.cu.cc
rename to paddle/fluid/operators/sequence_ops/sequence_softmax_cudnn_op.cu.cc
diff --git a/paddle/fluid/operators/sequence_softmax_op.cc b/paddle/fluid/operators/sequence_ops/sequence_softmax_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_softmax_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_softmax_op.cc
index ada3e0c8dbba38729c2b9c8b02335327835f2ef4..644a5bebc18886a2ac9210576f1c2251ad5ad0be 100644
--- a/paddle/fluid/operators/sequence_softmax_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_softmax_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_softmax_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_softmax_op.h"
 #include <string>
 
 namespace paddle {
diff --git a/paddle/fluid/operators/sequence_softmax_op.cu b/paddle/fluid/operators/sequence_ops/sequence_softmax_op.cu
similarity index 98%
rename from paddle/fluid/operators/sequence_softmax_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_softmax_op.cu
index e94ceaa170131e8bce7d1574b27f0baeaa8d1ffc..cc5e9821903fb7a726f52177df1d17757f697411 100644
--- a/paddle/fluid/operators/sequence_softmax_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_softmax_op.cu
@@ -14,7 +14,7 @@ limitations under the License. */
 
 #include <algorithm>
 #include <cub/cub.cuh>  // NOLINT
-#include "paddle/fluid/operators/sequence_softmax_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_softmax_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sequence_softmax_op.h b/paddle/fluid/operators/sequence_ops/sequence_softmax_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_softmax_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_softmax_op.h
diff --git a/paddle/fluid/operators/sequence_unpad_op.cc b/paddle/fluid/operators/sequence_ops/sequence_unpad_op.cc
similarity index 98%
rename from paddle/fluid/operators/sequence_unpad_op.cc
rename to paddle/fluid/operators/sequence_ops/sequence_unpad_op.cc
index e633e378a226ece8adea2e150cc6c1e9aa874331..2cf508e0b707ecc986886e72e5d42fde3c84894d 100644
--- a/paddle/fluid/operators/sequence_unpad_op.cc
+++ b/paddle/fluid/operators/sequence_ops/sequence_unpad_op.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_unpad_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_unpad_op.h"
 
 namespace paddle {
 namespace operators {
diff --git a/paddle/fluid/operators/sequence_unpad_op.cu b/paddle/fluid/operators/sequence_ops/sequence_unpad_op.cu
similarity index 95%
rename from paddle/fluid/operators/sequence_unpad_op.cu
rename to paddle/fluid/operators/sequence_ops/sequence_unpad_op.cu
index 75248372237ec2cb23122f6b16e64f6ce750ebf9..bf54f77f5b55cf7eb19873e352359c028207308a 100644
--- a/paddle/fluid/operators/sequence_unpad_op.cu
+++ b/paddle/fluid/operators/sequence_ops/sequence_unpad_op.cu
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/sequence_unpad_op.h"
+#include "paddle/fluid/operators/sequence_ops/sequence_unpad_op.h"
 
 namespace ops = paddle::operators;
 REGISTER_OP_CUDA_KERNEL(
diff --git a/paddle/fluid/operators/sequence_unpad_op.h b/paddle/fluid/operators/sequence_ops/sequence_unpad_op.h
similarity index 100%
rename from paddle/fluid/operators/sequence_unpad_op.h
rename to paddle/fluid/operators/sequence_ops/sequence_unpad_op.h
diff --git a/paddle/fluid/operators/tensorrt/CMakeLists.txt b/paddle/fluid/operators/tensorrt/CMakeLists.txt
new file mode 100644
index 0000000000000000000000000000000000000000..eee0b90fbae216e804e62993313796e914fcef5a
--- /dev/null
+++ b/paddle/fluid/operators/tensorrt/CMakeLists.txt
@@ -0,0 +1,5 @@
+op_library(tensorrt_engine_op DEPS tensorrt_engine tensorrt_converter)
+file(APPEND ${pybind_file} "USE_CUDA_ONLY_OP(tensorrt_engine);\n")
+nv_test(test_tensorrt_engine_op SRCS tensorrt_engine_op_test.cc
+  DEPS tensorrt_engine_op
+  analysis)
diff --git a/paddle/fluid/operators/tensorrt_engine_op.cc b/paddle/fluid/operators/tensorrt/tensorrt_engine_op.cc
similarity index 96%
rename from paddle/fluid/operators/tensorrt_engine_op.cc
rename to paddle/fluid/operators/tensorrt/tensorrt_engine_op.cc
index 41a5786fe8c3295390144732221280e152d0a15a..3cf2ce3c7ef87dcf75548f7d9c3a55d06ed765e8 100644
--- a/paddle/fluid/operators/tensorrt_engine_op.cc
+++ b/paddle/fluid/operators/tensorrt/tensorrt_engine_op.cc
@@ -17,7 +17,7 @@
 #include <string>
 #include <vector>
 
-#include "paddle/fluid/operators/tensorrt_engine_op.h"
+#include "paddle/fluid/operators/tensorrt/tensorrt_engine_op.h"
 
 namespace paddle {
 
diff --git a/paddle/fluid/operators/tensorrt_engine_op.cu.cc b/paddle/fluid/operators/tensorrt/tensorrt_engine_op.cu.cc
similarity index 93%
rename from paddle/fluid/operators/tensorrt_engine_op.cu.cc
rename to paddle/fluid/operators/tensorrt/tensorrt_engine_op.cu.cc
index e1ddfde6d51ef719ca0b89cf286b176195ee682a..cbe1b426f65386e722a7b02ec1fdfdf75bfd770c 100644
--- a/paddle/fluid/operators/tensorrt_engine_op.cu.cc
+++ b/paddle/fluid/operators/tensorrt/tensorrt_engine_op.cu.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/tensorrt_engine_op.h"
+#include "paddle/fluid/operators/tensorrt/tensorrt_engine_op.h"
 
 namespace ops = paddle::operators;
 
diff --git a/paddle/fluid/operators/tensorrt_engine_op.h b/paddle/fluid/operators/tensorrt/tensorrt_engine_op.h
similarity index 100%
rename from paddle/fluid/operators/tensorrt_engine_op.h
rename to paddle/fluid/operators/tensorrt/tensorrt_engine_op.h
diff --git a/paddle/fluid/operators/tensorrt_engine_op_test.cc b/paddle/fluid/operators/tensorrt/tensorrt_engine_op_test.cc
similarity index 99%
rename from paddle/fluid/operators/tensorrt_engine_op_test.cc
rename to paddle/fluid/operators/tensorrt/tensorrt_engine_op_test.cc
index e21101e8d12f210af08284dbcebe5c14c1af6dd3..56bdd6c2f2801967829f2baf889b5517a1d9d8d9 100644
--- a/paddle/fluid/operators/tensorrt_engine_op_test.cc
+++ b/paddle/fluid/operators/tensorrt/tensorrt_engine_op_test.cc
@@ -12,7 +12,7 @@ WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
 See the License for the specific language governing permissions and
 limitations under the License. */
 
-#include "paddle/fluid/operators/tensorrt_engine_op.h"
+#include "paddle/fluid/operators/tensorrt/tensorrt_engine_op.h"
 #include <gtest/gtest.h>
 #include "paddle/fluid/framework/block_desc.h"
 #include "paddle/fluid/framework/lod_tensor.h"
diff --git a/paddle/fluid/pybind/CMakeLists.txt b/paddle/fluid/pybind/CMakeLists.txt
index 6afa53cd36dee2004ada707a52995ce33b3650e9..6417da077e63dd78857d29ddd3484c646849daf4 100644
--- a/paddle/fluid/pybind/CMakeLists.txt
+++ b/paddle/fluid/pybind/CMakeLists.txt
@@ -10,12 +10,12 @@ if(WITH_PYTHON)
     hip_library(paddle_pybind SHARED
       SRCS ${PYBIND_SRCS}
       DEPS ${PYBIND_DEPS}
-      ${GLOB_OP_LIB})
+      ${GLOB_OP_LIB} ${GLOB_OPERATOR_DEPS})
   else()
     cc_library(paddle_pybind SHARED
       SRCS ${PYBIND_SRCS}
       DEPS ${PYBIND_DEPS}
-      ${GLOB_OP_LIB})
+      ${GLOB_OP_LIB} ${GLOB_OPERATOR_DEPS})
     if(NOT APPLE AND NOT ANDROID AND NOT WIN32)
       target_link_libraries(paddle_pybind rt)
     endif(NOT APPLE AND NOT ANDROID AND NOT WIN32)