diff --git a/.travis.yml b/.travis.yml index 0705baa1aca8b480b2a774076bd91fb9df401a53..162bebba091d84b295f929527de9804e65df5a65 100644 --- a/.travis.yml +++ b/.travis.yml @@ -25,9 +25,9 @@ addons: packages: - gcc-4.8 - g++-4.8 + - gfortran-4.8 - git - build-essential - - libatlas-base-dev - python - python-pip - python2.7-dev diff --git a/cmake/cblas.cmake b/cmake/cblas.cmake index 4e1ae7dc81231943c4bf3db4d4ac6f073f4fd1c4..26306f9849100d4463dde267acae5392cc81d7ac 100644 --- a/cmake/cblas.cmake +++ b/cmake/cblas.cmake @@ -16,7 +16,7 @@ set(CBLAS_FOUND OFF) ## Find MKL First. -set(MKL_ROOT $ENV{MKL_ROOT} CACHE PATH "Folder contains MKL") +set(MKL_ROOT $ENV{MKLROOT} CACHE PATH "Folder contains MKL") find_path(MKL_INCLUDE_DIR mkl.h PATHS ${MKL_ROOT}/include) diff --git a/cmake/external/glog.cmake b/cmake/external/glog.cmake index 71e20c85276b014c2e33735c3199c3772526c6c7..ab105611c812a4f4b642ac5b1213fdfe93fab97d 100644 --- a/cmake/external/glog.cmake +++ b/cmake/external/glog.cmake @@ -1,11 +1,11 @@ # Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. -# +# # Licensed under the Apache License, Version 2.0 (the "License"); # you may not use this file except in compliance with the License. # You may obtain a copy of the License at -# +# # http://www.apache.org/licenses/LICENSE-2.0 -# +# # Unless required by applicable law or agreed to in writing, software # distributed under the License is distributed on an "AS IS" BASIS, # WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. @@ -29,12 +29,14 @@ INCLUDE_DIRECTORIES(${GLOG_INCLUDE_DIR}) ExternalProject_Add( glog ${EXTERNAL_PROJECT_LOG_ARGS} + DEPENDS gflags GIT_REPOSITORY "https://github.com/google/glog.git" PREFIX ${GLOG_SOURCES_DIR} UPDATE_COMMAND "" CMAKE_ARGS -DCMAKE_INSTALL_PREFIX=${GLOG_INSTALL_DIR} CMAKE_ARGS -DCMAKE_POSITION_INDEPENDENT_CODE=ON - CMAKE_ARGS -DWITH_GFLAGS=OFF + CMAKE_ARGS -DWITH_GFLAGS=ON + CMAKE_ARGS -Dgflags_DIR=${GFLAGS_INSTALL_DIR}/lib/cmake/gflags CMAKE_ARGS -DBUILD_TESTING=OFF ) diff --git a/cmake/external/openblas.cmake b/cmake/external/openblas.cmake index 0e8c29c831c823f701d8eecd954d3b120085e495..29d17691db9f4575bae4372c61a0e1964e163fc9 100644 --- a/cmake/external/openblas.cmake +++ b/cmake/external/openblas.cmake @@ -15,7 +15,6 @@ INCLUDE(cblas) IF(NOT ${CBLAS_FOUND}) - MESSAGE(FATAL_ERROR "Please install OpenBlas, MKL or ATLAS.") INCLUDE(ExternalProject) SET(CBLAS_SOURCES_DIR ${THIRD_PARTY_PATH}/openblas) @@ -28,20 +27,40 @@ IF(NOT ${CBLAS_FOUND}) SET(CBLAS_LIBRARIES "${CBLAS_INSTALL_DIR}/lib/libopenblas.a" CACHE FILEPATH "openblas library" FORCE) ENDIF(WIN32) + IF(CMAKE_COMPILER_IS_GNUCC) + ENABLE_LANGUAGE(Fortran) + LIST(APPEND CBLAS_LIBRARIES gfortran pthread) + ENDIF(CMAKE_COMPILER_IS_GNUCC) + + IF(NOT CMAKE_Fortran_COMPILER) + MESSAGE(FATAL_ERROR "To build lapack in libopenblas, " + "you need to set gfortran compiler: cmake .. -DCMAKE_Fortran_COMPILER=...") + ENDIF(NOT CMAKE_Fortran_COMPILER) + ExternalProject_Add( openblas ${EXTERNAL_PROJECT_LOG_ARGS} - URL "https://github.com/xianyi/OpenBLAS/archive/v0.2.19.tar.gz" + GIT_REPOSITORY https://github.com/xianyi/OpenBLAS.git + GIT_TAG v0.2.19 PREFIX ${CBLAS_SOURCES_DIR} INSTALL_DIR ${CBLAS_INSTALL_DIR} BUILD_IN_SOURCE 1 - CONFIGURE_COMMAND "" - BUILD_COMMAND make CC=${CMAKE_C_COMPILER} FC=${CMAKE_Fortran_COMPILER} - INSTALL_COMMAND make install PREFIX= + BUILD_COMMAND ${CMAKE_MAKE_PROGRAM} FC=${CMAKE_Fortran_COMPILER} CC=${CMAKE_C_COMPILER} HOSTCC=${CMAKE_C_COMPILER} NO_SHARED=1 libs netlib + INSTALL_COMMAND ${CMAKE_MAKE_PROGRAM} install NO_SHARED=1 PREFIX= UPDATE_COMMAND "" + CONFIGURE_COMMAND "" + ) + + ExternalProject_Add_Step( + openblas lapacke_install + COMMAND ${CMAKE_COMMAND} -E copy "${CBLAS_SOURCES_DIR}/src/openblas/lapack-netlib/LAPACKE/include/lapacke_mangling_with_flags.h" "${CBLAS_INSTALL_DIR}/include/lapacke_mangling.h" + COMMAND ${CMAKE_COMMAND} -E copy "${CBLAS_SOURCES_DIR}/src/openblas/lapack-netlib/LAPACKE/include/lapacke.h" "${CBLAS_INSTALL_DIR}/include/lapacke.h" + COMMAND ${CMAKE_COMMAND} -E copy "${CBLAS_SOURCES_DIR}/src/openblas/lapack-netlib/LAPACKE/include/lapacke_config.h" "${CBLAS_INSTALL_DIR}/include/lapacke_config.h" + COMMAND ${CMAKE_COMMAND} -E copy "${CBLAS_SOURCES_DIR}/src/openblas/lapack-netlib/LAPACKE/include/lapacke_utils.h" "${CBLAS_INSTALL_DIR}/include/lapacke_utils.h" + DEPENDEES install ) LIST(APPEND external_project_dependencies openblas) -ENDIF() +ENDIF(NOT ${CBLAS_FOUND}) INCLUDE_DIRECTORIES(${CBLAS_INC_DIR}) diff --git a/cmake/external/protobuf.cmake b/cmake/external/protobuf.cmake index c0cf2719f9a7b3ae6be5cefffa3dbd2c3f712e82..84f459033f06f89d3b150317793c7e62274468b2 100644 --- a/cmake/external/protobuf.cmake +++ b/cmake/external/protobuf.cmake @@ -29,17 +29,12 @@ IF(WIN32) "${PROTOBUF_INSTALL_DIR}/lib/libprotoc.lib" CACHE FILEPATH "protoc library." FORCE) SET(PROTOBUF_PROTOC_EXECUTABLE "${PROTOBUF_INSTALL_DIR}/bin/protoc.exe" CACHE FILEPATH "protobuf executable." FORCE) ELSE(WIN32) - IF(${HOST_SYSTEM} STREQUAL "centos") - SET(LIB "lib64") - ELSE() - SET(LIB "lib") - ENDIF() SET(PROTOBUF_LITE_LIBRARY - "${PROTOBUF_INSTALL_DIR}/${LIB}/libprotobuf-lite.a" CACHE FILEPATH "protobuf lite library." FORCE) + "${PROTOBUF_INSTALL_DIR}/lib/libprotobuf-lite.a" CACHE FILEPATH "protobuf lite library." FORCE) SET(PROTOBUF_LIBRARY - "${PROTOBUF_INSTALL_DIR}/${LIB}/libprotobuf.a" CACHE FILEPATH "protobuf library." FORCE) + "${PROTOBUF_INSTALL_DIR}/lib/libprotobuf.a" CACHE FILEPATH "protobuf library." FORCE) SET(PROTOBUF_PROTOC_LIBRARY - "${PROTOBUF_INSTALL_DIR}/${LIB}/libprotoc.a" CACHE FILEPATH "protoc library." FORCE) + "${PROTOBUF_INSTALL_DIR}/lib/libprotoc.a" CACHE FILEPATH "protoc library." FORCE) SET(PROTOBUF_PROTOC_EXECUTABLE "${PROTOBUF_INSTALL_DIR}/bin/protoc" CACHE FILEPATH "protobuf executable." FORCE) ENDIF(WIN32) @@ -54,9 +49,11 @@ ExternalProject_Add( CONFIGURE_COMMAND ${CMAKE_COMMAND} ${PROTOBUF_SOURCES_DIR}/src/protobuf/cmake -Dprotobuf_BUILD_TESTS=OFF + -DZLIB_ROOT:FILEPATH=${ZLIB_ROOT} -DCMAKE_POSITION_INDEPENDENT_CODE=ON -DCMAKE_BUILD_TYPE=Release -DCMAKE_INSTALL_PREFIX=${PROTOBUF_INSTALL_DIR} + -DCMAKE_INSTALL_LIBDIR=lib ) LIST(APPEND external_project_dependencies protobuf) diff --git a/cmake/external/python.cmake b/cmake/external/python.cmake index 29247d5c3d474acaa5c65e450780f00b8885ee78..6372a9a768e580f74f837ccb6c57d4f4395eb779 100644 --- a/cmake/external/python.cmake +++ b/cmake/external/python.cmake @@ -26,11 +26,12 @@ IF(PYTHONLIBS_FOUND AND PYTHONINTERP_FOUND) find_python_module(wheel REQUIRED) find_python_module(google.protobuf REQUIRED) FIND_PACKAGE(NumPy REQUIRED) - IF(${PY_GOOGLE.PROTOBUF_VERSION} VERSION_LESS "3.0.0") + IF(${PY_GOOGLE.PROTOBUF_VERSION} AND ${PY_GOOGLE.PROTOBUF_VERSION} VERSION_LESS "3.0.0") MESSAGE(FATAL_ERROR "Found Python Protobuf ${PY_GOOGLE.PROTOBUF_VERSION} < 3.0.0, " - "please use pip to upgrade protobuf.") - ENDIF(${PY_GOOGLE.PROTOBUF_VERSION} VERSION_LESS "3.0.0") + "please use pip to upgrade protobuf. pip install -U protobuf") + ENDIF() ELSE(PYTHONLIBS_FOUND AND PYTHONINTERP_FOUND) + MESSAGE(FATAL_ERROR "Please install python 2.7 before building PaddlePaddle.") ##################################### PYTHON ######################################## SET(PYTHON_SOURCES_DIR ${THIRD_PARTY_PATH}/python) SET(PYTHON_INSTALL_DIR ${THIRD_PARTY_PATH}/install/python) diff --git a/cmake/external/swig.cmake b/cmake/external/swig.cmake index 63e8bd25462e50e2f78908899938468c989b3ac3..744c766ee7b067058b2cb4aa7f7b761cbb9778d4 100644 --- a/cmake/external/swig.cmake +++ b/cmake/external/swig.cmake @@ -38,14 +38,6 @@ IF(NOT SWIG_FOUND) SET(SWIG_DIR ${SWIG_SOURCES_DIR} CACHE FILEPATH "SWIG Directory" FORCE) SET(SWIG_EXECUTABLE ${SWIG_SOURCES_DIR}/swig.exe CACHE FILEPATH "SWIG Executable" FORCE) ELSE(WIN32) - # From PCRE configure - ExternalProject_Add(pcre - ${EXTERNAL_PROJECT_LOG_ARGS} - GIT_REPOSITORY https://github.com/svn2github/pcre.git - PREFIX ${SWIG_SOURCES_DIR}/pcre - CMAKE_ARGS -DCMAKE_INSTALL_PREFIX:PATH=${SWIG_INSTALL_DIR}/pcre - ) - # swig uses bison find it by cmake and pass it down FIND_PACKAGE(BISON) @@ -54,16 +46,11 @@ IF(NOT SWIG_FOUND) GIT_REPOSITORY https://github.com/swig/swig.git GIT_TAG rel-3.0.10 PREFIX ${SWIG_SOURCES_DIR} - CONFIGURE_COMMAND cd ${SWIG_SOURCES_DIR}/src/swig && ./autogen.sh - CONFIGURE_COMMAND cd ${SWIG_SOURCES_DIR}/src/swig && - env "PCRE_LIBS=${SWIG_INSTALL_DIR}/pcre/lib/libpcre.a ${SWIG_INSTALL_DIR}/pcre/lib/libpcrecpp.a ${SWIG_INSTALL_DIR}/pcre/lib/libpcreposix.a" - ./configure - --prefix=${SWIG_INSTALL_DIR} - --with-pcre-prefix=${SWIG_INSTALL_DIR}/pcre - BUILD_COMMAND cd ${SWIG_SOURCES_DIR}/src/swig && make - INSTALL_COMMAND cd ${SWIG_SOURCES_DIR}/src/swig && make install - UPDATE_COMMAND "" - DEPENDS pcre + CONFIGURE_COMMAND cd && ./autogen.sh && ./configure + --prefix=${SWIG_INSTALL_DIR} --without-pcre + BUILD_COMMAND cd && make + INSTALL_COMMAND cd && make install + UPDATE_COMMAND "" ) SET(SWIG_DIR ${SWIG_INSTALL_DIR}/share/swig/${SWIG_TARGET_VERSION}) diff --git a/cmake/external/warpctc.cmake b/cmake/external/warpctc.cmake index f5e4b3e1eb39acbe8dbcd0023956ca7e52c1ecd8..172c318b35d611d0432b78f2a18eb58a7d272b90 100644 --- a/cmake/external/warpctc.cmake +++ b/cmake/external/warpctc.cmake @@ -54,6 +54,7 @@ ExternalProject_Add( CMAKE_ARGS -DWITH_GPU=${WITH_GPU} CMAKE_ARGS -DWITH_OMP=${USE_OMP} CMAKE_ARGS -DWITH_TORCH=OFF + CMAKE_ARGS -DCMAKE_DISABLE_FIND_PACKAGE_Torch=TRUE CMAKE_ARGS -DBUILD_SHARED=ON ) diff --git a/cmake/flags.cmake b/cmake/flags.cmake index 0d1ef5cd8449bd31b4cfa4619f27bce7c1f55ebb..b76852fc6c50e80633c8294fb2724b83f15293a7 100644 --- a/cmake/flags.cmake +++ b/cmake/flags.cmake @@ -96,6 +96,7 @@ set(COMMON_FLAGS -Wno-unused-parameter -Wno-unused-function -Wno-error=literal-suffix + -Wno-error=sign-compare -Wno-error=unused-local-typedefs) set(GPU_COMMON_FLAGS @@ -105,6 +106,7 @@ set(GPU_COMMON_FLAGS -Wdelete-non-virtual-dtor -Wno-unused-parameter -Wno-unused-function + -Wno-error=sign-compare -Wno-error=literal-suffix -Wno-error=unused-local-typedefs -Wno-error=unused-function # Warnings in Numpy Header. diff --git a/cmake/system.cmake b/cmake/system.cmake index 788db404ebfb6facbaedf2910186f3b1afe775c1..3e472da7e0bd9c433f92f3e8b52970cd2cc6dcba 100644 --- a/cmake/system.cmake +++ b/cmake/system.cmake @@ -12,6 +12,14 @@ # See the License for the specific language governing permissions and # limitations under the License. +# Detects the OS and sets appropriate variables. +# CMAKE_SYSTEM_NAME only give us a coarse-grained name, +# but the name like centos is necessary in some scenes +# to distinguish system for customization. +# +# for instance, protobuf libs path is /lib64 +# on CentOS, but /lib on other systems. + IF(WIN32) SET(HOST_SYSTEM "win32") ELSE(WIN32) @@ -21,6 +29,7 @@ ELSE(WIN32) SET(MACOS_VERSION ${VERSION}) SET(HOST_SYSTEM "macosx") ELSE(APPLE) + IF(EXISTS "/etc/issue") FILE(READ "/etc/issue" LINUX_ISSUE) IF(LINUX_ISSUE MATCHES "CentOS") @@ -29,8 +38,24 @@ ELSE(WIN32) SET(HOST_SYSTEM "debian") ELSEIF(LINUX_ISSUE MATCHES "Ubuntu") SET(HOST_SYSTEM "ubuntu") + ELSEIF(LINUX_ISSUE MATCHES "Red Hat") + SET(HOST_SYSTEM "redhat") + ELSEIF(LINUX_ISSUE MATCHES "Fedora") + SET(HOST_SYSTEM "fedora") ENDIF() ENDIF(EXISTS "/etc/issue") + + IF(EXISTS "/etc/redhat-release") + FILE(READ "/etc/redhat-release" LINUX_ISSUE) + IF(LINUX_ISSUE MATCHES "CentOS") + SET(HOST_SYSTEM "centos") + ENDIF() + ENDIF(EXISTS "/etc/redhat-release") + + IF(NOT HOST_SYSTEM) + SET(HOST_SYSTEM ${CMAKE_SYSTEM_NAME}) + ENDIF() + ENDIF(APPLE) ENDIF(WIN32) @@ -47,7 +72,7 @@ SET(EXTERNAL_PROJECT_LOG_ARGS LOG_DOWNLOAD 0 # Wrap download in script to log output LOG_UPDATE 1 # Wrap update in script to log output LOG_CONFIGURE 1 # Wrap configure in script to log output - LOG_BUILD 1 # Wrap build in script to log output + LOG_BUILD 0 # Wrap build in script to log output LOG_TEST 1 # Wrap test in script to log output - LOG_INSTALL 1 # Wrap install in script to log output + LOG_INSTALL 0 # Wrap install in script to log output ) diff --git a/demo/traffic_prediction/predict.sh b/demo/traffic_prediction/predict.sh index cec35dce11d1c146a9e878ebab81abe904d6136c..2dbd5e8805dd97d35c7d58917f8ec6b5033bda03 100755 --- a/demo/traffic_prediction/predict.sh +++ b/demo/traffic_prediction/predict.sh @@ -25,6 +25,6 @@ paddle train \ --config_args=is_predict=1 \ --predict_output_dir=. -python gen_result.py > result.txt +python gen_result.py > result.csv rm -rf rank-00000 diff --git a/doc/api/trainer_config_helpers/layers.rst b/doc/api/trainer_config_helpers/layers.rst index 4e429650e545179eca2f947e4af660222ad7cda8..8b0e553eacc932bc59062103ac6e6ac4245d03cb 100644 --- a/doc/api/trainer_config_helpers/layers.rst +++ b/doc/api/trainer_config_helpers/layers.rst @@ -382,6 +382,15 @@ sampling_id_layer :members: sampling_id_layer :noindex: +Slicing and Joining Layers +========================== + +pad_layer +----------- +.. automodule:: paddle.trainer_config_helpers.layers + :members: pad_layer + :noindex: + .. _api_trainer_config_helpers_layers_cost_layers: Cost Layers diff --git a/doc/getstarted/build_and_install/build_from_source_en.md b/doc/getstarted/build_and_install/build_from_source_en.md index 6954be3b2bb956755c7820bf285addfd15226874..d9d54bff3096cb3520409971dbd1b2e179ac8be1 100644 --- a/doc/getstarted/build_and_install/build_from_source_en.md +++ b/doc/getstarted/build_and_install/build_from_source_en.md @@ -4,6 +4,8 @@ Installing from Sources * [1. Download and Setup](#download) * [2. Requirements](#requirements) * [3. Build on Ubuntu](#ubuntu) +* [4. Build on Centos](#centos) + ## Download and Setup You can download PaddlePaddle from the [github source](https://github.com/PaddlePaddle/Paddle). @@ -16,9 +18,10 @@ cd paddle To compile the source code, your computer must be equipped with the following dependencies. -- **Compiler**: GCC >= 4.8 or Clang >= 3.3 (AppleClang >= 5.1) -- **CMake**: version >= 3.0 (at least CMake 3.4 on Mac OS X) +- **Compiler**: GCC >= 4.8 or Clang >= 3.3 (AppleClang >= 5.1) and gfortran compiler +- **CMake**: CMake >= 3.0 (at least CMake 3.4 on Mac OS X) - **BLAS**: MKL, OpenBlas or ATLAS +- **Python**: only support Python 2.7 **Note:** For CUDA 7.0 and CUDA 7.5, GCC 5.0 and up are not supported! For CUDA 8.0, GCC versions later than 5.3 are not supported! @@ -64,7 +67,8 @@ As a simple example, consider the following: 1. **BLAS Dependencies(optional)** - Paddle will find BLAS from system's default path. But you can specify MKL, OpenBLAS or ATLAS via `MKL_ROOT`, `OPENBLAS_ROOT` or `ATLAS_ROOT`. + CMake will search BLAS libraries from system. If not found, OpenBLAS will be downloaded, built and installed automatically. + To utilize preinstalled BLAS， you can simply specify MKL, OpenBLAS or ATLAS via `MKL_ROOT`, `OPENBLAS_ROOT` or `ATLAS_ROOT`. ```bash # specify MKL @@ -94,12 +98,78 @@ As a simple example, consider the following: ### Install Dependencies -- **CPU Dependencies** +- **Paddle Dependencies** ```bash # necessary sudo apt-get update - sudo apt-get install -y g++ make cmake build-essential libatlas-base-dev python python-pip libpython-dev git + sudo apt-get install -y git curl gcc g++ gfortran make build-essential automake + sudo apt-get install -y python python-pip python-numpy libpython-dev bison + sudo pip install 'protobuf==3.1.0.post1' + + # install cmake 3.4 + curl -sSL https://cmake.org/files/v3.4/cmake-3.4.1.tar.gz | tar -xz && \ + cd cmake-3.4.1 && ./bootstrap && make -j4 && sudo make install && \ + cd .. && rm -rf cmake-3.4.1 + ``` + +- **GPU Dependencies (optional)** + + To build GPU version, you will need the following installed: + + 1. a CUDA-capable GPU + 2. A supported version of Linux with a gcc compiler and toolchain + 3. NVIDIA CUDA Toolkit (available at http://developer.nvidia.com/cuda-downloads) + 4. NVIDIA cuDNN Library (availabel at https://developer.nvidia.com/cudnn) + + The CUDA development environment relies on tight integration with the host development environment, + including the host compiler and C runtime libraries, and is therefore only supported on + distribution versions that have been qualified for this CUDA Toolkit release. + + After downloading cuDNN library, issue the following commands: + + ```bash + sudo tar -xzf cudnn-7.5-linux-x64-v5.1.tgz -C /usr/local + sudo chmod a+r /usr/local/cuda/include/cudnn.h /usr/local/cuda/lib64/libcudnn* + ``` + Then you need to set LD\_LIBRARY\_PATH, PATH environment variables in ~/.bashrc. + + ```bash + export LD_LIBRARY_PATH=/usr/local/cuda/lib64:$LD_LIBRARY_PATH + export PATH=/usr/local/cuda/bin:$PATH + ``` + +### Build and Install + +As usual, the best option is to create build folder under paddle project directory. + +```bash +mkdir build && cd build +``` + +Finally, you can build and install PaddlePaddle: + +```bash +# you can add build option here, such as: +cmake .. -DCMAKE_INSTALL_PREFIX= +# please use sudo make install, if you want to install PaddlePaddle into the system +make -j `nproc` && make install +# set PaddlePaddle installation path in ~/.bashrc +export PATH=/bin:$PATH +# install PaddlePaddle Python modules. +sudo pip install /opt/paddle/share/wheels/*.whl +``` +## Build on Centos 7 + +### Install Dependencies + +- **CPU Dependencies** + + ```bash + # necessary + sudo yum update + sudo yum install -y epel-release + sudo yum install -y make cmake3 python-devel python-pip gcc-gfortran swig git sudo pip install wheel numpy sudo pip install 'protobuf>=3.0.0' ``` @@ -142,7 +212,7 @@ Finally, you can build and install PaddlePaddle: ```bash # you can add build option here, such as: -cmake .. -DCMAKE_INSTALL_PREFIX= +cmake3 .. -DCMAKE_INSTALL_PREFIX= # please use sudo make install, if you want to install PaddlePaddle into the system make -j `nproc` && make install # set PaddlePaddle installation path in ~/.bashrc diff --git a/doc/howto/deep_model/rnn/hierarchical_layer_cn.rst b/doc/howto/deep_model/rnn/hierarchical_layer_cn.rst index 943b1d4bb84646d9f60de7790be166a83d10b1e0..4b328fc9d38bc5dfec35d5e0f0d46136aeeb41bc 100644 --- a/doc/howto/deep_model/rnn/hierarchical_layer_cn.rst +++ b/doc/howto/deep_model/rnn/hierarchical_layer_cn.rst @@ -32,7 +32,7 @@ pooling_layer 的使用示例如下，详细见 :ref:`api_trainer_config_helpers - `pooling_type` 目前支持两种，分别是：MaxPooling()和AvgPooling()。 -- `agg_level=AggregateLevel.TIMESTEP` 时（默认值）： +- `agg_level=AggregateLevel.EACH_TIMESTEP` 时（默认值）： - 作用：双层序列经过运算变成一个0层序列，或单层序列经过运算变成一个0层序列 - 输入：一个双层序列，或一个单层序列 @@ -54,7 +54,7 @@ last_seq 的使用示例如下（ :ref:`api_trainer_config_helpers_layers_first_ last = last_seq(input=layer, agg_level=AggregateLevel.EACH_SEQUENCE) -- `agg_level=AggregateLevel.TIMESTEP` 时（默认值）： +- `agg_level=AggregateLevel.EACH_TIMESTEP` 时（默认值）： - 作用：一个双层序列经过运算变成一个0层序列，或一个单层序列经过运算变成一个0层序列 - 输入：一个双层序列或一个单层序列 diff --git a/doc/howto/usage/k8s/k8s_aws_en.md b/doc/howto/usage/k8s/k8s_aws_en.md index b04bfba590de42956dfe99256cde325b24adbfab..a6422b9be00e210a6a305260585520acd72fb2f1 100644 --- a/doc/howto/usage/k8s/k8s_aws_en.md +++ b/doc/howto/usage/k8s/k8s_aws_en.md @@ -2,25 +2,16 @@ ## Create AWS Account and IAM Account -To use AWS, we need to sign up an AWS account on Amazon's Web site. -An AWS account allows us to login to the AWS Console Web interface to -create IAM users and user groups. Usually, we create a user group with -privileges required to run PaddlePaddle, and we create users for -those who are going to run PaddlePaddle and add these users into the -group. IAM users can identify themselves using password and tokens, -where passwords allows users to log in to the AWS Console, and tokens -make it easy for users to submit and inspect jobs from the command -line. +Under each AWS account, we can create multiple [IAM](http://docs.aws.amazon.com/IAM/latest/UserGuide/introduction.html) users. This allows us to grant some privileges to each IAM user and to create/operate AWS clusters as an IAM user. To sign up an AWS account, please follow [this guide](http://docs.aws.amazon.com/lambda/latest/dg/setting-up.html). -To create users and user groups under an AWS account, please +To create IAM users and user groups under an AWS account, please follow [this guide](http://docs.aws.amazon.com/IAM/latest/UserGuide/id_users_create.html). -Please be aware that this tutorial needs the following privileges in -the user group: +Please be aware that this tutorial needs the following privileges for the user in IAM: - AmazonEC2FullAccess - AmazonS3FullAccess @@ -31,14 +22,7 @@ the user group: - IAMUserSSHKeys - IAMFullAccess - NetworkAdministrator - - -By the time we write this tutorial, we noticed that Chinese AWS users -might suffer from authentication problems when running this tutorial. -Our solution is that we create a VM instance with the default Amazon -AMI and in the same zone as our cluster runs, so we can SSH to this VM -instance as a tunneling server and control our cluster and jobs from -it. +- AWSKeyManagementServicePowerUser ## PaddlePaddle on AWS @@ -46,9 +30,11 @@ it. Here we will show you step by step on how to run PaddlePaddle training on AWS cluster. -###Download kube-aws and kubectl +### Download kube-aws and kubectl + +#### kube-aws -####kube-aws +[kube-aws](https://github.com/coreos/kube-aws) is a CLI tool to automate cluster deployment to AWS. Import the CoreOS Application Signing Public Key: @@ -63,7 +49,7 @@ gpg2 --fingerprint FC8A365E ``` The correct key fingerprint is `18AD 5014 C99E F7E3 BA5F 6CE9 50BD D3E0 FC8A 365E` -Go to the [releases](https://github.com/coreos/kube-aws/releases) and download the latest release tarball and detached signature (.sig) for your architecture. +We can download `kube-aws` from its [release page](https://github.com/coreos/kube-aws/releases). In this tutorial, we use version 0.9.1 Validate the tarball's GPG signature: @@ -88,24 +74,30 @@ mv ${PLATFORM}/kube-aws /usr/local/bin ``` -####kubectl +#### kubectl -Go to the [releases](https://github.com/kubernetes/kubernetes/releases) and download the latest release tarball. +[kubectl](https://kubernetes.io/docs/user-guide/kubectl-overview/) is a command line interface for running commands against Kubernetes clusters. -Extract the tarball and then concate the kubernetes binaries directory into PATH: +Download `kubectl` from the Kubernetes release artifact site with the `curl` tool. ``` -export PATH=/platforms/linux/amd64:$PATH +# OS X +curl -O https://storage.googleapis.com/kubernetes-release/release/"$(curl -s https://storage.googleapis.com/kubernetes-release/release/stable.txt)"/bin/darwin/amd64/kubectl +# Linux +curl -O https://storage.googleapis.com/kubernetes-release/release/"$(curl -s https://storage.googleapis.com/kubernetes-release/release/stable.txt)"/bin/linux/amd64/kubectl ``` -User credentials and security tokens will be generated later in user directory, not in `~/.kube/config`, they will be necessary to use the CLI or the HTTP Basic Auth. +Make the kubectl binary executable and move it to your PATH (e.g. `/usr/local/bin`): +``` +chmod +x ./kubectl +sudo mv ./kubectl /usr/local/bin/kubectl +``` -###Configure AWS Credentials - -First check out [this](http://docs.aws.amazon.com/cli/latest/userguide/installing.html) for installing the AWS command line interface, if you use ec2 instance with default amazon AMI, the cli tool has already been installed on your machine. +### Configure AWS Credentials +First check out [this](http://docs.aws.amazon.com/cli/latest/userguide/installing.html) for installing the AWS command line interface. And then configure your AWS account information: @@ -115,44 +107,49 @@ aws configure ``` -Fill in the required fields (You can get your AWS aceess key id and AWS secrete access key by following [this](http://docs.aws.amazon.com/cli/latest/userguide/cli-chap-getting-started.html) instruction): +Fill in the required fields: ``` AWS Access Key ID: YOUR_ACCESS_KEY_ID AWS Secrete Access Key: YOUR_SECRETE_ACCESS_KEY -Default region name: us-west-2 +Default region name: us-west-1 Default output format: json - ``` -Test that your credentials work by describing any instances you may already have running on your account: +`YOUR_ACCESS_KEY_ID`, and `YOUR_SECRETE_ACCESS_KEY` is the IAM key and secret from [Create AWS Account and IAM Account](#create-aws-account-and-iam-account) + +Verify that your credentials work by describing any instances you may already have running on your account: ``` aws ec2 describe-instances ``` -###Define Cluster Parameters +### Define Cluster Parameters -####EC2 key pair +#### EC2 key pair The keypair that will authenticate SSH access to your EC2 instances. The public half of this key pair will be configured on each CoreOS node. -After creating a key pair, you will use the name you gave the keys to configure the cluster. Key pairs are only available to EC2 instances in the same region. More info in the [EC2 Keypair docs](http://docs.aws.amazon.com/AWSEC2/latest/UserGuide/ec2-key-pairs.html). +Follow [EC2 Keypair docs](http://docs.aws.amazon.com/AWSEC2/latest/UserGuide/ec2-key-pairs.html) to create a EC2 key pair + +After creating a key pair, you will use the key pair name to configure the cluster. -####KMS key +Key pairs are only available to EC2 instances in the same region. We are using us-west-1 in our tutorial, so make sure to creat key pairs in that region (N. California). + +#### KMS key Amazon KMS keys are used to encrypt and decrypt cluster TLS assets. If you already have a KMS Key that you would like to use, you can skip creating a new key and provide the Arn string for your existing key. You can create a KMS key in the AWS console, or with the aws command line tool: ``` -$ aws kms --region=us-west-2 create-key --description="kube-aws assets" +aws kms --region=us-west-1 create-key --description="kube-aws assets" { "KeyMetadata": { "CreationDate": 1458235139.724, "KeyState": "Enabled", - "Arn": "arn:aws:kms:us-west-2:xxxxxxxxx:key/xxxxxxxxxxxxxxxxxxx", + "Arn": "arn:aws:kms:us-west-1:aaaaaaaaaaaaa:key/xxxxxxxxxxxxxxxxxxx", "AWSAccountId": "xxxxxxxxxxxxx", "Enabled": true, "KeyUsage": "ENCRYPT_DECRYPT", @@ -162,11 +159,13 @@ $ aws kms --region=us-west-2 create-key --description="kube-aws assets" } ``` -You will use the `KeyMetadata.Arn` string to identify your KMS key in the init step. +We will need to use the value of `Arn` later. And then you need to add several inline policies in your user permission. -kms inline policy: +Go to IAM user page, click on `Add inline policy` button, and then select `Custom Policy` + +paste into following inline policies: ``` { @@ -180,18 +179,10 @@ kms inline policy: "kms:Encrypt" ], "Resource": [ - "arn:aws:kms:*:xxxxxxxxx:key/*" + "arn:aws:kms:*:AWS_ACCOUNT_ID:key/*" ] - } - ] -} -``` -cloudformation inline policy: - -``` -"Version": "2012-10-17", - "Statement": [ - { + }, + { "Sid": "Stmt1482205746000", "Effect": "Allow", "Action": [ @@ -200,26 +191,43 @@ cloudformation inline policy: "cloudformation:DeleteStack", "cloudformation:DescribeStacks", "cloudformation:DescribeStackResource", - "cloudformation:GetTemplate" + "cloudformation:GetTemplate", + "cloudformation:DescribeStackEvents" ], "Resource": [ - "arn:aws:cloudformation:us-west-2:xxxxxxxxx:stack/YOUR_CLUSTER_NAME/*" + "arn:aws:cloudformation:us-west-1:AWS_ACCOUNT_ID:stack/MY_CLUSTER_NAME/*" ] } ] } ``` +`AWS_ACCOUNT_ID`: You can get it from following command line: + +``` +aws sts get-caller-identity --output text --query Account +``` -####External DNS name +`MY_CLUSTER_NAME`: Pick a MY_CLUSTER_NAME that you like, you will use it later as well. -When the cluster is created, the controller will expose the TLS-secured API on a public IP address. You will need to create an A record for the external DNS hostname you want to point to this IP address. You can find the API external IP address after the cluster is created by invoking kube-aws status. +#### External DNS name -####S3 bucket +When the cluster is created, the controller will expose the TLS-secured API on a DNS name. + +The A record of that DNS name needs to be point to the cluster ip address. + +We will need to use DNS name later in tutorial. If you don't already own one, you can choose any DNS name (e.g., `paddle`) and modify `/etc/hosts` to associate cluster ip with that DNS name. + +#### S3 bucket You need to create an S3 bucket before startup the Kubernetes cluster. -####Initialize an asset directory +There are some bugs in aws cli in creating S3 bucket, so let's use the [Web console](https://console.aws.amazon.com/s3/home?region=us-west-1). + +Click on `Create Bucket`, fill in a unique BUCKET_NAME, and make sure region is us-west-1 (Northern California). + + +#### Initialize an asset directory Create a directory on your local machine to hold the generated assets: @@ -231,29 +239,44 @@ $ cd my-cluster Initialize the cluster CloudFormation stack with the KMS Arn, key pair name, and DNS name from the previous step: ``` -$ kube-aws init \ ---cluster-name=my-cluster-name \ ---external-dns-name=my-cluster-endpoint \ +kube-aws init \ +--cluster-name=MY_CLUSTER_NAME \ +--external-dns-name=MY_EXTERNAL_DNS_NAME \ --region=us-west-1 \ ---availability-zone=us-west-1c \ ---key-name=key-pair-name \ ---kms-key-arn="arn:aws:kms:us-west-2:xxxxxxxxxx:key/xxxxxxxxxxxxxxxxxxx" +--availability-zone=us-west-1a \ +--key-name=KEY_PAIR_NAME \ +--kms-key-arn="arn:aws:kms:us-west-1:xxxxxxxxxx:key/xxxxxxxxxxxxxxxxxxx" ``` +`MY_CLUSTER_NAME`: the one you picked in [KMS key](#kms-key) + +`MY_EXTERNAL_DNS_NAME`: see [External DNS name](#external-dns-name) + +`KEY_PAIR_NAME`: see [EC2 key pair](#ec2-key-pair) + +`--kms-key-arn`: the "Arn" in [KMS key](#kms-key) + +Here `us-west-1a` is used for parameter `--availability-zone`, but supported availability zone varies among AWS accounts. + +Please check if `us-west-1a` is supported by `aws ec2 --region us-west-1 describe-availability-zones`, if not switch to other supported availability zone. (e.g., `us-west-1a`, or `us-west-1b`) + +Note: please don't use `us-west-1c`. Subnets can currently only be created in the following availability zones: us-west-1b, us-west-1a. + There will now be a cluster.yaml file in the asset directory. This is the main configuration file for your cluster. -####Render contents of the asset directory + +#### Render contents of the asset directory In the simplest case, you can have kube-aws generate both your TLS identities and certificate authority for you. ``` -$ kube-aws render credentials --generate-ca +kube-aws render credentials --generate-ca ``` The next command generates the default set of cluster assets in your asset directory. ``` -sh $ kube-aws render stack +kube-aws render stack ``` Here's what the directory structure looks like: @@ -285,47 +308,62 @@ $ tree These assets (templates and credentials) are used to create, update and interact with your Kubernetes cluster. -###Kubernetes Cluster Start Up +### Kubernetes Cluster Start Up -####Create the instances defined in the CloudFormation template +#### Create the instances defined in the CloudFormation template -Now for the exciting part, creating your cluster: +Now let's create your cluster (choose any PREFIX for the command below): ``` -$ kube-aws up --s3-uri s3:/// +kube-aws up --s3-uri s3://BUCKET_NAME/PREFIX ``` -####Configure DNS +`BUCKET_NAME`: the bucket name that you used in [S3 bucket](#s3-bucket) -You can invoke `kube-aws status` to get the cluster API endpoint after cluster creation, if necessary. This command can take a while. And then dig the load balancer hostname to get the ip address, use this ip to setup an A record for your external dns name. -####Access the cluster +#### Configure DNS -Once the API server is running, you should see: +You can invoke `kube-aws status` to get the cluster API endpoint after cluster creation. ``` -$ kubectl --kubeconfig=kubeconfig get nodes -NAME STATUS AGE -ip-10-0-0-xxx.us-west-1.compute.internal Ready 5m -ip-10-0-0-xxx.us-west-1.compute.internal Ready 5m -ip-10-0-0-xx.us-west-1.compute.internal Ready,SchedulingDisabled 5m +$ kube-aws status +Cluster Name: paddle-cluster +Controller DNS Name: paddle-cl-ElbAPISe-EEOI3EZPR86C-531251350.us-west-1.elb.amazonaws.com ``` +Use command `dig` to check the load balancer hostname to get the ip address. -###Setup PaddlePaddle Environment on AWS +``` +$ dig paddle-cl-ElbAPISe-EEOI3EZPR86C-531251350.us-west-1.elb.amazonaws.com -Now, we've created a cluster with following network capability: +;; QUESTION SECTION: +;paddle-cl-ElbAPISe-EEOI3EZPR86C-531251350.us-west-1.elb.amazonaws.com. IN A -1. All Kubernetes nodes can communicate with each other. +;; ANSWER SECTION: +paddle-cl-ElbAPISe-EEOI3EZPR86C-531251350.us-west-1.elb.amazonaws.com. 59 IN A 54.241.164.52 +paddle-cl-ElbAPISe-EEOI3EZPR86C-531251350.us-west-1.elb.amazonaws.com. 59 IN A 54.67.102.112 +``` -1. All Docker containers on Kubernetes nodes can communicate with each other. +In the above output, both ip `54.241.164.52`, `54.67.102.112` will work. -1. All Kubernetes nodes can communicate with all Docker containers on Kubernetes nodes. +If you own a DNS name, set the A record to any of the above ip. Otherwise you can edit `/etc/hosts` to associate ip with the DNS name. -1. All other traffic loads from outside of Kubernetes nodes cannot reach to the Docker containers on Kubernetes nodes except for creating the services for containers. +#### Access the cluster +Once the API server is running, you should see: -For sharing the training data across all the Kubernetes nodes, we use EFS (Elastic File System) in AWS. Ceph might be a better solution, but it requires high version of Linux kernel that might not be stable enough at this moment. We haven't automated the EFS setup at this moment, so please do the following steps: +``` +$ kubectl --kubeconfig=kubeconfig get nodes +NAME STATUS AGE +ip-10-0-0-xxx.us-west-1.compute.internal Ready 5m +ip-10-0-0-xxx.us-west-1.compute.internal Ready 5m +ip-10-0-0-xx.us-west-1.compute.internal Ready,SchedulingDisabled 5m +``` + + +### Setup Elastic File System for Cluster + +Training data is usually served on a distributed filesystem, we use Elastic File System (EFS) on AWS. Ceph might be a better solution, but it requires high version of Linux kernel that might not be stable enough at this moment. We haven't automated the EFS setup at this moment, so please do the following steps: 1. Make sure you added AmazonElasticFileSystemFullAccess policy in your group. @@ -342,57 +380,71 @@ For sharing the training data across all the Kubernetes nodes, we use EFS (Elast
![](src/efs_mount.png)
-Before starting the training, you should place your user config and divided training data onto EFS. When the training start, each task will copy related files from EFS into container, and it will also write the training results back onto EFS, we will show you how to place the data later in this article. +We will place user config and divided training data onto EFS. Training task will cache related files by copying them from EFS into container. It will also write the training results back onto EFS. We will show you how to place the data later in this article. + + + +### Core Concepts of PaddlePaddle Training on AWS + +Now we've already setup a 3 nodes distributed Kubernetes cluster, and on each node we've attached the EFS volume. In this training demo, we will create three Kubernetes pods and schedule them on three nodes. Each pod contains a PaddlePaddle container. When container gets created, it will start parameter server (pserver) and trainer process, load the training data from EFS volume and start the distributed training task. +#### Distributed Training Job +A distributed training job is represented by a [kubernetes job](https://kubernetes.io/docs/user-guide/jobs/#what-is-a-job). -###Core Concept of PaddlePaddle Training on AWS +Each Kuberentes job is described by a job config file, which specifies the information like the number of pods in the job and environment variables. -Now we've already setup a 3 nodes distributed Kubernetes cluster, and on each node we've attached the EFS volume, in this training demo, we will create three Kubernetes pod and scheduling them on 3 node. Each pod contains a PaddlePaddle container. When container gets created, it will start pserver and trainer process, load the training data from EFS volume and start the distributed training task. +In a distributed training job, we would: -####Use Kubernetes Job +1. upload the partitioned training data and configuration file onto EFS volume, and +1. create and submit the Kubernetes job config to the Kubernetes cluster to start the training job. -We use Kubernetes job to represent one time of distributed training. After the job get finished, Kubernetes will destroy job container and release all related resources. +#### Parameter Servers and Trainers -We can write a yaml file to describe the Kubernetes job. The file contains lots of configuration information, for example PaddlePaddle's node number, `paddle pserver` open port number, the network card info etc., these information are passed into container for processes to use as environment variables. +There are two roles in a PaddlePaddle cluster: `parameter server` and `trainer`. Each parameter server process maintains a shard of the global model. Each trainer has its local copy of the model, and uses its local data to update the model. During the training process, trainers send model updates to parameter servers, parameter servers are responsible for aggregating these updates, so that trainers can synchronize their local copy with the global model. -In one time of distributed training, user will confirm the PaddlePaddle node number first. And then upload the pre-divided training data and configuration file onth EFS volume. And then create the Kubernetes job yaml file; submit to the Kubernetes cluster to start the training job. +
![Model is partitioned into two shards. Managed by two parameter servers respectively.](src/pserver_and_trainer.png)
-####Create PaddlePaddle Node +In order to communicate with pserver, trainer needs to know the ip address of each pserver. In kubernetes it's better to use a service discovery mechanism (e.g., DNS hostname) rather than static ip address, since any pserver's pod may be killed and a new pod could be schduled onto another node of different ip address. We will improve paddlepaddle's service discovery ability. For now we will use static ip. -After Kubernetes master gets the request, it will parse the yaml file and create several pods (defined by PaddlePaddle's node number), Kubernetes will allocate these pods onto cluster's node. A pod represents a PaddlePaddle node, when pod is successfully allocated onto one physical/virtual machine, Kubernetes will startup the container in the pod, and this container will use the environment variables in yaml file and start up `paddle pserver` and `paddle trainer` processes. +Parameter server and trainer are packaged into a same docker image. They will run once pod is scheduled by kubernetes job. +#### Trainer ID -####Start up Training +Each trainer process requires a trainer ID, a zero-based index value, passed in as a command-line parameter. The trainer process thus reads the data partition indexed by this ID. -After container gets started, it starts up the distributed training by using scripts. We know `paddle train` process need to know other node's ip address and it's own trainer_id, since PaddlePaddle currently don't have the ability to do the service discovery, so in the start up script, each node will use job pod's name to query all to pod info from Kubernetes apiserver (apiserver's endpoint is an environment variable in container by default). +#### Training -With pod information, we can assign each pod a unique trainer_id. Here we sort all the pods by pod's ip, and assign the index to each PaddlePaddle node as it's trainer_id. The workflow of starting up the script is as follows: +The entry-point of a container is a Python script. As it runs in a pod, it can see some environment variables pre-defined by Kubernetes. This includes one that gives the job's identity, which can be used in a remote call to the Kubernetes apiserver that lists all pods in the job. -1. Query the api server to get pod information, and assign the trainer_id by sorting the ip. +We rank each pod by sorting them by their ips. The rank of each pod could be the "pod ID". Because we run one trainer and one parameter server in each pod, we can use this "pod ID" as the trainer ID. A detailed workflow of the entry-point script is as follows: + +1. Query the api server to get pod information, and assign the `trainer_id` by sorting the ip. 1. Copy the training data from EFS sharing volume into container. -1. Parse the `paddle pserver` and 'paddle trainer' startup parameters from environment variables, and then start up the processes. -1. PaddlePaddle will automatically write the result onto the PaddlePaddle node with trainer_id:0, we set the output path to be the EFS volume to save the result data. +1. Parse the `paddle pserver` and `paddle trainer` startup parameters from environment variables, and then start up the processes. +1. Trainer with `train_id` 0 will automatically write results onto EFS volume. -###Start PaddlePaddle Training Demo on AWS +### Start PaddlePaddle Training Demo on AWS Now we'll start a PaddlePaddle training demo on AWS, steps are as follows: 1. Build PaddlePaddle Docker image. 1. Divide the training data file and upload it onto the EFS sharing volume. -1. Create the training job yaml file, and start up the job. +1. Create the training job config file, and start up the job. 1. Check the result after training. -####Build PaddlePaddle Docker Image +#### Build PaddlePaddle Docker Image -PaddlePaddle docker image need to provide the runtime environment for `paddle pserver` and `paddle train`, so the container use this image should have two main function: +PaddlePaddle docker image need to provide the runtime environment for `pserver` and `trainer`, so the container use this image should have two main function: 1. Copy the training data into container. -1. Generate the startup parameter for `paddle pserver` and `paddle train` process, and startup the training. +1. Generate the startup parameter for `pserver` and `trainer` process, and startup the training. + +We need to create a new image since official `paddledev/paddle:cpu-latest` only have PaddlePaddle binary, but lack of the above functionalities. -Since official `paddledev/paddle:cpu-latest` have already included the PaddlePaddle binary, but lack of the above functionalities, so we will create the startup script based on this image, to achieve the work above. the detailed Dockerfile is as follows: +Dockerfile for creating the new image is as follows: ``` FROM paddledev/paddle:cpu-latest @@ -481,7 +533,7 @@ And then push the built image onto docker registry. docker push your_repo/paddle:mypaddle ``` -####Upload Training Data File +#### Upload Training Data File Here we will use PaddlePaddle's official recommendation demo as the content for this training, we put the training data file into a directory named by job name, which located in EFS sharing volume, the tree structure for the directory looks like: @@ -498,10 +550,10 @@ efs └── recommendation ``` -The `paddle-cluster-job` directory is the job name for this training, this training includes 3 PaddlePaddle node, we store the pre-divided data under `paddle-cluster-job/data` directory, directory 0, 1, 2 each represent 3 nodes' trainer_id. the training data in in recommendation directory, the training results and logs will be in the output directory. +The `paddle-cluster-job` directory is the job name for this training, this training includes 3 PaddlePaddle node, we store the partitioned data under `paddle-cluster-job/data` directory, directory 0, 1, 2 each represent 3 nodes' trainer_id. the training data in in recommendation directory, the training results and logs will be in the output directory. -####Create Kubernetes Job +#### Create Kubernetes Job Kubernetes use yaml file to describe job details, and then use command line tool to create the job in Kubernetes cluster. @@ -583,7 +635,7 @@ After we execute the above command, Kubernetes will create 3 pods and then pull -####Check Training Results +#### Check Training Results During the training, we can see the logs and models on EFS sharing volume, the output directory contains the training results. (Caution: node_0, node_1, node_2 directories represents PaddlePaddle node and train_id, not the Kubernetes node) @@ -640,7 +692,7 @@ I1116 09:10:18.019836 50 ParameterClient2.cpp:122] pserver 5 192.168.129.71:7 It'll take around 8 hours to finish this PaddlePaddle recommendation training demo on three 2 core 8 GB EC2 machine (m3.large). -###Kubernetes Cluster Tear Down +### Kubernetes Cluster Tear Down If you want to tear down the whole Kubernetes cluster, make sure to *delete* the EFS volume first (otherwise, you will get stucked on following steps), and then use the following command: @@ -651,16 +703,3 @@ kube-aws destroy It's an async call, it might take 5 min to tear down the whole cluster. If you created any Kubernetes Services of type LoadBalancer, you must delete these first, as the CloudFormation cannot be fully destroyed if any externally-managed resources still exist. - - - -## For Experts with Kubernetes and AWS - -Sometimes we might need to create or manage the cluster on AWS manually with limited privileges, so here we will explain more on what’s going on with the Kubernetes setup script. - -### Some Presumptions - -* Instances run on CoreOS, the official IAM. -* Kubernetes node use instance storage, no EBS get mounted. Etcd is running on additional node. -* For networking, we use Flannel network at this moment, we will use Calico solution later on. -* When you create a service with Type=LoadBalancer, Kubernetes will create and ELB, and create a security group for the ELB. diff --git a/doc/howto/usage/k8s/src/pserver_and_trainer.png b/doc/howto/usage/k8s/src/pserver_and_trainer.png new file mode 100644 index 0000000000000000000000000000000000000000..f41fe48920590333ad332bb51eb18e03dc251541 Binary files /dev/null and b/doc/howto/usage/k8s/src/pserver_and_trainer.png differ diff --git a/paddle/function/BufferArg.cpp b/paddle/function/BufferArg.cpp index fde48a73b61c31d06225cc1763efbc6971c86f57..5d595deb12c6c8ea419dd1f31b3c131a2f6a587a 100644 --- a/paddle/function/BufferArg.cpp +++ b/paddle/function/BufferArg.cpp @@ -20,23 +20,27 @@ limitations under the License. */ namespace paddle { const SequenceArg& BufferArg::sequence() const { - // CHECK_EQ(bufferType_, TENSOR_SEQUENCE_DATA); + CHECK_EQ(bufferType_, TENSOR_SEQUENCE_DATA); return dynamic_cast(*this); } const SparseMatrixArg& BufferArg::sparse() const { - // CHECK_EQ(bufferType_, TENSOR_SPARSE); + CHECK_EQ(bufferType_, TENSOR_SPARSE); return dynamic_cast(*this); } SparseMatrixArg::SparseMatrixArg(const CpuSparseMatrix& sparse, ArgType argType) : BufferArg(sparse, argType), row_(reinterpret_cast(sparse.getRows()), VALUE_TYPE_INT32), - col_(reinterpret_cast(sparse.getCols()), VALUE_TYPE_INT32) {} + col_(reinterpret_cast(sparse.getCols()), VALUE_TYPE_INT32) { + bufferType_ = TENSOR_SPARSE; +} SparseMatrixArg::SparseMatrixArg(const GpuSparseMatrix& sparse, ArgType argType) : BufferArg(sparse, argType), row_(reinterpret_cast(sparse.getRows()), VALUE_TYPE_INT32), - col_(reinterpret_cast(sparse.getCols()), VALUE_TYPE_INT32) {} + col_(reinterpret_cast(sparse.getCols()), VALUE_TYPE_INT32) { + bufferType_ = TENSOR_SPARSE; +} } // namespace paddle diff --git a/paddle/function/BufferArg.h b/paddle/function/BufferArg.h index 12352ba29e33920ba65bd66088b6f7cc53517b52..84209265ce7634121e3e4dde609cd787093c45ec 100644 --- a/paddle/function/BufferArg.h +++ b/paddle/function/BufferArg.h @@ -23,10 +23,11 @@ limitations under the License. */ namespace paddle { enum BufferType { - TENSOR_NORMAL = 0, - TENSOR_SEQUENCE_ID = 1, - TENSOR_SEQUENCE_DATA = 2, - TENSOR_SPARSE = 3 + TENSOR_UNKNOWN = 0, + TENSOR_NORMAL = 1, + TENSOR_SEQUENCE_ID = 2, + TENSOR_SEQUENCE_DATA = 3, + TENSOR_SPARSE = 4 }; enum SparseDataType { @@ -39,7 +40,6 @@ enum SparseDataFormat { SPARSE_CSR_FORMAT = 0, SPARSE_CSC_FORMAT = 1 }; class BufferArg; class SequenceArg; class SparseMatrixArg; -typedef std::shared_ptr BufferArgPtr; /** * \brief BufferArg used as the argument type of Function. @@ -50,6 +50,11 @@ typedef std::shared_ptr BufferArgPtr; * 3. SequenceArg for a Buffer of sequence data. * 4. SparseMatrixArg for a Buffer of sparse matrix. * + * Buffer shape + * For most buffers, the first dimension `shape()[0]` represents + * the size of the mini-batch. + * + * Buffer argType * There is an ArgType property for the BufferArg used as Function Output. * Whether the result of the Function calculation is assigned to the * output Buffer or added to the output Buffer is determined by the @@ -71,6 +76,14 @@ public: ArgType getArgType() const { return argType_; } public: + BufferArg(ValueType valueType, + const TensorShape& shape, + ArgType argType = UNSPECIFIED) + : buf_(nullptr), + valueType_(valueType), + shape_(shape), + argType_(argType) {} + BufferArg(void* buf, ValueType valueType, const TensorShape& shape, @@ -86,6 +99,7 @@ public: valueType_(DataType::value), shape_(2), argType_(argType) { + bufferType_ = TENSOR_NORMAL; shape_.setDim(0, matrix.getHeight()); shape_.setDim(1, matrix.getWidth()); } @@ -98,6 +112,7 @@ public: valueType_(DataType::value), shape_(shape), argType_(argType) { + bufferType_ = TENSOR_NORMAL; CHECK_EQ(matrix.getElementCnt(), shape.getElements()); } @@ -107,6 +122,7 @@ public: valueType_(DataType::value), shape_(1), argType_(argType) { + bufferType_ = TENSOR_NORMAL; shape_.setDim(0, vector.getSize()); } @@ -116,6 +132,7 @@ public: valueType_(VALUE_TYPE_INT32), shape_(1), argType_(argType) { + bufferType_ = TENSOR_NORMAL; shape_.setDim(0, vector.getSize()); } @@ -150,6 +167,8 @@ public: ValueType valueType() const { return valueType_; } BufferType bufferType() const { return bufferType_; } const TensorShape& shape() const { return shape_; } + bool isSparse() const { return (TENSOR_SPARSE == bufferType_); } + bool isSequenceArg() const { return TENSOR_SEQUENCE_DATA == bufferType_; } const SequenceArg& sequence() const; const SparseMatrixArg& sparse() const; @@ -158,8 +177,8 @@ protected: void* buf_; ValueType valueType_; TensorShape shape_; - BufferType bufferType_; - ArgType argType_ = UNSPECIFIED; + BufferType bufferType_{TENSOR_UNKNOWN}; + ArgType argType_{UNSPECIFIED}; // leading dimensions. The size is dims_.size() // Dims lds_; }; @@ -170,15 +189,24 @@ protected: // if a < b then value_.buf_[a] < value_.buf_[b] class SequenceIdArg : public BufferArg { public: + SequenceIdArg(const TensorShape& shape, ArgType argType = UNSPECIFIED) + : BufferArg(VALUE_TYPE_INT32, shape, argType) { + CHECK_EQ(shape_.ndims(), (size_t)1); + CHECK_GT(shape_[0], 1); + numSeqs_ = shape_[0] - 1; + } + SequenceIdArg(void* buf, const TensorShape& shape, ArgType argType = UNSPECIFIED) : BufferArg(buf, VALUE_TYPE_INT32, shape, argType) { + bufferType_ = TENSOR_SEQUENCE_ID; CHECK_EQ(shape_.ndims(), (size_t)1); numSeqs_ = shape_[0] - 1; } SequenceIdArg(const IVector& vector) : BufferArg(vector) { + bufferType_ = TENSOR_SEQUENCE_ID; numSeqs_ = shape_[0] - 1; } @@ -190,26 +218,41 @@ private: size_t numSeqs_; }; -// sequence data +// sequences data +// For mini-batch calculate, +// one batch can contain more than one sequence of data. +// SequenceArg can be used to represent sequences that contain multiple +// unequal lengths. class SequenceArg : public BufferArg { public: + SequenceArg(ValueType valueType, + const TensorShape& shape, + ArgType argType = UNSPECIFIED) + : BufferArg(valueType, shape, argType), startPositions_(TensorShape()) {} + SequenceArg(void* buf, ValueType valueType, const TensorShape& shape, const SequenceIdArg& startPositions, ArgType argType = UNSPECIFIED) : BufferArg(buf, valueType, shape, argType), - startPositions_(startPositions) {} + startPositions_(startPositions) { + bufferType_ = TENSOR_SEQUENCE_DATA; + } SequenceArg(const Matrix& matrix, const IVector& vector, ArgType argType = UNSPECIFIED) - : BufferArg(matrix, argType), startPositions_(vector) {} + : BufferArg(matrix, argType), startPositions_(vector) { + bufferType_ = TENSOR_SEQUENCE_DATA; + } ~SequenceArg() {} void* getIdBuf() const { return startPositions_.data(); } size_t numSeqs() const { return startPositions_.numSeqs(); } + SequenceIdArg& getSequenceId() { return startPositions_; } + const SequenceIdArg& getSequenceId() const { return startPositions_; } private: SequenceIdArg startPositions_; @@ -235,6 +278,7 @@ public: nnz_(nnz), format_(format), type_(type) { + bufferType_ = TENSOR_SPARSE; CHECK((valueType == VALUE_TYPE_FLOAT) || (valueType == VALUE_TYPE_DOUBLE)); CHECK_EQ(shape_.ndims(), (size_t)2); CHECK_EQ(row_.shape().ndims(), (size_t)1); diff --git a/paddle/function/BufferArgTest.cpp b/paddle/function/BufferArgTest.cpp index b345597435c9911ce95b596f5f7f2add47f4cd03..1744f377808f137dcda4a28acce336dc22be3d01 100644 --- a/paddle/function/BufferArgTest.cpp +++ b/paddle/function/BufferArgTest.cpp @@ -14,9 +14,7 @@ limitations under the License. */ #include "BufferArg.h" #include -#include "Function.h" #include "paddle/math/MemoryHandle.h" -#include "paddle/math/SparseMatrix.h" namespace paddle { @@ -37,55 +35,4 @@ TEST(BufferTest, SequenceIdArg) { EXPECT_EQ(buffer.numSeqs(), 9); } -TEST(BufferTest, asArgument) { - MatrixPtr matrix = Matrix::create(100, 200); - VectorPtr vector = Vector::create(100, false); - CpuSparseMatrix sparse(200, 300, 50); - - // prepare arguments - BufferArgs argments; - argments.addArg(*matrix); - argments.addArg(*vector); - argments.addArg(sparse); - - // function - auto function = [=](const BufferArgs& inputs) { - EXPECT_EQ(inputs.size(), 3); - - // check inputs[0] - EXPECT_EQ(inputs[0].shape().ndims(), 2); - EXPECT_EQ(inputs[0].shape()[0], 100); - EXPECT_EQ(inputs[0].shape()[1], 200); - EXPECT_EQ(inputs[0].data(), matrix->getData()); - - EXPECT_EQ(inputs[0].matrix().getHeight(), - matrix->getHeight()); - EXPECT_EQ(inputs[0].matrix().getWidth(), - matrix->getWidth()); - EXPECT_EQ(inputs[0].matrix().getData(), matrix->getData()); - - // check inputs[1] - EXPECT_EQ(inputs[1].shape().ndims(), 1); - EXPECT_EQ(inputs[1].shape()[0], 100); - EXPECT_EQ(inputs[1].data(), vector->getData()); - CpuVector inVector = inputs[1].vector(); - EXPECT_EQ(inVector.getSize(), vector->getSize()); - EXPECT_EQ(inVector.getData(), vector->getData()); - - // check inputs[2] - EXPECT_EQ(inputs[2].shape().ndims(), 2); - EXPECT_EQ(inputs[2].shape()[0], 200); - EXPECT_EQ(inputs[2].shape()[1], 300); - EXPECT_EQ(inputs[2].data(), sparse.getData()); - // CHECK_EQ(inputs[2].sparse().nnz(), 50); - // CHECK_EQ(inputs[2].sparse().dataFormat(), SPARSE_CSR_FORMAT); - // CHECK_EQ(inputs[2].sparse().dataType(), SPARSE_FLOAT_VALUE); - EXPECT_EQ(inputs[2].sparse().getRowBuf(), sparse.getRows()); - EXPECT_EQ(inputs[2].sparse().getColBuf(), sparse.getCols()); - }; - - // call function - function(argments); -} - } // namespace paddle diff --git a/paddle/function/CMakeLists.txt b/paddle/function/CMakeLists.txt index 75a2acc55ec3d33687f96d2b0398e52b69e8680d..6d20868072c3acaab2c5f9381bad5ea99d841d26 100644 --- a/paddle/function/CMakeLists.txt +++ b/paddle/function/CMakeLists.txt @@ -19,12 +19,13 @@ if(WITH_TESTING) # TODO: # file(GLOB test_files . *OpTest.cpp) # add_executable(${test_bin} EXCLUDE_FROM_ALL ${test_files}) - # add_simple_unittest(CrossMapNormalOpTest) + add_simple_unittest(CrossMapNormalOpTest) add_simple_unittest(TensorShapeTest) add_simple_unittest(TensorTypeTest) add_simple_unittest(BufferArgTest) add_simple_unittest(FunctionTest) - # add_simple_unittest(ContextProjectionOpTest) + add_simple_unittest(ContextProjectionOpTest) + add_simple_unittest(PadOpTest) endif() endif() diff --git a/paddle/function/ContextProjectionOp.cpp b/paddle/function/ContextProjectionOp.cpp index cb448562ebb37022f727ee65024f06f69d63e9cb..6cd4e4abee8fccf3a4745b0bfc6701df4ddfa5c0 100644 --- a/paddle/function/ContextProjectionOp.cpp +++ b/paddle/function/ContextProjectionOp.cpp @@ -17,7 +17,10 @@ limitations under the License. */ #include "paddle/math/Vector.h" namespace paddle { - +/** + * Context Projection Forward with CPU Matrix Device. + * + */ template <> void ContextProjectionForward(CpuMatrix& out_mat, const CpuMatrix& input_mat, @@ -70,10 +73,30 @@ void ContextProjectionForward(CpuMatrix& out_mat, } /** - * \param inputs[0] input value. - * \param inputs[1] input weight. - * \param inputs[2] input sequence. - * \param outputs[0] output value. + * Paddle Function for Context Projection Forward. + * Calculate the output layer value sequence after context projection. + * + * What is Context Projection for a sequence? + * For example, assumed input (x) has 4 words and the dimension of each word + * representation is 2. If we use zero to pad instead of learned weight to pad, + * and the context_lenth is 3, the output (y) is: + * + * @code + * x = [a1, a2; + * b1, b2; + * c1, c2; + * d1, d2] + * y = [0, 0, a1, a2, b1, b2; + * a1, a2, b1, b2, c1, c2; + * b1, b2, c1, c2, d1, d2; + * c1, c2, d1, d2, 0, 0] + * @endcode + * + * \param outputs[0].matrix output layer value, n * (d * l) + * \param outputs[0].vector start position sequence, n * 1 + * \param inputs[0].matrix input layer value, n * d + * \param inputs[0].vector start position sequence, n * 1 + * \param inputs[1].matrix input layer weight, pad * d */ template class ContextProjectionForwardFunc : public FunctionBase { @@ -85,28 +108,37 @@ public: } void calc(const BufferArgs& inputs, const BufferArgs& outputs) override { - CHECK_EQ((size_t)3, inputs.size()); + CHECK(1 == inputs.size() || 2 == inputs.size()); CHECK_EQ((size_t)1, outputs.size()); + CHECK(inputs[0].isSequenceArg() && outputs[0].isSequenceArg()) + << "SequenceArg required here"; + const auto val_seqs = dynamic_cast(inputs[0]); + auto out_seq = dynamic_cast(outputs[0]); - CHECK(outputs[0].data() && inputs[0].data() && inputs[2].data()); - CHECK_EQ(outputs[0].shape().ndims(), (size_t)2); - CHECK_EQ(inputs[0].shape().ndims(), (size_t)2); - CHECK_EQ(inputs[1].shape().ndims(), (size_t)2); - CHECK_EQ(inputs[2].shape().ndims(), (size_t)1); + CHECK(out_seq.data() && val_seqs.data() && val_seqs.getSequenceId().data()); + CHECK_EQ(out_seq.shape().ndims(), (size_t)2); + CHECK_EQ(val_seqs.shape().ndims(), (size_t)2); + CHECK_EQ(val_seqs.getSequenceId().shape().ndims(), (size_t)1); + if (2 == inputs.size()) { + CHECK_EQ(inputs[1].shape().ndims(), (size_t)2); + } /// dim of output = dim of input * context_length - CHECK_EQ(outputs[0].shape()[1], inputs[0].shape()[1] * context_length_); - /// dim of input == dim of weight - CHECK_EQ(inputs[0].shape()[1], inputs[1].shape()[1]); + CHECK_EQ(out_seq.shape()[1], val_seqs.shape()[1] * context_length_); /// input and output has the same batch_size - CHECK_EQ(inputs[0].shape()[0], outputs[0].shape()[0]); + CHECK_EQ(val_seqs.shape()[0], out_seq.shape()[0]); + /// dim of input == dim of weight + if (2 == inputs.size()) { + CHECK_EQ(val_seqs.shape()[1], inputs[1].shape()[1]); + } - CHECK_EQ(outputs[0].getArgType(), ADD_TO); - auto out_mat = outputs[0].matrix(); - auto in_mat = inputs[0].matrix(); - auto w_mat = !inputs[1].data() - ? typename Tensor::Matrix(nullptr, 0, 0) - : inputs[1].matrix(); - auto seq_vec = inputs[2].vector(); + CHECK_EQ(out_seq.getArgType(), ADD_TO); + auto out_mat = out_seq.matrix(); + const auto in_mat = val_seqs.matrix(); + const auto w_mat = + (2 == inputs.size()) + ? inputs[1].matrix() + : typename Tensor::Matrix(nullptr, 0, 0); + const auto seq_vec = val_seqs.getSequenceId().vector(); ContextProjectionForward(out_mat, in_mat, w_mat, @@ -122,8 +154,12 @@ private: size_t begin_pad_; }; +/** + * Context Projection Backward with CPU Matrix Device. + * + */ template <> -void ContextProjectionBackward(CpuMatrix& out_grad_mat, +void ContextProjectionBackward(const CpuMatrix& out_grad_mat, CpuMatrix& in_grad_mat, CpuMatrix& w_grad_mat, const CpuIVector& seq_vec, @@ -146,7 +182,8 @@ void ContextProjectionBackward(CpuMatrix& out_grad_mat, int64_t pad_size = std::min(starts[i] - begin, starts[i + 1] - starts[i]); if (is_padding && w_grad_mat) { - MatrixPtr mat = out_grad_mat.subMatrix(starts[i], pad_size); + MatrixPtr mat = const_cast(out_grad_mat) + .subMatrix(starts[i], pad_size); MatrixPtr sub = w_grad_mat.subMatrix(j, pad_size); sub->addAtOffset(*mat, j * input_dim); } @@ -157,8 +194,8 @@ void ContextProjectionBackward(CpuMatrix& out_grad_mat, int64_t pad_size = std::min(end - starts[i + 1], starts[i + 1] - starts[i]); if (is_padding && w_grad_mat) { - MatrixPtr mat = - out_grad_mat.subMatrix(starts[i + 1] - pad_size, pad_size); + MatrixPtr mat = const_cast(out_grad_mat) + .subMatrix(starts[i + 1] - pad_size, pad_size); MatrixPtr sub = w_grad_mat.subMatrix( begin_pad + context_start + j - pad_size, pad_size); sub->addAtOffset(*mat, j * input_dim); @@ -169,17 +206,22 @@ void ContextProjectionBackward(CpuMatrix& out_grad_mat, if (end <= begin) continue; if (!in_grad_mat) continue; MatrixPtr src = in_grad_mat.subMatrix(begin, end - begin); - MatrixPtr dst = out_grad_mat.subMatrix(dst_begin, dst_end - dst_begin); + MatrixPtr dst = const_cast(out_grad_mat) + .subMatrix(dst_begin, dst_end - dst_begin); src->addAtOffset(*dst, j * input_dim); } } } /** - * \param inputs[0] input grad. - * \param inputs[1] weight grad. - * \param inputs[2] input sequence. - * \param outputs[0] output value. + * Context Projection Backward Function. + * Update the weight gradient and input layer gradient with backprop + * + * \param inputs[0].matrix output layer grad, n * (d * l) + * \param inputs[0].vector start position sequence, n * 1 + * \param outputs[0].matrix input layer grad, n * d + * \param outputs[0].vector start position sequence, n * 1 + * \param outputs[1] weight grad, pad * d */ template class ContextProjectionBackwardFunc : public FunctionBase { @@ -193,32 +235,36 @@ public: } void calc(const BufferArgs& inputs, const BufferArgs& outputs) override { - CHECK_EQ((size_t)3, inputs.size()); - CHECK_EQ((size_t)1, outputs.size()); + CHECK_EQ((size_t)1, inputs.size()); + CHECK_EQ((size_t)2, outputs.size()); + CHECK(inputs[0].isSequenceArg() && outputs[0].isSequenceArg()) + << "SequenceArg required here"; + const auto in_seq = dynamic_cast(inputs[0]); + auto out_seq = dynamic_cast(outputs[0]); + CHECK(in_seq.data() && in_seq.getSequenceId().data()); + CHECK_EQ(in_seq.shape().ndims(), (size_t)2); + CHECK_EQ(in_seq.getSequenceId().shape().ndims(), (size_t)1); + CHECK_EQ(out_seq.shape().ndims(), (size_t)2); + CHECK_EQ(out_seq.getSequenceId().shape().ndims(), (size_t)1); + CHECK_EQ(outputs[1].shape().ndims(), (size_t)2); - CHECK(outputs[0].data() && inputs[2].data()); - CHECK_EQ(outputs[0].shape().ndims(), (size_t)2); - CHECK_EQ(inputs[0].shape().ndims(), (size_t)2); - CHECK_EQ(inputs[1].shape().ndims(), (size_t)2); - CHECK_EQ(inputs[2].shape().ndims(), (size_t)1); + /// dim of input grad == dim of weight + CHECK_EQ(out_seq.shape()[1], outputs[1].shape()[1]); + /// input and output grad has the same batch_size + CHECK_EQ(out_seq.shape()[0], in_seq.shape()[0]); + /// dim of output grad = dim of input grad * context_length + CHECK_EQ(in_seq.shape()[1], out_seq.shape()[1] * context_length_); + CHECK_EQ(out_seq.getArgType(), ADD_TO); + CHECK_EQ(outputs[1].getArgType(), ADD_TO); - /// dim of input == dim of weight - CHECK_EQ(inputs[0].shape()[1], inputs[1].shape()[1]); - /// input and output has the same batch_size - CHECK_EQ(inputs[0].shape()[0], outputs[0].shape()[0]); - /// dim of output = dim of input * context_length - CHECK_EQ(outputs[0].shape()[1], inputs[0].shape()[1] * context_length_); - - CHECK_EQ(outputs[0].getArgType(), ADD_TO); - - auto out_grad_mat = outputs[0].matrix(); + const auto seq_vec = in_seq.getSequenceId().vector(); + const auto out_grad_mat = in_seq.matrix(); auto in_grad_mat = - !inputs[0].data() ? typename Tensor::Matrix(nullptr, 0, 0) - : inputs[0].matrix(); - auto w_grad_mat = !inputs[1].data() + !out_seq.data() ? typename Tensor::Matrix(nullptr, 0, 0) + : out_seq.matrix(); + auto w_grad_mat = !outputs[1].data() ? typename Tensor::Matrix(nullptr, 0, 0) - : inputs[1].matrix(); - auto seq_vec = inputs[2].vector(); + : outputs[1].matrix(); ContextProjectionBackward(out_grad_mat, in_grad_mat, w_grad_mat, @@ -238,11 +284,16 @@ private: size_t total_pad_; }; -#if 0 /** - * \param inputs[0] input grad. - * \param inputs[1] input sequence. - * \param outputs[0] output grad. + * Context Projection Backward Data Function + * Update input layer grad + * input: sequence of output layer grad + * output: sequence of input layer grad + * + * \param outputs[0].matrix input layer grad, n * d + * \param outputs[0].vector start position sequence, n * 1 + * \param inputs[0].matrix output layer grad, n * (d * l) + * \param inputs[0].vector start positon sequence, n * 1 */ template class ContextProjectionBackwardDataFunc : public FunctionBase { @@ -252,32 +303,30 @@ public: context_start_ = config.get("context_start"); } - void calc(const Arguments& inputs, - const Arguments& outputs, - const Arguments& inouts) override { - CHECK_EQ(2, static_cast(inputs.size())); + void calc(const BufferArgs& inputs, const BufferArgs& outputs) override { + CHECK_EQ(1, static_cast(inputs.size())); CHECK_EQ(1, static_cast(outputs.size())); - CHECK_EQ(0, static_cast(inouts.size())); - CHECK(inputs[0].getData() && outputs[0].getData() && inputs[1].getData()); - CHECK_EQ(static_cast(outputs[0].dims_.size()), 2); - CHECK_EQ(static_cast(inputs[0].dims_.size()), 2); - CHECK_EQ(static_cast(inputs[1].dims_.size()), 1); - CHECK_EQ(outputs[0].dims_[1], inputs[0].dims_[1] * context_length_); + CHECK(inputs[0].isSequenceArg() && outputs[0].isSequenceArg()) + << "SequenceArg required here"; + const auto in_seq = dynamic_cast(inputs[0]); + const auto out_seq = dynamic_cast(outputs[0]); + + CHECK(in_seq.data() && out_seq.data() && in_seq.getSequenceId().data()); + CHECK_EQ(static_cast(out_seq.shape().ndims()), 2); + CHECK_EQ(static_cast(in_seq.shape().ndims()), 2); + CHECK_EQ(static_cast(in_seq.getSequenceId().shape().ndims()), 1); + /// output layer grad dim == input layer grad dim * context_length_ + CHECK_EQ(in_seq.shape().ndims(), out_seq.shape().ndims() * context_length_); /// input and output has the same batch_size - CHECK_EQ(inputs[0].dims_[0], outputs[0].dims_[0]); + CHECK_EQ(in_seq.shape()[0], out_seq.shape()[0]); + CHECK_EQ(outputs[0].getArgType(), ASSIGN_TO); - auto out_grad_mat = std::make_shared::type>( - outputs[0].getData(), outputs[0].dims_[0], outputs[0].dims_[1]); - const auto in_grad_mat = std::make_shared::type>( - inputs[0].getData(), inputs[0].dims_[0], inputs[0].dims_[1]); - typename SequenceT::type seq_vec( - inputs[1].dims_[0], reinterpret_cast(inputs[1].getData())); + const auto out_grad_mat = in_seq.matrix(); + const auto seq_vec = in_seq.getSequenceId().vector(); + auto in_grad_mat = out_seq.matrix(); - ContextProjectionBackwardData(out_grad_mat.get(), - in_grad_mat.get(), - seq_vec, - context_length_, - context_start_); + ContextProjectionBackwardData( + out_grad_mat, in_grad_mat, seq_vec, context_length_, context_start_); } private: @@ -286,9 +335,14 @@ private: }; /** - * \param inputs[0] weight grad. - * \param inputs[1] input sequence. - * \param outputs[0] output grad. + * Context Projection Backward Weight Function + * Update weight grad by backprop + * input: sequence of output layer grad + * output: weight grad + * + * \param outputs[0] weight grad, pad * d + * \param inputs[0].matrix output layer grad, n * (d * l) + * \param inputs[0].vecotr start positon sequence, n * 1 */ template class ContextProjectionBackwardWeightFunc : public FunctionBase { @@ -300,28 +354,25 @@ public: total_pad_ = config.get("total_pad"); } - void calc(const Arguments& inputs, - const Arguments& outputs, - const Arguments& inouts) override { - CHECK_EQ(2, static_cast(inputs.size())); + void calc(const BufferArgs& inputs, const BufferArgs& outputs) override { + CHECK_EQ(1, static_cast(inputs.size())); CHECK_EQ(1, static_cast(outputs.size())); - CHECK_EQ(0, static_cast(inouts.size())); - - CHECK(inputs[0].getData() && outputs[0].getData() && inputs[1].getData()); - CHECK_EQ(static_cast(outputs[0].dims_.size()), 2); - CHECK_EQ(static_cast(inputs[0].dims_.size()), 2); - CHECK_EQ(static_cast(inputs[1].dims_.size()), 1); - CHECK_EQ(outputs[0].dims_[1], inputs[0].dims_[1] * context_length_); - - auto out_grad_mat = std::make_shared::type>( - outputs[0].getData(), outputs[0].dims_[0], outputs[0].dims_[1]); - auto w_grad_mat = std::make_shared::type>( - inputs[0].getData(), inputs[0].dims_[0], inputs[0].dims_[1]); - typename SequenceT::type seq_vec( - inputs[1].dims_[0], reinterpret_cast(inputs[1].getData())); + CHECK(inputs[0].isSequenceArg()) << "SequenceArg required here"; + const auto in_seq = dynamic_cast(inputs[0]); + CHECK(in_seq.data() && in_seq.getSequenceId().data() && outputs[0].data()); + CHECK_EQ(static_cast(outputs[0].shape().ndims()), 2); + CHECK_EQ(static_cast(in_seq.shape().ndims()), 2); + CHECK_EQ(static_cast(in_seq.getSequenceId().shape().ndims()), 1); + CHECK_EQ(in_seq.shape()[0], outputs[0].shape()[0]); + /// output layer grad dim == weight dim * context_length_ + CHECK_EQ(in_seq.shape()[1], outputs[0].shape()[1] * context_length_); + CHECK_EQ(outputs[0].getArgType(), ADD_TO); - ContextProjectionBackwardWeight(out_grad_mat.get(), - w_grad_mat.get(), + const auto seq_vec = in_seq.getSequenceId().vector(); + const auto out_grad_mat = in_seq.matrix(); + auto w_grad_mat = outputs[0].matrix(); + ContextProjectionBackwardWeight(out_grad_mat, + w_grad_mat, seq_vec, context_length_, context_start_, @@ -335,7 +386,6 @@ private: size_t begin_pad_; size_t total_pad_; }; -#endif REGISTER_TYPED_FUNC(ContextProjectionForward, CPU, @@ -350,7 +400,6 @@ REGISTER_TYPED_FUNC(ContextProjectionForward, REGISTER_TYPED_FUNC(ContextProjectionBackward, GPU, ContextProjectionBackwardFunc); -#if 0 REGISTER_TYPED_FUNC(ContextProjectionBackwardData, GPU, ContextProjectionBackwardDataFunc); @@ -358,5 +407,4 @@ REGISTER_TYPED_FUNC(ContextProjectionBackwardWeight, GPU, ContextProjectionBackwardWeightFunc); #endif -#endif } // namespace paddle diff --git a/paddle/function/ContextProjectionOp.h b/paddle/function/ContextProjectionOp.h index a558df5e072f2f4dcc5c45afa385b3cf88872d26..2bdd47e4e9b02483c2c5af82bf00c4e55d68f93e 100644 --- a/paddle/function/ContextProjectionOp.h +++ b/paddle/function/ContextProjectionOp.h @@ -21,14 +21,14 @@ namespace paddle { /** * \brief Context Projection Forward. * - * \param[out] outputs output data. - * \param[in] input input data. - * \param[in] weight input weight. - * \param[in] sequence input data. - * \param[in] context_length consecutive rows for concatenation. - * \param[in] context_start context start position. - * \param[in] begin_pad begining pad position. - * \param[in] is_padding whether padding 0 or not. + * \param[in/out] outputs output data. + * \param[in] input input data. + * \param[in] weight input weight. + * \param[in] sequence input data. + * \param[in] context_length consecutive rows for concatenation. + * \param[in] context_start context start position. + * \param[in] begin_pad begining pad position. + * \param[in] is_padding whether padding 0 or not. * */ template @@ -56,7 +56,7 @@ void ContextProjectionForward( */ template void ContextProjectionBackward( - typename Tensor::Matrix& out_grad, + const typename Tensor::Matrix& out_grad, typename Tensor::Matrix& in_grad, typename Tensor::Matrix& w_grad, const typename Tensor::Vector& seq_vec, @@ -68,7 +68,7 @@ void ContextProjectionBackward( template void ContextProjectionBackwardData( - typename Tensor::Matrix& out_grad, + const typename Tensor::Matrix& out_grad, typename Tensor::Matrix& in_grad, const typename Tensor::Vector& sequence, size_t context_length, @@ -76,7 +76,7 @@ void ContextProjectionBackwardData( template void ContextProjectionBackwardWeight( - typename Tensor::Matrix& out_grad, + const typename Tensor::Matrix& out_grad, typename Tensor::Matrix& w_grad, const typename Tensor::Vector& seq_vec, size_t context_length, diff --git a/paddle/function/ContextProjectionOpGpu.cu b/paddle/function/ContextProjectionOpGpu.cu index 6a4a01a6510416fc1f945305203f55ece7a28f11..1a5b4042402df3081a493962a5e080d72b7f40b2 100644 --- a/paddle/function/ContextProjectionOpGpu.cu +++ b/paddle/function/ContextProjectionOpGpu.cu @@ -138,10 +138,10 @@ void ContextProjectionForward(GpuMatrix& output, begin_pad); } -__global__ void KeContextProjectionBackwardData(real* out_grad, +__global__ void KeContextProjectionBackwardData(const real* out_grad, const int* sequence, real* in_grad, - int input_dim, + size_t input_dim, int context_length, int context_start) { int idx = threadIdx.x; @@ -152,7 +152,8 @@ __global__ void KeContextProjectionBackwardData(real* out_grad, real value = 0; int instances = seq_end - seq_start + context_length - 1; - out_grad += seq_start * input_dim * context_length; + auto out = const_cast(out_grad); + out += seq_start * input_dim * context_length; in_grad += seq_start * input_dim; for (int k = 0; k <= input_dim / block_size; k++) { if (idx < input_dim) { @@ -169,7 +170,7 @@ __global__ void KeContextProjectionBackwardData(real* out_grad, int outx = (i - context_length) < 0 ? i : (context_length - 1); int outy = (i - context_length) < 0 ? 0 : (i - (context_length - 1)); real* output_r = - out_grad + outy * input_dim * context_length + outx * input_dim; + out + outy * input_dim * context_length + outx * input_dim; for (int j = outy; j < seq_end - seq_start; j++) { value += output_r[idx]; if (j - outy == outx) break; @@ -194,7 +195,7 @@ __global__ void KeContextProjectionBackwardData(real* out_grad, * @param[in] context_start context start. * */ -void hl_context_projection_backward_data(real* out_grad, +void hl_context_projection_backward_data(const real* out_grad, const int* sequence, real* input_grad, size_t num_sequences, @@ -216,7 +217,7 @@ void hl_context_projection_backward_data(real* out_grad, } template <> -void ContextProjectionBackwardData(GpuMatrix& out_grad, +void ContextProjectionBackwardData(const GpuMatrix& out_grad, GpuMatrix& in_grad, const GpuIVector& sequence, size_t context_length, @@ -231,7 +232,7 @@ void ContextProjectionBackwardData(GpuMatrix& out_grad, } template -__global__ void KeContextProjectionBackwardWeight(real* out_grad, +__global__ void KeContextProjectionBackwardWeight(const real* out_grad, const int* sequence, real* w_grad, int num_sequences, @@ -254,7 +255,8 @@ __global__ void KeContextProjectionBackwardWeight(real* out_grad, for (int seqId = idy; seqId < num_sequences; seqId += THREADS_Y) { int seq_start = sequence[seqId]; int seq_end = sequence[seqId+1]; - output_r = out_grad + seq_start * w_dim * context_length; + output_r = const_cast(out_grad) + + seq_start * w_dim * context_length; if (context_start < 0) { if (padId + context_start < 0) { @@ -318,7 +320,7 @@ __global__ void KeContextProjectionBackwardWeight(real* out_grad, * beginning. * */ -void hl_context_projection_backward_weight(real* out_grad, +void hl_context_projection_backward_weight(const real* out_grad, const int* sequence, real* w_grad, size_t num_sequences, @@ -346,7 +348,7 @@ void hl_context_projection_backward_weight(real* out_grad, template <> void ContextProjectionBackwardWeight( - GpuMatrix& out_grad, + const GpuMatrix& out_grad, GpuMatrix& w_grad, const GpuIVector& seq_vec, size_t context_length, @@ -365,7 +367,7 @@ void ContextProjectionBackwardWeight( } template <> -void ContextProjectionBackward(GpuMatrix& out_grad, +void ContextProjectionBackward(const GpuMatrix& out_grad, GpuMatrix& in_grad, GpuMatrix& w_grad, const GpuIVector& sequence, diff --git a/paddle/function/ContextProjectionOpTest.cpp b/paddle/function/ContextProjectionOpTest.cpp index 6223d2fd23ac3bbb4fbcf51d37d22feaf3b1330b..c9db2ff8008e0bb0fa04370fb7b3ecd7641d2062 100644 --- a/paddle/function/ContextProjectionOpTest.cpp +++ b/paddle/function/ContextProjectionOpTest.cpp @@ -56,22 +56,25 @@ void testMatrixProjectionForward(int context_start, cpu_out.randomizeUniform(); gpu_out.copyFrom(cpu_out); - compare.getCpuFunction()->calc( - {Tensor(cpu_in.getData(), Dims{batch_size, input_dim}), - Tensor(cpu_weight ? cpu_weight->getData() : nullptr, - Dims{pad, input_dim}), - Tensor(reinterpret_cast(cpu_seq->getData()), - Dims{cpu_seq->getSize()})}, - {Tensor(cpu_out.getData(), Dims{batch_size, input_dim * context_length})}, - {}); - compare.getGpuFunction()->calc( - {Tensor(gpu_in.getData(), Dims{batch_size, input_dim}), - Tensor(gpu_weight ? gpu_weight->getData() : nullptr, - Dims{pad, input_dim}), - Tensor(reinterpret_cast(gpu_seq->getData()), - Dims{gpu_seq->getSize()})}, - {Tensor(gpu_out.getData(), Dims{batch_size, input_dim * context_length})}, - {}); + BufferArgs cpu_inputs; + BufferArgs cpu_outputs; + cpu_inputs.addArg(cpu_in, *cpu_seq); + if (cpu_weight) { + cpu_inputs.addArg(*cpu_weight, *cpu_seq); + } + cpu_outputs.addArg(cpu_out, *cpu_seq, ADD_TO); + + compare.getCpuFunction()->calc(cpu_inputs, cpu_outputs); + + BufferArgs gpu_inputs; + BufferArgs gpu_outputs; + gpu_inputs.addArg(gpu_in, *gpu_seq); + if (gpu_weight) { + gpu_inputs.addArg(*gpu_weight, *gpu_seq); + } + gpu_outputs.addArg(gpu_out, *gpu_seq, ADD_TO); + + compare.getGpuFunction()->calc(gpu_inputs, gpu_outputs); autotest::TensorCheckEqual(cpu_out, gpu_out); } @@ -117,25 +120,23 @@ void testMatrixProjectionBackward(int context_start, gpu_w_grad->copyFrom(*cpu_w_grad); } - compare.getCpuFunction()->calc( - {Tensor(cpu_in_grad.getData(), Dims{batch_size, input_dim}), - Tensor(cpu_w_grad ? cpu_w_grad->getData() : nullptr, - Dims{pad, input_dim}), - Tensor(reinterpret_cast(cpu_seq->getData()), - Dims{cpu_seq->getSize()})}, - {Tensor(cpu_out_grad.getData(), - Dims{batch_size, input_dim * context_length})}, - {}); - - compare.getGpuFunction()->calc( - {Tensor(gpu_in_grad.getData(), Dims{batch_size, input_dim}), - Tensor(gpu_w_grad ? gpu_w_grad->getData() : nullptr, - Dims{pad, input_dim}), - Tensor(reinterpret_cast(gpu_seq->getData()), - Dims{gpu_seq->getSize()})}, - {Tensor(gpu_out_grad.getData(), - Dims{batch_size, input_dim * context_length})}, - {}); + BufferArgs cpu_inputs; + BufferArgs cpu_outputs; + cpu_inputs.addArg(cpu_out_grad, *cpu_seq); + cpu_outputs.addArg(cpu_in_grad, *cpu_seq, ADD_TO); + cpu_outputs.addArg( + cpu_w_grad ? *cpu_w_grad : CpuMatrix(nullptr, 0, input_dim), ADD_TO); + + compare.getCpuFunction()->calc(cpu_inputs, cpu_outputs); + + BufferArgs gpu_inputs; + BufferArgs gpu_outputs; + gpu_inputs.addArg(gpu_out_grad, *gpu_seq); + gpu_outputs.addArg(gpu_in_grad, *gpu_seq, ADD_TO); + gpu_outputs.addArg( + gpu_w_grad ? *gpu_w_grad : GpuMatrix(nullptr, 0, input_dim), ADD_TO); + + compare.getGpuFunction()->calc(gpu_inputs, gpu_outputs); autotest::TensorCheckErr(cpu_in_grad, gpu_in_grad); if (is_padding) { diff --git a/paddle/function/CrossMapNormalOp.cpp b/paddle/function/CrossMapNormalOp.cpp index 92980c503fdaaaa9ac600070197dba6ba4bfb7a4..5c0bdd933b1e4a62e49981798c56c70907d16424 100644 --- a/paddle/function/CrossMapNormalOp.cpp +++ b/paddle/function/CrossMapNormalOp.cpp @@ -112,11 +112,51 @@ void CrossMapNormalGrad(real* inputsGrad, } /** - * \brief {o_0, o_1} = calc(i_0) + * \brief Normalization with across maps. * - * \param inputs[0] input value. - * \param outputs[0] output value. - * \param outputs[1] denoms. + * This Function comes from the paper + * "ImageNet Classification with Deep Convolutional Neural Networks". + * + * The original formula is: + * + * Input(i, x, y) + * Output(i, x, y) = ---------------------------------------------- + * -- upper + * (k + alpha * > (Input(j, x, y))^2) ^ (beta) + * -- j = lower + * + * upper is `min(C, c + N/2)` + * lower if `max(0, c - N/2)` + * + * Function implementation: + * + * inputs and outpus is NCHW format, while input.shape.ndims() is equal 4. + * And the meaning of each dimension(0-3) is respectively batch size, + * feature maps, rows and columns. + * + * Input and Output in the above formula is for each map(i) of one image, and + * Input(i, x, y), Output(i, x, y) represents an element in an image. + * + * C is the number of feature maps of one image, and N is a hyper-parameters + * is configured when Function is initialized. The sum in the denominator + * is the sum of the same position in the neighboring maps. + * + * In the implementation of Function, k is equal to 1, + * so Function has no argument for k. + * + * Function Arguments: + * + * \param size_ represent N + * \param scale_ represent alpha + * \param pow_ represent beta + * \param inputs[0] represent Input + * \param outputs[0] represent Output + * \param outputs[1] represent The denominator in the formula(except beta) + * + * Note: + * Save output[1] is to simplify the backward calculation. + * TODO, if only consider the forward calculation, we can optimize to + * remove the output[1]. */ template class CrossMapNormalFunc : public FunctionBase { @@ -161,13 +201,36 @@ private: }; /** - * \brief {o_0} = calc(i_0, i_1, i_2, i_3) + * \brief Backward calculation for normalization with across maps. + * + * Function implementation: + * + * The implementation of this Function is derived from the + * CrossMapNormalFunc implementation. + * + * InputGrad = OutputGrad * denoms ^ (-beta) + * -- upper + * + > (OutputGrad * OutputValue * (-2 * alpha * beta) / denoms) * InputValue + * -- lower * - * \param inputs[0] input value. - * \param inputs[1] output value. - * \param inputs[2] output grad. - * \param inputs[3] denoms. - * \param outputs[0] input grad. + * The data of inputs/outputs format is the same as the forward interface + * and is NCHW. + * + * The upper and lower is the same as forward. The logic of the sum + * is also the same as forward. + * + * Function Arguments: + * + * \param size_ represent N + * \param scale_ represent alpha + * \param pow_ represent beta + * \param inputs[0] represent InputValue, inputs[0] of CrossMapNormalFunc + * \param inputs[1] represent OutputValue, outputs[0] of CrossMapNormalFunc + * \param inputs[2] represent OutputGrad + * \param inputs[3] represent denoms, outputs[1] of CrossMapNormalFunc + * This is the intermediate result that is + * preserved in the forward calculation. + * \param outputs[0] represent InputGrad */ template class CrossMapNormalGradFunc : public FunctionBase { @@ -188,8 +251,13 @@ public: CHECK(inputs[0].shape() == inputs[3].shape()); CHECK(inputs[0].shape() == outputs[0].shape()); - // TODO(hedaoyuan): need support ASSIGN_TO mode. - CHECK_EQ(outputs[0].getArgType(), ADD_TO); + if (outputs[0].getArgType() != ADD_TO) { + // Currently, some algorithm implementations are ASSIGN_TO mode, + // if need to support the ADD_TO calculation, need to clear the output. + typename Tensor::Vector tmp( + outputs[0].shape().getElements(), outputs[0].data()); + tmp.zero(); + } size_t samples = inputs[0].shape()[0]; size_t channels = inputs[0].shape()[1]; diff --git a/paddle/function/CrossMapNormalOpTest.cpp b/paddle/function/CrossMapNormalOpTest.cpp index d65d9310affd7c9b7fee3118c79449870849c243..51f5da81bfc9ae870ac9949ba74da01a9449a04d 100644 --- a/paddle/function/CrossMapNormalOpTest.cpp +++ b/paddle/function/CrossMapNormalOpTest.cpp @@ -27,15 +27,19 @@ TEST(CrossMapNormal, real) { << " imgSizeH=" << imgSizeH << " imgSizeW=" << imgSizeW << " size=" << size; - FunctionCompare compare("CrossMapNormal", - FuncConfig() - .set("size", size) - .set("scale", (real)1.5) - .set("pow", (real)0.5)); - Dims dims{numSamples, channels, imgSizeH, imgSizeW}; - compare.cmpWithArg({Tensor(nullptr, dims)}, - {Tensor(nullptr, dims), Tensor(nullptr, dims)}, - {}); + // init Test object + FunctionCompare test("CrossMapNormal", + FuncConfig() + .set("size", size) + .set("scale", (real)1.5) + .set("pow", (real)0.5)); + // prepare input arguments + TensorShape shape{numSamples, channels, imgSizeH, imgSizeW}; + test.addInputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + test.addOutputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + test.addOutputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + // run Function + test.run(); } } } @@ -53,18 +57,19 @@ TEST(CrossMapNormalGrad, real) { << " imgSizeH=" << imgSizeH << " imgSizeW=" << imgSizeW << " size=" << size; - FunctionCompare compare("CrossMapNormalGrad", - FuncConfig() - .set("size", size) - .set("scale", (real)1.5) - .set("pow", (real)0.5)); - Dims dims{numSamples, channels, imgSizeH, imgSizeW}; - compare.cmpWithArg({Tensor(nullptr, dims), - Tensor(nullptr, dims), - Tensor(nullptr, dims), - Tensor(nullptr, dims)}, - {Tensor(nullptr, dims)}, - {}); + FunctionCompare test("CrossMapNormalGrad", + FuncConfig() + .set("size", size) + .set("scale", (real)1.5) + .set("pow", (real)0.5)); + TensorShape shape{numSamples, channels, imgSizeH, imgSizeW}; + test.addInputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + test.addInputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + test.addInputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + test.addInputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + test.addOutputs(BufferArg(VALUE_TYPE_FLOAT, shape)); + // run Function + test.run(); } } } diff --git a/paddle/function/Function.cpp b/paddle/function/Function.cpp index dbe3a4e9f608df6333a5637f2d962a555b04d7c3..f47d55a4ade97d76e0f1940a2234e34e20efade6 100644 --- a/paddle/function/Function.cpp +++ b/paddle/function/Function.cpp @@ -79,15 +79,25 @@ FuncConfig& FuncConfig::set(const std::string& key, bool v) { void BufferArgs::addArg(const Matrix& arg, const TensorShape& shape, ArgType argType) { - args_.push_back(std::make_shared(arg, shape, argType)); + _args_.push_back(new BufferArg(arg, shape, argType)); + addArg(*_args_.back()); } void BufferArgs::addArg(const CpuSparseMatrix& arg, ArgType argType) { - args_.push_back(std::make_shared(arg, argType)); + _args_.push_back(new SparseMatrixArg(arg, argType)); + addArg(*_args_.back()); } void BufferArgs::addArg(const GpuSparseMatrix& arg, ArgType argType) { - args_.push_back(std::make_shared(arg, argType)); + _args_.push_back(new SparseMatrixArg(arg, argType)); + addArg(*_args_.back()); +} + +void BufferArgs::addArg(const Matrix& matrix, + const IVector& vector, + ArgType argType) { + _args_.push_back(new SequenceArg(matrix, vector, argType)); + addArg(*_args_.back()); } ClassRegistrar FunctionBase::funcRegistrar_; diff --git a/paddle/function/Function.h b/paddle/function/Function.h index 249f8f9cfad58bf596e8cdce9188409b5690f969..9215c137eb8e85a9a03575104d7f89bbce441eba 100644 --- a/paddle/function/Function.h +++ b/paddle/function/Function.h @@ -50,19 +50,44 @@ protected: * Argument type for Function::calc(). * A BufferArgs contains a set of BufferArg, * because Function can have multiple inputs and outputs. + * + * addArg() with Matix object used to adapt Layer Argument. + * Will create a BufferArg object in addArg(), + * and free in destructor of BufferArgs. + * + * addArg() with BufferArg object, just save BufferArg object address, + * and the caller needs to guarantee the validity of the BufferArg object + * in the BufferArgs life time. */ class BufferArgs { public: BufferArgs() {} + + ~BufferArgs() { + for (auto arg : _args_) { + delete arg; + } + } + size_t size() const { return args_.size(); } // add argument into BufferArgs // Tensor can be Matrix, Vector, IVector. // For inputs, do not need argType. // For outputs, the argType needs to be specified as ASSIGN_TO or ADD_TO. - template - void addArg(const Tensor& arg, ArgType argType = UNSPECIFIED) { - args_.push_back(std::make_shared(arg, argType)); + void addArg(const Matrix& arg, ArgType argType = UNSPECIFIED) { + _args_.push_back(new BufferArg(arg, argType)); + addArg(*_args_.back()); + } + + void addArg(const Vector& arg, ArgType argType = UNSPECIFIED) { + _args_.push_back(new BufferArg(arg, argType)); + addArg(*_args_.back()); + } + + void addArg(const IVector& arg, ArgType argType = UNSPECIFIED) { + _args_.push_back(new BufferArg(arg, argType)); + addArg(*_args_.back()); } // Add arg into BufferArgs and reshape the arg. @@ -77,20 +102,37 @@ public: void addArg(const CpuSparseMatrix& arg, ArgType argType = UNSPECIFIED); void addArg(const GpuSparseMatrix& arg, ArgType argType = UNSPECIFIED); + void addArg(const Matrix& matrix, + const IVector& vector, + ArgType argType = UNSPECIFIED); + // get argument const BufferArg& operator[](size_t num) const { CHECK_LT(num, args_.size()); return *args_[num]; } + void addArg(BufferArg& arg) { args_.push_back(&arg); } + + void addArg(SequenceIdArg& arg) { args_.push_back(&arg); } + + void addArg(SequenceArg& arg) { args_.push_back(&arg); } + + void addArg(SparseMatrixArg& arg) { args_.push_back(&arg); } + private: - std::vector args_; + std::vector args_; + // The BufferArg object is constructed and freed by BufferArgs. + std::vector _args_; }; /** * \brief Base class for Function. * The basic Function implementation requires override init and calc interfaces. * + * The caller needs to ensure the validity of the arguments + * during Function execution. + * * Function inputs are readonly, Function outputs have two modes: ASSIGN_TO * and ADD_TO. * If output.getArgType() == ASSIGN_TO, this is assign mode, and the calculation diff --git a/paddle/function/FunctionTest.cpp b/paddle/function/FunctionTest.cpp index 7ce908320a6f6f764e8fdacc96432aca78d7b2df..fdf7e631e5ab8c67eb5cf906bd0af49740d60112 100644 --- a/paddle/function/FunctionTest.cpp +++ b/paddle/function/FunctionTest.cpp @@ -14,6 +14,7 @@ limitations under the License. */ #include "Function.h" #include +#include "paddle/math/SparseMatrix.h" namespace paddle { @@ -56,4 +57,110 @@ TEST(Function, BufferArgs) { Function(gpuArgments); } +/** + * Some tests case are used to check the consistency between the BufferArg type + * argument received by Function and the original type argument. + * + * Use Case: + * TEST() { + * Matrix matrix(...); + * CheckBufferArg lambda = [=](const BufferArg& arg) { + * // check matrix and arg are equivalent + * EXPECT_EQ(matrix, arg); + * } + * + * BufferArgs argments{matrix...}; + * std::vector checkFunc{lambda...}; + * testBufferArgs(argments, checkFunc); + * } + */ +typedef std::function CheckBufferArg; + +void testBufferArgs(const BufferArgs& inputs, + const std::vector& check) { + EXPECT_EQ(inputs.size(), check.size()); + for (size_t i = 0; i < inputs.size(); i++) { + check[i](inputs[i]); + } +} + +void testBufferArgs(const BufferArgs& inputs, const CheckBufferArg& check) { + EXPECT_EQ(inputs.size(), 1); + check(inputs[0]); +} + +TEST(Arguments, Matrix) { + MatrixPtr matrix = Matrix::create(100, 200); + CheckBufferArg check = [=](const BufferArg& arg) { + EXPECT_EQ(arg.shape().ndims(), 2); + EXPECT_EQ(arg.shape()[0], 100); + EXPECT_EQ(arg.shape()[1], 200); + EXPECT_EQ(arg.data(), matrix->getData()); + + EXPECT_EQ(arg.matrix().getHeight(), matrix->getHeight()); + EXPECT_EQ(arg.matrix().getWidth(), matrix->getWidth()); + EXPECT_EQ(arg.matrix().getData(), matrix->getData()); + }; + + BufferArgs argments; + argments.addArg(*matrix); + std::vector checkFunc; + checkFunc.push_back(check); + testBufferArgs(argments, checkFunc); +} + +TEST(Arguments, Vector) { + VectorPtr vector = Vector::create(100, false); + CheckBufferArg check = [=](const BufferArg& arg) { + EXPECT_EQ(arg.shape().ndims(), 1); + EXPECT_EQ(arg.shape()[0], 100); + EXPECT_EQ(arg.data(), vector->getData()); + + CpuVector inVector = arg.vector(); + EXPECT_EQ(inVector.getSize(), vector->getSize()); + EXPECT_EQ(inVector.getData(), vector->getData()); + }; + + BufferArgs argments; + argments.addArg(*vector); + std::vector checkFunc; + checkFunc.push_back(check); + testBufferArgs(argments, checkFunc); +} + +TEST(Arguments, CpuSparseMatrix) { + CpuSparseMatrix sparse(200, 300, 50); + CheckBufferArg check = [=](const BufferArg& arg) { + EXPECT_EQ(arg.shape().ndims(), 2); + EXPECT_EQ(arg.shape()[0], 200); + EXPECT_EQ(arg.shape()[1], 300); + EXPECT_EQ(arg.data(), sparse.getData()); + // CHECK_EQ(arg.sparse().nnz(), 50); + // CHECK_EQ(arg.sparse().dataFormat(), SPARSE_CSR_FORMAT); + // CHECK_EQ(arg.sparse().dataType(), SPARSE_FLOAT_VALUE); + EXPECT_EQ(arg.sparse().getRowBuf(), sparse.getRows()); + EXPECT_EQ(arg.sparse().getColBuf(), sparse.getCols()); + }; + + BufferArgs argments; + argments.addArg(sparse); + std::vector checkFunc; + checkFunc.push_back(check); + testBufferArgs(argments, checkFunc); +} + +TEST(Arguments, BufferArg) { + BufferArg arg(nullptr, VALUE_TYPE_FLOAT, {1, 2, 3}); + CheckBufferArg check = [=](const BufferArg& arg) { + EXPECT_EQ(arg.shape().ndims(), 3); + EXPECT_EQ(arg.shape()[0], 1); + EXPECT_EQ(arg.shape()[1], 2); + EXPECT_EQ(arg.shape()[2], 3); + }; + + BufferArgs argments; + argments.addArg(arg); + testBufferArgs(argments, check); +} + } // namespace paddle diff --git a/paddle/function/FunctionTest.h b/paddle/function/FunctionTest.h index 32131037f6de4a9f7a3ebf8f5773eccd65dc2cdb..24e7a36a43cfa8941535cb778aa1557ec5a0a6f4 100644 --- a/paddle/function/FunctionTest.h +++ b/paddle/function/FunctionTest.h @@ -15,95 +15,186 @@ limitations under the License. */ #include "Function.h" #include "paddle/math/Vector.h" #include "paddle/math/tests/TensorCheck.h" +#include "paddle/testing/TestUtil.h" namespace paddle { +typedef std::shared_ptr BufferArgPtr; + +/** + * \brief A class for comparing CPU and GPU implementations of Function. + * + * + * Use case: + * // Initializes a test object, the corresponding cpu and gpu Function + * // are constructed according to FunctionName and FuncConfig. + * FunctionCompare test(FunctionName, FuncConfig); + * // Prepare inputs and outputs arguments. + * // Here the input and output can not contain real data, + * // only contains the argument type and shape. + * test.addInputs(input1); + * test.addInputs(input2); + * test.addOutputs(output1); + * test.addOutputs(output2); + * // Run. + * // Will according to the type and shape of arguments(inputs_/outputs_), + * // automatic initialization cpu and gpu function required arguments + * // (cpuInputs_/cpuOutputs_/gpuInputs_/gpuOutputs_). + * // Call the CPU and GPU Function calculation results. + * // Compares CPU and GPU calculation results for consistency. + * test.run(); + */ class FunctionCompare { public: FunctionCompare(const std::string& name, const FuncConfig& config) - : cpu(FunctionBase::funcRegistrar_.createByType(name + "-CPU")), - gpu(FunctionBase::funcRegistrar_.createByType(name + "-GPU")) { - cpu->init(config); - gpu->init(config); + : cpuFunc_(FunctionBase::funcRegistrar_.createByType(name + "-CPU")), + gpuFunc_(FunctionBase::funcRegistrar_.createByType(name + "-GPU")) { + cpuFunc_->init(config); + gpuFunc_->init(config); + } + + ~FunctionCompare() {} + + // input need only contains shape, do not contains data. + void addInputs(const BufferArg& input) { + size_t size = + input.shape().getElements() * sizeOfValuType(input.valueType()); + cpuMemory_.emplace_back(std::make_shared(size)); + gpuMemory_.emplace_back(std::make_shared(size)); + + cpuInputs_.emplace_back(std::make_shared( + cpuMemory_.back()->getBuf(), input.valueType(), input.shape())); + gpuInputs_.emplace_back(std::make_shared( + gpuMemory_.back()->getBuf(), input.valueType(), input.shape())); + } + + // output need only contains shape, do not contains data. + void addOutputs(const BufferArg& output) { + size_t size = + output.shape().getElements() * sizeOfValuType(output.valueType()); + cpuMemory_.emplace_back(std::make_shared(size)); + gpuMemory_.emplace_back(std::make_shared(size)); + + cpuOutputs_.emplace_back( + std::make_shared(cpuMemory_.back()->getBuf(), + output.valueType(), + output.shape(), + ASSIGN_TO)); + gpuOutputs_.emplace_back( + std::make_shared(gpuMemory_.back()->getBuf(), + output.valueType(), + output.shape(), + ASSIGN_TO)); } - void cmpWithArg(const Arguments& inputs, - const Arguments& outputs, - const Arguments& inouts) { - // init cpu and gpu arguments - auto initArgs = [=]( - Arguments& cpuArgs, Arguments& gpuArgs, const Arguments& inArgs) { - for (const auto arg : inArgs) { - size_t size = sizeof(real); - for (const auto dim : arg.dims_) { - size *= dim; - } - if (arg.getData()) { - // todo(tianbing), waste unnecessary mem here - cpuMemory.emplace_back(std::make_shared(size)); - gpuMemory.emplace_back(std::make_shared(size)); - cpuArgs.emplace_back(Tensor((real*)arg.getData(), arg.dims_)); - gpuArgs.emplace_back(Tensor((real*)arg.getData(), arg.dims_)); - // already init outside - } else { - cpuMemory.emplace_back(std::make_shared(size)); - gpuMemory.emplace_back(std::make_shared(size)); - cpuArgs.emplace_back( - Tensor((real*)cpuMemory.back()->getBuf(), arg.dims_)); - gpuArgs.emplace_back( - Tensor((real*)gpuMemory.back()->getBuf(), arg.dims_)); - // will use an api to refactor this code. - CpuVector cpuVector(size / sizeof(real), - (real*)cpuArgs.back().getData()); - GpuVector gpuVector(size / sizeof(real), - (real*)gpuArgs.back().getData()); - cpuVector.uniform(0.001, 1); - gpuVector.copyFrom(cpuVector); - } + void addInputs(const SequenceArg& input) { + size_t batchSize = input.shape()[0]; + size_t numSeqs = batchSize / 10 + 1; + + size_t sizeId = (numSeqs + 1) * sizeOfValuType(VALUE_TYPE_INT32); + cpuMemory_.emplace_back(std::make_shared(sizeId)); + gpuMemory_.emplace_back(std::make_shared(sizeId)); + + TensorShape seqsId({numSeqs + 1}); + // void* cpuBuffer = cpuMemory_.back()->getBuf(); + // void* gpuBuffer = gpuMemory_.back()->getBuf(); + + size_t size = + input.shape().getElements() * sizeOfValuType(input.valueType()); + cpuMemory_.emplace_back(std::make_shared(size)); + gpuMemory_.emplace_back(std::make_shared(size)); + + // TODO: need be implemented. + } + + void run() { + // prepare cpu/gpu arguments + initInputs(); + + // function calculate + auto callFunction = [](FunctionBase* function, + std::vector& inputs, + std::vector& outputs) { + BufferArgs inArgs; + BufferArgs outArgs; + for (auto arg : inputs) { + inArgs.addArg(*arg); } + for (auto arg : outputs) { + outArgs.addArg(*arg); + } + function->calc(inArgs, outArgs); }; - initArgs(cpuInputs, gpuInputs, inputs); - initArgs(cpuOutputs, gpuOutputs, outputs); - initArgs(cpuInouts, gpuInouts, inouts); - // function calculate - cpu->calc(cpuInputs, cpuOutputs, cpuInouts); - gpu->calc(gpuInputs, gpuOutputs, gpuInouts); + callFunction(cpuFunc_.get(), cpuInputs_, cpuOutputs_); + callFunction(gpuFunc_.get(), gpuInputs_, gpuOutputs_); // check outputs and inouts - auto checkArgs = [=](const Arguments& cpuArgs, const Arguments& gpuArgs) { - for (size_t i = 0; i < cpuArgs.size(); i++) { - auto cpu = cpuArgs[i]; - auto gpu = gpuArgs[i]; - size_t size = 1; - for (auto dim : cpu.dims_) { - size *= dim; - } - CpuVector cpuVector(size, (real*)cpu.getData()); - GpuVector gpuVector(size, (real*)gpu.getData()); - - autotest::TensorCheckErr(cpuVector, gpuVector); - } - }; - checkArgs(cpuOutputs, gpuOutputs); - checkArgs(cpuInouts, gpuInouts); + compareOutputs(); + } + + std::shared_ptr getCpuFunction() const { return cpuFunc_; } + + std::shared_ptr getGpuFunction() const { return gpuFunc_; } + +protected: + void initInputs() { + for (size_t i = 0; i < cpuInputs_.size(); i++) { + initArg(*cpuInputs_[i]); + + // TODO: Need a BufferCopy used to copy from one BufferArg to another. + CpuVector cpuVector(cpuInputs_[i]->shape().getElements(), + (real*)cpuInputs_[i]->data()); + GpuVector gpuVector(gpuInputs_[i]->shape().getElements(), + (real*)gpuInputs_[i]->data()); + + gpuVector.copyFrom(cpuVector); + } + } + + void compareOutputs() { + for (size_t i = 0; i < cpuOutputs_.size(); i++) { + // TODO, Need a BufferCheck used to compare the two buffers. + auto cpu = cpuOutputs_[i]; + auto gpu = gpuOutputs_[i]; + CpuVector cpuVector(cpu->shape().getElements(), (real*)cpu->data()); + GpuVector gpuVector(cpu->shape().getElements(), (real*)gpu->data()); + + autotest::TensorCheckErr(cpuVector, gpuVector); + } } - std::shared_ptr getCpuFunction() const { return cpu; } + // only init cpu argument, gpu argument copy from cpu argument. + void initArg(BufferArg& arg) { + CpuVector vector(arg.shape().getElements(), (real*)arg.data()); + vector.uniform(0.001, 1); + } - std::shared_ptr getGpuFunction() const { return gpu; } + void initArg(SequenceIdArg& arg, size_t batchSize) { + size_t numSeqs = arg.numSeqs(); + int* buf = reinterpret_cast(arg.data()); + int pos = 0; + size_t maxLen = 2 * batchSize / numSeqs; + for (int i = 0; i < (int)numSeqs; ++i) { + int len = uniformRandom( + std::min(maxLen, batchSize - pos - numSeqs + i)) + + 1; + buf[i] = pos; + pos += len; + VLOG(1) << " len=" << len; + } + buf[numSeqs] = batchSize; + } protected: - std::shared_ptr cpu; - std::shared_ptr gpu; - std::vector cpuMemory; - std::vector gpuMemory; - Arguments cpuInputs; - Arguments cpuOutputs; - Arguments cpuInouts; - Arguments gpuInputs; - Arguments gpuOutputs; - Arguments gpuInouts; + std::shared_ptr cpuFunc_; + std::shared_ptr gpuFunc_; + std::vector cpuMemory_; + std::vector gpuMemory_; + std::vector cpuInputs_; + std::vector cpuOutputs_; + std::vector gpuInputs_; + std::vector gpuOutputs_; }; } // namespace paddle diff --git a/paddle/function/PadOp.cpp b/paddle/function/PadOp.cpp new file mode 100644 index 0000000000000000000000000000000000000000..f1a0d2a1a96f24ddff8cd120681a8bc8cddaf40a --- /dev/null +++ b/paddle/function/PadOp.cpp @@ -0,0 +1,223 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#include "PadOp.h" +#include "paddle/math/Vector.h" + +namespace paddle { + +template <> +void Pad(real* outputs, + const real* inputs, + const int num, + const int inC, + const int inH, + const int inW, + const PadConf& pad) { + int cstart = pad.channelStart, cend = pad.channelEnd; + int hstart = pad.heightStart, hend = pad.heightEnd; + int wstart = pad.widthStart, wend = pad.widthEnd; + int outC = inC + cstart + cend; + int outH = inH + hstart + hend; + int outW = inW + wstart + wend; + for (int i = 0; i < num; i++) { + for (int c = 0; c < inC; c++) { + for (int h = 0; h < inH; h++) { + int inoff = ((i * inC + c) * inH + h) * inW; + int outoff = + ((i * outC + c + cstart) * outH + h + hstart) * outW + wstart; + memcpy(outputs + outoff, inputs + inoff, inW * sizeof(real)); + } + } + } +} + +template <> +void PadGrad(real* inGrad, + const real* outGrad, + const int num, + const int inC, + const int inH, + const int inW, + const PadConf& pad) { + int cstart = pad.channelStart, cend = pad.channelEnd; + int hstart = pad.heightStart, hend = pad.heightEnd; + int wstart = pad.widthStart, wend = pad.widthEnd; + int outC = inC + cstart + cend; + int outH = inH + hstart + hend; + int outW = inW + wstart + wend; + for (int i = 0; i < num; i++) { + for (int c = 0; c < inC; c++) { + for (int h = 0; h < inH; h++) { + int inoff = ((i * inC + c) * inH + h) * inW; + int outoff = + ((i * outC + c + cstart) * outH + h + hstart) * outW + wstart; + CpuVector inG = CpuVector(inW, inGrad + inoff); + CpuVector outG = CpuVector(inW, const_cast(outGrad + outoff)); + inG += outG; + } + } + } +} + +/** + * \brief Padding zeros to input according to the specify dimension. + * The struct pad_ contains the padding size in each dimension. + * The input and output is a 4D tensor. In PadFunc, we only + * pad zeros to the 2nd to 4th dimension. + * + * Argument in this Function: + * \param pad_ A struct object contains the padding size in each dimension. + * It has six integers. The channelStart and channelEnd indicate + * how many zeros to add before and after the input in channel + * dimension. And the heightStart and heightEnd indicate padding + * in height dimension. The widthStart and widthEnd indicate the + * padding in width dimension. + * \param inputs A 4D tensor, only one input. + * \param outputs A 4D tensor, the output value after padding. + * + * For example, + * Input(2,2,2,3) = [ + * [ [[1,2,3], [3,4,5]], + * [[2,3,5], [1,6,7]] ], + * [ [[4,3,1], [1,8,7]], + * [[3,8,9], [2,3,5]] ] + * ] # the shape is (1,2,2,3) + * + * pad_: if channelStart = channelEnd = 1, others are 0. + * Output(2,4,2,3) = [ + * [ [[0,0,0], [0,0,0]], + * [[1,2,3], [3,4,5]], + * [[2,3,5], [1,6,7]], + * [[0,0,0], [0,0,0]] ], + * [ [[0,0,0], [0,0,0]], + * [[4,3,1], [1,8,7]], + * [[3,8,9], [2,3,5]], + * [[0,0,0], [0,0,0]] ] + * ] # the shape is (2,4,2,3) + * + * pad_: if widthStart = 1, widthEnd = 2, others are 0. + * Output(2,2,2,6) = [ + * [ [[0,1,2,3,0,0], [0,3,4,5,0,0]], + * [[0,2,3,5,0,0], [0,1,6,7,0,0]] ], + * [ [[0,4,3,1,0,0], [0,1,8,7,0,0]], + * [[0,3,8,9,0,0], [0,2,3,5,0,0]] ], + * ] # the shape is (2,2,2,6) + * + * pad_: if heightStart = 1, heightEnd = 1, others are 0. + * Output(2,2,4,3) = [ + * [ [[0,0,0], [1,2,3], [3,4,5], [0,0,0]], + * [[0,0,0], [2,3,5], [1,6,7], [0,0,0]] ], + * [ [[0,0,0], [4,3,1], [1,8,7], [0,0,0]], + * [[0,0,0], [3,8,9], [2,3,5], [0,0,0]] ], + * ] # the shape is (2,2,4,3) + */ + +template +class PadFunc : public FunctionBase { +public: + void init(const FuncConfig& config) override { + pad_.channelStart = config.get("cstart"); + pad_.channelEnd = config.get("cend"); + pad_.heightStart = config.get("hstart"); + pad_.heightEnd = config.get("hend"); + pad_.widthStart = config.get("wstart"); + pad_.widthEnd = config.get("wend"); + } + + void calc(const BufferArgs& inputs, const BufferArgs& outputs) override { + CHECK_EQ(1UL, inputs.size()); + CHECK_EQ(1UL, outputs.size()); + CHECK_EQ(outputs[0].getArgType(), ASSIGN_TO); + + size_t num = inputs[0].shape()[0]; + size_t inC = inputs[0].shape()[1]; + size_t inH = inputs[0].shape()[2]; + size_t inW = inputs[0].shape()[3]; + typename Tensor::Vector vec(outputs[0].shape().getElements(), + outputs[0].data()); + vec.zero(); + + Pad(outputs[0].data(), + inputs[0].data(), + num, + inC, + inH, + inW, + pad_); + } + +private: + PadConf pad_; +}; + +/** + * \brief The backward propagation of padding Function. Remove the elements + * in the padding positions of forward. + * + * Argument in this Function: + * \param pad_ The same meaning as it in PadFunc. + * \param inputs The gradient with respect to the output value of PadFunc. + * \param outputs The gradient with respect to the input value of PadFunc. + */ + +template +class PadGradFunc : public FunctionBase { +public: + void init(const FuncConfig& config) override { + pad_.channelStart = config.get("cstart"); + pad_.channelEnd = config.get("cend"); + pad_.heightStart = config.get("hstart"); + pad_.heightEnd = config.get("hend"); + pad_.widthStart = config.get("wstart"); + pad_.widthEnd = config.get("wend"); + } + + void calc(const BufferArgs& inputs, const BufferArgs& outputs) override { + CHECK_EQ(1UL, inputs.size()); + CHECK_EQ(1UL, outputs.size()); + + size_t num = outputs[0].shape()[0]; + size_t inC = outputs[0].shape()[1]; + size_t inH = outputs[0].shape()[2]; + size_t inW = outputs[0].shape()[3]; + + if (outputs[0].getArgType() != ADD_TO) { + // for unit test + typename Tensor::Vector tmp( + outputs[0].shape().getElements(), outputs[0].data()); + tmp.zero(); + } + + PadGrad(outputs[0].data(), + inputs[0].data(), + num, + inC, + inH, + inW, + pad_); + } + +private: + PadConf pad_; +}; + +REGISTER_TYPED_FUNC(Pad, CPU, PadFunc); +REGISTER_TYPED_FUNC(PadGrad, CPU, PadGradFunc); +#ifndef PADDLE_ONLY_CPU +REGISTER_TYPED_FUNC(Pad, GPU, PadFunc); +REGISTER_TYPED_FUNC(PadGrad, GPU, PadGradFunc); +#endif + +} // namespace paddle diff --git a/paddle/function/PadOp.h b/paddle/function/PadOp.h new file mode 100644 index 0000000000000000000000000000000000000000..7b5c730a6a0fa57833e63beba085cb17054ae2f5 --- /dev/null +++ b/paddle/function/PadOp.h @@ -0,0 +1,79 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#pragma once + +#include "Function.h" + +namespace paddle { + +struct PadConf { + /// how many values to add before the data along channel dimension. + int channelStart; + /// how many values to add after the data along channel dimension. + int channelEnd; + /// how many values to add before the data along height dimension. + int heightStart; + /// how many values to add after the data along height dimension. + int heightEnd; + /// how many values to add before the data along width dimension. + int widthStart; + /// how many values to add after the data along width dimension. + int widthEnd; +}; + +/** + * \brief This funtion pads zeros to inputs according to the specify dimension. + * The input and output is a 4D tensor. Padding zeros from the 2nd to + * the 4th dimenstion according argument of pad. + * + * \param[out] outputs save results. + * \param[in] inputs input data. + * \param[in] num batch size of input data. + * \param[in] inC channel number of input data. + * \param[in] inH height of input data. + * \param[in] inH with of input data. + * \param[in] pad the padding config, contains the size along the + * specify dimension. + */ +template +void Pad(real* outputs, + const real* inputs, + const int num, + const int inC, + const int inH, + const int inW, + const PadConf& pad); + +/** + * \brief Padding operation backward. + * + * \param[out] inGrad gradients of previous layer. + * \param[in] outGrad output gradients. + * \param[in] num batch size of input data. + * \param[in] inC channel number of input data. + * \param[in] inH height of input data. + * \param[in] inH with of input data. + * \param[in] pad the padding config, contains the size along the + * specify dimension. + */ +template +void PadGrad(real* inGrad, + const real* outGrad, + const int num, + const int inC, + const int inH, + const int inW, + const PadConf& pad); +} // namespace paddle diff --git a/paddle/function/PadOpGpu.cu b/paddle/function/PadOpGpu.cu new file mode 100644 index 0000000000000000000000000000000000000000..9104b1aca507c526858c2117e0a5db59f535091e --- /dev/null +++ b/paddle/function/PadOpGpu.cu @@ -0,0 +1,98 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#include "hl_base.h" +#include "PadOp.h" + +namespace paddle { + +__global__ void KePad(real* outputs, const real* inputs, + int inC, int inH, int inW, + int padc, int padh, int padw, + int outC, int outH, int outW, int nthreads) { + const int idx = threadIdx.x + blockIdx.x * blockDim.x; + if (idx < nthreads) { + const int w = idx % inW; + const int h = (idx / inW) % inH; + const int c = (idx / inW / inH) % inC; + const int n = idx / inW / inH / inC; + + const int off = ((n * outC + c + padc) * outH + h + padh) * outW + padw + w; + outputs[off] = inputs[idx]; + } +} + +template <> +void Pad(real* outputs, + const real* inputs, + const int num, + const int inC, + const int inH, + const int inW, + const PadConf& pad) { + size_t nth = num * inC * inH * inW; + int blockSize = 1024; + int gridSize = (nth + 1024 - 1) / 1024; + int cstart = pad.channelStart, cend = pad.channelEnd; + int hstart = pad.heightStart, hend = pad.heightEnd; + int wstart = pad.widthStart, wend = pad.widthEnd; + int outC = inC + cstart + cend; + int outH = inH + hstart + hend; + int outW = inW + wstart + wend; + KePad<<>> + (outputs, inputs, inC, inH, inW, cstart, hstart, wstart, + outC, outH, outW, nth); + CHECK_SYNC("Pad"); +} + +__global__ void KePadDiff(real* inGrad, const real* outGrad, + int inC, int inH, int inW, + int padc, int padh, int padw, + int outC, int outH, int outW, int nthreads) { + const int idx = threadIdx.x + blockIdx.x * blockDim.x; + if (idx < nthreads) { + const int w = idx % inW; + const int h = (idx / inW) % inH; + const int c = (idx / inW / inH) % inC; + const int n = idx / inW / inH / inC; + + const int off = ((n * outC + c + padc) * outH + h + padh) * outW + padw + w; + inGrad[idx] += outGrad[off]; + } +} + +template <> +void PadGrad(real* inGrad, + const real* outGrad, + const int num, + const int inC, + const int inH, + const int inW, + const PadConf& pad) { + int nth = num * inC * inH * inW; + int blockSize = 1024; + int gridSize = (nth + 1024 - 1) / 1024; + int cstart = pad.channelStart, cend = pad.channelEnd; + int hstart = pad.heightStart, hend = pad.heightEnd; + int wstart = pad.widthStart, wend = pad.widthEnd; + int outC = inC + cstart + cend; + int outH = inH + hstart + hend; + int outW = inW + wstart + wend; + KePadDiff <<>> + (inGrad, outGrad, inC, inH, inW, cstart, hstart, wstart, + outC, outH, outW, nth); + CHECK_SYNC("PadGrad"); +} + +} // namespace paddle diff --git a/paddle/function/PadOpTest.cpp b/paddle/function/PadOpTest.cpp new file mode 100644 index 0000000000000000000000000000000000000000..cd22d9113567912f7694e05e5d631e49d940e3ac --- /dev/null +++ b/paddle/function/PadOpTest.cpp @@ -0,0 +1,75 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#include +#include "FunctionTest.h" + +namespace paddle { + +TEST(Pad, real) { + for (size_t numSamples : {5, 32}) { + for (size_t channels : {1, 5, 32}) { + for (size_t imgSizeH : {5, 33, 100}) { + for (size_t imgSizeW : {5, 32, 96}) { + VLOG(3) << " numSamples=" << numSamples << " channels=" << channels + << " imgSizeH=" << imgSizeH << " imgSizeW=" << imgSizeW; + + FunctionCompare compare("Pad", + FuncConfig() + .set("cstart", 2) + .set("cend", 3) + .set("hstart", 1) + .set("hend", 2) + .set("wstart", 3) + .set("wend", 2)); + TensorShape inDims{numSamples, channels, imgSizeH, imgSizeW}; + TensorShape outDims{ + numSamples, channels + 5, imgSizeH + 3, imgSizeW + 5}; + compare.addInputs(BufferArg(VALUE_TYPE_FLOAT, inDims)); + compare.addOutputs(BufferArg(VALUE_TYPE_FLOAT, outDims, ASSIGN_TO)); + compare.run(); + } + } + } + } +} + +TEST(PadGrad, real) { + for (size_t numSamples : {5, 32}) { + for (size_t channels : {1, 5, 32}) { + for (size_t imgSizeH : {5, 33, 100}) { + for (size_t imgSizeW : {5, 32, 96}) { + VLOG(3) << " numSamples=" << numSamples << " channels=" << channels + << " imgSizeH=" << imgSizeH << " imgSizeW=" << imgSizeW; + FunctionCompare compare("PadGrad", + FuncConfig() + .set("cstart", 2) + .set("cend", 3) + .set("hstart", 1) + .set("hend", 2) + .set("wstart", 3) + .set("wend", 2)); + TensorShape inDims{numSamples, channels, imgSizeH, imgSizeW}; + TensorShape outDims{ + numSamples, channels + 5, imgSizeH + 3, imgSizeW + 5}; + compare.addInputs(BufferArg(VALUE_TYPE_FLOAT, outDims)); + compare.addOutputs(BufferArg(VALUE_TYPE_FLOAT, inDims, ASSIGN_TO)); + compare.run(); + } + } + } + } +} + +} // namespace paddle diff --git a/paddle/function/TensorShape.h b/paddle/function/TensorShape.h index e491e3f1d6b26e14a5273b3b5a38aec941f5a9e5..cda58f19dfa4a8b80efc97570c83ca38fd7adf27 100644 --- a/paddle/function/TensorShape.h +++ b/paddle/function/TensorShape.h @@ -55,6 +55,15 @@ public: numElements(); } + void reshape(std::initializer_list dims) { + ndims_ = dims.size(); + if (ndims_ > kMinDims) { + dims_.resize(ndims_); + } + dims_.assign(dims); + numElements(); + } + // number of dimensions of the tensor size_t ndims() const { return ndims_; } @@ -82,7 +91,7 @@ private: // init dims_ void initDims(size_t ndims) { - size_t count = ndims < 4 ? 4 : ndims; + size_t count = ndims < kMinDims ? kMinDims : ndims; dims_.assign(count, 1); } @@ -92,6 +101,7 @@ private: // number of elements size_t nelements_; std::vector dims_; + static const size_t kMinDims = 4; }; } // namespace paddle diff --git a/paddle/gserver/activations/ActivationFunction.cpp b/paddle/gserver/activations/ActivationFunction.cpp index f8c4bcac2f8eb41400659dc24ba81768e7ae3640..c541b72e104bf2b81e2ac222d4af13ea2f90d289 100644 --- a/paddle/gserver/activations/ActivationFunction.cpp +++ b/paddle/gserver/activations/ActivationFunction.cpp @@ -69,8 +69,14 @@ static ClassRegistrar gActivationRegistrar; class IdentityActivation : public ActivationFunction { public: static const std::string name; - void forward(Argument& act) { (void)act; } - void backward(Argument& act) { (void)act; } + Error __must_check forward(Argument& act) { + (void)act; + return Error(); + } + Error __must_check backward(Argument& act) { + (void)act; + return Error(); + } const std::string& getName() const { return name; } }; const std::string IdentityActivation::name = ""; @@ -86,8 +92,14 @@ static InitFunction __reg_activation__identity([] { * \f] */ BEGIN_DEFINE_ACTIVATION(sigmoid) -void forward(Argument& act) { act.value->sigmoid(*act.value); } -void backward(Argument& act) { act.grad->sigmoidDerivative(*act.value); } +Error __must_check forward(Argument& act) { + act.value->sigmoid(*act.value); + return Error(); +} +Error __must_check backward(Argument& act) { + act.grad->sigmoidDerivative(*act.value); + return Error(); +} END_DEFINE_ACTIVATION(sigmoid) /** @@ -103,9 +115,12 @@ MatrixPtr sftMaxDot_; MatrixPtr one_; public: -void forward(Argument& act) { act.value->softmax(*act.value); } +Error __must_check forward(Argument& act) { + act.value->softmax(*act.value); + return Error(); +} -void backward(Argument& act) { +Error __must_check backward(Argument& act) { MatrixPtr outputV = act.value; MatrixPtr outputG = act.grad; @@ -137,6 +152,7 @@ void backward(Argument& act) { act.grad->softmaxDerivative(*act.value, *sftMaxSum_); } + return Error(); } END_DEFINE_ACTIVATION(softmax) @@ -151,8 +167,11 @@ ACTIVATION_CLASS_NAME(softmax) softmax_; Argument argument_; public: -void forward(Argument& act) { - CHECK_EQ(act.value->getWidth(), 1UL); +Error __must_check forward(Argument& act) { + if (act.value->getWidth() != 1UL) { + return Error( + "Input width for each timestep of sequence softmax should be 1"); + } if (!argument_.value) { argument_.value = Matrix::create(nullptr, @@ -169,10 +188,14 @@ void forward(Argument& act) { auto starts = act.sequenceStartPositions->getVector(useGpu(act.deviceId)); act.value->sequenceSoftmax(*act.value, *starts); + return Error(); } -void backward(Argument& act) { - CHECK_EQ(act.grad->getWidth(), 1UL); +Error __must_check backward(Argument& act) { + if (act.value->getWidth() != 1UL) { + return Error( + "Input width for each timestep of sequence softmax should be 1"); + } size_t numSequences = act.getNumSequences(); const int* starts = act.sequenceStartPositions->getData(false); @@ -184,8 +207,10 @@ void backward(Argument& act) { argument_.value->setData(act.value->getData() + offset, 1UL, size); argument_.grad->setData(act.grad->getData() + offset, 1UL, size); - softmax_.backward(argument_); + Error status = softmax_.backward(argument_); + if (!status) return status; } + return Error(); } END_DEFINE_ACTIVATION(sequence_softmax) @@ -200,9 +225,15 @@ END_DEFINE_ACTIVATION(sequence_softmax) * 0 otherwise. */ BEGIN_DEFINE_ACTIVATION(relu) -void forward(Argument& act) { act.value->relu(*act.value); } +Error __must_check forward(Argument& act) { + act.value->relu(*act.value); + return Error(); +} -void backward(Argument& act) { act.grad->reluDerivative(*act.value); } +Error __must_check backward(Argument& act) { + act.grad->reluDerivative(*act.value); + return Error(); +} END_DEFINE_ACTIVATION(relu) /** @@ -219,9 +250,15 @@ END_DEFINE_ACTIVATION(relu) * TODO(yuyang18): Remove magic number 24 or make it configuable. */ BEGIN_DEFINE_ACTIVATION(brelu) -void forward(Argument& act) { act.value->brelu(*act.value); } +Error __must_check forward(Argument& act) { + act.value->brelu(*act.value); + return Error(); +} -void backward(Argument& act) { act.grad->breluDerivative(*act.value); } +Error __must_check backward(Argument& act) { + act.grad->breluDerivative(*act.value); + return Error(); +} END_DEFINE_ACTIVATION(brelu) /** @@ -231,9 +268,15 @@ END_DEFINE_ACTIVATION(brelu) * \f] */ BEGIN_DEFINE_ACTIVATION(tanh) -void forward(Argument& act) { act.value->tanh(*act.value); } +Error __must_check forward(Argument& act) { + act.value->tanh(*act.value); + return Error(); +} -void backward(Argument& act) { act.grad->tanhDerivative(*act.value); } +Error __must_check backward(Argument& act) { + act.grad->tanhDerivative(*act.value); + return Error(); +} END_DEFINE_ACTIVATION(tanh) /** @@ -248,10 +291,14 @@ real a, b; public: ACTIVATION_CLASS_NAME(stanh)() : a(1.7159), b(2. / 3.) {} -void forward(Argument& act) { act.value->scaledTanh(*act.value, a, b); } +Error __must_check forward(Argument& act) { + act.value->scaledTanh(*act.value, a, b); + return Error(); +} -void backward(Argument& act) { +Error __must_check backward(Argument& act) { act.grad->scaledTanhDerivative(*act.value, a, b); + return Error(); } END_DEFINE_ACTIVATION(stanh) @@ -262,9 +309,15 @@ END_DEFINE_ACTIVATION(stanh) * \f] */ BEGIN_DEFINE_ACTIVATION(softrelu) -void forward(Argument& act) { act.value->softrelu(*act.value); } +Error __must_check forward(Argument& act) { + act.value->softrelu(*act.value); + return Error(); +} -void backward(Argument& act) { act.grad->softreluDerivative(*act.value); } +Error __must_check backward(Argument& act) { + act.grad->softreluDerivative(*act.value); + return Error(); +} END_DEFINE_ACTIVATION(softrelu) /** @@ -280,7 +333,7 @@ END_DEFINE_ACTIVATION(softrelu) * 0 if z=0 */ BEGIN_DEFINE_ACTIVATION(abs) -void forward(Argument& act) { +Error __must_check forward(Argument& act) { SetDevice device(act.deviceId); Matrix::resizeOrCreate(act.in, act.value->getHeight(), @@ -290,9 +343,13 @@ void forward(Argument& act) { act.in->copyFrom(*act.value); act.value->abs2(*act.value); + return Error(); } -void backward(Argument& act) { act.grad->absDerivative(*act.in); } +Error __must_check backward(Argument& act) { + act.grad->absDerivative(*act.in); + return Error(); +} END_DEFINE_ACTIVATION(abs) /** @@ -302,7 +359,7 @@ END_DEFINE_ACTIVATION(abs) * \f] */ BEGIN_DEFINE_ACTIVATION(square) -void forward(Argument& act) { +Error __must_check forward(Argument& act) { SetDevice device(act.deviceId); Matrix::resizeOrCreate(act.in, act.value->getHeight(), @@ -312,9 +369,13 @@ void forward(Argument& act) { act.in->copyFrom(*act.value); act.value->square2(*act.value); + return Error(); } -void backward(Argument& act) { act.grad->squareDerivative(*act.in); } +Error __must_check backward(Argument& act) { + act.grad->squareDerivative(*act.in); + return Error(); +} END_DEFINE_ACTIVATION(square) /** @@ -324,9 +385,15 @@ END_DEFINE_ACTIVATION(square) * \f] */ BEGIN_DEFINE_ACTIVATION(exponential) -void forward(Argument& act) { act.value->exp2(*act.value); } +Error __must_check forward(Argument& act) { + act.value->exp2(*act.value); + return Error(); +} -void backward(Argument& act) { act.grad->expDerivative(*act.value); } +Error __must_check backward(Argument& act) { + act.grad->expDerivative(*act.value); + return Error(); +} END_DEFINE_ACTIVATION(exponential) /** @@ -336,7 +403,7 @@ END_DEFINE_ACTIVATION(exponential) * \f] */ BEGIN_DEFINE_ACTIVATION(log) -void forward(Argument& act) { +Error __must_check forward(Argument& act) { SetDevice device(act.deviceId); Matrix::resizeOrCreate(act.in, act.value->getHeight(), @@ -346,9 +413,13 @@ void forward(Argument& act) { act.in->copyFrom(*act.value); act.value->log2(*act.value); + return Error(); } -void backward(Argument& act) { act.grad->dotDiv(*act.grad, *act.in); } +Error __must_check backward(Argument& act) { + act.grad->dotDiv(*act.grad, *act.in); + return Error(); +} END_DEFINE_ACTIVATION(log) ActivationFunction* ActivationFunction::create(const std::string& type) { diff --git a/paddle/gserver/activations/ActivationFunction.h b/paddle/gserver/activations/ActivationFunction.h index 601e3b6c0cd401ec007e8cf51e44416f82832e58..f208224e304a79125679c6f3a5c0be09552465ef 100644 --- a/paddle/gserver/activations/ActivationFunction.h +++ b/paddle/gserver/activations/ActivationFunction.h @@ -15,6 +15,7 @@ limitations under the License. */ #pragma once #include #include +#include "paddle/utils/Error.h" namespace paddle { @@ -48,7 +49,7 @@ public: * * Usually, act is Layer::output_ */ - virtual void forward(Argument& act) = 0; + virtual Error __must_check forward(Argument& act) = 0; /** * @brief Backward propagaion @@ -57,7 +58,7 @@ public: * - Before calling backward(), act.grad = dE / dy, where E is the error/cost * - After backward() returns, act.grad = dE / dx = (dE/dy) * (dy/dx) */ - virtual void backward(Argument& act) = 0; + virtual Error __must_check backward(Argument& act) = 0; virtual const std::string& getName() const = 0; }; diff --git a/paddle/gserver/layers/ContextProjection.cpp b/paddle/gserver/layers/ContextProjection.cpp index ebcc87cbf48a3c34a4e625e67f872fed69cdf44f..d7042af1c25e7432e5b1efbb89cd8fd3f63fb4ae 100644 --- a/paddle/gserver/layers/ContextProjection.cpp +++ b/paddle/gserver/layers/ContextProjection.cpp @@ -118,16 +118,15 @@ void ContextProjection::forward() { /// first use state_, otherwise use weight_(padding false === w nullptr) auto w_ptr = state_ ? state_.get() : is_padding ? weight_->getW().get() : nullptr; - auto start_pos = in_->sequenceStartPositions; - + const auto start_pos = in_->sequenceStartPositions->getVector(useGpu_); BufferArgs inputs; BufferArgs outputs; - inputs.addArg(*in_->value); - inputs.addArg(CpuMatrix(w_ptr ? w_ptr->getData() : nullptr, - w_ptr ? w_ptr->getHeight() : 0, - input_dim)); - inputs.addArg(*in_->sequenceStartPositions->getVector(useGpu_)); - outputs.addArg(*out_->value, ADD_TO); + inputs.addArg(*in_->value, *start_pos); + if (w_ptr) { + inputs.addArg(CpuMatrix(w_ptr->getData(), w_ptr->getHeight(), input_dim), + *start_pos); + } + outputs.addArg(*out_->value, *start_pos, ADD_TO); forward_[0]->calc(inputs, outputs); if (state_ && config_.context_start() < 0) { @@ -166,13 +165,16 @@ void ContextProjection::backward(const UpdateCallback& callback) { BufferArgs inputs; BufferArgs outputs; - inputs.addArg(CpuMatrix( - in_->grad ? in_->grad->getData() : nullptr, batch_size, input_dim)); - inputs.addArg(CpuMatrix(w_ptr ? w_ptr->getData() : nullptr, - w_ptr ? w_ptr->getHeight() : 0, - input_dim)); - inputs.addArg(*in_->sequenceStartPositions->getVector(useGpu_)); - outputs.addArg(*out_->grad, ADD_TO); + inputs.addArg(*out_->grad, *in_->sequenceStartPositions->getVector(useGpu_)); + outputs.addArg( + CpuMatrix( + in_->grad ? in_->grad->getData() : nullptr, batch_size, input_dim), + *in_->sequenceStartPositions->getVector(useGpu_), + ADD_TO); + outputs.addArg(CpuMatrix(w_ptr ? w_ptr->getData() : nullptr, + w_ptr ? w_ptr->getHeight() : 0, + input_dim), + ADD_TO); backward_[0]->calc(inputs, outputs); if (config_.trainable_padding()) { diff --git a/paddle/gserver/layers/GatedRecurrentLayer.cpp b/paddle/gserver/layers/GatedRecurrentLayer.cpp index 930d9a056164e7c677adb53b7b67901364da1309..d3aeea921801da301b2829736059130aec14cef6 100644 --- a/paddle/gserver/layers/GatedRecurrentLayer.cpp +++ b/paddle/gserver/layers/GatedRecurrentLayer.cpp @@ -314,13 +314,13 @@ void GatedRecurrentLayer::forwardBatch(int batchSize, batchValue_->resizeOrCreate(*output_.value); batchValue_->copy(*inputValue, *gate_.value, /* seq2batch */ true); - if (bias_ && bias_->getWGrad()) { + if (bias_) { gate_.value->addBias(*(bias_->getW()), 1); } { int numBatch = batchValue_->getNumBatch(); - int batchSize = 0; + int curBatchSize = 0; AsyncGpuBlock asyncGpuBlock; for (int n = 0; n < numBatch; n++) { MatrixPtr outputValueTmp = batchValue_->getBatchValue(n); @@ -330,16 +330,17 @@ void GatedRecurrentLayer::forwardBatch(int batchSize, gruValue.resetOutputValue = (batchValue_->getBatchValue(*resetOutput_.value, n))->getData(); - batchSize = outputValueTmp->getHeight(); + curBatchSize = outputValueTmp->getHeight(); gruValue.prevOutValue = - (n == 0 ? nullptr - : (batchValue_->getBatchValue(n - 1, batchSize))->getData()); + (n == 0 + ? nullptr + : (batchValue_->getBatchValue(n - 1, curBatchSize))->getData()); { if (useGpu_) { - GruCompute::forward<1>(gruValue, getSize(), batchSize); + GruCompute::forward<1>(gruValue, getSize(), curBatchSize); } else { - GruCompute::forward<0>(gruValue, getSize(), batchSize); + GruCompute::forward<0>(gruValue, getSize(), curBatchSize); } } } diff --git a/paddle/gserver/layers/Layer.cpp b/paddle/gserver/layers/Layer.cpp index c47943f81c01589eada4b825d54be5c69314b6fa..f76d41ad3e8a3b1730f9d50c0773ee4f61ddb541 100644 --- a/paddle/gserver/layers/Layer.cpp +++ b/paddle/gserver/layers/Layer.cpp @@ -15,6 +15,7 @@ limitations under the License. */ #include "paddle/utils/Util.h" #include "paddle/math/SparseMatrix.h" +#include "paddle/utils/Error.h" #include "paddle/utils/Logging.h" #include "AddtoLayer.h" @@ -334,7 +335,8 @@ void Layer::showOutputStats() { void Layer::forwardActivation() { /* activation */ - activation_->forward(output_); + auto status = activation_->forward(output_); + status.check(); /* dropout */ if (config_.drop_rate() > 0) { @@ -372,7 +374,8 @@ void Layer::backwardActivation() { oGrad->dotMul(*oGrad, *dropOutMask_); } - activation_->backward(output_); + auto status = activation_->backward(output_); + status.check(); } void Layer::forwardDropOut() { diff --git a/paddle/gserver/layers/MDLstmLayer.cpp b/paddle/gserver/layers/MDLstmLayer.cpp index fb41af563195496a57eafcc52b49eadac697fa0a..88d934d782b549a984f1d7798e54bcc4436ea0cf 100644 --- a/paddle/gserver/layers/MDLstmLayer.cpp +++ b/paddle/gserver/layers/MDLstmLayer.cpp @@ -506,9 +506,12 @@ void MDLstmLayer::forwardGate2OutputSequence(int start, *frameState_[start + preOffsetV[i]].value, *checkFgOneDim, 1.0, 1.0); } } - activationGate_->forward(frameInputGate_[idxCurr]); - activationGate_->forward(frameForgetGate_[idxCurr]); - activation_->forward(frameInputNode_[idxCurr]); + auto status = activationGate_->forward(frameInputGate_[idxCurr]); + status.check(); + status = activationGate_->forward(frameForgetGate_[idxCurr]); + status.check(); + status = activation_->forward(frameInputNode_[idxCurr]); + status.check(); frameState_[idxCurr].value->zeroMem(); for (int i = 0; i < numDims_; i++) { @@ -530,10 +533,12 @@ void MDLstmLayer::forwardGate2OutputSequence(int start, frameOutputGate_[idxCurr].value->addDotMul( *frameState_[idxCurr].value, *checkOg_, 1.0, 1.0); - activationGate_->forward(frameOutputGate_[idxCurr]); + status = activationGate_->forward(frameOutputGate_[idxCurr]); + status.check(); framePreOutput_[idxCurr].value->copyFrom(*(frameState_[idxCurr].value)); - activationState_->forward(framePreOutput_[idxCurr]); + status = activationState_->forward(framePreOutput_[idxCurr]); + status.check(); frameOutput_[idxCurr].value->dotMul(*framePreOutput_[idxCurr].value, *frameOutputGate_[idxCurr].value); @@ -640,12 +645,12 @@ void MDLstmLayer::backwardGate2OutputSequence(int start, framePreOutput_[idxCurr].grad->dotMul(*frameOutput_[idxCurr].grad, *frameOutputGate_[idxCurr].value); - activationState_->backward(framePreOutput_[idxCurr]); + activationState_->backward(framePreOutput_[idxCurr]).check(); frameState_[idxCurr].grad->copyFrom(*(framePreOutput_[idxCurr].grad)); frameOutputGate_[idxCurr].grad->dotMul(*frameOutput_[idxCurr].grad, *framePreOutput_[idxCurr].value); - activationGate_->backward(frameOutputGate_[idxCurr]); + activationGate_->backward(frameOutputGate_[idxCurr]).check(); frameState_[idxCurr].grad->addDotMul( *frameOutputGate_[idxCurr].grad, *checkOg_, 1.0, 1.0); @@ -702,9 +707,9 @@ void MDLstmLayer::backwardGate2OutputSequence(int start, } } - activationGate_->backward(frameInputGate_[idxCurr]); - activationGate_->backward(frameForgetGate_[idxCurr]); - activation_->backward(frameInputNode_[idxCurr]); + activationGate_->backward(frameInputGate_[idxCurr]).check(); + activationGate_->backward(frameForgetGate_[idxCurr]).check(); + activation_->backward(frameInputNode_[idxCurr]).check(); if (bias_->getWGrad()) { for (int i = 0; i < numDims_; i++) { diff --git a/paddle/gserver/layers/NCELayer.cpp b/paddle/gserver/layers/NCELayer.cpp index 5ab765247f63dfe6e6651ca4d27dc7183a9f33e1..3542e739df8d03470bf2c455b4f3492a7f9e973a 100644 --- a/paddle/gserver/layers/NCELayer.cpp +++ b/paddle/gserver/layers/NCELayer.cpp @@ -193,7 +193,8 @@ public: forwardOneInput(l); } - activation_->forward(sampleOut_); + auto status = activation_->forward(sampleOut_); + status.check(); forwardCost(); } @@ -207,7 +208,8 @@ public: backwardCost(); - activation_->backward(sampleOut_); + auto status = activation_->backward(sampleOut_); + status.check(); if (biases_->getWGrad()) { backwardBias(callback); diff --git a/paddle/gserver/layers/PadLayer.cpp b/paddle/gserver/layers/PadLayer.cpp new file mode 100644 index 0000000000000000000000000000000000000000..bb618c09f9777785d93995fa7140dd4a5383cd1b --- /dev/null +++ b/paddle/gserver/layers/PadLayer.cpp @@ -0,0 +1,115 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#include "PadLayer.h" +#include "paddle/utils/Stat.h" + +namespace paddle { + +REGISTER_LAYER(pad, PadLayer); + +bool PadLayer::init(const LayerMap& layerMap, + const ParameterMap& parameterMap) { + /* Initialize the basic parent class */ + Layer::init(layerMap, parameterMap); + + auto& pad_conf = config_.inputs(0).pad_conf(); + auto& img_conf = pad_conf.image_conf(); + CHECK_EQ(config_.inputs_size(), 1); + inDims_ = TensorShape( + {0, + img_conf.channels(), + img_conf.has_img_size_y() ? img_conf.img_size_y() : img_conf.img_size(), + img_conf.img_size()}); + + CHECK_EQ(2, pad_conf.pad_c_size()); + CHECK_EQ(2, pad_conf.pad_h_size()); + CHECK_EQ(2, pad_conf.pad_w_size()); + padc_.push_back(pad_conf.pad_c(0)); + padc_.push_back(pad_conf.pad_c(1)); + padh_.push_back(pad_conf.pad_h(0)); + padh_.push_back(pad_conf.pad_h(1)); + padw_.push_back(pad_conf.pad_w(0)); + padw_.push_back(pad_conf.pad_w(1)); + + outDims_ = TensorShape(4); + setOutDims(0); + + createFunction(forward_, + "Pad", + FuncConfig() + .set("cstart", padc_[0]) + .set("cend", padc_[1]) + .set("hstart", padh_[0]) + .set("hend", padh_[1]) + .set("wstart", padw_[0]) + .set("wend", padw_[1])); + createFunction(backward_, + "PadGrad", + FuncConfig() + .set("cstart", padc_[0]) + .set("cend", padc_[1]) + .set("hstart", padh_[0]) + .set("hend", padh_[1]) + .set("wstart", padw_[0]) + .set("wend", padw_[1])); + + return true; +} + +void PadLayer::setOutDims(const size_t batchSize) { + outDims_.reshape({batchSize, + inDims_[1] + padc_[0] + padc_[1], + inDims_[2] + padh_[0] + padh_[1], + inDims_[3] + padw_[0] + padw_[1]}); +} + +void PadLayer::setTensorDim(const size_t batchSize) { + CHECK_EQ(static_cast(inputLayers_.size()), 1); + inDims_.setDim(0, batchSize); + int h = inputLayers_[0]->getOutput().getFrameHeight(); + if (h != 0) inDims_.setDim(2, h); + int w = inputLayers_[0]->getOutput().getFrameWidth(); + if (w != 0) inDims_.setDim(3, w); + setOutDims(batchSize); +} + +void PadLayer::forward(PassType passType) { + Layer::forward(passType); + MatrixPtr input = inputLayers_[0]->getOutputValue(); + size_t batchSize = input->getHeight(); + setTensorDim(batchSize); + int size = outDims_[1] * outDims_[2] * outDims_[3]; + resetOutput(batchSize, size); + MatrixPtr outV = getOutputValue(); + REGISTER_TIMER_INFO("PadForward", getName().c_str()); + + BufferArgs inputs; + BufferArgs outputs; + inputs.addArg(*getInputValue(0), inDims_); + outputs.addArg(*getOutputValue(), outDims_, ASSIGN_TO); + forward_[0]->calc(inputs, outputs); +} + +void PadLayer::backward(const UpdateCallback& callback) { + (void)callback; + REGISTER_TIMER_INFO("PadBackward", getName().c_str()); + + BufferArgs inputs; + BufferArgs outputs; + inputs.addArg(*getOutputGrad(), outDims_); + outputs.addArg(*getInputGrad(0), inDims_, ADD_TO); + backward_[0]->calc(inputs, outputs); +} +} // namespace paddle diff --git a/paddle/gserver/layers/PadLayer.h b/paddle/gserver/layers/PadLayer.h new file mode 100644 index 0000000000000000000000000000000000000000..3e3a21a9970370c2bc9c2ac656af776719dfca24 --- /dev/null +++ b/paddle/gserver/layers/PadLayer.h @@ -0,0 +1,46 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#pragma once + +#include "Layer.h" + +namespace paddle { + +/** + * \brief This layer pads zeros to inputs according to the specify dimension. + * The input and output is a 4D tensor. Padding zeros from the 2nd to + * the 4th dimenstion according padc_, padh_ and padw_. + */ +class PadLayer : public Layer { +public: + explicit PadLayer(const LayerConfig& config) : Layer(config) {} + + ~PadLayer() {} + + bool init(const LayerMap& layerMap, const ParameterMap& parameterMap); + void forward(PassType passType); + void backward(const UpdateCallback& callback = nullptr); + +protected: + void setOutDims(const size_t batchSize); + void setTensorDim(const size_t batchSize); + + std::vector padc_; + std::vector padh_; + std::vector padw_; + TensorShape inDims_; + TensorShape outDims_; +}; +} // namespace paddle diff --git a/paddle/gserver/layers/RecurrentLayer.cpp b/paddle/gserver/layers/RecurrentLayer.cpp index 55e0fdfb9048c02b2dcd474c6887eee180328260..b843fa1265cf3c0ad0814fb90f69e245ee5ab4ad 100644 --- a/paddle/gserver/layers/RecurrentLayer.cpp +++ b/paddle/gserver/layers/RecurrentLayer.cpp @@ -217,21 +217,22 @@ void RecurrentLayer::forwardOneSequence(int start, int length) { if (prevOutput_) { frameOutput_[start].value->mul(*prevOutput_, *weight_->getW(), 1, 1); } - activation_->forward(frameOutput_[start]); + activation_->forward(frameOutput_[start]).check(); + for (int i = 1; i < length; ++i) { frameOutput_[start + i].value->mul( *frameOutput_[start + i - 1].value, *weight_->getW(), 1, 1); - activation_->forward(frameOutput_[start + i]); + activation_->forward(frameOutput_[start + i]).check(); } if (prevOutput_) { prevOutput_->assign(*frameOutput_[start + length - 1].value); } } else { - activation_->forward(frameOutput_[start + length - 1]); + activation_->forward(frameOutput_[start + length - 1]).check(); for (int i = length - 2; i >= 0; --i) { frameOutput_[start + i].value->mul( *frameOutput_[start + i + 1].value, *weight_->getW(), 1, 1); - activation_->forward(frameOutput_[start + i]); + activation_->forward(frameOutput_[start + i]).check(); } } } @@ -280,11 +281,11 @@ void RecurrentLayer::backwardOneSequence(int start, int length) { MatrixPtr weightT = weight_->getW()->getTranspose(); if (!reversed_) { for (int i = length - 1; i > 0; --i) { - activation_->backward(frameOutput_[start + i]); + activation_->backward(frameOutput_[start + i]).check(); frameOutput_[start + i - 1].grad->mul( *frameOutput_[start + i].grad, *weightT, 1, 1); } - activation_->backward(frameOutput_[start]); + activation_->backward(frameOutput_[start]).check(); if (weight_->getWGrad()) { weight_->getWGrad()->mul( *output_.value->subMatrix(start, length - 1)->getTranspose(), @@ -294,11 +295,11 @@ void RecurrentLayer::backwardOneSequence(int start, int length) { } } else { for (int i = 0; i < length - 1; ++i) { - activation_->backward(frameOutput_[start + i]); + activation_->backward(frameOutput_[start + i]).check(); frameOutput_[start + i + 1].grad->mul( *frameOutput_[start + i].grad, *weightT, 1, 1); } - activation_->backward(frameOutput_[start + length - 1]); + activation_->backward(frameOutput_[start + length - 1]).check(); if (weight_->getWGrad()) { weight_->getWGrad()->mul( *output_.value->subMatrix(start + 1, length - 1)->getTranspose(), @@ -333,7 +334,7 @@ void RecurrentLayer::forwardBatch(int batchSize, } Argument arg; arg.value = batch2; - activation_->forward(arg); + activation_->forward(arg).check(); } } batchValue_->copyBackSeq(*output_.value); @@ -363,7 +364,7 @@ void RecurrentLayer::backwardBatch(int batchSize, Argument arg; arg.value = batch1; arg.grad = batch2; - activation_->backward(arg); + activation_->backward(arg).check(); if (n != 0) { batch1 = batchGrad_->getBatchValue(n - 1, batch2->getHeight()); diff --git a/paddle/gserver/layers/SelectiveFullyConnectedLayer.cpp b/paddle/gserver/layers/SelectiveFullyConnectedLayer.cpp index 5eacff6b7143996130bea64766ef42c66f4c7310..d9a91de8a6f4daf514f089a3d63cb519223bfdd0 100644 --- a/paddle/gserver/layers/SelectiveFullyConnectedLayer.cpp +++ b/paddle/gserver/layers/SelectiveFullyConnectedLayer.cpp @@ -192,7 +192,8 @@ void SelectiveFullyConnectedLayer::forward(PassType passType) { nnz, /*trans=*/false, /*useGpu=*/useGpu_); - activation_->forward(arg); + //! TODO(yuyang18): Why we cannot invoke forwardActivation here? + activation_->forward(arg).check(); } else /* train and test in train, not generating */ { // during training, this layer output value is *Matrix*, which is input of // eg. multi-class-cross-entropy diff --git a/paddle/gserver/tests/test_LayerGrad.cpp b/paddle/gserver/tests/test_LayerGrad.cpp index 66a70ecd41091b9590038dab3194dd2a0c59dd03..8c8e876bd64fb97e11bc04c26ec45358f3f808a1 100644 --- a/paddle/gserver/tests/test_LayerGrad.cpp +++ b/paddle/gserver/tests/test_LayerGrad.cpp @@ -310,7 +310,11 @@ TEST(Layer, CTCLayer) { config.layerConfig.add_inputs(); for (auto useGpu : {false, true}) { - testLayerGrad(config, "ctc", 100, /* trans */ false, /* useGpu */ useGpu); + testLayerGrad(config, + "ctc", + 100, + /* trans */ false, /* useGpu */ + useGpu); } } @@ -587,7 +591,11 @@ TEST(Layer, hsigmoidLayer) { config.layerConfig.add_inputs(); // Not support GPU now - testLayerGrad(config, "hsigmoid", 100, /* trans */ false, /* useGpu */ false); + testLayerGrad(config, + "hsigmoid", + 100, + /* trans */ false, /* useGpu */ + false); } TEST(Layer, multi_cross) { @@ -1022,8 +1030,12 @@ void testNormLayer(const string& normType, bool trans, bool useGpu) { } TEST(Layer, NormLayer) { - testNormLayer("cmrnorm-projection", /* trans= */ false, /* useGpu= */ true); - testNormLayer("cmrnorm-projection", /* trans= */ false, /* useGpu= */ false); + testNormLayer("cmrnorm-projection", + /* trans= */ false, /* useGpu= */ + true); + testNormLayer("cmrnorm-projection", + /* trans= */ false, /* useGpu= */ + false); } void setPoolConfig(TestConfig* config, @@ -1563,6 +1575,35 @@ TEST(Layer, MultiplexLayer) { } } +TEST(Layer, PadLayer) { + TestConfig config; + config.biasSize = 0; + config.layerConfig.set_type("pad"); + + int c = 4; + int h = 31; + int w = 36; + size_t size = c * h * w; + config.inputDefs.push_back({INPUT_DATA, "layer_0", size, 0}); + LayerInputConfig* input = config.layerConfig.add_inputs(); + PadConfig* pad = input->mutable_pad_conf(); + ImageConfig* image = pad->mutable_image_conf(); + + image->set_channels(c); + image->set_img_size(h); + image->set_img_size_y(w); + pad->add_pad_c(1); + pad->add_pad_c(2); + pad->add_pad_h(2); + pad->add_pad_h(3); + pad->add_pad_w(3); + pad->add_pad_w(5); + + for (auto useGpu : {false, true}) { + testLayerGrad(config, "pad", 10, false, useGpu); + } +} + int main(int argc, char** argv) { testing::InitGoogleTest(&argc, argv); initMain(argc, argv); diff --git a/paddle/gserver/tests/test_WarpCTCLayer.cpp b/paddle/gserver/tests/test_WarpCTCLayer.cpp index 23ae95852e84216c9065f1b123d35ce868fbb90f..55427e2f12fd7b77c6eea1f65b3229e6fd29d71d 100644 --- a/paddle/gserver/tests/test_WarpCTCLayer.cpp +++ b/paddle/gserver/tests/test_WarpCTCLayer.cpp @@ -148,11 +148,11 @@ LayerPtr createCTCLayer(string name, ActivationFunction* softmaxActivation = ActivationFunction::create("softmax"); - softmaxActivation->forward(dataLayer->getOutput()); + softmaxActivation->forward(dataLayer->getOutput()).check(); layer->forward(PASS_GC); layer->backward(); - softmaxActivation->backward(dataLayer->getOutput()); + softmaxActivation->backward(dataLayer->getOutput()).check(); return layer; } diff --git a/paddle/math/RowBuffer.h b/paddle/math/RowBuffer.h new file mode 100644 index 0000000000000000000000000000000000000000..dbb829c4e24a659e4a97c0a3ba4c5c78b68815d3 --- /dev/null +++ b/paddle/math/RowBuffer.h @@ -0,0 +1,135 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#pragma once +#include +#include "MemoryHandle.h" +#include "paddle/utils/Util.h" + +namespace paddle { + +/** + * @brief The RowBuffer class + * Represent the SparseRow Matrix Data. + * + * If not set memory handler, then the data could be auto growth. + */ +class RowBuffer { +public: + /** + * @brief RowBuffer create a auto-growth row buffer. The row length is width. + * @param width the length of each row, a.k.a matrix width. + */ + explicit RowBuffer(size_t width) : width_(width) {} + + /** + * @brief RowBuffer create a row buffer, which cannot be auto-growth. + * @param mem the pre-allocated memory. + * @param width the length of each row, a.k.a matrix width. + */ + RowBuffer(const CpuMemHandlePtr& mem, size_t width) + : preallocatedBuf_(mem), width_(width) {} + + /** + * @brief resize resize the buffer with rowCount + * @param rowCnt number of row. matrix height. + */ + inline void resize(int rowCnt) { + if (preallocatedBuf_) { + CHECK(preallocatedBuf_->getSize() >= rowCnt * width_ * sizeof(real)); + } else { + rowStore_.resize(rowCnt * width_); + } + } + + /** + * @brief get a row buffer with row index. + * @param row the index of row. + * @return row buffer. + */ + inline real* get(int row) const { + if (preallocatedBuf_) { + CHECK_LE((row + 1) * width_ * sizeof(real), preallocatedBuf_->getSize()); + return reinterpret_cast(preallocatedBuf_->getBuf()) + row * width_; + } else { + CHECK_LE((row + 1) * width_, rowStore_.size()); + return const_cast(rowStore_.data() + row * width_); + } + } + + /** + * @brief get a row buffer with row index. If row index is larger than local + * buffer, the size of local buffer will grow. + * @param row the index of row. + * @return row buffer. + */ + inline real* getWithAutoGrowth(int row) { + if (preallocatedBuf_) { + return get(row); + } else { + if ((rowStore_.size() <= row * width_)) { + rowStore_.resize((row + 1) * width_); + } + return rowStore_.data() + row * width_; + } + } + + /** + * @return raw data buffer. + */ + inline real* data() { + if (preallocatedBuf_) { + return reinterpret_cast(preallocatedBuf_->getBuf()); + } else { + return rowStore_.data(); + } + } + + /** + * @brief clear local buffer. It only affect auto-growth buffer. + */ + inline void clear() { rowStore_.clear(); } + + /** + * @brief get current number of rows. + * @return number of rows. + */ + inline size_t getRowCount() const { + if (preallocatedBuf_) { + return preallocatedBuf_->getSize() / sizeof(real) / width_; + } else { + return rowStore_.size() / width_; + } + } + + /** + * @brief get is this buffer can automatically grow or not. + * @return ture if can automacitally grow. + */ + inline bool isAutoGrowth() const { return !preallocatedBuf_; } + + /** + * @brief return the width of matrix. a.k.a length of row. + * @return width of matrix + */ + inline size_t getWidth() const { return width_; } + +private: + //! TODO(yuyang18): Add resize method to CpuMemHandlePtr, then we can get rid + //! of std::vector here. + CpuMemHandlePtr preallocatedBuf_; + std::vector> rowStore_; + size_t width_; +}; +} // namespace paddle diff --git a/paddle/math/SparseRowMatrix.h b/paddle/math/SparseRowMatrix.h index 778a9bd845661849261b52dcbeb519809d0c6306..c05fc98ff9fe739688ed3c21466fb29b70e36854 100644 --- a/paddle/math/SparseRowMatrix.h +++ b/paddle/math/SparseRowMatrix.h @@ -18,6 +18,7 @@ limitations under the License. */ #include #include #include "Matrix.h" +#include "RowBuffer.h" #include "paddle/utils/Util.h" DECLARE_bool(allow_inefficient_sparse_update); @@ -45,12 +46,9 @@ public: IndexDictPtr indexDictHandle = nullptr, bool trans = false) : CpuMatrix(nullptr, height, width, trans), - storeMat_(dataHandle, - dataHandle ? dataHandle->getSize() / sizeof(real) / width : 0, - width, - trans), indexDictHandle_(indexDictHandle) { init(height, width); + buf_.reset(new RowBuffer(dataHandle, width)); } virtual ~SparseRowCpuMatrix() {} @@ -71,25 +69,16 @@ public: * * @param row row id in local storage */ - real* getLocalRow(size_t row) { - if (storeMat_.getData()) return storeMat_.rowBuf(row); - if (rowStore_.size() <= row * width_) { - rowStore_.resize((row + 1) * width_); - } - return rowStore_.data() + row * width_; - } + real* getLocalRow(size_t row) { return buf_->getWithAutoGrowth(row); } /** - * reserve the storage for rows according to current size of indexDictHandle. + * reserve the storage for rows according to current size of + * indexDictHandle. * * This is only used when SparseRowCpuMatrix is constructed with * indexDictHandle. */ - void reserveStore() { - if (!storeMat_.getData() && !localIndices_->empty()) { - rowStore_.resize(localIndices_->size() * width_); - } - } + void reserveStore() { buf_->resize(localIndices_->size()); } // row is the row id in the original matrix virtual real* getRowBuf(size_t row) { return getRow(row); } @@ -117,7 +106,8 @@ public: * * If L1 decay set use L1, else if L2 set use L2, otherwise no decay atall. * - * t0 is a int vector used by L1/L2 decay, size = height of parameter matrix, + * t0 is a int vector used by L1/L2 decay, size = height of parameter + * matrix, * store the time that each weight row last updated. * * Time is batchId, currentTime is current batchId. @@ -176,8 +166,7 @@ public: protected: template void apply(Func f) { - real* data = storeMat_.getData() ? storeMat_.getData() : rowStore_.data(); - f(data, localIndices_->size() * width_); + f(buf_->data(), localIndices_->size() * width_); } void init(size_t height, size_t width); @@ -188,25 +177,24 @@ protected: globalIndices_[id] = kUnusedId_; } localIndices_->clear(); - rowStore_.clear(); + buf_->clear(); } inline void checkStoreSize() { - if (storeMat_.getData()) { - CHECK_LE(localIndices_->size(), storeMat_.getHeight()); - } else if (!FLAGS_allow_inefficient_sparse_update) { - if (localIndices_->size() > 0.5 * height_) { + if (buf_->isAutoGrowth()) { + if (buf_->getRowCount() > 0.5 * height_) { LOG(WARNING) << "There are more than 0.5*height (" << localIndices_->size() << ") rows are used for sparse " << "update, which is not efficient. Considering not use " << "sparse_update or set --allow_inefficient_sparse_update=true"; } + } else { + CHECK_LE(localIndices_->size(), buf_->getRowCount()); } } - CpuMatrix storeMat_; - std::vector> rowStore_; + std::unique_ptr buf_; IndexDictPtr indexDictHandle_; std::vector* localIndices_; // =&indexDictHandle_->localIndices unsigned int* globalIndices_; // =indexDictHandle_->globalIndices.data(); diff --git a/paddle/math/tests/CMakeLists.txt b/paddle/math/tests/CMakeLists.txt index 06fc10bae7232fb1278e89e8d9cbdf477fc27b60..ceb96b2e250d8e04ffb2b1d8c77ad498dca91cf3 100644 --- a/paddle/math/tests/CMakeLists.txt +++ b/paddle/math/tests/CMakeLists.txt @@ -4,6 +4,7 @@ add_simple_unittest(test_ExecViaCpu) add_simple_unittest(test_SIMDFunctions) add_simple_unittest(test_TrainingAlgorithm) add_simple_unittest(test_SparseMatrix) +add_simple_unittest(test_RowBuffer) # TODO(yuyang18): Refactor TestUtil.cpp. Remove this cross module reference. add_unittest(test_matrixCompare diff --git a/paddle/math/tests/test_RowBuffer.cpp b/paddle/math/tests/test_RowBuffer.cpp new file mode 100644 index 0000000000000000000000000000000000000000..5f66f22ef73dcff1868c1a3e03139a680b1ce2b5 --- /dev/null +++ b/paddle/math/tests/test_RowBuffer.cpp @@ -0,0 +1,65 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#include +#include "paddle/math/RowBuffer.h" + +TEST(RowBuffer, testAutoGrow) { + paddle::RowBuffer buf(128); + ASSERT_EQ(128, buf.getWidth()); + ASSERT_TRUE(buf.isAutoGrowth()); + buf.resize(2); + ASSERT_EQ(2, buf.getRowCount()); + for (size_t i = 0; i < buf.getWidth() * 2; ++i) { + buf.data()[i] = i; + } + for (size_t i = 0; i < buf.getRowCount(); ++i) { + for (size_t j = 0; j < buf.getWidth(); ++j) { + ASSERT_NEAR(i * buf.getWidth() + j, buf.get(i)[j], 1e-5); + } + } + + auto data = buf.getWithAutoGrowth(2); + for (size_t i = 0; i < buf.getWidth(); ++i) { + data[i] = i; + } + + ASSERT_EQ(3, buf.getRowCount()); + for (size_t i = 0; i < buf.getRowCount() - 1; ++i) { + for (size_t j = 0; j < buf.getWidth(); ++j) { + ASSERT_NEAR(i * buf.getWidth() + j, buf.get(i)[j], 1e-5); + } + } + for (size_t i = 0; i < buf.getWidth(); ++i) { + ASSERT_NEAR(i, buf.get(2)[i], 1e-5); + } +} + +TEST(RowBuffer, testWithMemBuf) { + paddle::CpuMemHandlePtr mem = + std::make_shared(128 * 2 * sizeof(real)); + paddle::RowBuffer buf(mem, 128); + ASSERT_TRUE(!buf.isAutoGrowth()); + ASSERT_EQ(2, buf.getRowCount()); + for (size_t i = 0; i < buf.getWidth() * 2; ++i) { + buf.data()[i] = i; + } + for (size_t i = 0; i < buf.getRowCount(); ++i) { + for (size_t j = 0; j < buf.getWidth(); ++j) { + ASSERT_NEAR(i * buf.getWidth() + j, buf.getWithAutoGrowth(i)[j], 1e-5); + } + } + + ASSERT_DEATH_IF_SUPPORTED(buf.getWithAutoGrowth(3), ".*"); +} diff --git a/paddle/parameter/ParameterUpdaterBase.h b/paddle/parameter/ParameterUpdaterBase.h index b230e170c15f1b004c5357fb7d0ad2204d01f44b..6265c828a1a254d01dc975b0155e7ac69df49a31 100644 --- a/paddle/parameter/ParameterUpdaterBase.h +++ b/paddle/parameter/ParameterUpdaterBase.h @@ -55,7 +55,7 @@ public: // between startBatch() and finishBatch(), update() will be called // by the trainer multiple times, each time for updating one Parameter // with its gradient in PARAMETER_GRADIENT - virtual void update(Parameter* para) { + void update(Parameter* para) { SetDevice setDevice(para->getDeviceId()); para->updateHook(); this->updateImpl(para); diff --git a/paddle/pserver/test/test_ProtoServer.cpp b/paddle/pserver/test/test_ProtoServer.cpp index 9f86ee80f4e5cc99ea3597b3ed37a387578f032a..04236fda2fb62b928b5c06ff38acfd3eb7217b08 100644 --- a/paddle/pserver/test/test_ProtoServer.cpp +++ b/paddle/pserver/test/test_ProtoServer.cpp @@ -12,14 +12,13 @@ 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/utils/Util.h" - #include - +#include #include "ParameterService.pb.h" #include "paddle/math/Vector.h" #include "paddle/pserver/ProtoServer.h" #include "paddle/utils/Stat.h" +#include "paddle/utils/Util.h" DEFINE_string(server_addr, "127.0.0.1", "Server address"); DEFINE_int64(dim, 50000000, "Data size"); @@ -162,18 +161,9 @@ TEST(ProtoServer, extended) { int main(int argc, char** argv) { paddle::initMain(argc, argv); testing::InitGoogleTest(&argc, argv); - - MyServer* server; - if (FLAGS_rdma_tcp == "rdma") { - server = new MyServer(FLAGS_port, 0); - } else { - server = new MyServer(FLAGS_port); - } - - server->start(); + MyServer server(FLAGS_port, FLAGS_rdma_tcp == "rdma" ? 0 : -1); + server.start(); usleep(10000); - int ret = RUN_ALL_TESTS(); - - exit(ret); + return RUN_ALL_TESTS(); } diff --git a/paddle/py_paddle/dataprovider_converter.py b/paddle/py_paddle/dataprovider_converter.py index 981d10afda2671be9e8f0da1a4bee755f7aa9d61..21d1cb75f4d40e6ed011b33c6366c9d31c0fcc7c 100644 --- a/paddle/py_paddle/dataprovider_converter.py +++ b/paddle/py_paddle/dataprovider_converter.py @@ -34,6 +34,10 @@ class IScanner(object): class DenseScanner(IScanner): + """ + :type __mat__: numpy.ndarray + """ + def __init__(self, input_type, pos): IScanner.__init__(self, input_type, pos) self.__mat__ = None @@ -47,6 +51,8 @@ class DenseScanner(IScanner): def finish_scan(self, argument): assert isinstance(argument, swig_paddle.Arguments) assert isinstance(self.input_type, dp2.InputType) + if self.__mat__.dtype != numpy.float32: + self.__mat__ = self.__mat__.astype(numpy.float32) m = swig_paddle.Matrix.createDenseFromNumpy(self.__mat__, True, False) argument.setSlotValue(self.pos, m) diff --git a/paddle/scripts/docker/Dockerfile b/paddle/scripts/docker/Dockerfile index 1522be023f6de32f86fc8a367867bbe2f1c9aeb6..79c4efbed0b8562e6f50b27d6ac297da0929de79 100644 --- a/paddle/scripts/docker/Dockerfile +++ b/paddle/scripts/docker/Dockerfile @@ -4,28 +4,32 @@ MAINTAINER PaddlePaddle Authors ARG DEBIAN_FRONTEND=noninteractive ARG UBUNTU_MIRROR RUN /bin/bash -c 'if [[ -n ${UBUNTU_MIRROR} ]]; then sed -i 's#http://archive.ubuntu.com#${UBUNTU_MIRROR}#g' /etc/apt/sources.list; fi' -RUN apt-get update \ - && apt-get install -y cmake libprotobuf-dev protobuf-compiler git \ - libgoogle-glog-dev libgflags-dev libgtest-dev \ - libatlas-dev libatlas3-base g++ m4 python-pip \ - python-protobuf python-numpy python-dev swig openssh-server \ - wget unzip python-matplotlib tar xz-utils bzip2 gzip coreutils \ - sed grep graphviz libjpeg-dev zlib1g-dev doxygen \ - clang-3.8 llvm-3.8 libclang-3.8-dev \ - && apt-get clean -y -RUN cd /usr/src/gtest && cmake . && make && cp *.a /usr/lib -RUN pip install -U BeautifulSoup docopt PyYAML pillow \ - sphinx sphinx_rtd_theme recommonmark jupyter + +RUN apt-get update && \ + apt-get install -y git python-pip python-dev openssh-server bison && \ + apt-get install -y wget unzip tar xz-utils bzip2 gzip coreutils && \ + apt-get install -y curl sed grep graphviz libjpeg-dev zlib1g-dev && \ + apt-get install -y python-numpy python-matplotlib gcc g++ gfortran && \ + apt-get install -y automake clang-3.8 llvm-3.8 libclang-3.8-dev && \ + apt-get clean -y + +RUN pip install --upgrade pip && \ + pip install 'protobuf==3.1.0.post1' && \ + pip install -U wheel pillow BeautifulSoup && \ + pip install -U docopt PyYAML sphinx && \ + pip install -U sphinx_rtd_theme recommonmark jupyter + +RUN curl -sSL https://cmake.org/files/v3.4/cmake-3.4.1.tar.gz | tar -xz && \ + cd cmake-3.4.1 && ./bootstrap && make -j4 && make install && \ + cd .. && rm -rf cmake-3.4.1 ARG WITH_AVX ARG WITH_DOC -ARG WITH_SWIG_PY ARG WITH_STYLE_CHECK ENV WITH_GPU=OFF ENV WITH_AVX=${WITH_AVX:-ON} ENV WITH_DOC=${WITH_DOC:-ON} -ENV WITH_SWIG_PY=${WITH_SWIG_PY:-ON} ENV WITH_STYLE_CHECK=${WITH_STYLE_CHECK:-OFF} RUN mkdir /paddle diff --git a/paddle/scripts/docker/Dockerfile.gpu b/paddle/scripts/docker/Dockerfile.gpu index 09f07043e2172319de257cc952fb81ba53ce89a5..6c1c2225d1a304f234a940584c6c33502eaabbb8 100644 --- a/paddle/scripts/docker/Dockerfile.gpu +++ b/paddle/scripts/docker/Dockerfile.gpu @@ -4,28 +4,32 @@ MAINTAINER PaddlePaddle Authors ARG DEBIAN_FRONTEND=noninteractive ARG UBUNTU_MIRROR RUN /bin/bash -c 'if [[ -n ${UBUNTU_MIRROR} ]]; then sed -i 's#http://archive.ubuntu.com#${UBUNTU_MIRROR}#g' /etc/apt/sources.list; fi' -RUN apt-get update \ - && apt-get install -y cmake libprotobuf-dev protobuf-compiler git \ - libgoogle-glog-dev libgflags-dev libgtest-dev \ - libatlas-dev libatlas3-base g++ m4 python-pip \ - python-protobuf python-numpy python-dev swig openssh-server \ - wget unzip python-matplotlib tar xz-utils bzip2 gzip coreutils \ - sed grep graphviz libjpeg-dev zlib1g-dev doxygen \ - clang-3.8 llvm-3.8 libclang-3.8-dev \ - && apt-get clean -y -RUN cd /usr/src/gtest && cmake . && make && cp *.a /usr/lib -RUN pip install -U BeautifulSoup docopt PyYAML pillow \ - sphinx sphinx_rtd_theme recommonmark jupyter + +RUN apt-get update && \ + apt-get install -y git python-pip python-dev openssh-server bison && \ + apt-get install -y wget unzip tar xz-utils bzip2 gzip coreutils && \ + apt-get install -y curl sed grep graphviz libjpeg-dev zlib1g-dev && \ + apt-get install -y python-numpy python-matplotlib gcc g++ gfortran && \ + apt-get install -y automake clang-3.8 llvm-3.8 libclang-3.8-dev && \ + apt-get clean -y + +RUN pip install --upgrade pip && \ + pip install 'protobuf==3.1.0.post1' && \ + pip install -U wheel pillow BeautifulSoup && \ + pip install -U docopt PyYAML sphinx && \ + pip install -U sphinx_rtd_theme recommonmark jupyter + +RUN curl -sSL https://cmake.org/files/v3.4/cmake-3.4.1.tar.gz | tar -xz && \ + cd cmake-3.4.1 && ./bootstrap && make -j4 && make install && \ + cd .. && rm -rf cmake-3.4.1 ARG WITH_AVX ARG WITH_DOC -ARG WITH_SWIG_PY ARG WITH_STYLE_CHECK ENV WITH_GPU=ON ENV WITH_AVX=${WITH_AVX:-ON} ENV WITH_DOC=${WITH_DOC:-ON} -ENV WITH_SWIG_PY=${WITH_SWIG_PY:-ON} ENV WITH_STYLE_CHECK=${WITH_STYLE_CHECK:-OFF} RUN mkdir /paddle diff --git a/paddle/scripts/travis/before_install.osx.sh b/paddle/scripts/travis/before_install.osx.sh index 7036f971fdd7bac68b67c7b5a92e50c352e214c1..80f031a74e7052d183b5ef21d432476ff1cce722 100755 --- a/paddle/scripts/travis/before_install.osx.sh +++ b/paddle/scripts/travis/before_install.osx.sh @@ -1,6 +1,4 @@ #!/bin/bash brew update brew tap homebrew/science -brew install python -sudo pip install --upgrade protobuf -brew install swig openblas md5sha1sum protobuf +brew install openblas swig md5sha1sum diff --git a/paddle/scripts/travis/build_and_test.sh b/paddle/scripts/travis/build_and_test.sh index fd3aeb02b21d659f783702905117fc838b93eafd..5e6350b57458594163f23cca41a546d7bd9b1eda 100755 --- a/paddle/scripts/travis/build_and_test.sh +++ b/paddle/scripts/travis/build_and_test.sh @@ -6,7 +6,7 @@ if [[ "$TRAVIS_OS_NAME" == "linux" ]]; then export PYTHONPATH=/opt/python/2.7.12/lib/python2.7/site-packages export PYTHONHOME=/opt/python/2.7.12 export PATH=/opt/python/2.7.12/bin:${PATH} - cmake .. -DON_TRAVIS=ON -DON_COVERALLS=ON -DCOVERALLS_UPLOAD=ON ${EXTRA_CMAKE_OPTS} + cmake .. -DCMAKE_Fortran_COMPILER=/usr/bin/gfortran-4.8 -DON_TRAVIS=ON -DON_COVERALLS=ON -DCOVERALLS_UPLOAD=ON ${EXTRA_CMAKE_OPTS} NRPOC=`nproc` make -j $NPROC make coveralls diff --git a/paddle/scripts/travis/docs.sh b/paddle/scripts/travis/docs.sh index bdafb145bcd4e5990f382bb890f804687c474f7c..6b43cad20b76e9abeb3cb10a726d3d8e3da5f8e2 100755 --- a/paddle/scripts/travis/docs.sh +++ b/paddle/scripts/travis/docs.sh @@ -4,7 +4,7 @@ source ./common.sh # Compile Documentation only. -cmake .. -DCMAKE_BUILD_TYPE=Debug -DWITH_GPU=OFF -DWITH_DOC=ON ${EXTRA_CMAKE_OPTS} +cmake .. -DCMAKE_BUILD_TYPE=Debug -DCMAKE_Fortran_COMPILER=/usr/bin/gfortran-4.8 -DWITH_GPU=OFF -DWITH_DOC=ON ${EXTRA_CMAKE_OPTS} make paddle_docs paddle_docs_cn # check websites for broken links diff --git a/paddle/trainer/ParameterUpdater.h b/paddle/trainer/ParameterUpdater.h index c3207e63ce72b73a57c2e40c72c5259f0ae61bc9..9e9e948b8856d2712f8894b3d14db9c795d5f694 100644 --- a/paddle/trainer/ParameterUpdater.h +++ b/paddle/trainer/ParameterUpdater.h @@ -184,7 +184,6 @@ protected: * @param para */ virtual void updateImpl(Parameter* para) {} - virtual void update(Parameter* para) {} }; /** diff --git a/paddle/utils/Compiler.h b/paddle/utils/Compiler.h new file mode 100644 index 0000000000000000000000000000000000000000..cebca5a2a3766110b83231eb0705e48800a7bda6 --- /dev/null +++ b/paddle/utils/Compiler.h @@ -0,0 +1,33 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserved. +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + http://www.apache.org/licenses/LICENSE-2.0 +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#pragma once +/** + * This header defines some useful attribute by each compiler. It is the + * abstract layer of compilers. + */ +#ifdef __GNUC__ +#define GCC_VERSION \ + (__GNUC__ * 10000 + __GNUC_MINOR__ * 100 + __GNUC_PATCHLEVEL__) +#else +#define GCC_VERSION +#endif + +/** + * __must_check macro. It make the function's return value must be used, + * otherwise it will raise a compile warning. And also Paddle treat all compile + * warnings as errors. + */ +#if GCC_VERSION >= 30400 +#define __must_check __attribute__((warn_unused_result)) +#else +#define __must_check +#endif diff --git a/paddle/utils/Error.h b/paddle/utils/Error.h new file mode 100644 index 0000000000000000000000000000000000000000..2b4fbef4e015e7c6895745f220bd444f3883c121 --- /dev/null +++ b/paddle/utils/Error.h @@ -0,0 +1,130 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#pragma once + +#include +#include +#include +#include +#include +#include "Compiler.h" + +namespace paddle { + +/** + * Error is Paddle error code. It only contain a std::string as error message. + * + * + * There are two styles to return error in Paddle. + * + * 1. Return Error + * When method return a status, the return must use `__must_check` attribute. + * Example as below. + * @code{cpp} + * Error __must_check foo(); + * + * Error __must_check bar() { + * // do something. + * Status s = foo(); // invoke other method return status. + * if (!s) return s; + * // do something else. + * return Status(); + * } + * @endcode{cpp} + * + * 2. Return by parameter. + * It is another way to return an error, by using a pointer parameter. + * Example as below. + * + * @code{cpp} + * Error bar(); + * + * int foo(Error* error) { + * // Do something. + * Error s = bar(); + * if (!s) { + * *error = s; + * return 0; + * } + * // Do something else. + * if (someInternalErrorHappend) { + * *error = Error("Some dimension is too large, %d", dimension); + * return 0; + * } + * // End of method. + * return someValue; + * } + * + * Error foobar() { + * Error s; + * // do something. + * foo(&s); + * if (!s) return s; + * } + * @endcode{cpp} + * + * + * Currently there is a helper method 'check' in status, because Paddle always + * use log(FATAL) or CHECK to make program exit before. When we clean all + * log(FATAL) and CHECK in Paddle, 'check' method will be removed. + */ +class Error { +public: + /** + * Construct a no-error value. + */ + Error() {} + + /** + * @brief Create an Error use printf syntax. + */ + explicit Error(const char* fmt, ...) { + va_list ap; + va_start(ap, fmt); + constexpr size_t kBufferSize = 1024; + char buffer[kBufferSize]; + vsnprintf(buffer, kBufferSize, fmt, ap); + this->msg_.reset(new std::string(buffer)); + va_end(ap); + } + + /** + * @brief msg will return the error message. If no error, return nullptr. + */ + const char* msg() const { + if (msg_) { + return msg_->c_str(); + } else { + return nullptr; + } + } + + /** + * @brief operator bool, return True if there is no error. + */ + operator bool() const { return msg_ == nullptr; } + + /** + * @brief check this status by glog. + * @note It is a temp method used during cleaning Paddle code. It will be + * removed later. + */ + void check() const { CHECK(*this) << msg(); } + +private: + std::shared_ptr msg_; +}; + +} // namespace paddle diff --git a/paddle/utils/Util.cpp b/paddle/utils/Util.cpp index 411a64aa8d0737a8d57e62fbd0788ffaacfbc9f7..220aac1ff11e0ff263df8459f539237944b94c81 100644 --- a/paddle/utils/Util.cpp +++ b/paddle/utils/Util.cpp @@ -144,20 +144,20 @@ void runInitFunctions() { } void initMain(int argc, char** argv) { - initializeLogging(argc, argv); installLayerStackTracer(); std::string line; for (int i = 0; i < argc; ++i) { line += argv[i]; line += ' '; } - LOG(INFO) << "commandline: " << line; #ifndef GFLAGS_GFLAGS_H_ namespace gflags = google; #endif gflags::ParseCommandLineFlags(&argc, &argv, true); + initializeLogging(argc, argv); + LOG(INFO) << "commandline: " << line; CHECK_EQ(argc, 1) << "Unknown commandline argument: " << argv[1]; installProfilerSwitch(); diff --git a/paddle/utils/tests/CMakeLists.txt b/paddle/utils/tests/CMakeLists.txt index 26fafbd1ab3f2967b765b8bcb973fb745c0e6422..aa923b355377752f9b297a125f5c43c364ba9b06 100644 --- a/paddle/utils/tests/CMakeLists.txt +++ b/paddle/utils/tests/CMakeLists.txt @@ -4,6 +4,7 @@ add_simple_unittest(test_CustomStackTrace) add_simple_unittest(test_ThreadBarrier) add_simple_unittest(test_SpinLock) add_simple_unittest(test_SIMDFlags) +add_simple_unittest(test_Error) add_executable( test_CustomStackTracePrint diff --git a/paddle/utils/tests/test_Error.cpp b/paddle/utils/tests/test_Error.cpp new file mode 100644 index 0000000000000000000000000000000000000000..85156466e2cafd36d49941836c066a542dbbd60e --- /dev/null +++ b/paddle/utils/tests/test_Error.cpp @@ -0,0 +1,34 @@ +/* Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserve. + +Licensed under the Apache License, Version 2.0 (the "License"); +you may not use this file except in compliance with the License. +You may obtain a copy of the License at + + http://www.apache.org/licenses/LICENSE-2.0 + +Unless required by applicable law or agreed to in writing, software +distributed under the License is distributed on an "AS IS" BASIS, +WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +See the License for the specific language governing permissions and +limitations under the License. */ + +#include "paddle/utils/Error.h" + +#include + +TEST(Error, testAll) { + paddle::Error error; + ASSERT_TRUE(error); + error = paddle::Error("I'm the error"); + ASSERT_FALSE(error); + ASSERT_STREQ("I'm the error", error.msg()); + + error = paddle::Error("error2"); + ASSERT_FALSE(error); + ASSERT_STREQ("error2", error.msg()); + + int i = 3; + auto error3 = paddle::Error("error%d", i); + ASSERT_FALSE(error3); + ASSERT_STREQ("error3", error3.msg()); +} diff --git a/proto/ModelConfig.proto b/proto/ModelConfig.proto index 3a9d339976fff91d79e7459ad5984cf78ea8990a..0456404832c301d8ceb9338d32da0cea9eae5234 100644 --- a/proto/ModelConfig.proto +++ b/proto/ModelConfig.proto @@ -255,6 +255,13 @@ message PriorBoxConfig { repeated float variance = 4; } +message PadConfig { + required ImageConfig image_conf = 1; + repeated uint32 pad_c = 2; + repeated uint32 pad_h = 3; + repeated uint32 pad_w = 4; +} + message LayerInputConfig { required string input_layer_name = 1; optional string input_parameter_name = 2; @@ -271,6 +278,7 @@ message LayerInputConfig { optional MaxOutConfig maxout_conf = 11; optional SppConfig spp_conf = 12; optional PriorBoxConfig priorbox_conf = 13; + optional PadConfig pad_conf = 14; } message LayerConfig { diff --git a/python/CMakeLists.txt b/python/CMakeLists.txt index 1cda4762eb2a55175d6c9faee98aaeaa1f763890..ee7a5bff84ca96ef1010fa7430356722f807fb0f 100644 --- a/python/CMakeLists.txt +++ b/python/CMakeLists.txt @@ -4,11 +4,13 @@ set(OUTPUT_DIR file(GLOB TRAINER_PY_FILES . ./paddle/trainer/*.py) file(GLOB HELPERS_PY_FILES . ./paddle/trainer_config_helpers/*.py) file(GLOB UTILS_PY_FILES . ./paddle/utils/*.py) +file(GLOB V2_PY_FILES . ./paddle/v2/*.py) set(PY_FILES paddle/__init__.py ${TRAINER_PY_FILES} ${HELPERS_PY_FILES} - ${UTILS_PY_FILES}) + ${UTILS_PY_FILES} + ${V2_PY_FILES}) configure_file(${CMAKE_CURRENT_SOURCE_DIR}/setup.py.in ${CMAKE_CURRENT_BINARY_DIR}/setup.py) diff --git a/python/paddle/trainer/config_parser.py b/python/paddle/trainer/config_parser.py index 674b5ac58b6febd914cb36c75356d8aa70a908b1..6701eced60d068312a1a866a6312002f9f5207f7 100644 --- a/python/paddle/trainer/config_parser.py +++ b/python/paddle/trainer/config_parser.py @@ -493,6 +493,7 @@ class Input(Cfg): block_expand=None, maxout=None, spp=None, + pad=None, format=None, nnz=None, is_static=None, @@ -844,6 +845,12 @@ class SpatialPyramidPool(Cfg): self.add_keys(locals()) +@config_class +class Pad(Cfg): + def __init__(self, channels, pad_c, pad_h, pad_w): + self.add_keys(locals()) + + @config_class class Norm(Cfg): def __init__(self, @@ -1102,7 +1109,7 @@ def parse_bilinear(bilinear, input_layer_name, bilinear_conf): bilinear_conf.out_size_y = bilinear.out_size_y -def parse_pool(pool, input_layer_name, pool_conf): +def parse_pool(pool, input_layer_name, pool_conf, ceil_mode): pool_conf.pool_type = pool.pool_type config_assert(pool.pool_type in [ 'max-projection', 'avg-projection', 'cudnn-max-pool', 'cudnn-avg-pool' @@ -1127,10 +1134,10 @@ def parse_pool(pool, input_layer_name, pool_conf): pool_conf.padding_y = default(pool.padding_y, pool_conf.padding) pool_conf.output_x = cnn_output_size(pool_conf.img_size, pool_conf.size_x, pool_conf.padding, pool_conf.stride, - False) + not ceil_mode) pool_conf.output_y = cnn_output_size(pool_conf.img_size_y, pool_conf.size_y, pool_conf.padding_y, - pool_conf.stride_y, False) + pool_conf.stride_y, not ceil_mode) def parse_spp(spp, input_layer_name, spp_conf): @@ -1803,9 +1810,8 @@ class ConvTransLayer(ConvTransLayerBase): @config_layer('norm') class NormLayer(LayerBase): - def __init__(self, name, inputs, device=None): - super(NormLayer, self).__init__( - name, 'norm', 0, inputs=inputs, device=device) + def __init__(self, name, inputs, **xargs): + super(NormLayer, self).__init__(name, 'norm', 0, inputs=inputs, **xargs) for input_index in xrange(len(self.inputs)): input_layer = self.get_input_layer(input_index) norm_conf = self.config.inputs[input_index].norm_conf @@ -1817,23 +1823,22 @@ class NormLayer(LayerBase): @config_layer('pool') class PoolLayer(LayerBase): - def __init__(self, name, inputs, device=None): - super(PoolLayer, self).__init__( - name, 'pool', 0, inputs=inputs, device=device) + def __init__(self, name, inputs, ceil_mode=True, **xargs): + super(PoolLayer, self).__init__(name, 'pool', 0, inputs=inputs, **xargs) for input_index in xrange(len(self.inputs)): input_layer = self.get_input_layer(input_index) pool_conf = self.config.inputs[input_index].pool_conf parse_pool(self.inputs[input_index].pool, input_layer.name, - pool_conf) + pool_conf, ceil_mode) self.set_cnn_layer(name, pool_conf.output_y, pool_conf.output_x, pool_conf.channels) @config_layer('spp') class SpatialPyramidPoolLayer(LayerBase): - def __init__(self, name, inputs, device=None): + def __init__(self, name, inputs, **xargs): super(SpatialPyramidPoolLayer, self).__init__( - name, 'spp', 0, inputs=inputs, device=device) + name, 'spp', 0, inputs=inputs, **xargs) for input_index in xrange(len(self.inputs)): input_layer = self.get_input_layer(input_index) spp_conf = self.config.inputs[input_index].spp_conf @@ -1842,6 +1847,25 @@ class SpatialPyramidPoolLayer(LayerBase): self.set_cnn_layer(name, 1, output_x, spp_conf.image_conf.channels) +@config_layer('pad') +class PadLayer(LayerBase): + def __init__(self, name, inputs, **xargs): + super(PadLayer, self).__init__(name, 'pad', 0, inputs=inputs, **xargs) + pad = self.inputs[0].pad + self.config.inputs[0].pad_conf.pad_c.extend(pad.pad_c) + self.config.inputs[0].pad_conf.pad_h.extend(pad.pad_h) + self.config.inputs[0].pad_conf.pad_w.extend(pad.pad_w) + + input_layer = self.get_input_layer(0) + image_conf = self.config.inputs[0].pad_conf.image_conf + parse_image(pad, input_layer.name, image_conf) + out_ch = pad.channels + pad.pad_c[0] + pad.pad_c[1] + out_h = image_conf.img_size_y + pad.pad_h[0] + pad.pad_h[1] + out_w = image_conf.img_size + pad.pad_w[0] + pad.pad_w[1] + self.set_cnn_layer(name, out_h, out_w, out_ch) + self.config.size = out_ch * out_h * out_w + + @config_layer('batch_norm') class BatchNormLayer(LayerBase): layer_type = 'batch_norm' @@ -1851,7 +1875,6 @@ class BatchNormLayer(LayerBase): inputs, active_type="linear", bias=True, - device=None, use_global_stats=True, moving_average_fraction=0.9, batch_norm_type=None, @@ -1893,7 +1916,6 @@ class BatchNormLayer(LayerBase): 0, active_type=active_type, inputs=inputs, - device=device, **xargs) if use_global_stats is not None: @@ -1927,9 +1949,9 @@ class BatchNormLayer(LayerBase): @config_layer('trans') class TransLayer(LayerBase): - def __init__(self, name, inputs, device=None): + def __init__(self, name, inputs, **xargs): super(TransLayer, self).__init__( - name, 'trans', 0, inputs=inputs, device=device) + name, 'trans', 0, inputs=inputs, **xargs) config_assert( len(self.inputs) == 1, 'TransLayer must have one and only one input') @@ -1938,9 +1960,9 @@ class TransLayer(LayerBase): @config_layer('resize') class ResizeLayer(LayerBase): - def __init__(self, name, size, inputs, device=None): + def __init__(self, name, size, inputs, **xargs): super(ResizeLayer, self).__init__( - name, 'resize', size=size, inputs=inputs, device=device) + name, 'resize', size=size, inputs=inputs, **xargs) config_assert( len(self.inputs) == 1, 'ResizeLayer must have one and only one input') @@ -1948,9 +1970,9 @@ class ResizeLayer(LayerBase): @config_layer('blockexpand') class BlockExpandLayer(LayerBase): - def __init__(self, name, inputs, device=None): + def __init__(self, name, inputs, **xargs): super(BlockExpandLayer, self).__init__( - name, 'blockexpand', 0, inputs=inputs, device=device) + name, 'blockexpand', 0, inputs=inputs, **xargs) for input_index in xrange(len(self.inputs)): input_layer = self.get_input_layer(input_index) parse_block_expand( @@ -2628,7 +2650,7 @@ class AverageLayer(LayerBase): @config_layer('cos') class CosSimLayer(LayerBase): - def __init__(self, name, inputs, cos_scale=5, device=None): + def __init__(self, name, inputs, cos_scale=1, device=None): super(CosSimLayer, self).__init__( name, 'cos', 1, inputs=inputs, device=device) config_assert(len(self.inputs) == 2, 'CosSimLayer must have 2 inputs') diff --git a/python/paddle/trainer_config_helpers/layers.py b/python/paddle/trainer_config_helpers/layers.py index 9b6e5774bc82dc05e14a2565fa9cce98764adf04..85a28e14aeb8ebbddb5247f9ae3c3afee075465d 100644 --- a/python/paddle/trainer_config_helpers/layers.py +++ b/python/paddle/trainer_config_helpers/layers.py @@ -108,6 +108,7 @@ __all__ = [ 'print_layer', 'priorbox_layer', 'spp_layer', + 'pad_layer', ] @@ -170,6 +171,7 @@ class LayerType(object): BLOCK_EXPAND = "blockexpand" MAXOUT = "maxout" SPP_LAYER = "spp" + PAD_LAYER = "pad" PRINT_LAYER = "print" PRIORBOX_LAYER = "priorbox" @@ -1673,7 +1675,7 @@ def trans_layer(input, name=None, layer_attr=None): @wrap_name_default() @layer_support() -def cos_sim(a, b, scale=5, size=1, name=None, layer_attr=None): +def cos_sim(a, b, scale=1, size=1, name=None, layer_attr=None): """ Cosine Similarity Layer. The cosine similarity equation is here. @@ -1979,7 +1981,8 @@ def img_pool_layer(input, layer_attr=None, pool_size_y=None, stride_y=None, - padding_y=None): + padding_y=None, + ceil_mode=True): """ Image pooling Layer. @@ -2010,6 +2013,23 @@ def img_pool_layer(input, :type stride_y: int|None :param layer_attr: Extra Layer attribute. :type layer_attr: ExtraLayerAttribute + :param ceil_mode: Wether to use ceil mode to calculate output height and with. + Defalut is True. If set false, Otherwise use floor. + + - ceil_mode=True: + + .. math:: + + w = 1 + int(ceil(input_width + 2 * padding - pool_size) / float(stride)) + h = 1 + int(ceil(input_height + 2 * padding_y - pool_size_y) / float(stride_y)) + + - ceil_mode=False: + + .. math:: + + w = 1 + int(floor(input_width + 2 * padding - pool_size) / float(stride)) + h = 1 + int(floor(input_height + 2 * padding_y - pool_size_y) / float(stride_y)) + :type ceil_mode: bool :return: LayerOutput object. :rtype: LayerOutput """ @@ -2047,6 +2067,7 @@ def img_pool_layer(input, stride_y=stride_y, padding_y=padding_y)) ], + ceil_mode=ceil_mode, **ExtraLayerAttribute.to_kwargs(layer_attr)) return LayerOutput( name, @@ -3488,9 +3509,6 @@ def conv_projection(input, groups=1, param_attr=None): """ - ConvProjection with a layer as input. - It performs element-wise multiplication with weight. - Different from img_conv_layer and conv_op, conv_projection is an Projection, which can be used in mixed_layer and conat_layer. It use cudnn to implement conv and only support GPU mode. @@ -3499,7 +3517,7 @@ def conv_projection(input, .. code-block:: python - proj = conv_projection(img=input1, + proj = conv_projection(input=input1, filter_size=3, num_filters=64, num_channels=64) @@ -3582,6 +3600,109 @@ def conv_projection(input, return proj +@wrap_name_default("pad") +@layer_support() +def pad_layer(input, + pad_c=None, + pad_h=None, + pad_w=None, + name=None, + layer_attr=None): + """ + This operation pads zeros to the input data according to pad_c,pad_h + and pad_w. pad_c, pad_h, pad_w specifies the which dimension and size + of padding. And the input data shape is NCHW. + + For example, pad_c=[2,3] means padding 2 zeros before the + input data and 3 zeros after the input data in channel dimension. + pad_h means padding zeros in height dimension. pad_w means padding zeros + in width dimension. + + For example, + + .. code-block:: + + input(2,2,2,3) = [ + [ [[1,2,3], [3,4,5]], + [[2,3,5], [1,6,7]] ], + [ [[4,3,1], [1,8,7]], + [[3,8,9], [2,3,5]] ] + ] + + pad_c=[1,1], pad_h=[0,0], pad_w=[0,0] + output(2,4,2,3) = [ + [ [[0,0,0], [0,0,0]], + [[1,2,3], [3,4,5]], + [[2,3,5], [1,6,7]], + [[0,0,0], [0,0,0]] ], + [ [[0,0,0], [0,0,0]], + [[4,3,1], [1,8,7]], + [[3,8,9], [2,3,5]], + [[0,0,0], [0,0,0]] ] + ] + + The simply usage is: + + .. code-block:: python + + pad = pad_layer(input=ipt, + pad_c=[4,4], + pad_h=[0,0], + pad_w=[2,2]) + + :param input: layer's input. + :type input: LayerOutput + :param pad_c: padding size in channel dimension. + :type pad_c: list|None + :param pad_h: padding size in height dimension. + :type pad_h: list|None + :param pad_w: padding size in width dimension. + :type pad_w: list|None + :param layer_attr: Extra Layer Attribute. + :type layer_attr: ExtraLayerAttribute + :param name: layer name. + :type name: basestring + :return: LayerOutput object. + :rtype: LayerOutput + """ + if pad_c is not None: + assert isinstance(pad_c, collections.Sequence) and len(pad_c) == 2 + else: + pad_c = [0, 0] + + if pad_h is not None: + assert isinstance(pad_h, collections.Sequence) and len(pad_h) == 2 + else: + pad_h = [0, 0] + + if pad_w is not None: + assert isinstance(pad_w, collections.Sequence) and len(pad_w) == 2 + else: + pad_w = [0, 0] + + assert input.num_filters is not None + in_ch = input.num_filters + out_ch = in_ch + pad_c[0] + pad_c[1] + + l = Layer( + name=name, + type=LayerType.PAD_LAYER, + inputs=Input( + input.name, + pad=Pad( + channels=in_ch, + pad_c=pad_c, + pad_h=pad_h, + pad_w=pad_w, )), + **ExtraLayerAttribute.to_kwargs(layer_attr)) + return LayerOutput( + name, + layer_type=LayerType.PAD_LAYER, + parents=[input], + num_filters=out_ch, + size=l.config.size) + + @wrap_name_default() @layer_support() def conv_shift_layer(a, b, name=None, layer_attr=None): diff --git a/python/paddle/trainer_config_helpers/tests/CMakeLists.txt b/python/paddle/trainer_config_helpers/tests/CMakeLists.txt index 403aafabe9143472dd2f0857ecd25f7acf515b6c..93dd7796c246ae81a146759df7e0c19e334375f1 100644 --- a/python/paddle/trainer_config_helpers/tests/CMakeLists.txt +++ b/python/paddle/trainer_config_helpers/tests/CMakeLists.txt @@ -9,17 +9,10 @@ add_test(NAME test_reset_hook ${PYTHON_EXECUTABLE} ${PROJ_ROOT}/python/paddle/trainer_config_helpers/tests/test_reset_hook.py WORKING_DIRECTORY ${PROJ_ROOT}/python/paddle) -if (PROTOBUF_3) - add_paddle_exe(protobuf_equal - ProtobufEqualMain.cpp) - add_test(NAME test_layerHelpers - COMMAND - ${PROJ_ROOT}/python/paddle/trainer_config_helpers/tests/configs/run_tests.sh ${PYTHON_EXECUTABLE} - ${CMAKE_CURRENT_BINARY_DIR}/protobuf_equal - ) -else() - add_test(NAME test_layerHelpers - COMMAND - ${PROJ_ROOT}/python/paddle/trainer_config_helpers/tests/configs/run_tests.sh ${PYTHON_EXECUTABLE} - ) -endif() +add_paddle_exe(protobuf_equal + ProtobufEqualMain.cpp) +add_test(NAME test_layerHelpers + COMMAND + ${PROJ_ROOT}/python/paddle/trainer_config_helpers/tests/configs/run_tests.sh ${PYTHON_EXECUTABLE} + ${CMAKE_CURRENT_BINARY_DIR}/protobuf_equal +) diff --git a/python/paddle/trainer_config_helpers/tests/configs/protostr/test_ntm_layers.protostr b/python/paddle/trainer_config_helpers/tests/configs/protostr/test_ntm_layers.protostr index b30bbb2a4e24d74ebe1d6c8eda8be5aa09217f6d..c1bfdf1b19c61d096c25af061c6fbb3bbfc50265 100644 --- a/python/paddle/trainer_config_helpers/tests/configs/protostr/test_ntm_layers.protostr +++ b/python/paddle/trainer_config_helpers/tests/configs/protostr/test_ntm_layers.protostr @@ -79,7 +79,7 @@ layers { inputs { input_layer_name: "b" } - cos_scale: 5 + cos_scale: 1 } layers { name: "__cos_sim_1__" @@ -92,7 +92,7 @@ layers { inputs { input_layer_name: "c" } - cos_scale: 5 + cos_scale: 1 } layers { name: "__sum_to_one_norm_layer_0__" diff --git a/python/paddle/trainer_config_helpers/tests/configs/run_tests.sh b/python/paddle/trainer_config_helpers/tests/configs/run_tests.sh index a37eb6439e6d2803a417883f0aed2a5d56d059b9..c8a3b190b19148ddb701020f5be55c4c29a17079 100755 --- a/python/paddle/trainer_config_helpers/tests/configs/run_tests.sh +++ b/python/paddle/trainer_config_helpers/tests/configs/run_tests.sh @@ -2,16 +2,18 @@ cd `dirname $0` set -e +PYTHON_EXEC=$1 +COMPARE_PROTO_UTIL=$2 protostr=`dirname $0`/protostr files=`ls $protostr | grep -v "unittest"` -./generate_protostr.sh $1 +./generate_protostr.sh ${PYTHON_EXEC} . ./file_list.sh -if [ -z $1 ]; then +if [ -z ${COMPARE_PROTO_UTIL} ]; then for file in $files do base_protostr=$protostr/$file @@ -22,20 +24,20 @@ if [ -z $1 ]; then else for file in ${configs[*]} do - if ! $1 $protostr/$file.protostr $protostr/$file.protostr.unittest; then + if ! ${COMPARE_PROTO_UTIL} $protostr/$file.protostr $protostr/$file.protostr.unittest; then diff $protostr/$file.protostr $protostr/$file.protostr.unittest -u fi - if ! $1 $protostr/$file.protostr $protostr/$file.protostr.non_file_config.unittest; then + if ! ${COMPARE_PROTO_UTIL} $protostr/$file.protostr $protostr/$file.protostr.non_file_config.unittest; then diff $protostr/$file.protostr $protostr/$file.protostr.non_file_config.unittest -u fi done for file in ${whole_configs[*]} do - if ! $1 $protostr/$file.protostr $protostr/$file.protostr.unittest --whole; then + if ! ${COMPARE_PROTO_UTIL} $protostr/$file.protostr $protostr/$file.protostr.unittest --whole; then diff $protostr/$file.protostr $protostr/$file.protostr.unittest -u fi - if ! $1 $protostr/$file.protostr $protostr/$file.protostr.non_file_config.unittest --whole; then + if ! ${COMPARE_PROTO_UTIL} $protostr/$file.protostr $protostr/$file.protostr.non_file_config.unittest --whole; then diff $protostr/$file.protostr $protostr/$file.protostr.non_file_config.unittest -u fi done diff --git a/python/paddle/trainer_config_helpers/tests/configs/test_pad.py b/python/paddle/trainer_config_helpers/tests/configs/test_pad.py new file mode 100644 index 0000000000000000000000000000000000000000..bb5f13410dbbbaeea9e28c271d33a15fb3000dcf --- /dev/null +++ b/python/paddle/trainer_config_helpers/tests/configs/test_pad.py @@ -0,0 +1,21 @@ +from paddle.trainer_config_helpers import * + +settings(batch_size=1000, learning_rate=1e-5) + +data = data_layer(name='data', size=2304, height=48, width=42) + +conv = img_conv_layer( + input=data, + filter_size=3, + num_channels=1, + num_filters=16, + padding=1, + act=LinearActivation(), + bias_attr=True) + +pool = img_pool_layer( + input=conv, num_channels=8, pool_size=2, stride=2, pool_type=MaxPooling()) + +pad = pad_layer(input=pool, pad_c=[2, 3], pad_h=[1, 2], pad_w=[3, 1]) + +outputs(pad) diff --git a/python/paddle/v2/__init__.py b/python/paddle/v2/__init__.py new file mode 100644 index 0000000000000000000000000000000000000000..f662d6826321eb840739382558f76327d27b5847 --- /dev/null +++ b/python/paddle/v2/__init__.py @@ -0,0 +1,13 @@ +# Copyright (c) 2016 PaddlePaddle Authors. All Rights Reserved +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. diff --git a/python/setup.py.in b/python/setup.py.in index b66a42e87c78701e9eb26b1b7dc8f46a95035a76..1e1324eea825ab1945a38cb43eceec29a4ebb1a1 100644 --- a/python/setup.py.in +++ b/python/setup.py.in @@ -4,7 +4,8 @@ packages=['paddle', 'paddle.proto', 'paddle.trainer', 'paddle.trainer_config_helpers', - 'paddle.utils'] + 'paddle.utils', + 'paddle.v2'] setup(name='paddle', version='${PADDLE_VERSION}',