diff --git a/Makefile_linux b/Makefile_linux index bc77923a..2b1255e4 100644 --- a/Makefile_linux +++ b/Makefile_linux @@ -32,7 +32,6 @@ OPENCV_INCLUDEDIR = -I $(OPENCV_INCLUDEPATH) OPENCV_LIBDIR = -L $(OPENCV_LIBPATH) OPENCV_LIBRARIES = -lopencv_highgui -lopencv_core -lopencv_imgproc -lopencv_legacy - WITH_CUDA = 0 WITH_OPENCL = 0 @@ -132,6 +131,7 @@ LINUXAMD64_DEMO_BENCHMARK_CV = $(CC) $(COMPILER_FLAGS) \ LINUXAMD64_DEMO_BENCHMARK_CVOCL = $(CC) $(COMPILER_FLAGS) \ -o $(OUTPUT_BINPATH)/demo_$(BENCHMARK_CVOCL_NAME) \ src/demo/$(BENCHMARK_CVOCL_NAME).cpp src/demo/timer.cpp \ +>>>>>>> upstream/master -lvisiongl -lopencv_ocl \ $(LINUXAMD64_DIRS_LIBS) \ diff --git a/src/cl2cpp_shaders.cpp b/src/cl2cpp_shaders.cpp index 777bfdce..32a5de76 100644 --- a/src/cl2cpp_shaders.cpp +++ b/src/cl2cpp_shaders.cpp @@ -424,3 +424,92 @@ void vglClThreshold(VglImage* src, VglImage* dst, float thresh){ vglSetContext(dst, VGL_CL_CONTEXT); } +/** Erosion of src image by mask. Result is stored in dst image. + + */ +void vglClErosion(VglImage* img_input, VglImage* img_output, float* convolution_window, int window_size_x, int window_size_y){ + + vglCheckContext(img_input, VGL_CL_CONTEXT); + vglCheckContext(img_output, VGL_CL_CONTEXT); + + cl_int err; + + cl_mem mobj_convolution_window = NULL; + mobj_convolution_window = clCreateBuffer(cl.context, CL_MEM_READ_ONLY, (window_size_x*window_size_y)*sizeof(float), NULL, &err); + vglClCheckError( err, (char*) "clCreateBuffer convolution_window" ); + err = clEnqueueWriteBuffer(cl.commandQueue, mobj_convolution_window, CL_TRUE, 0, (window_size_x*window_size_y)*sizeof(float), convolution_window, 0, NULL, NULL); + vglClCheckError( err, (char*) "clEnqueueWriteBuffer convolution_window" ); + + cl_mem mobj_window_size_x = NULL; + mobj_window_size_x = clCreateBuffer(cl.context, CL_MEM_READ_ONLY, (1)*sizeof(int), NULL, &err); + vglClCheckError( err, (char*) "clCreateBuffer window_size_x" ); + err = clEnqueueWriteBuffer(cl.commandQueue, mobj_window_size_x, CL_TRUE, 0, (1)*sizeof(int), &window_size_x, 0, NULL, NULL); + vglClCheckError( err, (char*) "clEnqueueWriteBuffer window_size_x" ); + + cl_mem mobj_window_size_y = NULL; + mobj_window_size_y = clCreateBuffer(cl.context, CL_MEM_READ_ONLY, (1)*sizeof(int), NULL, &err); + vglClCheckError( err, (char*) "clCreateBuffer window_size_y" ); + err = clEnqueueWriteBuffer(cl.commandQueue, mobj_window_size_y, CL_TRUE, 0, (1)*sizeof(int), &window_size_y, 0, NULL, NULL); + vglClCheckError( err, (char*) "clEnqueueWriteBuffer window_size_y" ); + + static cl_program program = NULL; + if (program == NULL) + { + char* file_path = (char*) "CL/vglClErosion.cl"; + printf("Compiling %s\n", file_path); + std::ifstream file(file_path); + if(file.fail()) + { + std::string str("File not found: "); + str.append(file_path); + vglClCheckError(-1, (char*)str.c_str()); + } + std::string prog( std::istreambuf_iterator( file ), ( std::istreambuf_iterator() ) ); + const char *source_str = prog.c_str(); +#ifdef __DEBUG__ + printf("Kernel to be compiled:\n%s\n", source_str); +#endif + program = clCreateProgramWithSource(cl.context, 1, (const char **) &source_str, 0, &err ); + vglClCheckError(err, (char*) "clCreateProgramWithSource" ); + err = clBuildProgram(program, 1, cl.deviceId, NULL, NULL, NULL ); + vglClBuildDebug(err, program); + } + + static cl_kernel kernel = NULL; + if (kernel == NULL) + { + kernel = clCreateKernel( program, "vglClErosion", &err ); + vglClCheckError(err, (char*) "clCreateKernel" ); + } + + + err = clSetKernelArg( kernel, 0, sizeof( cl_mem ), (void*) &img_input->oclPtr ); + vglClCheckError( err, (char*) "clSetKernelArg 0" ); + + err = clSetKernelArg( kernel, 1, sizeof( cl_mem ), (void*) &img_output->oclPtr ); + vglClCheckError( err, (char*) "clSetKernelArg 1" ); + + err = clSetKernelArg( kernel, 2, sizeof( cl_mem ), (float*) &mobj_convolution_window ); + vglClCheckError( err, (char*) "clSetKernelArg 2" ); + + err = clSetKernelArg( kernel, 3, sizeof( cl_mem ), (int*) &mobj_window_size_x ); + vglClCheckError( err, (char*) "clSetKernelArg 3" ); + + err = clSetKernelArg( kernel, 4, sizeof( cl_mem ), (int*) &mobj_window_size_y ); + vglClCheckError( err, (char*) "clSetKernelArg 4" ); + + size_t worksize[] = { img_input->width, img_input->height, 1 }; + clEnqueueNDRangeKernel( cl.commandQueue, kernel, 2, NULL, worksize, 0, 0, 0, 0 ); + vglClCheckError( err, (char*) "clEnqueueNDRangeKernel" ); + + err = clReleaseMemObject( mobj_convolution_window ); + vglClCheckError(err, (char*) "clReleaseMemObject mobj_convolution_window"); + + err = clReleaseMemObject( mobj_window_size_x ); + vglClCheckError(err, (char*) "clReleaseMemObject mobj_window_size_x"); + + err = clReleaseMemObject( mobj_window_size_y ); + vglClCheckError(err, (char*) "clReleaseMemObject mobj_window_size_y"); + + vglSetContext(img_output, VGL_CL_CONTEXT); +} diff --git a/src/cl2cpp_shaders.h b/src/cl2cpp_shaders.h index d5dcfff8..e80c5626 100644 --- a/src/cl2cpp_shaders.h +++ b/src/cl2cpp_shaders.h @@ -37,3 +37,5 @@ void vglClInvert(VglImage* img_input, VglImage* img_output); */ void vglClThreshold(VglImage* src, VglImage* dst, float thresh); +void vglClErosion(VglImage* img_input, VglImage* img_output, float* convolution_window, int window_size_x, int window_size_y); + diff --git a/src/demo/OpenCVCLBenchmark.cpp b/src/demo/OpenCVCLBenchmark.cpp new file mode 100644 index 00000000..49e36c4b --- /dev/null +++ b/src/demo/OpenCVCLBenchmark.cpp @@ -0,0 +1,193 @@ +#include +#include +#include + +#include "demo/timer.h" +#include + +using namespace std; +using namespace cv; + + +/* + argv[1] = input_image_path + argv[2] = number of operations to execute +*/ +int main(int argc, char* argv[]) +{ + if (argc != 3) + { + printf("Usage: OpenCVCLBenchmark lena_1024.tif 1000\n"); + printf("for this example, will run the program for lena_1024.tif using 1000 operations each\n"); + printf("bad arguments, read usage again\n"); + exit(1); + } + + int limite = atoi(argv[2]); + char* image_path = argv[1]; + printf("OpenCVCLBenchmark on %s for %d operations\n\n",image_path,limite); + cv::Mat imgxm = imread(image_path,1); + + cvtColor(imgxm,imgxm,CV_BGR2RGBA); + ocl::oclMat img(imgxm), out(imgxm); + + vector saveparams; + saveparams.push_back(CV_IMWRITE_PNG_COMPRESSION); + saveparams.push_back(1); + + if (img.data == NULL) + { + std::string str("cvLoadImage/File not found: "); + str.append(image_path); + printf("%s",str.c_str()); + } + + //Primeira chamada a blurSq3 + Size ksize(3,3); + TimerStart(); + ocl::GaussianBlur(img,out,Size(3,3),0); + //ocl::blur(img,out,ksize); + printf("Primeira chamada de Blur: %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para limite blur 3x3 sem a criação da operação + int p = 0; + TimerStart(); + while (p < limite) + { + p++; + //ocl::blur(img,out,ksize); + ocl::GaussianBlur(img,out,Size(3,3),0); + } + printf("Tempo gasto para fazer %d Blur 3x3: %s\n", limite, getTimeElapsedInSeconds()); + + cv::Mat saveout(out); + cvtColor(saveout,saveout,CV_RGBA2BGR); + cv::imwrite("../images/lenaout_blur33.png", saveout,saveparams); + + //Primeira chamada a Convolution + + Mat cvkernel33 = (Mat_(3,3) << 1/9.0, 1/9.0, 1/9.0, + 1/9.0, 1/9.0, 1/9.0, + 1/9.0, 1/9.0, 1/9.0 ); + + Mat cvkernel55 = (Mat_(5,5) << 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0 ); + + + TimerStart(); + cv::Ptr filter33 = ocl::createLinearFilter_GPU(CV_8UC4,CV_8UC4, cvkernel33); + cv::Ptr filter55 = ocl::createLinearFilter_GPU(CV_8UC4,CV_8UC4, cvkernel55); + filter33->apply(img,out); + printf("Primeira chamada da Convolucao com kernel 3x3: %s\n", getTimeElapsedInSeconds()); + //Mede o tempo para limite convoluções 3x3 sem a criação da operação + p = 0; + TimerStart(); + while (p < limite) + { + p++; + filter33->apply(img,out); + } + printf("Tempo gasto para fazer %d convolucoes 3x3: %s \n", limite, getTimeElapsedInSeconds()); + + saveout = *new cv::Mat(out); + cvtColor(saveout,saveout,CV_RGBA2BGR); + cv::imwrite("../images/lenaout_conv33.png", saveout,saveparams); + + //Mede o tempo para limite convoluções 5x5 sem a criação da operação + p = 0; + filter55->apply(img,out); + TimerStart(); + while (p < limite) + { + p++; + filter55->apply(out,out); + } + printf("Tempo gasto para fazer %d convolucoes 5x5: %s\n", limite, getTimeElapsedInSeconds()); + + saveout = *new cv::Mat(out); + cvtColor(saveout,saveout,CV_RGBA2BGR); + cv::imwrite("../images/lenaout_conv55.png", saveout,saveparams); + + //Primeira chamada a threshold + TimerStart(); + Mat cverode33 = getStructuringElement(MORPH_CROSS, Size( 3, 3 )); + ocl::erode(img,out,cverode33); + + printf("Primeira chamada da Erode: %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para limite thresholds sem a criação da operação + p = 0; + TimerStart(); + while (p < limite) + { + p++; + ocl::erode(img,out,cverode33); + } + printf("Tempo gasto para fazer %d Erosions: %s\n", limite, getTimeElapsedInSeconds()); + + saveout = *new cv::Mat(out); + cvtColor(saveout,saveout,CV_RGBA2BGR); + cv::imwrite("../images/lenaout_erosion.png", saveout,saveparams); + + //Primeira chamada a vglClInvert + TimerStart(); + ocl::bitwise_not(img,out); + printf("Primeira chamada da Invert: %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para limite invert sem a criação da operação + p = 0; + TimerStart(); + while (p < limite) + { + p++; + ocl::bitwise_not(img,out); + } + printf("Tempo gasto para fazer %d Invert: %s\n",limite, getTimeElapsedInSeconds()); + + saveout = *new cv::Mat(out); + cvtColor(saveout,saveout,CV_RGBA2BGR); + cv::imwrite("../images/lenaout_invert.png", saveout,saveparams); + + TimerStart(); + + img.copyTo(out); + printf("Primeira chamada da Copy: %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para limite copia GPU->GPU + p = 0; + TimerStart(); + ocl::oclMat x1(img.size(), CV_8UC4); + while (p < limite) + { + p++; + img.copyTo(x1); + } + printf("Tempo gasto para fazer %d copia GPU->GPU: %s\n", limite, getTimeElapsedInSeconds()); + + saveout = *new cv::Mat(out); + cvtColor(saveout,saveout,CV_RGBA2BGR); + cv::imwrite("../images/lenaout_copy.png", saveout,saveparams); + + //Mede o tempo para limite copia CPU->GPU + + ocl::oclMat x; + p = 0; + TimerStart(); + while (p < limite) + { + p++; + img.upload(imgxm); + } + printf("Tempo gasto para fazer %d copia CPU->GPU(Upload) %s\n", limite, getTimeElapsedInSeconds()); + + //Mede o tempo para limite copia GPU->CPU + cv::Mat x2; + p = 0; + TimerStart(); + while (p < limite) + { + p++; + img.download(x2); + } + printf("Tempo gasto para fazer %d copia GPU->CPU(Download) %s\n", limite, getTimeElapsedInSeconds()); + return 0; +} diff --git a/src/demo/camcl.cpp b/src/demo/camcl.cpp new file mode 100644 index 00000000..b3a8bb33 --- /dev/null +++ b/src/demo/camcl.cpp @@ -0,0 +1,241 @@ + +//GL +#define __OPENCL__ + +#include +#include + +// visiongl +#include +#include +#include +#include +#include + +//IplImage, cvLoadImage +#include + +//time +#include + +//begin cameras +#define MAX_CAM 4 +CvCapture *capture[MAX_CAM]; +//end Cameras + +void display(void); +void reshape(int w, int h); +void keyboard(unsigned char key, int x, int y); + +using namespace std; + +#define NUM_IMG 20 +VglImage* img[NUM_IMG]; +VglImage* img_gray[NUM_IMG]; + +GLuint tex = (unsigned int)-1; + +int width, height; +int img_width, img_height; + +int i_frame = 0; +int disparity = 0; +int last_i_frame = -1; + + +// indicates number of rows and columns of images in main window +int refreshall = 2; + +char* winname[VGL_MAX_WINDOWS]; + + +VglNamedWindowList* window_list; + + +int num_cam = 0; + +float zoomfactor = 1.0; + +int showstats() { + static int lasttime; + static int fpscounter; + static int fps; + int curtime; + curtime = time(NULL); + if( lasttime != curtime) { + fps=fpscounter; + fpscounter=1; + lasttime = curtime; + fprintf(stderr, "fps = %d, %f msecs\n",fps,1.0/((float)fps)*1000); + } else { + fpscounter++; + } + return fps; +} + +void display_cam(void) +{ + + for (int i = 0; i < num_cam; i++) + { + cvGrabFrame(capture[i]); + cvCopy(cvRetrieveFrame(capture[i]), img[i]->ipl); + vglSetContext(img[i], VGL_RAM_CONTEXT); + vglUpload(img[i]); + vglClUpload(img[i]); + } + i_frame++; +} + + +void display(void) +{ + + display_cam(); + + if (num_cam == 1) + { + vglClCopy(img[0], img[1]); + } + + vglClCopy(img[0], img[2]); + vglClCopy(img[0], img[3]); + vglClCopy(img[0], img[4]); + + vglClInvert(img[0], img[2]); + vglClCopy(img[2], img[4]); + + float erosion[9] = {0,1,0,1,1,1,0,1,0}; + + vglClErosion(img[2], img[3],erosion,3,3); + vglClErosion(img[3], img[2],erosion,3,3); + + vglClErosion(img[2], img[3],erosion,3,3); + vglClErosion(img[3], img[2],erosion,3,3); + + vglClErosion(img[2], img[3],erosion,3,3); + vglClErosion(img[3], img[2],erosion,3,3); + + vglClThreshold(img[4], img[5],0.5); + vglClErosion(img[5], img[6],erosion,3,3); + + vglClInvert(img[5], img[7]); + vglClCopy(img[7], img[3]); + + window_list->RefreshAll(refreshall); + + //CHECK_FRAMEBUFFER_STATUS() + //ERRCHECK() + + int fps = showstats(); +} + +void reshape(int w, int h) +{ + glViewport(0, 0, (GLsizei) w, (GLsizei) h); + width = w, height = h; + glMatrixMode(GL_PROJECTION); + glLoadIdentity(); + glMatrixMode(GL_MODELVIEW); + glLoadIdentity(); +} + +void keyboard(unsigned char key, int x, int y) +{ + switch (key) { + case 'q': + case 'Q': + case 27: + glutLeaveMainLoop(); + break; + default: + break; + } +} + +void PrintCaptureProperty(CvCapture *capture){ + float result; + IplImage *img_in = cvQueryFrame(capture); + printf("CV_CAP_PROP_POS_MSEC %f\n", cvGetCaptureProperty(capture, CV_CAP_PROP_POS_MSEC)); + printf("CV_CAP_PROP_POS_FRAMES %f\n", cvGetCaptureProperty(capture, CV_CAP_PROP_POS_FRAMES)); + printf("CV_CAP_PROP_POS_AVI_RATIO %f\n", cvGetCaptureProperty(capture, CV_CAP_PROP_POS_AVI_RATIO)); + + result = cvGetCaptureProperty(capture, CV_CAP_PROP_FRAME_WIDTH); + if (result == 0.0) result = img_in->width; + printf("CV_CAP_PROP_FRAME_WIDTH %f\n", result); + + result = cvGetCaptureProperty(capture, CV_CAP_PROP_FRAME_HEIGHT); + if (result == 0.0) result = img_in->height; + printf("CV_CAP_PROP_FRAME_HEIGHT %f\n", result); + + printf("CV_CAP_PROP_FPS %f\n", cvGetCaptureProperty(capture, CV_CAP_PROP_FPS)); + printf("CV_CAP_PROP_FOURCC %f\n", cvGetCaptureProperty(capture, CV_CAP_PROP_FOURCC)); +} + +int main(int argc, char** argv) +{ + for (int i = 0; i < MAX_CAM; i++){ + capture[i] = 0; + capture[i] = cvCaptureFromCAM(i); + cvSetCaptureProperty(capture[i], CV_CAP_PROP_FPS, 30.0f); + if (capture[i]){ + printf("Camera %d at %p\n", i, capture[i]); + PrintCaptureProperty(capture[i]); + num_cam++; + printf("\n\n"); + } + } + + window_list = new VglNamedWindowList(); + window_list->main_window_id = vglInit(1280, 960); + vglClInit(); + glutDisplayFunc(display); + glutIdleFunc(display); + + glutReshapeFunc(reshape); + glutKeyboardFunc(keyboard); + + for (int i = 0; i < VGL_MAX_WINDOWS; i++){ + winname[i] = (char*) malloc(255); + sprintf(winname[i], "win[%d]", i); + } + + for (int i = 0; i < num_cam; i++){ + const int HAS_MIPMAP = 0; + img[i] = vglCreateImage(cvSize(640, 480), IPL_DEPTH_8U, 3, 1, HAS_MIPMAP); + img_gray[i] = vglCreateImage(cvSize(640, 480), IPL_DEPTH_8U, 1, 1, HAS_MIPMAP); + + if (!img[i]){ + fprintf(stderr, "Error loading img[%d] from file\n", i); + exit(1); + } + //vglPrintImageInfo(img[i]); + } + + for (int i = num_cam; i < NUM_IMG; i++){ + img[i] = vglCreateImage(img[0]); + img_gray[i] = vglCreateImage(img_gray[0]); + if (!img[i]){ + fprintf(stderr, "Error creating img[%d]\n", i); + exit(1); + } + vglClear(img[i], 0.0, 0.0, 0.0, 0.0); + //vglPrintImageInfo(img[i]); + char* img_name = (char*)malloc(255); + sprintf(img_name, "img[%d]", i); + vglPrintContext(img[i], img_name); + } + + for (int i = 0; i < 4; i++){ + window_list->NamedWindow(winname[i]); + window_list->ShowImage(winname[i], img[i]); + } + + GLint iUnits; + glGetIntegerv(GL_MAX_TEXTURE_UNITS, &iUnits); + printf("Max number of textures = %d\n", iUnits); + + glutMainLoop(); + + return 0; +} diff --git a/src/demo/opencvbenchmark.cpp b/src/demo/opencvbenchmark.cpp new file mode 100644 index 00000000..f81e24d6 --- /dev/null +++ b/src/demo/opencvbenchmark.cpp @@ -0,0 +1,177 @@ +#define __OPENCL__ +#include +#include + +#include "demo/timer.h" +#include + +#include + +using namespace cv; + + +int main(int argc, char* argv[]) +{ + + if (argc != 3) + { + printf("Usage: OpenCVCLBenchmark lena_1024.tiff 1000\n"); + printf("for this example, will run the program for lena_1024.tiff using 1000 operations each\n"); + printf("bad arguments, read usage again %d \n", argc); + exit(1); + } + + int limite = atoi(argv[2]); + char* image_path = argv[1]; + printf("OpenCV on %s for %d operations\n\n",image_path,limite); + cv::Mat img = imread(image_path,1); + + vector saveparams; + saveparams.push_back(CV_IMWRITE_PNG_COMPRESSION); + saveparams.push_back(1); + + if (img.data == NULL) + { + std::string str("cvLoadImage/File not found: "); + str.append(image_path); + printf("%s",str.c_str()); + } + cv::Mat out(img.cols,img.rows,CV_8UC3); + + //Primeira chamada a blurSq3 + TimerStart(); + cv::GaussianBlur(img,out,Size(3,3),0); + printf("Primeira chamada da Blur: %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para "limite" blur 3x3 sem a criação da operação + int p = 0; + TimerStart(); + while (p < limite) + { + p++; + cv::GaussianBlur(img,out,Size(3,3),0); + } + printf("Tempo gasto para fazer %d Blur 3x3: %s\n",limite, getTimeElapsedInSeconds()); + + cv::imwrite("../images/lenaout_blur33.png", out,saveparams); + + // Kernels para convolucao + + Mat cvkernel33 = (Mat_(3,3) << 1/9.0, 1/9.0, 1/9.0, + 1/9.0, 1/9.0, 1/9.0, + 1/9.0, 1/9.0, 1/9.0 ); + + Mat cvkernel55 = (Mat_(5,5) << 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0, + 1/25.0, 1/25.0, 1/25.0, 1/25.0, 1/25.0 ); + + + TimerStart(); + cv::filter2D(img,out,-1, cvkernel33); + printf("Primeira chamada da Filter2D com kernel 3x3: %s\n", getTimeElapsedInSeconds()); + //Mede o tempo para "limite" convoluções 3x3 sem a criação da operação + p = 0; + TimerStart(); + while (p < limite) + { + p++; + cv::filter2D(img,out,-1, cvkernel33); + } + printf("Tempo gasto para fazer %d convolucoes 3x3: %s \n",limite, getTimeElapsedInSeconds()); + + + cv::imwrite("../images/lenaout_conv33.png", out,saveparams); + + //Mede o tempo para "limite" convoluções 5x5 sem a criação da operação + p = 0; + TimerStart(); + while (p < limite) + { + p++; + cv::filter2D(img,out,-1, cvkernel55); + } + printf("Tempo gasto para fazer %d convolucoes 5x5: %s\n",limite, getTimeElapsedInSeconds()); + + cv::imwrite("../images/lenaout_conv55.png", out,saveparams); + + //Primeira chamada a threshold + TimerStart(); + Mat cverode33 = getStructuringElement(MORPH_CROSS, Size( 3, 3 )); + cv::erode(img,out,cverode33); + printf("Primeira chamada da cv::Erode : %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para "limite" thresholds sem a criação da operação + p = 0; + TimerStart(); + while (p < limite) + { + p++; + cv::erode(img,out,cverode33); + } + printf("Tempo gasto para fazer %d Erosions: %s\n",limite, getTimeElapsedInSeconds()); + + + cv::imwrite("../images/lenaout_erode33.png", out,saveparams); + + + //invert + TimerStart(); + + cv::bitwise_not(img,out); + printf("Primeira chamada da Invert: %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para "limite" invert sem a criação da operação + p = 0; + TimerStart(); + while (p < limite) + { + p++; + cv::bitwise_not(img,out); + } + printf("Tempo gasto para fazer %d Invert: %s\n",limite, getTimeElapsedInSeconds()); + + //vglCheckContext(out, VGL_RAM_CONTEXT); + cv::imwrite("../images/lenaout_invert33.png", out,saveparams); + + //Primeira chamada a vglClCopy + TimerStart(); + img.copyTo(out); + printf("Primeira chamada da cvCopy: %s \n", getTimeElapsedInSeconds()); + //Mede o tempo para "limite" copia GPU->GPU + p = 0; + TimerStart(); + while (p < limite) + { + p++; + img.copyTo(out); + } + printf("Tempo gasto para fazer %d copia CPU->CPU: %s\n",limite, getTimeElapsedInSeconds()); + + cv::imwrite("../images/lenaout_copy.png", out,saveparams); + + cv::Mat gray(img.cols,img.rows,CV_8UC1); + //Mede o tempo para "limite" conversão RGB->Grayscale na CPU + p = 0; + TimerStart(); + while (p < limite) + { + p++; + cvtColor(img,gray, CV_BGR2GRAY); + } + printf("Tempo gasto para fazer %d conversões BGR->Gray: %s\n",limite, getTimeElapsedInSeconds()); + + cv::imwrite("../images/lenaout_rgbtogray.png", gray,saveparams); + + cv::Mat RGBA(img.cols,img.rows,CV_8UC4); + //Mede o tempo para "limite" conversão RGB->RGBA na CPU + p = 0; + TimerStart(); + while (p < limite) + { + p++; + cvtColor(img,RGBA, CV_BGR2RGBA); + } + printf("Tempo gasto para fazer %d conversões BGR->RGBA: %s\n",limite, getTimeElapsedInSeconds()); + + cv::imwrite("../images/lenaout_rgbtorgba.png", RGBA,saveparams); + return 0; +} diff --git a/src/vglClImage.cpp b/src/vglClImage.cpp index 831aa4fd..78b73911 100644 --- a/src/vglClImage.cpp +++ b/src/vglClImage.cpp @@ -57,7 +57,21 @@ void vglClInit() vglClCheckError(err, (char*) "clGetDeviceIDs get devices id"); //precisa adicionar a propriedade CL_KHR_gl_sharing no contexto e pra isso precisará do id do contexto do GL que deverá ser o parametro window //cl_context_properties props[] = {CL_GL_CONTEXT_KHR, (cl_context_properties) wglGetCurrentContext()}; - cl.context = clCreateContext(NULL,1,cl.deviceId,NULL, NULL, &err ); +#ifdef linux + cl_context_properties properties[] = { + CL_GL_CONTEXT_KHR, (cl_context_properties) glXGetCurrentContext(), + CL_GLX_DISPLAY_KHR, (cl_context_properties) glXGetCurrentDisplay(), + CL_CONTEXT_PLATFORM, (cl_context_properties) cl.platformId[0], + 0}; +#elif defined WIN32 + cl_context_properties properties[] = { + CL_GL_CONTEXT_KHR, (cl_context_properties) wglGetCurrentContext(), + CL_WGL_HDC_KHR, (cl_context_properties) wglGetCurrentDC(), + CL_CONTEXT_PLATFORM, (cl_context_properties) cl.platformId[0], + 0}; +#endif + cl.context = clCreateContext(properties,1,cl.deviceId,NULL, NULL, &err ); + //cl.context = clCreateContext(NULL,1,cl.deviceId,NULL, NULL, &err ); vglClCheckError(err, (char*) "clCreateContext GPU"); cl.commandQueue = clCreateCommandQueue( cl.context, *cl.deviceId, 0, &err ); diff --git a/src/vglClImage.h b/src/vglClImage.h index 6dd3eda1..d28dc49e 100644 --- a/src/vglClImage.h +++ b/src/vglClImage.h @@ -13,6 +13,7 @@ #ifdef __OPENCL__ #include +#include #include "vglImage.h"