util/X86MAC64/cuda/samples/3_Imaging/SobelFilter/SobelFilter.cpp

   1 /*
   2  * Copyright 1993-2013 NVIDIA Corporation.  All rights reserved.
   3  *
   4  * Please refer to the NVIDIA end user license agreement (EULA) associated
   5  * with this source code for terms and conditions that govern your use of
   6  * this software. Any use, reproduction, disclosure, or distribution of
   7  * this software and related documentation outside the terms of the EULA
   8  * is strictly prohibited.
   9  *
  10  */
  11
  12 // OpenGL Graphics includes
  13 #include <GL/glew.h>
  14 #if defined(__APPLE__) || defined(MACOSX)
  15 #include <GLUT/glut.h>
  16 #else
  17 #include <GL/freeglut.h>
  18 #endif
  19
  20 // CUDA utilities and system includes
  21 #include <cuda_runtime.h>
  22 #include <cuda_gl_interop.h>
  23
  24 // Includes
  25 #include <stdlib.h>
  26 #include <stdio.h>
  27 #include <string.h>
  28
  29 #include "SobelFilter_kernels.h"
  30
  31 // includes, project
  32 #include <helper_functions.h> // includes for SDK helper functions
  33 #include <helper_cuda.h>      // includes for cuda initialization and error checking
  34
  35 const char *filterMode[] =
  36 {
  37     "No Filtering",
  38     "Sobel Texture",
  39     "Sobel SMEM+Texture",
  40     NULL
  41 };
  42
  43 //
  44 // Cuda example code that implements the Sobel edge detection
  45 // filter. This code works for 8-bit monochrome images.
  46 //
  47 // Use the '-' and '=' keys to change the scale factor.
  48 //
  49 // Other keys:
  50 // I: display image
  51 // T: display Sobel edge detection (computed solely with texture)
  52 // S: display Sobel edge detection (computed with texture and shared memory)
  53
  54 void cleanup(void);
  55 void initializeData(char *file) ;
  56
  57 #define MAX_EPSILON_ERROR 5.0f
  58 #define REFRESH_DELAY     10 //ms
  59
  60 const char *sSDKsample = "CUDA Sobel Edge-Detection";
  61
  62 static int wWidth   = 512; // Window width
  63 static int wHeight  = 512; // Window height
  64 static int imWidth  = 0;   // Image width
  65 static int imHeight = 0;   // Image height
  66
  67 // Code to handle Auto verification
  68 const int frameCheckNumber = 4;
  69 int fpsCount = 0;      // FPS count for averaging
  70 int fpsLimit = 8;      // FPS limit for sampling
  71 unsigned int frameCount = 0;
  72 unsigned int g_TotalErrors = 0;
  73 StopWatchInterface *timer = NULL;
  74 unsigned int g_Bpp;
  75 unsigned int g_Index = 0;
  76
  77 bool g_bQAReadback = false;
  78
  79 // Display Data
  80 static GLuint pbo_buffer = 0;  // Front and back CA buffers
  81 struct cudaGraphicsResource *cuda_pbo_resource; // CUDA Graphics Resource (to transfer PBO)
  82
  83 static GLuint texid = 0;       // Texture for display
  84 unsigned char *pixels = NULL;  // Image pixel data on the host
  85 float imageScale = 1.f;        // Image exposure
  86 enum SobelDisplayMode g_SobelDisplayMode;
  87
  88 int *pArgc   = NULL;
  89 char **pArgv = NULL;
  90
  91 extern "C" void runAutoTest(int argc, char **argv);
  92
  93 #define OFFSET(i) ((char *)NULL + (i))
  94 #define MAX(a,b) ((a > b) ? a : b)
  95
  96 void computeFPS()
  97 {
  98     frameCount++;
  99     fpsCount++;
 100
 101     if (fpsCount == fpsLimit)
 102     {
 103         char fps[256];
 104         float ifps = 1.f / (sdkGetAverageTimerValue(&timer) / 1000.f);
 105         sprintf(fps, "CUDA Edge Detection (%s): %3.1f fps", filterMode[g_SobelDisplayMode], ifps);
 106
 107         glutSetWindowTitle(fps);
 108         fpsCount = 0;
 109
 110         sdkResetTimer(&timer);
 111     }
 112 }
 113
 114
 115 // This is the normal display path
 116 void display(void)
 117 {
 118     sdkStartTimer(&timer);
 119
 120     // Sobel operation
 121     Pixel *data = NULL;
 122
 123     // map PBO to get CUDA device pointer
 124     checkCudaErrors(cudaGraphicsMapResources(1, &cuda_pbo_resource, 0));
 125     size_t num_bytes;
 126     checkCudaErrors(cudaGraphicsResourceGetMappedPointer((void **)&data, &num_bytes,
 127                                                          cuda_pbo_resource));
 128     //printf("CUDA mapped PBO: May access %ld bytes\n", num_bytes);
 129
 130     sobelFilter(data, imWidth, imHeight, g_SobelDisplayMode, imageScale);
 131     checkCudaErrors(cudaGraphicsUnmapResources(1, &cuda_pbo_resource, 0));
 132
 133     glClear(GL_COLOR_BUFFER_BIT);
 134
 135     glBindTexture(GL_TEXTURE_2D, texid);
 136     glBindBuffer(GL_PIXEL_UNPACK_BUFFER, pbo_buffer);
 137     glTexSubImage2D(GL_TEXTURE_2D, 0, 0, 0, imWidth, imHeight,
 138                     GL_LUMINANCE, GL_UNSIGNED_BYTE, OFFSET(0));
 139     glBindBuffer(GL_PIXEL_UNPACK_BUFFER, 0);
 140
 141     glDisable(GL_DEPTH_TEST);
 142     glEnable(GL_TEXTURE_2D);
 143     glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_LINEAR);
 144     glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_LINEAR);
 145     glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_WRAP_S, GL_REPEAT);
 146     glTexParameterf(GL_TEXTURE_2D, GL_TEXTURE_WRAP_T, GL_REPEAT);
 147
 148     glBegin(GL_QUADS);
 149     glVertex2f(0, 0);
 150     glTexCoord2f(0, 0);
 151     glVertex2f(0, 1);
 152     glTexCoord2f(1, 0);
 153     glVertex2f(1, 1);
 154     glTexCoord2f(1, 1);
 155     glVertex2f(1, 0);
 156     glTexCoord2f(0, 1);
 157     glEnd();
 158     glBindTexture(GL_TEXTURE_2D, 0);
 159     glutSwapBuffers();
 160
 161     sdkStopTimer(&timer);
 162
 163     computeFPS();
 164 }
 165
 166 void timerEvent(int value)
 167 {
 168     glutPostRedisplay();
 169     glutTimerFunc(REFRESH_DELAY, timerEvent, 0);
 170 }
 171
 172 void keyboard(unsigned char key, int /*x*/, int /*y*/)
 173 {
 174     char temp[256];
 175
 176     switch (key)
 177     {
 178         case 27:
 179         case 'q':
 180         case 'Q':
 181             printf("Shutting down...\n");
 182             exit(EXIT_SUCCESS);
 183             break;
 184
 185         case '-':
 186             imageScale -= 0.1f;
 187             printf("brightness = %4.2f\n", imageScale);
 188             break;
 189
 190         case '=':
 191             imageScale += 0.1f;
 192             printf("brightness = %4.2f\n", imageScale);
 193             break;
 194
 195         case 'i':
 196         case 'I':
 197             g_SobelDisplayMode = SOBELDISPLAY_IMAGE;
 198             sprintf(temp, "CUDA Edge Detection (%s)", filterMode[g_SobelDisplayMode]);
 199             glutSetWindowTitle(temp);
 200             break;
 201
 202         case 's':
 203         case 'S':
 204             g_SobelDisplayMode = SOBELDISPLAY_SOBELSHARED;
 205             sprintf(temp, "CUDA Edge Detection (%s)", filterMode[g_SobelDisplayMode]);
 206             glutSetWindowTitle(temp);
 207             break;
 208
 209         case 't':
 210         case 'T':
 211             g_SobelDisplayMode = SOBELDISPLAY_SOBELTEX;
 212             sprintf(temp, "CUDA Edge Detection (%s)", filterMode[g_SobelDisplayMode]);
 213             glutSetWindowTitle(temp);
 214             break;
 215
 216         default:
 217             break;
 218     }
 219 }
 220
 221 void reshape(int x, int y)
 222 {
 223     glViewport(0, 0, x, y);
 224     glMatrixMode(GL_PROJECTION);
 225     glLoadIdentity();
 226     glOrtho(0, 1, 0, 1, 0, 1);
 227     glMatrixMode(GL_MODELVIEW);
 228     glLoadIdentity();
 229 }
 230
 231 void cleanup(void)
 232 {
 233     cudaGraphicsUnregisterResource(cuda_pbo_resource);
 234
 235     glBindBuffer(GL_PIXEL_UNPACK_BUFFER, 0);
 236     glDeleteBuffers(1, &pbo_buffer);
 237     glDeleteTextures(1, &texid);
 238     deleteTexture();
 239
 240     sdkDeleteTimer(&timer);
 241 }
 242
 243 void initializeData(char *file)
 244 {
 245     GLint bsize;
 246     unsigned int w, h;
 247     size_t file_length= strlen(file);
 248
 249     if (!strcmp(&file[file_length-3], "pgm"))
 250     {
 251         if (sdkLoadPGM<unsigned char>(file, &pixels, &w, &h) != true)
 252         {
 253             printf("Failed to load PGM image file: %s\n", file);
 254             exit(EXIT_FAILURE);
 255         }
 256
 257         g_Bpp = 1;
 258     }
 259     else if (!strcmp(&file[file_length-3], "ppm"))
 260     {
 261         if (sdkLoadPPM4(file, &pixels, &w, &h) != true)
 262         {
 263             printf("Failed to load PPM image file: %s\n", file);
 264             exit(EXIT_FAILURE);
 265         }
 266
 267         g_Bpp = 4;
 268     }
 269     else
 270     {
 271         cudaDeviceReset();
 272         exit(EXIT_FAILURE);
 273     }
 274
 275     imWidth = (int)w;
 276     imHeight = (int)h;
 277     setupTexture(imWidth, imHeight, pixels, g_Bpp);
 278
 279     memset(pixels, 0x0, g_Bpp * sizeof(Pixel) * imWidth * imHeight);
 280
 281     if (!g_bQAReadback)
 282     {
 283         // use OpenGL Path
 284         glGenBuffers(1, &pbo_buffer);
 285         glBindBuffer(GL_PIXEL_UNPACK_BUFFER, pbo_buffer);
 286         glBufferData(GL_PIXEL_UNPACK_BUFFER,
 287                      g_Bpp * sizeof(Pixel) * imWidth * imHeight,
 288                      pixels, GL_STREAM_DRAW);
 289
 290         glGetBufferParameteriv(GL_PIXEL_UNPACK_BUFFER, GL_BUFFER_SIZE, &bsize);
 291
 292         if ((GLuint)bsize != (g_Bpp * sizeof(Pixel) * imWidth * imHeight))
 293         {
 294             printf("Buffer object (%d) has incorrect size (%d).\n", (unsigned)pbo_buffer, (unsigned)bsize);
 295             cudaDeviceReset();
 296             exit(EXIT_FAILURE);
 297         }
 298
 299         glBindBuffer(GL_PIXEL_UNPACK_BUFFER, 0);
 300
 301         // register this buffer object with CUDA
 302         checkCudaErrors(cudaGraphicsGLRegisterBuffer(&cuda_pbo_resource, pbo_buffer, cudaGraphicsMapFlagsWriteDiscard));
 303
 304         glGenTextures(1, &texid);
 305         glBindTexture(GL_TEXTURE_2D, texid);
 306         glTexImage2D(GL_TEXTURE_2D, 0, ((g_Bpp==1) ? GL_LUMINANCE : GL_BGRA),
 307                      imWidth, imHeight,  0, GL_LUMINANCE, GL_UNSIGNED_BYTE, NULL);
 308         glBindTexture(GL_TEXTURE_2D, 0);
 309
 310         glPixelStorei(GL_UNPACK_ALIGNMENT, 1);
 311         glPixelStorei(GL_PACK_ALIGNMENT, 1);
 312     }
 313 }
 314
 315 void loadDefaultImage(char *loc_exec)
 316 {
 317
 318     printf("Reading image: lena.pgm\n");
 319     const char *image_filename = "lena.pgm";
 320     char *image_path = sdkFindFilePath(image_filename, loc_exec);
 321
 322     if (image_path == NULL)
 323     {
 324         printf("Failed to read image file: <%s>\n", image_filename);
 325         exit(EXIT_FAILURE);
 326     }
 327
 328     initializeData(image_path);
 329     free(image_path);
 330 }
 331
 332
 333 void initGL(int *argc, char **argv)
 334 {
 335     glutInit(argc, argv);
 336     glutInitDisplayMode(GLUT_RGBA | GLUT_DOUBLE);
 337     glutInitWindowSize(wWidth, wHeight);
 338     glutCreateWindow("CUDA Edge Detection");
 339
 340     glewInit();
 341
 342     if (!glewIsSupported("GL_VERSION_1_5 GL_ARB_vertex_buffer_object GL_ARB_pixel_buffer_object"))
 343     {
 344         fprintf(stderr, "Error: failed to get minimal extensions for demo\n");
 345         fprintf(stderr, "This sample requires:\n");
 346         fprintf(stderr, "  OpenGL version 1.5\n");
 347         fprintf(stderr, "  GL_ARB_vertex_buffer_object\n");
 348         fprintf(stderr, "  GL_ARB_pixel_buffer_object\n");
 349         exit(EXIT_FAILURE);
 350     }
 351 }
 352
 353 void runAutoTest(int argc, char *argv[])
 354 {
 355     printf("[%s] (automated testing w/ readback)\n", sSDKsample);
 356     int devID = findCudaDevice(argc, (const char **)argv);
 357
 358     loadDefaultImage(argv[0]);
 359
 360     Pixel *d_result;
 361     checkCudaErrors(cudaMalloc((void **)&d_result, imWidth*imHeight*sizeof(Pixel)));
 362
 363     char *ref_file = NULL;
 364     char  dump_file[256];
 365
 366     int mode = 0;
 367     mode = getCmdLineArgumentInt(argc, (const char **)argv, "mode");
 368     getCmdLineArgumentString(argc, (const char **)argv, "file", &ref_file);
 369
 370     switch (mode)
 371     {
 372         case 0:
 373             g_SobelDisplayMode = SOBELDISPLAY_IMAGE;
 374             sprintf(dump_file, "lena_orig.pgm");
 375             break;
 376
 377         case 1:
 378             g_SobelDisplayMode = SOBELDISPLAY_SOBELTEX;
 379             sprintf(dump_file, "lena_tex.pgm");
 380             break;
 381
 382         case 2:
 383             g_SobelDisplayMode = SOBELDISPLAY_SOBELSHARED;
 384             sprintf(dump_file, "lena_shared.pgm");
 385             break;
 386
 387         default:
 388             printf("Invalid Filter Mode File\n");
 389             exit(EXIT_FAILURE);
 390             break;
 391     }
 392
 393     printf("AutoTest: %s <%s>\n", sSDKsample, filterMode[g_SobelDisplayMode]);
 394     sobelFilter(d_result, imWidth, imHeight, g_SobelDisplayMode, imageScale);
 395     checkCudaErrors(cudaDeviceSynchronize());
 396
 397     unsigned char *h_result = (unsigned char *)malloc(imWidth*imHeight*sizeof(Pixel));
 398     checkCudaErrors(cudaMemcpy(h_result, d_result, imWidth*imHeight*sizeof(Pixel), cudaMemcpyDeviceToHost));
 399     sdkSavePGM(dump_file, h_result, imWidth, imHeight);
 400
 401     if (!sdkComparePGM(dump_file, sdkFindFilePath(ref_file, argv[0]), MAX_EPSILON_ERROR, 0.15f, false))
 402     {
 403         g_TotalErrors++;
 404     }
 405
 406     checkCudaErrors(cudaFree(d_result));
 407     free(h_result);
 408
 409     if (g_TotalErrors != 0)
 410     {
 411         printf("Test failed!\n");
 412         exit(EXIT_FAILURE);
 413     }
 414
 415     printf("Test passed!\n");
 416     exit(EXIT_SUCCESS);
 417 }
 418
 419 int main(int argc, char **argv)
 420 {
 421     pArgc = &argc;
 422     pArgv = argv;
 423
 424     printf("%s Starting...\n\n", sSDKsample);
 425
 426     if (checkCmdLineFlag(argc, (const char **)argv, "help"))
 427     {
 428         printf("\nUsage: SobelFilter <options>\n");
 429         printf("\t\t-mode=n (0=original, 1=texture, 2=smem + texture)\n");
 430         printf("\t\t-file=ref_orig.pgm (ref_tex.pgm, ref_shared.pgm)\n\n");
 431         exit(EXIT_SUCCESS);
 432     }
 433
 434     if (checkCmdLineFlag(argc, (const char **)argv, "file"))
 435     {
 436         g_bQAReadback = true;
 437         runAutoTest(argc, argv);
 438     }
 439
 440     // use command-line specified CUDA device, otherwise use device with highest Gflops/s
 441     if (checkCmdLineFlag(argc, (const char **)argv, "device"))
 442     {
 443         printf("   This SDK does not explicitly support -device=n when running with OpenGL.\n");
 444         printf("   When specifying -device=n (n=0,1,2,....) the sample must not use OpenGL.\n");
 445         printf("   See details below to run without OpenGL:\n\n");
 446         printf(" > %s -device=n\n\n", argv[0]);
 447         printf("exiting...\n");
 448         exit(EXIT_SUCCESS);
 449     }
 450
 451     // First initialize OpenGL context, so we can properly set the GL for CUDA.
 452     // This is necessary in order to achieve optimal performance with OpenGL/CUDA interop.
 453     initGL(&argc, argv);
 454     cudaGLSetGLDevice(gpuGetMaxGflopsDeviceId());
 455
 456     sdkCreateTimer(&timer);
 457     sdkResetTimer(&timer);
 458
 459     glutDisplayFunc(display);
 460     glutKeyboardFunc(keyboard);
 461     glutReshapeFunc(reshape);
 462
 463     loadDefaultImage(argv[0]);
 464
 465     // If code is not printing the USage, then we execute this path.
 466     printf("I: display Image (no filtering)\n");
 467     printf("T: display Sobel Edge Detection (Using Texture)\n");
 468     printf("S: display Sobel Edge Detection (Using SMEM+Texture)\n");
 469     printf("Use the '-' and '=' keys to change the brightness.\n");
 470     fflush(stdout);
 471     atexit(cleanup);
 472     glutTimerFunc(REFRESH_DELAY, timerEvent,0);
 473     glutMainLoop();
 474
 475     cudaDeviceReset();
 476     exit(EXIT_SUCCESS);
 477 }