Programar en CUDA (IV); math.h desde Kernel CUDA y llamadas a función desde Kernel

Converted document

Funciones matemáticas desde librerías invocadas dentro de los Kernels:


Esta parte es la más simple, os ahorro (aunque os recomiendo) leer la guía de programación de nvidia…
Si habéis intentado llamar a funciones de la librería math.h estándar de C/C++ habréis visto el error al tratar de compilar el código. Esto es debido a que, por el momento, la arquitectura de estas tarjetas no es capaz de trabajar con doble precisión. Llegados a este punto podemos hacer dos cosas en función de las exigencias en cuanto a precisión:
  • Si se requiere una alta precisión en nuestros cálculos (y no consideramos otra alternativa que no sea trabajar con CUDA) deberemos desarrollar nosotros mismos las funciones por distintos métodos. Por ejemplo, una función cos(x) puede desarrollarse hasta la precisión deseada mediante series de McLaurin. Sin embargo esto no queda aquí, pues deberemos ingeniárnoslas para “partir” nuestro número, puesto que no cabe dentro de un float, y después “unirlo”. El tema es interesante y hay varias discusiones abiertas que incluyen análisis de errores y pseudocódigo…
  • En caso de conformarnos con la capacidad de un float usaremos la librería math optimizada para cuda, id est, math_functions.h . No tenemos más que importarla con un include al principio del programa. Las funciones de esta librería se llaman añadiendo una “f” al final de la función, me explico, si quieremos usar la función “pow(base,exponente)” escribiremos “powf(base,exponente)”. Sencillo, ¿verdad?.

Llamadas y definición de funciones para los Kernels:


Esta parte realmente da mucho juego una vez hemos empezado a jugar con este lenguaje. Si se os ha planteado la situación tendréis la duda de cómo llamar o definir una función para usarla dentro de un kernel… que no cunda el pánico, es posible. Y además muy sencillo.
Un poco de teoría; las funciones que utiliza el código del host no pueden ser usadas por el kernel a no ser que cuenten con una definición especial. Esta definición se hace anteponiendo los prefijos __host__ o __device__ .
Veamos un ejemplo:
__device__ int suma(int a, int b)
{
return a+b;
}
Esta sería una función que podríamos llamar de forma normal desde nuestro kernel.
__global__ funciongpu1(float *a…)
{
resultado = suma(a,b); //esta línea llama a la función suma definida para el device.😉
}
Muy intuitivo. Podemos tener también la necesidad de llamar a la misma función desde el host o desde el kernel de manera indistinta, esto puede hacerse declarando la función como sigue (también hay otras maneras):
__host__ __device__ int suma(int a, int b)
{
cuerpo de la funcion
}
Desde mi ignorancia no recomiendo esto último y se me ocurren varias razones para ello. Una de ellas ha quedado implícita en el apartado anterior; la precisión. Mientras que la GPU solo puede trabajar con precisión simple, la CPU lo hace con doble. Esto puede suponer o no un problema, mucho cuidado. La segunda es la compilación. Si trabajamos con un fichero para las funciones de cuda y otro para las del host ¿dónde colocamos la función?.

Esta entrada fue publicada en C/C++, CUDA. Guarda el enlace permanente.

Responder

Introduce tus datos o haz clic en un icono para iniciar sesión:

Logo de WordPress.com

Estás comentando usando tu cuenta de WordPress.com. Cerrar sesión / Cambiar )

Imagen de Twitter

Estás comentando usando tu cuenta de Twitter. Cerrar sesión / Cambiar )

Foto de Facebook

Estás comentando usando tu cuenta de Facebook. Cerrar sesión / Cambiar )

Google+ photo

Estás comentando usando tu cuenta de Google+. Cerrar sesión / Cambiar )

Conectando a %s