CUDA pow function with integer arguments

Question

I'm new in CUDA, and cannot understand what I'm doing wrong.

I'm trying to calculate the distance of object it has id in array, axis x in array and axis y in array to find neighbors for each object

__global__ 
void dist(int *id_d, int *x_d, int *y_d, 
              int *dist_dev, int dimBlock, int i)
{
    int idx = threadIdx.x + blockIdx.x*blockDim.x;

    while(idx < dimBlock){
        int i;
        for(i= 0; i< dimBlock; i++){
            if (idx == i)continue;
            dist_dev[idx] = pow(x_d[idx] - x_d[i], 2) + pow(y_d[idx] - y_d[i], 2); // error here
        }
    }
}

Is pow not defined in kernel code?

Answer 1

Your problem is that while pow is defined in the CUDA math API (see here), it is not template specialised for integer arguments, ie. there is no version like this:

__device__ ​ int pow ( int  x, int  y ) 

This is why you are getting an error. You will need to explicitly cast the base argument to a floating point type like this:

dist_dev[idx] = pow((double)(x_d[idx] - x_d[i]), 2.0) + 
                    pow((double)(y_d[idx] - y_d[i]), 2.0); 

Having said that, using double precision floating point exponential in your example for a integer square will be poor from an efficiency point of view. It would be preferable to perform the calculation using integer multiplication instead:

int dx = x_d[idx] - x_d[i];
int dy = y_d[idx] - y_d[i];
dist_dev[idx] = (dx * dx) + (dy * dy);