TEST(Projection, Forward) { cudaDeviceReset(); //Case 1: { int3 dimsvolume = {8, 8, 8}; int3 dimsimage = {8, 8, 1}; tfloat3 angles = tfloat3(PI / 2.0f, PI / 2.0f, 0.0f); tfloat2 shifts = tfloat2(0, 0); tfloat weight = (tfloat)1; tfloat* d_inputvolume = (tfloat*)CudaMallocFromBinaryFile("Data\\Projection\\Input_Forward_1.bin"); tfloat* d_volumepsf = CudaMallocValueFilled(ElementsFFT(dimsvolume), (tfloat)1); tfloat* d_proj = (tfloat*)CudaMallocValueFilled(Elements(dimsimage), (tfloat)0); tfloat* desired_output = (tfloat*)MallocFromBinaryFile("Data\\Projection\\Output_Forward_1.bin"); tfloat* d_projpsf = CudaMallocValueFilled(ElementsFFT(dimsimage), (tfloat)0); d_ProjForward(d_inputvolume, d_volumepsf, dimsvolume, d_proj, d_projpsf, &angles, &shifts, T_INTERP_CUBIC, 1); tfloat* h_output = (tfloat*)MallocFromDeviceArray(d_proj, Elements(dimsimage) * sizeof(tfloat)); double MeanAbsolute = GetMeanAbsoluteError((tfloat*)desired_output, (tfloat*)h_output, Elements(dimsimage)); ASSERT_LE(MeanAbsolute, 1e-5); cudaFree(d_inputvolume); cudaFree(d_proj); free(desired_output); free(h_output); } cudaDeviceReset(); }
TEST(Transformation, Warp2D) { cudaDeviceReset(); //Case 1: { int2 dimsimage = toInt2(1024, 1024); int nframes = 40; tfloat* d_input = (tfloat*)CudaMallocFromBinaryFile("Data\\Transformation\\Input_Warp2D.bin"); tfloat* d_output = CudaMallocValueFilled(Elements2(dimsimage) * nframes, (tfloat)0); int2 dimsgrid = toInt2(4, 4); tfloat2* h_grid = (tfloat2*)MallocValueFilled(Elements2(dimsgrid) * 2, (tfloat)0); for (uint n = 0; n < nframes; n++) { h_grid[1 * 4 + 0] = tfloat2(n * 1.0, 0.0); h_grid[1 * 4 + 1] = tfloat2(n * 1.0, 0.0); h_grid[1 * 4 + 2] = tfloat2(n * 1.0, 0.0); h_grid[1 * 4 + 3] = tfloat2(n * 1.0, 0.0); h_grid[2 * 4 + 0] = tfloat2(n * 1.0, 0.0); h_grid[2 * 4 + 1] = tfloat2(n * 1.0, 0.0); h_grid[2 * 4 + 2] = tfloat2(n * 1.0, 0.0); h_grid[2 * 4 + 3] = tfloat2(n * 1.0, 0.0); tfloat2* d_grid = (tfloat2*)CudaMallocFromHostArray(h_grid, Elements2(dimsgrid) * sizeof(tfloat2)); d_Warp2D(d_input, dimsimage, d_grid, dimsgrid, d_output + Elements2(dimsimage) * n); cudaFree(d_grid); } d_WriteMRC(d_output, toInt3(dimsimage.x, dimsimage.y, nframes), "d_warped.mrc"); } cudaDeviceReset(); }
TEST(Correlation, Peak) { cudaDeviceReset(); //Case 1: { int3 dims = {8, 8, 1}; tfloat* d_input = (tfloat*)CudaMallocFromBinaryFile("Data\\Correlation\\Input_Peak_1.bin"); tfloat3* desired_output = (tfloat3*)MallocFromBinaryFile("Data\\Correlation\\Output_Peak_1.bin"); tfloat3* d_positions; cudaMalloc((void**)&d_positions, sizeof(tfloat3)); tfloat* d_values; cudaMalloc((void**)&d_values, sizeof(tfloat)); d_Peak(d_input, d_positions, d_values, dims, T_PEAK_MODE::T_PEAK_INTEGER); tfloat3* h_positions = (tfloat3*)MallocFromDeviceArray(d_positions, sizeof(tfloat3)); tfloat* h_values = (tfloat*)MallocFromDeviceArray(d_values, sizeof(tfloat)); double MeanRelative = GetMeanRelativeError((tfloat*)desired_output, (tfloat*)h_positions, 3); ASSERT_LE(MeanRelative, 1e-5); cudaFree(d_input); cudaFree(d_positions); cudaFree(d_values); free(desired_output); free(h_positions); free(h_values); } //Case 2: { int3 dims = {20, 20, 1}; tfloat* d_input = (tfloat*)CudaMallocFromBinaryFile("Data\\Correlation\\Input_Peak_2.bin"); tfloat3* desired_output = (tfloat3*)MallocFromBinaryFile("Data\\Correlation\\Output_Peak_2.bin"); tfloat3* d_positions; cudaMalloc((void**)&d_positions, sizeof(tfloat3)); tfloat* d_values; cudaMalloc((void**)&d_values, sizeof(tfloat)); d_Peak(d_input, d_positions, d_values, dims, T_PEAK_MODE::T_PEAK_SUBCOARSE); tfloat3* h_positions = (tfloat3*)MallocFromDeviceArray(d_positions, sizeof(tfloat3)); tfloat* h_values = (tfloat*)MallocFromDeviceArray(d_values, sizeof(tfloat)); double MeanRelative = GetMeanAbsoluteError((tfloat*)desired_output, (tfloat*)h_positions, DimensionCount(dims)); ASSERT_LE(MeanRelative, 0.05); cudaFree(d_input); cudaFree(d_positions); cudaFree(d_values); free(desired_output); free(h_positions); free(h_values); } //Case 3: { int3 dims = {20, 20, 1}; tfloat* d_input = (tfloat*)CudaMallocFromBinaryFile("Data\\Correlation\\Input_Peak_2.bin"); tfloat3* desired_output = (tfloat3*)MallocFromBinaryFile("Data\\Correlation\\Output_Peak_2.bin"); tfloat3* d_positions; cudaMalloc((void**)&d_positions, sizeof(tfloat3)); tfloat* d_values; cudaMalloc((void**)&d_values, sizeof(tfloat)); d_Peak(d_input, d_positions, d_values, dims, T_PEAK_MODE::T_PEAK_SUBFINE); tfloat3* h_positions = (tfloat3*)MallocFromDeviceArray(d_positions, sizeof(tfloat3)); tfloat* h_values = (tfloat*)MallocFromDeviceArray(d_values, sizeof(tfloat)); double MeanRelative = GetMeanAbsoluteError((tfloat*)desired_output, (tfloat*)h_positions, DimensionCount(dims)); ASSERT_LE(MeanRelative, 0.02); cudaFree(d_input); cudaFree(d_positions); cudaFree(d_values); free(desired_output); free(h_positions); free(h_values); } cudaDeviceReset(); }