From 0fdb4ed4a7be8cf86aa6f4dffbd412ccdb3e09e7 Mon Sep 17 00:00:00 2001 From: Pavel Govyadinov Date: Mon, 25 May 2015 15:09:12 -0500 Subject: [PATCH] fixed the issue with the improperly visualized textures as well as the cost function. The cost and gl/cuda/gl ping-pong is fully functional --- cuda/cost.h | 14 ++++++++------ gl/gl_spider.h | 18 ++++++++++++++---- 2 files changed, 22 insertions(+), 10 deletions(-) diff --git a/cuda/cost.h b/cuda/cost.h index 5aaeb2f..fd8a121 100644 --- a/cuda/cost.h +++ b/cuda/cost.h @@ -46,7 +46,7 @@ float get_sum(float *diff) __device__ float Template(int x) { - if(x < DIM_X/3 || x > DIM_X*2/3){ + if(x < DIM_X/6 || x > DIM_X*5/6 || (x > DIM_X*2/6 && x < DIM_X*4/6)){ return 1.0; }else{ return 0.0; @@ -61,18 +61,20 @@ void get_diff (float *result) int x = threadIdx.x + blockIdx.x * blockDim.x; int y = threadIdx.y + blockIdx.y * blockDim.y; int idx = y*DIM_X+x; + //idx = x*DIM_Y+y; //int idx = x*DIM_X+y; //uchar4 color = tex2D(texIn, x, y); //float3 tempcolor = make_float3(color.x, color.y, color.z); //float valIn = tempcolor.x + tempcolor.y + tempcolor.z; - float valIn = tex2D(texIn, x, y); + float valIn = tex2D(texIn, x, y)/255.0; float valTemp = Template(x); - result[idx] = valIn; + result[idx] = abs(valIn-valTemp); + //-valTemp; // - valTemp; - //#if __CUDA_ARCH__>=200 - // printf("Value is : %f\n and the result is : %f\n", valIn, result[idx]); - //#endif +// #if __CUDA_ARCH__>=200 +// printf("Value is : %f\n and the result is : %f\n", valIn, result[idx]); +// #endif //cuPrintf("Value is : %f\n and the result is : %f\n", valIn, result[idx]); } diff --git a/gl/gl_spider.h b/gl/gl_spider.h index a22aac5..d1c65be 100644 --- a/gl/gl_spider.h +++ b/gl/gl_spider.h @@ -312,6 +312,8 @@ class gl_spider : public virtual gl_texture ); glVertex2f(1.0, 2.0); glEnd(); + //glClearColor(1,1,1,1); + //glClear(GL_COLOR_BUFFER_BIT); glBindTexture(GL_TEXTURE_3D, 0); glDisable(GL_TEXTURE_3D); glBindFramebuffer(GL_FRAMEBUFFER,0); @@ -336,6 +338,7 @@ class gl_spider : public virtual gl_texture width, height, 0, GL_LUMINANCE, GL_UNSIGNED_BYTE, texels); delete[] texels; glBindFramebuffer(GL_FRAMEBUFFER, 0); + glBindTexture(GL_TEXTURE_2D, 0); } @@ -353,7 +356,7 @@ class gl_spider : public virtual gl_texture printf(" device name: %s\n", prop.name); printf(" device maxTexture3D: %d %d %d\n", prop.maxTexture3D[0] ,prop.maxTexture3D[0], prop.maxTexture3D[2]) ; HANDLE_ERROR( cudaGLSetGLDevice (device)); */ - stim::cudaSetDevice(); + //stim::cudaSetDevice(); /* HANDLE_ERROR( cudaGraphicsMapResources(1, &resource, 0) ); @@ -384,10 +387,14 @@ class gl_spider : public virtual gl_texture HANDLE_ERROR( cudaGraphicsGLRegisterImage( &resource, - fboID, + texbufferID, GL_TEXTURE_2D, - CU_GRAPHICS_REGISTER_FLAGS_NONE) + //CU_GRAPHICS_REGISTER_FLAGS_NONE) + cudaGraphicsMapFlagsReadOnly) ); + //HANDLE_ERROR( + // cudaGraphicsMapResources(1, &resource) + //); } void @@ -396,11 +403,14 @@ class gl_spider : public virtual gl_texture HANDLE_ERROR( cudaGraphicsUnregisterResource(resource) ); + //HANDLE_ERROR( + // cudaGraphicsUnmapResources(1,&resource) + //); } float getCost() - { + { createResource(); float cost = get_cost(resource); destroyResource(); -- libgit2 0.21.4