Make it compile on cuda < 6.0 and add fix 0-size cases.

theano/sandbox/cuda/cuda_ndarray.cu

@@ -4190,7 +4190,8 @@ int CudaNdarray_sgemv(float alpha, const CudaNdarray * A, const CudaNdarray * B,
     if (sa_1 == 0)
         sa_1 = 1;
 
-    cublasStatus_t err;
+    // This is important because we can end up not calling Sgemv at all
+    cublasStatus_t err = CUBLAS_STATUS_SUCCESS;
     if (CudaNdarray_SIZE(C)) {
         if ((CudaNdarray_HOST_DIMS(A)[0] <= 1)
             || ((CudaNdarray_HOST_STRIDES(A)[0] == 1)
@@ -4306,7 +4307,8 @@ int CudaNdarray_sger(float alpha, const CudaNdarray * x, const CudaNdarray * y,
     int sa_1 = (CudaNdarray_HOST_DIMS(A)[1] > 1) ? CudaNdarray_HOST_STRIDES(A)[1]
                                                  : CudaNdarray_HOST_DIMS(A)[0];
 
-    cublasStatus_t err;
+    // This is important because we can end up not calling Sger at all
+    cublasStatus_t err = CUBLAS_STATUS_SUCCESS;
     if(CudaNdarray_SIZE(A)){
         // If A is in col-major
         if ((CudaNdarray_HOST_DIMS(A)[0] <= 1)

theano/sandbox/cuda/cuda_ndarray.cuh

@@ -596,16 +596,20 @@ DllExport inline const char* ALWAYS_INLINE cublasGetErrorString(cublasStatus_t e
         return "the resource allocation failed";
     case CUBLAS_STATUS_INVALID_VALUE:
         return "the parameters n<0 or incx,incy=0";
+#ifdef CUBLAS_STATUS_ARCH_MISMATCH
     case CUBLAS_STATUS_ARCH_MISMATCH:
         return "required device feature not present";
+#endif
     case CUBLAS_STATUS_MAPPING_ERROR:
         return "an access to GPU memory space failed";
     case CUBLAS_STATUS_EXECUTION_FAILED:
         return "the function failed to launch on the GPU";
     case CUBLAS_STATUS_INTERNAL_ERROR:
         return "an internal operation failed";
+#ifdef CUBLAS_STATUS_NOT_SUPPORTED
     case CUBLAS_STATUS_NOT_SUPPORTED:
         return "unsupported function";
+#endif
     default:
         return "unknow code";
     }