elarmarioinformatico

CUDA: modelo de programación

2010-04-05T00:32:00.000-07:00

En los últimos años se está explotando el potencial de las tarjetas gráficas para otros propósitos, a parte del que le hemos dado hasta el momento.
Estamos hablando de un procesador capaz de mover millones de polígonos y realizar operaciones sobre matrices enormes en un tiempo mínimo, ¿porqué limitarnos a usarlo únicamente con ese propósito?
Esto se ha denominado como GPUGP (Graphic Process Unit General Purpose) y existen diversas extensiones y lenguajes de programación que explotan algunas arquitecturas de algunas GPUs.

CUDA

En Noviembre de 2006, NVIDIA crea CUDA (Compute Unified Device Architecture), una arquitectura de procesamiento paralelo de propósito general, con un nuevo modelo de programación paralela.
Con esta extesión se puede obtener un código escalable a 100s cores y a 1000s hilos paralelos, independientemente de cuál sea la arquitectura del procesador CUDA.
Empezó como una pequeña extensión de C, pero ya es soportada por OpenCL, Fortran...
Antes de comenzar, vamos a ver una serie de equivalencias y aspectos fundamentales inherentes
del comportamiento de CUDA:

Device = GPU
Host = CPU
Kernel = Función llamada desde el Host que se ejecuta en Device
1 CUDA Kernel se ejecuta mediante un array de Threads.
Todos los Threads ejecutan el mismo código.
Cada Thread tiene un ID que se usa para direccionar la memoria y tomar las decisiones de control.
Unidad básica de operación es el Thread.
Los Threads están organizados en bloques de Threads. (Blocks)
Los bloques están organizados en mallas de bloques. (Grids).
Un Grid solo puede ejecutar un Kernel.

Los Threads van identificados mediante threadIdx, que es un array de
elementos 3D ( tiene 3 componentes, x, y y z) .
Cada Thread puede venir identificado por un índice, es decir, puede tener 1, 2 ó 3 dimensiones, que dependará de cómo vayamos a explotar el paralelismo y del uso del grid.
Los Threads de un bloque pueden cooperar entre sí mediante el uso de memoria compartida dentro del bloque y sincronizando su ejecución para coordinar los accesos a memoria, para lo que usaremos el identificador mencionado anteriormente.
Los grids pueden ser de 1 o 2 dimensiones, luego cada bloque dentro de un grid puede ser direccionado por un índice de 1 o 2 dimensiones mediante blockIdx.
Asimismo, la dimensión del bloque también se puede obtener desde dentro del kernel mediante blockDim.
Los Cuda Threads pueden acceder a los datos de múltiples espacios de memoria durante su ejecución, ya que:

Cada Thread posee su propia memoria local.
Cada bloque su propia memoria compartida por todos los threads del bloque y con el mismo tiempo de vida que los Threads que lo componen.

Todos los Trheads tienen acceso a la memoria global.

El modelo de programación de CUDA asume que los CUDA threads se ejecutan en un device que actúa como coprocesador de un host que ejecuta un programa, proporcionando instrucciones para reservar, liberar, copiar memoria en la memoria del device, así como transferir datos entre el host y el device.
Este modelo asume que host y device poseen su propia DRAM, host memory y device memory.
CUDA.

Kernel

Escalabilidad y sincronización

El tamaño del bloque es elegido aparentemente de forma “arbitraria”, y el grid es creado con suficientes bloques para tener un Thread por un elemento de la matriz.
Todos los Threads de un bloque se ejecutan dentro del mismo core. El número de threads por bloque está limitado por los recursos de memoria del core:
En la misma GPU, actualmente(?) un bloque puede contener 512 threads.
El tamaño de los datos suele ser más grande que el de los bloques:

Independencia de ejecución entre bloques:

Se busca que sea indiferente el orden en el que se ejecutan, y si se ejecutan en paralelo o en serie.
Si no, usar __syncthreads().

Concluyendo, los bloques son necearios para permitir la escalabilidad a diferentes números de cores.
Accesos a Memoria:

CUDA asume que device y host tienen su propia memoria. En principio, device trabaja con la host memory. Para que trabaje con su propia memoria, CUDA proporciona, entre otros:

cudaMalloc(void **, size_t);
cudaMemcpy(void *,void *,
size_t,cudaMemcpyHostToDevice|
cudaMemcpyDeviceToHost);
cudaFree(void *);

Compilación:

Apéndice

Interoperatividad con Directx y OpenGL.
Versión 2.3.1 (26/08/2009)
Arquitectura actual(?): nvidia FERMI: 512 cuda cores.
228 universidades enseñan cuda actualmente. (4 de ellas Españolas

Código genérico en C

2010-04-01T03:32:00.000-07:00

En C, voy a explicaros cómo introducir algo de genericidad en vuestros programas C.
Todos sabemos que C no es un lenguaje orientado a objetos y que no conoce los conceptos de polimorfismo, sobrecarga... entonces, ¿Como se puede escribir código genérico con este lenguaje? La respuesta es sencilla: jugando con (void *).
El puntero a void puede representar nada o cualquier cosa, así que jugando con los castings podemos introducir un cierto grado de genericidad en nuestro código.
De esta manera, si tenemos un void *elemento, podemos hacer que ese elemento sea una cadena, un entero, un float... o cualquier TDA.
Veamos un ejemplo ilustrativo de una idea básica de cómo hacerlo.

En el ejemplo encuadro en rojo lo que sería nuestro TDA. Este TDA va a ser muy sencillito, solo incluye un elemento ( o muchos, si hacemos que elemento sea un array ).
Lo que quería mostraros es como insertar en ese TDA un entero y una cadena.
Realmente esto debería estar en 3 ficheros distintos. Por un lado tenemos el genérico, que sería el que implementa el TDA CeldaGenerica y contendría todas las operaciones que se pueden realizar sobre ella, es decir, todas las que se puedan realizar sobre void.
Y por otro los otros 2, que serían los específicos del tipo de datos en concreto (cadena, entero, flotante...).
Así, podemos tener un fichero genérico, por ejemplo, que implmente una lista de elementos, y otro más específico que haga uso del genérico que implemente una lista de enteros.

Esto es todo, espero que os sirva de ayuda.
SALUDOS!!

Algoritmos Bioinspirados

2010-03-30T07:21:00.000-07:00

En nuestro entorno se dan una serie de acontencimientos que han inspirado algunos algoritmos de búsqueda de soluciones. Algoritmos genéticos y colonia de hormigas son algunos de ellos.
Estos algoritmos ofrecen soluciones óptimas y pseudoóptimas, y su estudio ha demostrado que se pueden obtener muy buenos resultados en un tiempo en el que otros algoritmos de búsqueda en profundidad no se acercaría ni por asomo a una solución semejante.

Genético:

El funcionamiento del genético se basa en la generación de estados mediante un conjunto de individuos, que forman una población.
Cada individuo se compone de un conjunto de genes, que suelen ser 0 y 1s, aunque pueden ser enteros, caracteres... todo es cuestión de saber interpretar el estado y la solución obtenida según la representación del individuo escogida.
Esta población se va modificando mediante la reproducción de unos individuos que han pasado un proceso de selección, mezclando sus genes. La probabilidad de que un individuo sea escogido para la reproducción es proporcional a lo bueno que sea ese individuo, para lo que deberemos implementar una función, denominada fitness, que nos da dicho valor.
Una vez pasada la etapa de selección, se inicia la etapa de reproducción. Podemos elegir que los individuos se procreen 2 a 2 por la mitad, o con un random... escogiendo los genes que conservará el hijo de cada individuo padre y formando la siguiente generación.
Ahora solo nos queda la etapa de mutación. En esta etapa se procede a mutar los genes, según cierta probabilidad, que suele ser muy baja. Una probabilidad alta nos da una convergencia lenta, pero una alta amplia el espacio de búsqueda cuando el algoritmo se ha estancado en una solución debido a que hemos alcanzado un máximo local.
Este proceso se repite durante un número de iteraciones definido por el usuario.
Una vez implementado el algoritmo, se trata de jugar con los parámetros probabilidad de selección, probabilidad de mutación, numero de individuos y número de iteraciones.

Colonia de hormigas:

En la naturaleza, las hormigas encuentran el camino óptimo desde el alimento hasta el hormiguero ayudandose de una sustancia denominada feromona que es depositada por las hormigas y se evapora con el tiempo. A mayor cantidad de feromona mejor es el camino. Cada hormiga escogerá los caminos de mejor feromona. Así, inicialmente se mueven de forma aleatoria hasta encontrar el alimento. Es en la vuelta al hormiguero cuando depositan el rastro de feromona para que otras hormigas lo sigan.
En esto es básicamente en lo que se basa este algoritmo.
Las hormigas son agentes computacionales, que colaboran entre sí para encontrar la solución óptima, almacenando en cada paso cómo de bueno es el camino escogido.
Existen varios tipos de algoritmos que implementan este comportamiento, pero solo me voy a centrar en explicar a nivel general en qué consiste la idea.
Cada agente puede explorar el espacio de estados o explotar la feromona, según un valor probabilístico.
Explorar el espacio de estados significará que avanzará al siguiente estado vecino, sin importarle la distancia ni la feromona.
Al explotar la feromona el agente eligirá el siguiente estado en función de una probabilidad dada por la distancia al siguiente estado y por la feromona que hay en ese camino. Esta feromona como ya he comentado se decrementa con el tiempo.
A la vuelta del recorrido el agente actualiza los niveles de feromona, indicando como de bueno ha sido el recorrido.

Lo curioso de estos algoritmos es lo bien que funcionan basándose simplemente en parámetros de probabilidad, en el que escoger el siguiente estado es "pseudoaleatorio" en cierto modo.

Sobre punteros c

2010-03-26T05:18:00.000-07:00

Empty your memory,
with a free()…
like a pointer!

If you cast a pointer to a integer,
it becomes the integer,
if you cast a pointer to a struct,
it becomes the struct…
The pointer can crash…,
and can Overflow…
Be a pointer my friend…

¿Quién no ha tenido nunca más de un quebradero de cabeza con una violación de segmento, o un core? A veces te vale la excusa de "es que en casa sí funciona...", pero créeme, en la mayoría de los casos no.

Trabajar con punteros es relativamente sencillo si tienes una idea clara de cómo referenciar a memoria y si sigues una metodología sistemática a la hora de hacerlo.
Intenta entender los siguientes aspectos, es posible que una vez entendidos disminuyas considerablemente el número de cores o fallos provocados por accesos a memoria:

A la hora de realizar alguna comparación, ten cuidado cuando las variables implicadas sean punteros, es posible que no quieras comparar el VALOR DEL PUNTERO (es decir, la dirección de memoria) y que lo que quieras comparar realmente es el VALOR al que apunta el puntero (es decir, el valor contenido en memoria).
Asegurate de que liberas los punteros para evitar fugas de memoria (o memory leaks). Esto es un problema sobretodo en proyectos grandes. Imagínate que en tu programa realizas una llamada a una función que reserva 1 KB, que olvidas liberar. Esto no es problema si tu programa terminase, pero si actua como demonio, en segundo plano... y además no realiza esa llamada una vez, sino varias, tu programa estaría consumiendo de forma innecesaria memoria del sistema de forma incremental.
De igual modo que en 1.-, ten cuidado a la hora de asignar valores a punteros. Tenemos una tendencia a copiar un puntero en otro puntero con el operador igual. Esto solo es válido si lo que quieres copiar son las referencias a memoria. Si no, deberías copiar elemento a elemento o mediante un memcpy...
A la hora de trabajar con paso por referencia, es más comodo reservar la variable que vas a pasar por referencia de forma estática y en la llamada al procedimiento/función pasarle la dirección con el operador &.
Recuerda que si reservas memoria dentro de un procedimiento/función la estás reservando en pila, por lo que cuando llames a otras 1 o 2 funciones más se machacará, probocando un core. Si quieres reservar memoria dentro de un procedimiento para una variable que le pasas por referencia debes pasar el puntero por referencia, y no la variable. Es decir, la signatura del parámetro debe ser "puntero a puntero" y no "puntero"
Ten en cuenta que la mayoría de las veces que declaras un puntero debes reservar memoria, a no ser que lo vayas a usar como un puntero auxiliar o algo por el estilo.
Presta especial atención a no acceder a posiciones del puntero que sobrepasen lo que has reservado para el. Recuerda siempre que debes indexar entre 0 y TAMAÑO-1.

Veamos un ejemplo del caso 1.-

Este ejemplo está testeado en un ubuntu 9.10.
Los resultados son los esperados:

La primera comparación, en la que comparamos los punteros, nos dice que no son iguales.
La segunda, en la que comparamos el contenido de cada posición, nos dice que sí.

Si te fijas en cómo comparo elemento a elemento, sumo -i- al puntero. Hubiera dado lo mismo si lo hubiese hecho con -[i]-, pero es por diferenciar un poco entre memoria estática y dinámica.
Efectos del caso 2.-

Veamos un ejemplo de como producir una fuga de forma "involuntaria".

Podemos observar en el código que la memoria reservada en foo() no se libera, y que se llama a dicha función en el código de forma indefinida mediante un while(1), por lo que si este proceso se ejecuta en segundo plano provocará un incremento del uso de memoria a medida que pase el tiempo.

Una forma de monitorizar el uso de memoria en linux es a través de ps aux, observando la columna stat:

Efectos del caso 3.-

En rojo encuadro el error habitual y en azul una forma correcta de hacerlo.

La parte comentada sería la forma incorrecta de hacerlo. Está así porque da un core. Los printfs están colocados a posta para que se machaque la pila, si no estuvieran es posible que no se pudiera reproducir el core.
Si te fijas, al llamar a foo2, el parámetro viene precedido del símbolo &. Esto significa que es la dirección de esa variable que le pasamos como parámetro, que sería la dirección del puntero, es decir, un puntero a puntero, cumpliendo con la signatura de la cabecera de la función.
Esto es todo de momento...
SALUDOS!

Vuelos Baratos
Anuns