Merge pull request #663 from gangliao/docker

Refine docker install doc and FAQ for gpu driver

Author: reyoung

doc_cn/faq/index.rst

@@ -203,7 +203,23 @@ PaddlePaddle的参数使用名字 :code:`name` 作为参数的ID，相同名字
 
 * 卸载PaddlePaddle包 :code:`pip uninstall paddle`, 清理掉老旧的PaddlePaddle安装包，使得单元测试有一个干净的环境。如果PaddlePaddle包已经在python的site-packages里面，单元测试会引用site-packages里面的python包，而不是源码目录里 :code:`/python` 目录下的python包。同时，即便设置 :code:`PYTHONPATH` 到 :code:`/python` 也没用，因为python的搜索路径是优先已经安装的python包。
 
-9. CMake源码编译, 找到的PythonLibs和PythonInterp版本不一致
+
+9. 运行Docker GPU镜像出现 "CUDA driver version is insufficient"
+----------------------------------------------------------------
+
+用户在使用PaddlePaddle GPU的Docker镜像的时候，常常出现 `Cuda Error: CUDA driver version is insufficient for CUDA runtime version`, 原因在于没有把机器上CUDA相关的驱动和库映射到容器内部。
+具体的解决方法是：
+
+..  code-block:: bash
+
+    $ export CUDA_SO="$(\ls usr/lib64/libcuda* | xargs -I{} echo '-v {}:{}') $(\ls /usr/lib64/libnvidia* | xargs -I{} echo '-v {}:{}')"
+    $ export DEVICES=$(\ls /dev/nvidia* | xargs -I{} echo '--device {}:{}')
+    $ docker run ${CUDA_SO} ${DEVICES} -it paddledev/paddlepaddle:latest-gpu
+
+更多关于Docker的安装与使用, 请参考 `PaddlePaddle Docker 文档 <http://www.paddlepaddle.org/doc_cn/build_and_install/install/docker_install.html>`_ 。
+
+
+10. CMake源码编译, 找到的PythonLibs和PythonInterp版本不一致
 ----------------------------------------------------------
 
 这是目前CMake寻找Python的逻辑存在缺陷，如果系统安装了多个Python版本，CMake找到的Python库和Python解释器版本可能有不一致现象，导致编译PaddlePaddle失败。正确的解决方法是，

doc_cn/howto/build_docker_image.rst

@@ -1,35 +1,35 @@
-构建PaddlePaddle的Docker Image
-==============================
-PaddlePaddle的Docker Image构建源码放置在 ``${源码根目录}/paddle/scripts/docker/`` 目录下。该目录有三类文件：
-
-- Dockerfile：Docker Image的描述文件，包括构建步骤、各种参数和维护人员等。
-  
-  - 一共维护了12个Dockerfile，Dockerfile.m4是它们的模板。
-  - PaddlePaddle中所有的Image都基于ubuntu 14.04。
-
-- build.sh：Docker Image的构建脚本，使用方式见下一小节。
-- generate.sh：通过Dockerfile.m4模板生成不同的Dockerfile。
-
-使用脚本构建Docker Image
-------------------------
-
-进入源码目录，执行 ``docker build`` 命令，即可在本地编译出PaddlePaddle的镜像。简单的使用样例为
-
-..  code-block:: bash
-
-    cd ${源码根目录}/paddle/scripts/docker/
-    docker build --build-arg LOWEST_DL_SPEED=50K \
-                 --build-arg WITH_GPU=ON \
-                 --tag  paddle_gpu:latest .
-
-其中，``--build-arg`` 传入的配置参数包括:
-
-- LOWEST\_DL\_SPEED\: 在多线程下载过程中，设置下载线程的最低速度。
-
-  - 默认单位是Bytes，但可以传入10K、10M、或10G等这样的单位。
-  - 如果小于这个速度，那么这个线程将会关闭。当所有的线程都关闭了，那么下载进程将会重启。
--  WITH\_GPU\: ON or OFF，是否开启GPU功能。注意，
-  - **编译** PaddlePaddle的GPU版本 **不一定** 要在具有GPU的机器上进行。
-  - **运行** PaddlePaddle的GPU版本 **一定** 要在具有GPU的机器上运行。
-
-注意：所有Image的构建在Docker 1.12版本测试通过, 低于1.12的版本并没有测试。原因是旧版本可能缺乏 ``--build-arg`` 参数，从而不能在运行编译命令的时候接受参数。
+构建PaddlePaddle的Docker Image
+==============================
+PaddlePaddle的Docker Image构建源码放置在 ``${源码根目录}/paddle/scripts/docker/`` 目录下。该目录有三类文件：
+
+- Dockerfile：Docker Image的描述文件，包括构建步骤、各种参数和维护人员等。
+  
+  - 一共维护了12个Dockerfile，Dockerfile.m4是它们的模板。
+  - PaddlePaddle中所有的Image都基于ubuntu 14.04。
+
+- build.sh：Docker Image的构建脚本，使用方式见下一小节。
+- generate.sh：通过Dockerfile.m4模板生成不同的Dockerfile。
+
+使用脚本构建Docker Image
+------------------------
+
+进入源码目录，执行 ``docker build`` 命令，即可在本地编译出PaddlePaddle的镜像。简单的使用样例为
+
+..  code-block:: bash
+
+    cd ${源码根目录}/paddle/scripts/docker/
+    docker build --build-arg LOWEST_DL_SPEED=50K \
+                 --build-arg WITH_GPU=ON \
+                 --tag  paddle_gpu:latest .
+
+其中，``--build-arg`` 传入的配置参数包括:
+
+- LOWEST\_DL\_SPEED\: 在多线程下载过程中，设置下载线程的最低速度。
+
+  - 默认单位是Bytes，但可以传入10K、10M、或10G等这样的单位。
+  - 如果小于这个速度，那么这个线程将会关闭。当所有的线程都关闭了，那么下载进程将会重启。
+-  WITH\_GPU\: ON or OFF，是否开启GPU功能。注意，
+  - **编译** PaddlePaddle的GPU版本 **不一定** 要在具有GPU的机器上进行。
+  - **运行** PaddlePaddle的GPU版本 **一定** 要在具有GPU的机器上运行。
+
+注意：所有Image的构建在Docker 1.12版本测试通过, 低于1.12的版本并没有测试。原因是旧版本可能缺乏 ``--build-arg`` 参数，从而不能在运行编译命令的时候接受参数。

paddle/cuda/src/hl_cuda_cnn.cu

@@ -16,6 +16,7 @@ limitations under the License. */
 #include <float.h>
 #include "hl_base.h"
 #include "hl_cnn.h"
+#include "hl_device_functions.cuh"
 
 __global__ void KeFeature2col(size_t n, size_t height, const real* data_im,
                               size_t blockH, size_t blockW, size_t width,
@@ -641,10 +642,10 @@ __global__ void KeBilinearInterpBw(real* in,
     real* inPos =
       &in[outIdH * inputW + channelId * inImgSize + inImgIdy * inImgW + inImgIdx];
     const real* outPos = &out[outIdH * outputW + outIdW];
-    atomicAdd(&inPos[0], h2lambda * w2lambda * outPos[0]);
-    atomicAdd(&inPos[wId], h2lambda * w1lambda * outPos[0]);
-    atomicAdd(&inPos[hId * inImgW], h1lambda * w2lambda * outPos[0]);
-    atomicAdd(&inPos[hId * inImgW + wId], h1lambda * w1lambda * outPos[0]);
+    paddle::paddleAtomicAdd(&inPos[0], h2lambda * w2lambda * outPos[0]);
+    paddle::paddleAtomicAdd(&inPos[wId], h2lambda * w1lambda * outPos[0]);
+    paddle::paddleAtomicAdd(&inPos[hId * inImgW], h1lambda * w2lambda * outPos[0]);
+    paddle::paddleAtomicAdd(&inPos[hId * inImgW + wId], h1lambda * w1lambda * outPos[0]);
   }
 }