Skip to content
Snippets Groups Projects
Commit e9f11e3b authored by Tobias Winchen's avatar Tobias Winchen
Browse files

Removed unnecessary includes

parent 6118c967
No related branches found
No related tags found
No related merge requests found
...@@ -2,17 +2,8 @@ ...@@ -2,17 +2,8 @@
#define PSRDADA_CPP_EFFELSBERG_EDD_SKRFIREPLACEMENTCUDA_CUH #define PSRDADA_CPP_EFFELSBERG_EDD_SKRFIREPLACEMENTCUDA_CUH
#include "psrdada_cpp/common.hpp" #include "psrdada_cpp/common.hpp"
#include <thrust/host_vector.h>
#include <thrust/device_vector.h> #include <thrust/device_vector.h>
#include <thrust/complex.h> #include <thrust/complex.h>
#include <thrust/reduce.h>
#include <thrust/transform.h>
#include <thrust/copy.h>
#include <thrust/functional.h>
#include <thrust/execution_policy.h>
#include <thrust/random/normal_distribution.h>
#include <thrust/random/linear_congruential_engine.h>
#include <nvToolsExt.h>
namespace psrdada_cpp { namespace psrdada_cpp {
namespace effelsberg { namespace effelsberg {
......
...@@ -3,15 +3,7 @@ ...@@ -3,15 +3,7 @@
#include "psrdada_cpp/common.hpp" #include "psrdada_cpp/common.hpp"
#include <thrust/device_vector.h> #include <thrust/device_vector.h>
#include <thrust/host_vector.h>
#include <thrust/complex.h> #include <thrust/complex.h>
#include <thrust/copy.h>
#include <thrust/transform.h>
#include <thrust/functional.h>
#include <thrust/iterator/transform_iterator.h>
#include <thrust/iterator/discard_iterator.h>
#include <thrust/execution_policy.h>
#include <nvToolsExt.h>
namespace psrdada_cpp { namespace psrdada_cpp {
namespace effelsberg { namespace effelsberg {
......
#include "psrdada_cpp/effelsberg/edd/SKRfiReplacementCuda.cuh" #include "psrdada_cpp/effelsberg/edd/SKRfiReplacementCuda.cuh"
#include <thrust/reduce.h>
#include <thrust/transform.h>
#include <thrust/copy.h>
#include <thrust/random/normal_distribution.h>
#include <thrust/random/linear_congruential_engine.h>
#include <nvToolsExt.h>
namespace psrdada_cpp { namespace psrdada_cpp {
namespace effelsberg { namespace effelsberg {
namespace edd { namespace edd {
struct mean_subtraction_square{ struct mean_subtraction_square{
const float mean; const float mean;
mean_subtraction_square(float _mean) :mean(_mean) {} mean_subtraction_square(float _mean) :mean(_mean) {}
...@@ -14,6 +24,7 @@ struct mean_subtraction_square{ ...@@ -14,6 +24,7 @@ struct mean_subtraction_square{
} }
}; };
struct generate_replacement_data{ struct generate_replacement_data{
float normal_dist_mean, normal_dist_std; float normal_dist_mean, normal_dist_std;
generate_replacement_data(float mean, float std) generate_replacement_data(float mean, float std)
...@@ -29,21 +40,20 @@ struct generate_replacement_data{ ...@@ -29,21 +40,20 @@ struct generate_replacement_data{
} }
}; };
SKRfiReplacementCuda::SKRfiReplacementCuda()
{ SKRfiReplacementCuda::SKRfiReplacementCuda() {
BOOST_LOG_TRIVIAL(debug) << "Creating new SKRfiReplacementCuda instance..\n"; BOOST_LOG_TRIVIAL(debug) << "Creating new SKRfiReplacementCuda instance..\n";
} }
SKRfiReplacementCuda::~SKRfiReplacementCuda()
{ SKRfiReplacementCuda::~SKRfiReplacementCuda() {
BOOST_LOG_TRIVIAL(debug) << "Destroying SKRfiReplacementCuda instance..\n"; BOOST_LOG_TRIVIAL(debug) << "Destroying SKRfiReplacementCuda instance..\n";
} }
void SKRfiReplacementCuda::replace_rfi_data(const thrust::device_vector<int> &rfi_status, void SKRfiReplacementCuda::replace_rfi_data(const thrust::device_vector<int> &rfi_status,
thrust::device_vector<thrust::complex<float>> &data, thrust::device_vector<thrust::complex<float>> &data,
std::size_t clean_windows) std::size_t clean_windows) {
{
nvtxRangePushA("replace_rfi_data"); nvtxRangePushA("replace_rfi_data");
thrust::device_vector<thrust::complex<float>> replacement_data; thrust::device_vector<thrust::complex<float>> replacement_data;
//initialize data members of the class //initialize data members of the class
......
#include "psrdada_cpp/effelsberg/edd/SpectralKurtosisCuda.cuh" #include "psrdada_cpp/effelsberg/edd/SpectralKurtosisCuda.cuh"
#include <thrust/transform.h>
#include <thrust/iterator/discard_iterator.h>
#include <nvToolsExt.h>
namespace psrdada_cpp { namespace psrdada_cpp {
namespace effelsberg { namespace effelsberg {
namespace edd { namespace edd {
...@@ -71,7 +78,7 @@ __global__ void compute_sk_kernel(const thrust::complex<float>* __restrict__ dat ...@@ -71,7 +78,7 @@ __global__ void compute_sk_kernel(const thrust::complex<float>* __restrict__ dat
int g_index = thread_offset + blockIdx.x * window_size; int g_index = thread_offset + blockIdx.x * window_size;
pow = thrust::abs(data[g_index]) * thrust::abs(data[g_index]); pow = thrust::abs(data[g_index]) * thrust::abs(data[g_index]);
pow_sq = pow * pow; pow_sq = pow * pow;
s1[l_index] += pow; s1[l_index] += pow;
s2[l_index] += pow_sq; s2[l_index] += pow_sq;
} }
__syncthreads(); __syncthreads();
...@@ -127,18 +134,18 @@ void SpectralKurtosisCuda::compute_sk_thrust(const thrust::device_vector<thrust: ...@@ -127,18 +134,18 @@ void SpectralKurtosisCuda::compute_sk_thrust(const thrust::device_vector<thrust:
_d_s2.resize(_nwindows); _d_s2.resize(_nwindows);
//computing _d_s1 for all windows //computing _d_s1 for all windows
nvtxRangePushA("compute_sk_reduce_by_key_call"); nvtxRangePushA("compute_sk_reduce_by_key_call");
thrust::reduce_by_key(thrust::device, thrust::reduce_by_key(thrust::device,
thrust::make_transform_iterator(thrust::counting_iterator<int> (0), (thrust::placeholders::_1 / _window_size)), thrust::make_transform_iterator(thrust::counting_iterator<int> (0), (thrust::placeholders::_1 / _window_size)),
thrust::make_transform_iterator(thrust::counting_iterator<int> (_sample_size - 1), (thrust::placeholders::_1 / _window_size)), thrust::make_transform_iterator(thrust::counting_iterator<int> (_sample_size - 1), (thrust::placeholders::_1 / _window_size)),
thrust::make_transform_iterator(data.begin(), compute_power()), thrust::make_transform_iterator(data.begin(), compute_power()),
thrust::discard_iterator<int>(), thrust::discard_iterator<int>(),
_d_s1.begin()); _d_s1.begin());
//computing _d_s2 for all windows //computing _d_s2 for all windows
thrust::reduce_by_key(thrust::device, thrust::reduce_by_key(thrust::device,
thrust::make_transform_iterator(thrust::counting_iterator<int> (0), (thrust::placeholders::_1 / _window_size)), thrust::make_transform_iterator(thrust::counting_iterator<int> (0), (thrust::placeholders::_1 / _window_size)),
thrust::make_transform_iterator(thrust::counting_iterator<int> (_sample_size - 1), (thrust::placeholders::_1 / _window_size)), thrust::make_transform_iterator(thrust::counting_iterator<int> (_sample_size - 1), (thrust::placeholders::_1 / _window_size)),
thrust::make_transform_iterator(data.begin(), power_square()), thrust::make_transform_iterator(data.begin(), power_square()),
thrust::discard_iterator<int>(), thrust::discard_iterator<int>(),
_d_s2.begin()); _d_s2.begin());
nvtxRangePop(); nvtxRangePop();
//computes SK and checks the threshold to detect RFI. //computes SK and checks the threshold to detect RFI.
......
0% Loading or .
You are about to add 0 people to the discussion. Proceed with caution.
Please register or to comment