Las pruebas para el infinito en CUDA

Question

En un programa CUDA, recientemente he cambiado de pruebas para inifinity usando

return x==INFINITY || x==-INFINITY; 

donde INFINITY es de math.h, a

return !isfinite(x); 

y se sorprendió al llegar diferentes resultados gnu.org sugiere que en realidad deberían comportarse de manera similar. ¿Me estoy perdiendo de algo? ¿No está permitido usar INFINITY en un kernel CUDA?

Editar: acabo de descubrir isinf y se dio cuenta que la comprobación utilizando

return isinf(x); 

da el mismo resultado que el cheque infinito. ¿Por qué no está isfinite(x)==!isinf(x)?

Answer 1

isfinite(a) es lo mismo que !isnan(a) && !isinf(a). Si x es NaN, ambos isfinite(x) y isinf(x) son falsos.

Answer 2

Las comparaciones de punto flotante no son necesariamente válidas. Por ejemplo, es posible que (1.0f + 3.0f != 2.0f + 2.0f). Es perfectamente posible que isfinite considere que los valores menores que una constante específica aparte sean iguales a INFINITE o -INFINITE, mientras que usted escribió una igualdad literal.

Answer 3

Muchas GPU y unidades SIMD no son totalmente compatibles con IEEE754, especialmente para casos de borde alrededor de infinitos y NaN. Anoche noté que un procesador vectorial particular con el que trabajé reclamó que ∞ +1 ≠ ∞ yx == x incluso para x ∈ NaN.

Answer 4

isinf() verifica solo para +INFINITY o -INFINITY.

!isfinite() comprueba +INFINITY, -INFINITY o NaN.

Answer 5

En la publicación reciente Checking if a matrix contains nans or infinite values in CUDA, Robert Crovella sugirió usar isinf() para verificar valores infinitos en CUDA.

A continuación proporciono un ejemplo de comprobación de valores infinitos en una matriz usando isinf() y explotando CUDA Thrust. Tal vez podría ser útil como referencia para otros usuarios. El siguiente ejemplo es equivalente al d_result=isinf(d_data); de Matlab. Es diferente del ejemplo que publiqué para la pregunta citada anteriormente en que el presente revisa para cada elemento individual es infinito mientras que el otro revisó si el conjunto completo contenía al menos un NaN y era equivalente a sum(isnan(d_data)); de Matlab.

#include <thrust/sequence.h> 

#include <thrust/device_vector.h> 
#include <thrust/host_vector.h> 
#include <thrust\device_vector.h> 
#include <thrust\reduce.h> 

#include <float.h> 

// --- Operator for testing inf values 
struct isinf_test { 
    __host__ __device__ bool operator()(const float a) const { 
     return isinf(a); 
    } 
}; 

void main(){ 

    const int N = 10; 

    thrust::host_vector<float> h_data(N); 
    for (int i=0; i<N; i++) 
     h_data[i] = rand()/RAND_MAX; 

    h_data[0] = FLT_MAX/FLT_MIN; 

    thrust::device_vector<float> d_data(h_data); 
    thrust::device_vector<float> d_result(h_data); 

    thrust::transform(d_data.begin(), d_data.end(), d_result.begin(), isinf_test()); 

    for (int i=0; i<N; i++) { 
     float val = d_result[i]; 
     printf("Isinf test for element number %i equal to %f\n",i,val); 
    } 

    getchar(); 

}