Commit 7e099e805883a52ba4abc4450f0fe3198bc4cbb0
1 parent
a39577bf
"lots of stuff, testing atm"
Showing
1 changed file
with
133 additions
and
0 deletions
Show diff stats
1 | +#include <assert.h> | ||
2 | +#include <cuda.h> | ||
3 | +#include <cuda_runtime.h> | ||
4 | +#include <cublas_v2.h> | ||
5 | +#include "cuPrintf.cu" | ||
6 | +#include "cuPrintf.cuh" | ||
7 | +#include <stdio.h> | ||
8 | +#include "../visualization/colormap.h" | ||
9 | + | ||
10 | +#define DIM_X 400 | ||
11 | +#define DIM_Y 200 | ||
12 | +typedef unsigned char uchar; | ||
13 | +//surface<void, 2> texOut; ///// maybe just do a normal array instead of a surface. | ||
14 | + //we may not need a surface at all. | ||
15 | +//texture<float, cudaTextureType2D, cudaReadModeElementType> texTemplate | ||
16 | +texture<float, cudaTextureType2D, cudaReadModeElementType> texIn; | ||
17 | +float *result; | ||
18 | +float* v_dif; | ||
19 | +cudaArray* srcArray; | ||
20 | + | ||
21 | +inline void checkCUDAerrors(const char *msg) | ||
22 | +{ | ||
23 | + cudaError_t err = cudaGetLastError(); | ||
24 | + if (cudaSuccess != err){ | ||
25 | + fprintf(stderr, "Cuda error: %s: %s.\n", msg, cudaGetErrorString(err) ); | ||
26 | + exit(1); | ||
27 | + } | ||
28 | +} | ||
29 | + | ||
30 | + | ||
31 | +float get_sum(float *diff) | ||
32 | +{ | ||
33 | + | ||
34 | + cublasStatus_t ret; | ||
35 | + cublasHandle_t handle; | ||
36 | + ret = cublasCreate(&handle); | ||
37 | + | ||
38 | + ret = cublasSetVector(DIM_X*DIM_Y, sizeof(*diff), diff, 1, v_dif, 1); | ||
39 | + //float* out = (float*) malloc(sizeof(float)); | ||
40 | + float out; | ||
41 | + ret = cublasSasum(handle, DIM_X*DIM_Y, v_dif, 1, &out); | ||
42 | + cublasDestroy(handle); | ||
43 | + cudaFree(v_dif); | ||
44 | + return out; | ||
45 | +} | ||
46 | + | ||
47 | +__device__ float Template(int x) | ||
48 | +{ | ||
49 | + if(x < DIM_X/3 || x > DIM_X*2/3){ | ||
50 | + return 1.0; | ||
51 | + }else{ | ||
52 | + return 0.0; | ||
53 | + } | ||
54 | + | ||
55 | +} | ||
56 | + | ||
57 | +__global__ | ||
58 | +void get_diff (float *result) | ||
59 | +{ | ||
60 | + //cuPrintf("Hello"); | ||
61 | + int x = threadIdx.x + blockIdx.x * blockDim.x; | ||
62 | + int y = threadIdx.y + blockIdx.y * blockDim.y; | ||
63 | + int idx = y*DIM_Y+x; | ||
64 | + //int idx = x*DIM_X+y; | ||
65 | + | ||
66 | + //uchar4 color = tex2D(texIn, x, y); | ||
67 | + //float3 tempcolor = make_float3(color.x, color.y, color.z); | ||
68 | + //float valIn = tempcolor.x + tempcolor.y + tempcolor.z; | ||
69 | + float valIn = x;//tex2D(texIn, x, y); | ||
70 | + float valTemp = Template(x); | ||
71 | + result[idx] = valIn; | ||
72 | +// - valTemp; | ||
73 | + //#if __CUDA_ARCH__>=200 | ||
74 | + // printf("Value is : %f\n and the result is : %f\n", valIn, result[idx]); | ||
75 | + //#endif | ||
76 | + //cuPrintf("Value is : %f\n and the result is : %f\n", valIn, result[idx]); | ||
77 | +} | ||
78 | + | ||
79 | + | ||
80 | + | ||
81 | + | ||
82 | +void initArray(cudaGraphicsResource_t src) | ||
83 | +{ | ||
84 | + //cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<uchar> (); | ||
85 | + //cudaMallocArray(&result, &channelDesc, DIM_X, DIM_Y, 0); | ||
86 | + //HANDLE_ERROR( | ||
87 | + // cudaGraphicsGLRegisterImage(&src, | ||
88 | + // fboID, | ||
89 | + // GL_TEXTURE_2D, | ||
90 | + HANDLE_ERROR( | ||
91 | + cudaGraphicsMapResources(1, &src) | ||
92 | + ); | ||
93 | + HANDLE_ERROR( | ||
94 | + cudaGraphicsSubResourceGetMappedArray(&srcArray, src,0,0) | ||
95 | + ); | ||
96 | + HANDLE_ERROR( | ||
97 | + cudaBindTextureToArray(texIn, srcArray) | ||
98 | + ); | ||
99 | + cudaMalloc( (void**) &result, DIM_X*DIM_Y*sizeof(float)); | ||
100 | + checkCUDAerrors("Memory Allocation Issue 1"); | ||
101 | + cudaMalloc((void **) &v_dif, DIM_X*DIM_Y*sizeof(*result)); | ||
102 | + checkCUDAerrors("Memory Allocation Issue 2"); | ||
103 | + //HANDLE_ERROR( | ||
104 | + // cudaBindTextureToArray(texIn, ptr, &channelDesc) | ||
105 | + // ); | ||
106 | +} | ||
107 | + | ||
108 | +void cleanUP(cudaGraphicsResource_t src) | ||
109 | +{ | ||
110 | + HANDLE_ERROR( | ||
111 | + cudaUnbindTexture(texIn) | ||
112 | + ); | ||
113 | + HANDLE_ERROR( | ||
114 | + cudaFree(result) | ||
115 | + ); | ||
116 | + HANDLE_ERROR( | ||
117 | + cudaGraphicsUnmapResources(1,&src) | ||
118 | + ); | ||
119 | +} | ||
120 | + | ||
121 | +extern "C" | ||
122 | +float get_cost(cudaGraphicsResource_t src) | ||
123 | +{ | ||
124 | + initArray(src); | ||
125 | + dim3 grid(400, 200); | ||
126 | + dim3 block(1, 1); | ||
127 | + //texIn.normalized = 1; | ||
128 | + get_diff <<< grid, block >>> (result); | ||
129 | + stim::gpu2image<float>(result, "test2.bmp", 400,200,0,255); | ||
130 | + float output = get_sum(result); | ||
131 | + cleanUP(src); | ||
132 | + return output; | ||
133 | +} |