cudaMalloc, unspecified launch failure

#include <cuda.h>
#include <cutil.h>
#include <stdio.h>

#define BLOCKS 32
#define IND(x,y,w) (y * w + x)

typedef unsigned char uchar;

const int conv_r = 2;
float conv_m[5][5] = {
  { 0.01, 0.01, 0.01, 0.01, 0.01 },
  { 0.01, 0.1, 0.1, 0.1, 0.01 },
  { 0.01, 0.1, 0, 0.1, 0.01 },
  { 0.01, 0.1, 0.1, 0.1, 0.01 },
  { 0.01, 0.01, 0.01, 0.01, 0.01 }
};

__global__ void imageConvoluter ( uchar* in, uchar* out, int w, int h, float conv_m[5][5] ) {
  int idx = blockIdx.x*blockDim.x + threadIdx.x;
  int idy = blockIdx.y*blockDim.y + threadIdx.y;
  int index = idx * w + idy;
  out[index] = 0;
  for ( int idx_ = -conv_r; idx_ <= conv_r; idx_++ ) {
    for ( int idy_ = -conv_r; idy_ <= conv_r; idy_ ++ ) {
      int local_index_x = idx + idx_;
      int local_index_y = idy + idy_;
      if ( (local_index_x > 0) && (local_index_x < w) && (local_index_y > 0) && (local_index_y < h) )
        out[index] += in[IND(local_index_x, local_index_y, w)] * conv_m[idy_+conv_r][idx_+conv_r];
    }
  }
}

int main () {
  unsigned int image_w, image_h;
  uchar *data_original = NULL,
        *data_original_dev = NULL,
        *data_convoluted_dev = NULL;

  CUT_SAFE_CALL( cutLoadPGMub( "pic.pgm", &data_original, &image_w, &image_h ) );
  printf( "W=%d, H=%d\n", image_w, image_h );
  int image_size = image_w * image_h * sizeof(uchar);
  uchar *data_convoluted = new uchar[image_w * image_h],
        *data_convoluted_cpu = new uchar[image_w * image_h];

  cudaMalloc( &data_original_dev, image_size );
  cudaMalloc( &data_convoluted_dev, image_size );
  CUDA_SAFE_CALL( cudaMemcpy( data_original_dev, data_original, image_size, cudaMemcpyHostToDevice ) );

  dim3 blocks = dim3( BLOCKS, BLOCKS );
  dim3 threads = dim3( image_w / BLOCKS, image_h / BLOCKS );

  imageConvoluter <<< blocks, threads >>> ( data_original_dev, data_convoluted_dev, image_w, image_h, conv_m );

  for ( int idx = conv_r; idx < image_h-conv_r; idx++ ) {
    for ( int idy = conv_r; idy < image_w-conv_r; idy++ ) {
      int index = IND( idx, idy, image_w );
      data_convoluted_cpu[index] = 0;
      for ( int idx_ = -conv_r; idx_ <= conv_r; idx_++ ) {
        for ( int idy_ = -conv_r; idy_ <= conv_r; idy_ ++ ) {
          int index_around = IND((idx + idx_), (idy + idy_), image_w);
          data_convoluted_cpu[index] += data_original[index_around] * conv_m[idy_+conv_r][idx_+conv_r];
        }
      }
    }
  }

  CUDA_SAFE_CALL( cudaMemcpy( data_convoluted, data_convoluted_dev, image_size, cudaMemcpyDeviceToHost ) );

  // CUT_SAFE_CALL( cutCompareub( data_convoluted, data_convoluted_cpu, image_w * image_h ) );

  CUT_SAFE_CALL( cutSavePGMub( "pic_convoluted.pgm", data_convoluted, image_w, image_h ) );
  CUT_SAFE_CALL( cutSavePGMub( "pic_convoluted_cpu.pgm", data_convoluted_cpu, image_w, image_h ) );

  cutFree( data_original );
  cudaFree( data_original_dev );
  delete data_convoluted;
  delete data_convoluted_cpu;
}

Cuda error in file 'task00.cu' in line 68 : unspecified launch failure.

all CUDA-capable devices are busy or unavailable.