test_dnn are passing. fixed PoolingNdDescriptor

theano/sandbox/gpuarray/dnn.py

@@ -1048,9 +1048,13 @@ class GpuDnnPoolDesc(Op):
   static const int win[%(nd)d] = {%(win)s};
   static const int pad[%(nd)d] = {%(pad)s};
   static const int str[%(nd)d] = {%(str)s};
-  err = cudnnSetPoolingNdDescriptor(
-    %(desc)s, %(mode_flag)s, %(nd)d,
-    win, pad, str);
+
+#if CUDNN_VERSION >= 5000
+    err = cudnnSetPoolingNdDescriptor(%(desc)s, %(mode_flag)s, CUDNN_PROPAGATE_NAN, %(nd)d, win, pad, str);
+#else
+    err = cudnnSetPoolingNdDescriptor(%(desc)s, %(mode_flag)s, %(nd)d, win, pad, str);
+#endif
+
   if (err != CUDNN_STATUS_SUCCESS) {
     PyErr_Format(PyExc_RuntimeError, "could not set op descriptor: %%s",
                  cudnnGetErrorString(err));
@@ -1063,7 +1067,7 @@ class GpuDnnPoolDesc(Op):
            str=', '.join(map(str, self.stride)))
 
     def c_code_cache_version(self):
-        return (3, version())
+        return (4, version())
 
 
 class GpuDnnPool(DnnBase):

theano/sandbox/gpuarray/dnn_base.c

@@ -51,6 +51,8 @@ c_set_tensorNd(PyGpuArrayObject *var, cudnnTensorDescriptor_t desc) {
 static int
 c_set_filter(PyGpuArrayObject *var, cudnnFilterDescriptor_t desc) {
   cudnnDataType_t dt;
+  cudnnStatus_t err;
+
   if (!GpuArray_IS_C_CONTIGUOUS(&var->ga)) {
     PyErr_SetString(PyExc_ValueError,
 		    "Only contiguous filters (kernels) are supported.");
@@ -86,10 +88,10 @@ c_set_filter(PyGpuArrayObject *var, cudnnFilterDescriptor_t desc) {
     dims[i] = PyGpuArray_DIM(var, i);
   }
 
-#if CUDNN_VERSION > 5000
-    cudnnStatus_t err = cudnnSetFilterNdDescriptor(desc, dt, CUDNN_TENSOR_NCHW, nd, dims);
+#if CUDNN_VERSION >= 5000
+    err = cudnnSetFilterNdDescriptor(desc, dt, CUDNN_TENSOR_NCHW, nd, dims);
 #else
-    cudnnStatus_t err = cudnnSetFilterNdDescriptor(desc, dt, nd, dims);
+    err = cudnnSetFilterNdDescriptor(desc, dt, nd, dims);
 #endif
 
   if (err != CUDNN_STATUS_SUCCESS) {

theano/sandbox/gpuarray/dnn_pool.c

@@ -24,7 +24,7 @@ if ((APPLY_SPECIFIC(err) = cudnnCreateTensorDescriptor(&APPLY_SPECIFIC(output)))
 }
 if ((APPLY_SPECIFIC(err) = cudnnCreatePoolingDescriptor(&APPLY_SPECIFIC(pool))) != CUDNN_STATUS_SUCCESS) {
   PyErr_Format(PyExc_MemoryError, "could not allocate pooling descriptor"
-                "(pool): %s", cudnnGetErrorString(APPLY_SPECIFIC(err)));  
+                "(pool): %s", cudnnGetErrorString(APPLY_SPECIFIC(err)));
   FAIL;
 }
 
@@ -38,7 +38,7 @@ if (APPLY_SPECIFIC(pool) != NULL) { cudnnDestroyPoolingDescriptor(APPLY_SPECIFIC
 #section support_code_struct
 
 int APPLY_SPECIFIC(dnn_pool)(PyGpuArrayObject *img,
-                             PyArrayObject *ws, 
+                             PyArrayObject *ws,
                              PyArrayObject *stride,
                              PyArrayObject *pad,
                              PyGpuArrayObject **out,
@@ -69,7 +69,12 @@ int APPLY_SPECIFIC(dnn_pool)(PyGpuArrayObject *img,
   for(int i = 0; i < ndims; i++) {
      s[i] = *((npy_intp*)PyArray_GETPTR1(stride, i));
   }
+
+#if CUDNN_VERSION >= 5000
+  err = cudnnSetPoolingNdDescriptor(APPLY_SPECIFIC(pool), MODE_FLAG, CUDNN_PROPAGATE_NAN, ndims, w, p, s);
+#else
   err = cudnnSetPoolingNdDescriptor(APPLY_SPECIFIC(pool), MODE_FLAG, ndims, w, p, s);
+#endif
 
   if (err != CUDNN_STATUS_SUCCESS) {
     PyErr_Format(PyExc_RuntimeError, "could not set op descriptor %s", cudnnGetErrorString(err));