OpenCL, double precision, GeForce 8500 GT

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi!
I can’t use double in kernels for GeForce 8500 GT. If I use only float or compile and perform kernels on Intel CPU, problem doesh’t appear. I tried to specify ‘#pragma OPENCL EXTENSION cl_khr_fp64: enable’, but it didn’t help. For example, i have a kernel like this:

#pragma OPENCL EXTENSION cl_khr_fp64: enable

__kernel void DelaunayRadius( __global double *x, __global double *y,
__global double * t )
{
int tid;
tid = get_global_id(0);
t[tid] = x[tid] + y[tid];
}

And when I checked this code in OpenCLCodeChecker, I recived build log

ptxas application ptx input, line 42; : error : Instruction ‘ld’ requires SM 1.3 or higher, or map_f64_to_f32 directive
ptxas application ptx input, line 43; : error : Instruction ‘ld’ requires SM 1.3 or higher, or map_f64_to_f32 directive
ptxas application ptx input, line 44; : error : Instruction ‘add’ requires SM 1.3 or higher, or map_f64_to_f32 directive
ptxas application ptx input, line 46; : error : Instruction ‘st’ requires SM 1.3 or higher, or map_f64_to_f32 directive
ptxas : fatal error : Ptx assembly aborted due to errors
ptxas application ptx input, line 42; : warning : Double is not supported. Demoting to float

Can somebody give me any advise?
I have lastest version of drivers and CUDA SDK.

2

Your 8500 GT is compute capability 1.1 (see this list of GPUs), which does not support doubles. The only solution is to use floats, or to buy a new GPU.

3

Thanks a lot!
And maybe I missed something, but does ‘compute capability’ have some matching with version of OpenCL? I mean can I use inforamtion about version of OpenCL to define where ‘double’ is supported?

4

“Instruction ‘ld’ requires SM 1.3 or higher”

GeForce 8/9 GPUs do no support FP64, it’s clearly shown in the error message and this is a hardware limitation.

Consider upgrading to Nvidia Maxwell GeForce GTX 750 Ti or GTX 750 which does support FP64.

5

“Compute capability” is a term that NVIDIA uses to categorize the hardware features of their GPUs. The very first GeForce 8000 cards were compute capability 1.0 (from 2007!), and now the latest Maxwell cards are compute capability 5.0.

OpenCL versions change much more slowly, and (being a multi-vendor standard) have no connection to NVIDIA compute capability numbering. (Similarly, CUDA version numbering also has no connection to NVIDIA compute capability numbering.) Since OpenCL is designed to be compatible with a wide range of hardware, the latest OpenCL version will still work with older devices, so the OpenCL version number doesn’t tell you much.

I am not an OpenCL developer, but the API might have a way to query the features supported by the OpenCL device which would allow you to determine if double precision is supported.