Skip to content

Commit

Permalink
Formatting
Browse files Browse the repository at this point in the history
  • Loading branch information
@jeremykubica @DinoBektesevic
jeremykubica authored and DinoBektesevic committed Jan 10, 2024
1 parent 76bc3f9 commit 997ece5
Show file tree
Hide file tree
Showing 2 changed files with 21 additions and 27 deletions.
29 changes: 13 additions & 16 deletions src/kbmod/search/kernels.cu
View file
Original file line number Diff line number Diff line change
Expand Up @@ -23,22 +23,18 @@

namespace search {

extern "C" void device_allocate_psi_phi_array(PsiPhiArray* data) {
if (!data->cpu_array_allocated())
throw std::runtime_error("CPU data is not allocated.");
if (data->gpu_array_allocated())
throw std::runtime_error("GPU data is already allocated.");
extern "C" void device_allocate_psi_phi_array(PsiPhiArray *data) {
if (!data->cpu_array_allocated()) throw std::runtime_error("CPU data is not allocated.");
if (data->gpu_array_allocated()) throw std::runtime_error("GPU data is already allocated.");

void* device_array_ptr;
void *device_array_ptr;
checkCudaErrors(cudaMalloc((void **)&device_array_ptr, data->get_total_array_size()));
checkCudaErrors(cudaMemcpy(device_array_ptr,
data->get_cpu_array_ptr(),
data->get_total_array_size(),
checkCudaErrors(cudaMemcpy(device_array_ptr, data->get_cpu_array_ptr(), data->get_total_array_size(),
cudaMemcpyHostToDevice));
data->set_gpu_array_ptr(device_array_ptr);
}

extern "C" void device_free_psi_phi_array(PsiPhiArray* data) {
extern "C" void device_free_psi_phi_array(PsiPhiArray *data) {
if (data->gpu_array_allocated()) {
checkCudaErrors(cudaFree(data->get_gpu_array_ptr()));
data->set_gpu_array_ptr(nullptr);
Expand Down Expand Up @@ -312,8 +308,9 @@ extern "C" void deviceSearchFilter(PsiPhiArray &psi_phi_array, float *image_time
dim3 threads(THREAD_DIM_X, THREAD_DIM_Y);

// Launch Search
searchFilterImages<<<blocks, threads>>>(psi_phi_array.get_meta_data(), psi_phi_array.get_gpu_array_ptr(), device_img_times,
params, num_trajectories, device_tests, device_search_results);
searchFilterImages<<<blocks, threads>>>(psi_phi_array.get_meta_data(), psi_phi_array.get_gpu_array_ptr(),
device_img_times, params, num_trajectories, device_tests,
device_search_results);

// Read back results
checkCudaErrors(cudaMemcpy(best_results, device_search_results, sizeof(Trajectory) * num_results,
Expand All @@ -326,7 +323,7 @@ extern "C" void deviceSearchFilter(PsiPhiArray &psi_phi_array, float *image_time
}

__global__ void deviceGetCoaddStamp(int num_images, int width, int height, float *image_vect,
float* image_times, int num_trajectories, Trajectory *trajectories,
float *image_times, int num_trajectories, Trajectory *trajectories,
StampParameters params, int *use_index_vect, float *results) {
// Get the trajectory that we are going to be using.
const int trj_index = blockIdx.x * blockDim.x + threadIdx.x;
Expand Down Expand Up @@ -422,7 +419,7 @@ __global__ void deviceGetCoaddStamp(int num_images, int width, int height, float
}

void deviceGetCoadds(const unsigned int num_images, const unsigned int width, const unsigned int height,
std::vector<float *> data_refs, std::vector<float>& image_times, int num_trajectories,
std::vector<float *> data_refs, std::vector<float> &image_times, int num_trajectories,
Trajectory *trajectories, StampParameters params,
std::vector<std::vector<bool>> &use_index_vect, float *results) {
// Allocate Device memory
Expand Down Expand Up @@ -465,8 +462,8 @@ void deviceGetCoadds(const unsigned int num_images, const unsigned int width, co

// Allocate and copy the times.
checkCudaErrors(cudaMalloc((void **)&device_times, sizeof(float) * num_images));
checkCudaErrors(cudaMemcpy(device_times, image_times.data(), sizeof(float) * num_images,
cudaMemcpyHostToDevice));
checkCudaErrors(
cudaMemcpy(device_times, image_times.data(), sizeof(float) * num_images, cudaMemcpyHostToDevice));

// Allocate and copy the images.
checkCudaErrors(cudaMalloc((void **)&device_img, sizeof(float) * num_image_pixels));
Expand Down
19 changes: 8 additions & 11 deletions src/kbmod/search/stamp_creator.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -3,9 +3,9 @@
namespace search {
#ifdef HAVE_CUDA
void deviceGetCoadds(const unsigned int num_images, const unsigned int width, const unsigned int height,
std::vector<float *> data_refs, std::vector<float>& image_times, int num_trajectories,
Trajectory *trajectories, StampParameters params,
std::vector<std::vector<bool>> &use_index_vect, float *results);
std::vector<float*> data_refs, std::vector<float>& image_times, int num_trajectories,
Trajectory* trajectories, StampParameters params,
std::vector<std::vector<bool>>& use_index_vect, float* results);
#endif

StampCreator::StampCreator() {}
Expand Down Expand Up @@ -43,24 +43,21 @@ std::vector<RawImage> StampCreator::get_stamps(ImageStack& stack, const Trajecto
// NO_DATA tagged (so we can filter it out of mean/median).
RawImage StampCreator::get_median_stamp(ImageStack& stack, const Trajectory& trj, int radius,
const std::vector<bool>& use_index) {
return create_median_image(
create_stamps(stack, trj, radius, true /*=keep_no_data*/, use_index));
return create_median_image(create_stamps(stack, trj, radius, true /*=keep_no_data*/, use_index));
}

// For creating coadded stamps, we do not interpolate the pixel values and keep
// NO_DATA tagged (so we can filter it out of mean/median).
RawImage StampCreator::get_mean_stamp(ImageStack& stack, const Trajectory& trj, int radius,
const std::vector<bool>& use_index) {
return create_mean_image(
create_stamps(stack, trj, radius, true /*=keep_no_data*/, use_index));
return create_mean_image(create_stamps(stack, trj, radius, true /*=keep_no_data*/, use_index));
}

// For creating summed stamps, we do not interpolate the pixel values and replace NO_DATA
// with zero (which is the same as filtering it out for the sum).
RawImage StampCreator::get_summed_stamp(ImageStack& stack, const Trajectory& trj, int radius,
const std::vector<bool>& use_index) {
return create_summed_image(
create_stamps(stack, trj, radius, false /*=keep_no_data*/, use_index));
return create_summed_image(create_stamps(stack, trj, radius, false /*=keep_no_data*/, use_index));
}

std::vector<RawImage> StampCreator::get_coadded_stamps(ImageStack& stack, std::vector<Trajectory>& t_array,
Expand Down Expand Up @@ -191,8 +188,8 @@ std::vector<RawImage> StampCreator::get_coadded_stamps_gpu(ImageStack& stack,
data_refs[t] = sci.data();
}

deviceGetCoadds(num_images, width, height, data_refs, image_times, num_trajectories, t_array.data(), params,
use_index_vect, stamp_data.data());
deviceGetCoadds(num_images, width, height, data_refs, image_times, num_trajectories, t_array.data(),
params, use_index_vect, stamp_data.data());
#else
throw std::runtime_error("Non-GPU co-adds is not implemented.");
#endif
Expand Down

0 comments on commit 997ece5

Please sign in to comment.