Mandelbrot program with SDL

Accelerated Computing CUDA CUDA Programming and Performance
1

Hello !

I’ve a problem with my Mandelbrot program in CUDA, using the SDL library : i’ve always a black window, and never any draw in. I’ve done this program in

C in CPU and it’s works perfectly.

I think the problem is in the kernel, but I can’t find the way to solve it.

So I need some help ! :D

The’re my C program :

[codebox]#include <stdlib.h>

#include <stdio.h>

#include <SDL/SDL.h>

#include <SDL/SDL_ttf.h>

void fractale (SDL_Surface *screen, SDL_Surface *pixel, SDL_Surface *localisation, TTF_Font *font,

		   double x, double y, double zoom, int interation);

int main(int argc, char *argv)

{

int continuer = 1, iteration = 80;

double zoom = 1, x = 0, y = 0;



SDL_Event event;

TTF_Font *font = NULL;

SDL_Surface *screen = NULL, *pixel = NULL, *localisation = NULL;

SDL_Init(SDL_INIT_VIDEO);

TTF_Init();

screen = SDL_SetVideoMode(640, 640, 32, SDL_HWSURFACE | SDL_DOUBLEBUF);

SDL_WM_SetCaption("Fractale de Mandelbrot", NULL);

pixel = SDL_CreateRGBSurface(SDL_HWSURFACE, 1, 1, 32, 0, 0, 0, 0);

font = TTF_OpenFont("lucon.ttf", 10);

	

SDL_WarpMouse(screen->w / 2, screen->h / 2);

SDL_ShowCursor(0);

// Appel de la méthode fractale

fractale(screen, pixel, localisation, font, 0, 0, zoom, iteration);

while (continuer)

{

    SDL_WaitEvent(&event);

    switch(event.type)

    {

        case SDL_QUIT:

            continuer = 0;

            break;

        case SDL_KEYDOWN:

            switch(event.key.keysym.sym)

            {

                case SDLK_ESCAPE:

                    continuer = 0;

                    break;

		case SDLK_KP5:

			fractale(screen, pixel, localisation, font, x, y, zoom, iteration += 30);

                    break;

		case SDLK_KP2:

			fractale(screen, pixel, localisation, font, x, y, zoom, iteration -= 30);

                    break;

		case SDLK_KP_PLUS:

                    	fractale(screen, pixel, localisation, font, x, y, zoom += zoom /2, iteration);

                    break;

		case SDLK_KP_MINUS:

			fractale(screen, pixel, localisation, font, x, y, zoom -= zoom /2, iteration);

		break;

		case SDLK_RIGHT:

                    	fractale(screen, pixel, localisation, font, x+=0.2/zoom, y, zoom, iteration);

                    break;

               	case SDLK_LEFT:

                    	fractale(screen, pixel, localisation, font, x-=0.2/zoom, y, zoom, iteration);

                    break;

                    case SDLK_UP:

                    	fractale(screen, pixel, localisation, font, x, y-=0.2/zoom, zoom, iteration);

                    break;

                	case SDLK_DOWN:

                    	fractale(screen, pixel, localisation, font, x, y+=0.2/zoom, zoom, iteration);

                    break;

		default:

		break;

            }

		case SDL_MOUSEBUTTONUP:

			if (event.button.button == SDL_BUTTON_WHEELUP)

				fractale(screen, pixel, localisation, font, x, y, zoom += zoom /2, iteration);

			else if (event.button.button == SDL_BUTTON_WHEELDOWN)

				fractale(screen, pixel, localisation, font, x, y, zoom -= zoom /2, iteration);

			break;

    }

}

SDL_SaveBMP(screen, "Mandelbrot.bmp");

TTF_CloseFont(font);

TTF_Quit();

SDL_FreeSurface(pixel);

SDL_FreeSurface(localisation);

SDL_Quit();

return EXIT_SUCCESS;

}

void fractale (SDL_Surface *screen, SDL_Surface *pixel, SDL_Surface *localisation, TTF_Font *font,

		   double x, double y, double zoom, int iteration){

int i, j, n;

double zNa, zNb, copie, zModule2, a, b;

char coordonnees[256] = {0};

SDL_Color couleur = {255, 255, 255};

SDL_Rect position;

for (i = -screen->w / 2; i < screen ->w / 2; i++){

	position.x = i + screen->w / 2;

	a = 4 * i / ((double)screen->w * zoom) + x; 

	for (j = -screen->h /2; j < screen->h /2; j++){

		n = zNa = zNb = 0;

		position.y = j + screen->h / 2;

		b = 4 * j / ((double)screen->h * zoom) + y;

		if (( -5 < i && i < 5 && j == 0 ) || ( -5 < j && j < 5 && i == 0 ))

			SDL_FillRect (pixel, NULL, SDL_MapRGB(pixel->format, 255, 255, 255));

		else {

			do {

				n ++;

				copie = zNa;

				zNa=zNa*zNa-zNb*zNb+a;

                zNb=2*copie*zNb+b;

                zModule2=zNa*zNa+zNb*zNb;

			}

			while (n < iteration && zModule2 <= 4);

			

			if ( n == iteration)

				SDL_FillRect (pixel, NULL, SDL_MapRGB(pixel->format, 0, 0, 0));

			else 

				SDL_FillRect (pixel, NULL, SDL_MapRGB(pixel->format, 64*n/iteration, 255*n/iteration, 0));

		}

		SDL_BlitSurface(pixel, NULL, screen, &position);

	}

}

position.x = 10;

position.y = 10;

sprintf(coordonnees, "(%.4f; %.4f)", x, -y);

localisation=TTF_RenderText_Solid(font, coordonnees, couleur);

SDL_BlitSurface(localisation, NULL, screen, &position);

SDL_Flip(screen);

return;

}

[/codebox]

And my CUDA program :

[codebox]#include “cutil_inline.h”

#include <shrUtils.h>

#include <stdio.h>

#include <stdlib.h>

#include <time.h>

#include <cuda.h>

#include <conio.h>

#include <SDL/SDL.h>

#include <SDL/SDL_ttf.h>

#define SCREEN_WIDTH 640

#define SCREEN_HEIGHT 640

#define SCREEN_BPP 32

#define XBLOCK_SIZE 16

#define YBLOCK_SIZE 16

#define MY_MAPRGB(r,g,b) ((b) + ((g) << 8) + (® << 16))

global void mandelbrot(float *pixelColor, double x, double y, double zoom, int iteration) {

int i = 0, j = 0, n = 0;

double zNa = 0, zNb = 0, copie = 0, zModule2 = 0, constA = 0, constB = 0;

// Thread index

const int tx = threadIdx.x;

const int ty = threadIdx.y;

// Block index

const int bx = blockIdx.x;

const int by = blockIdx.y;

int pt = blockDim.x * blockIdx.x + threadIdx.x;

const int screenX = tx + bx * XBLOCK_SIZE;

const int screenY = ty + by * YBLOCK_SIZE;

__syncthreads();

for (i = -screenX / 2; i < screenX / 2; i++){

	x = i + screenX / 2;

	constA = 4 * i / ((double)screenX * zoom) + x; 

	for (j = -screenY /2; j < screenY /2; j++){

		n = zNa = zNb = 0;

		y = j + screenY / 2;

		constB = 4 * j / ((double)screenY * zoom) + y;

		if (( -5 < i && i < 5 && j == 0 ) || ( -5 < j && j < 5 && i == 0 ))

			pixelColor[pt] = MY_MAPRGB(255, 255, 255);

		else {

			do {

				n ++;

				copie = zNa;

				zNa=zNa*zNa-zNb*zNb+constA;

                zNb=2*copie*zNb+constB;

                zModule2=zNa*zNa+zNb*zNb;

			}

			while (n < iteration && zModule2 <= 4);

			

			if ( n == iteration)

			pixelColor[pt] = MY_MAPRGB(0, 0, 0);

			else

			pixelColor[pt] = MY_MAPRGB(64 *n / iteration, 255 * n / iteration, 255);

		}

	}

}

x = 10;

y = 10;

__syncthreads();

}

int main( int argc, char** argv){

double zoom = 1, x = 0, y = 0;

int iteration = 80;

int dimA = SCREEN_WIDTH * SCREEN_HEIGHT;

char coordonnees[256] = {0};

int numThreadsPerBlock = 256;

int numBlocks = dimA / numThreadsPerBlock;

size_t memSize = numBlocks * numThreadsPerBlock * sizeof(int);

float *pixel_h = (float*)malloc(memSize);

float *pixel_d;

cutilSafeCall( cudaMalloc((void **)&pixel_d, memSize) );

for (int i = 0; i < dimA; i++)

	pixel_h[i] = 0;

cudaMemcpy( pixel_d, pixel_h, memSize, cudaMemcpyHostToDevice );

int continuer = 1;

SDL_Event event;

SDL_Surface *ecran = NULL, *localisation = NULL, *pixel = NULL;

TTF_Font *font = NULL;

SDL_Color couleur = {255, 255, 255};

SDL_Rect position;

SDL_Init(SDL_INIT_VIDEO);

TTF_Init( );

ecran = SDL_SetVideoMode( SCREEN_WIDTH, SCREEN_HEIGHT, SCREEN_BPP, SDL_HWSURFACE | SDL_DOUBLEBUF );

SDL_WM_SetCaption( "Fractale de Mandelbrot", NULL );

font = TTF_OpenFont("lucon.ttf", 10 );

pixel = SDL_CreateRGBSurface(SDL_HWSURFACE, 1, 1, 32, 0, 0, 0, 0);

SDL_WarpMouse(SCREEN_HEIGHT / 2, SCREEN_WIDTH / 2);



dim3 dimGrid(numBlocks);

dim3 dimBlock(numThreadsPerBlock);

mandelbrot <<< dimGrid, dimBlock >>> (pixel_d, x, y, zoom, iteration);

cudaMemcpy( pixel_h, pixel_d, memSize, cudaMemcpyDeviceToHost );

cudaThreadSynchronize();

SDL_FillRect (pixel, NULL, Uint32(pixel_d));

SDL_BlitSurface(pixel, NULL, ecran, &position);

sprintf(coordonnees, "(%.4f; %.4f)", x, -y);

localisation=TTF_RenderText_Solid(font, coordonnees, couleur);

SDL_BlitSurface(localisation, NULL, ecran, &position);

SDL_Flip(ecran);

while (continuer)

{

    SDL_WaitEvent(&event);

    switch(event.type)

    {

        case SDL_QUIT:

            continuer = 0;

            break;

        case SDL_KEYDOWN:

            switch(event.key.keysym.sym)

            {

                case SDLK_ESCAPE:

	    continuer = 0;

                break;

	    case SDLK_KP5:

		iteration +=30;

                break;

	    case SDLK_KP2:

		iteration -=30;

                break;

	    case SDLK_KP_PLUS:

                    zoom += zoom / 2;

                break;

	    case SDLK_KP_MINUS:

		zoom -= zoom / 2;

	    break;

	    case SDLK_RIGHT:

                    x += 0.2 / zoom;

                break;

                case SDLK_LEFT:

                    x -= 0.2 / zoom;

                break;

                case SDLK_UP:

                    y -= 0.2 / zoom;

                break;

                case SDLK_DOWN:

                    y += 0.2 / zoom;

                 break;

	     default:

	     break;

            }

		case SDL_MOUSEBUTTONUP:

			if (event.button.button == SDL_BUTTON_WHEELUP)

				 zoom += zoom / 2;

			else if (event.button.button == SDL_BUTTON_WHEELDOWN)

				 zoom -= zoom / 2;

			break;

    }

	mandelbrot <<< dimGrid, dimBlock >>> (pixel_d, x, y, zoom, iteration);

	cudaMemcpy( pixel_h, pixel_d, memSize, cudaMemcpyDeviceToHost );

	cudaThreadSynchronize();

	sprintf(coordonnees, "(%.4f; %.4f)", x, -y);

	localisation=TTF_RenderText_Solid(font, coordonnees, couleur);

	SDL_BlitSurface(localisation, NULL, ecran, &position);

	SDL_Flip(ecran);

}

SDL_SaveBMP (ecran, "mandelbrot.bmp");

// Libération mémoire et exit

cudaFree(pixel_d);

free(pixel_h);

TTF_CloseFont(font);

TTF_Quit();

SDL_FreeSurface(ecran);

SDL_Quit();

return EXIT_SUCCESS;

}

[/codebox]

Thanks a lot ! :)