mirror of
https://github.com/gnss-sdr/gnss-sdr
synced 2025-07-01 17:43:02 +00:00
Merge branch 'next' of https://github.com/gnss-sdr/gnss-sdr into next
This commit is contained in:
commit
2b6c145a2e
@ -220,7 +220,7 @@ void GpsL1CaPcpsOpenClAcquisition::set_local_code()
|
|||||||
{
|
{
|
||||||
if (item_type_ == "gr_complex")
|
if (item_type_ == "gr_complex")
|
||||||
{
|
{
|
||||||
std::complex<float>* code = new std::complex<float>[code_length_];
|
auto* code = new std::complex<float>[code_length_];
|
||||||
|
|
||||||
gps_l1_ca_code_gen_complex_sampled(code, gnss_synchro_->PRN, fs_in_, 0);
|
gps_l1_ca_code_gen_complex_sampled(code, gnss_synchro_->PRN, fs_in_, 0);
|
||||||
|
|
||||||
@ -261,9 +261,9 @@ float GpsL1CaPcpsOpenClAcquisition::calculate_threshold(float pfa)
|
|||||||
unsigned int ncells = vector_length_ * frequency_bins;
|
unsigned int ncells = vector_length_ * frequency_bins;
|
||||||
double exponent = 1 / static_cast<double>(ncells);
|
double exponent = 1 / static_cast<double>(ncells);
|
||||||
double val = pow(1.0 - pfa, exponent);
|
double val = pow(1.0 - pfa, exponent);
|
||||||
double lambda = double(vector_length_);
|
auto lambda = double(vector_length_);
|
||||||
boost::math::exponential_distribution<double> mydist(lambda);
|
boost::math::exponential_distribution<double> mydist(lambda);
|
||||||
float threshold = static_cast<float>(quantile(mydist, val));
|
auto threshold = static_cast<float>(quantile(mydist, val));
|
||||||
|
|
||||||
return threshold;
|
return threshold;
|
||||||
}
|
}
|
||||||
|
@ -61,13 +61,14 @@
|
|||||||
#include <fstream>
|
#include <fstream>
|
||||||
#include <iostream>
|
#include <iostream>
|
||||||
#include <sstream>
|
#include <sstream>
|
||||||
|
#include <utility>
|
||||||
|
|
||||||
|
|
||||||
using google::LogMessage;
|
using google::LogMessage;
|
||||||
|
|
||||||
pcps_opencl_acquisition_cc_sptr pcps_make_opencl_acquisition_cc(
|
pcps_opencl_acquisition_cc_sptr pcps_make_opencl_acquisition_cc(
|
||||||
unsigned int sampled_ms, unsigned int max_dwells,
|
uint32_t sampled_ms, uint32_t max_dwells,
|
||||||
unsigned int doppler_max, long fs_in,
|
uint32_t doppler_max, int64_t fs_in,
|
||||||
int samples_per_ms, int samples_per_code,
|
int samples_per_ms, int samples_per_code,
|
||||||
bool bit_transition_flag,
|
bool bit_transition_flag,
|
||||||
bool dump,
|
bool dump,
|
||||||
@ -75,15 +76,15 @@ pcps_opencl_acquisition_cc_sptr pcps_make_opencl_acquisition_cc(
|
|||||||
{
|
{
|
||||||
return pcps_opencl_acquisition_cc_sptr(
|
return pcps_opencl_acquisition_cc_sptr(
|
||||||
new pcps_opencl_acquisition_cc(sampled_ms, max_dwells, doppler_max, fs_in, samples_per_ms,
|
new pcps_opencl_acquisition_cc(sampled_ms, max_dwells, doppler_max, fs_in, samples_per_ms,
|
||||||
samples_per_code, bit_transition_flag, dump, dump_filename));
|
samples_per_code, bit_transition_flag, dump, std::move(dump_filename)));
|
||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
pcps_opencl_acquisition_cc::pcps_opencl_acquisition_cc(
|
pcps_opencl_acquisition_cc::pcps_opencl_acquisition_cc(
|
||||||
unsigned int sampled_ms,
|
uint32_t sampled_ms,
|
||||||
unsigned int max_dwells,
|
uint32_t max_dwells,
|
||||||
unsigned int doppler_max,
|
uint32_t doppler_max,
|
||||||
long fs_in,
|
int64_t fs_in,
|
||||||
int samples_per_ms,
|
int samples_per_ms,
|
||||||
int samples_per_code,
|
int samples_per_code,
|
||||||
bool bit_transition_flag,
|
bool bit_transition_flag,
|
||||||
@ -114,7 +115,7 @@ pcps_opencl_acquisition_cc::pcps_opencl_acquisition_cc(
|
|||||||
d_cl_fft_batch_size = 1;
|
d_cl_fft_batch_size = 1;
|
||||||
|
|
||||||
d_in_buffer = new gr_complex *[d_max_dwells];
|
d_in_buffer = new gr_complex *[d_max_dwells];
|
||||||
for (unsigned int i = 0; i < d_max_dwells; i++)
|
for (uint32_t i = 0; i < d_max_dwells; i++)
|
||||||
{
|
{
|
||||||
d_in_buffer[i] = static_cast<gr_complex *>(volk_gnsssdr_malloc(d_fft_size * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
d_in_buffer[i] = static_cast<gr_complex *>(volk_gnsssdr_malloc(d_fft_size * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
||||||
}
|
}
|
||||||
@ -122,7 +123,7 @@ pcps_opencl_acquisition_cc::pcps_opencl_acquisition_cc(
|
|||||||
d_fft_codes = static_cast<gr_complex *>(volk_gnsssdr_malloc(d_fft_size_pow2 * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
d_fft_codes = static_cast<gr_complex *>(volk_gnsssdr_malloc(d_fft_size_pow2 * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
||||||
d_zero_vector = static_cast<gr_complex *>(volk_gnsssdr_malloc((d_fft_size_pow2 - d_fft_size) * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
d_zero_vector = static_cast<gr_complex *>(volk_gnsssdr_malloc((d_fft_size_pow2 - d_fft_size) * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
||||||
|
|
||||||
for (unsigned int i = 0; i < (d_fft_size_pow2 - d_fft_size); i++)
|
for (uint32_t i = 0; i < (d_fft_size_pow2 - d_fft_size); i++)
|
||||||
{
|
{
|
||||||
d_zero_vector[i] = gr_complex(0.0, 0.0);
|
d_zero_vector[i] = gr_complex(0.0, 0.0);
|
||||||
}
|
}
|
||||||
@ -140,7 +141,7 @@ pcps_opencl_acquisition_cc::pcps_opencl_acquisition_cc(
|
|||||||
|
|
||||||
// For dumping samples into a file
|
// For dumping samples into a file
|
||||||
d_dump = dump;
|
d_dump = dump;
|
||||||
d_dump_filename = dump_filename;
|
d_dump_filename = std::move(dump_filename);
|
||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
@ -148,14 +149,14 @@ pcps_opencl_acquisition_cc::~pcps_opencl_acquisition_cc()
|
|||||||
{
|
{
|
||||||
if (d_num_doppler_bins > 0)
|
if (d_num_doppler_bins > 0)
|
||||||
{
|
{
|
||||||
for (unsigned int i = 0; i < d_num_doppler_bins; i++)
|
for (uint32_t i = 0; i < d_num_doppler_bins; i++)
|
||||||
{
|
{
|
||||||
volk_gnsssdr_free(d_grid_doppler_wipeoffs[i]);
|
volk_gnsssdr_free(d_grid_doppler_wipeoffs[i]);
|
||||||
}
|
}
|
||||||
delete[] d_grid_doppler_wipeoffs;
|
delete[] d_grid_doppler_wipeoffs;
|
||||||
}
|
}
|
||||||
|
|
||||||
for (unsigned int i = 0; i < d_max_dwells; i++)
|
for (uint32_t i = 0; i < d_max_dwells; i++)
|
||||||
{
|
{
|
||||||
volk_gnsssdr_free(d_in_buffer[i]);
|
volk_gnsssdr_free(d_in_buffer[i]);
|
||||||
}
|
}
|
||||||
@ -193,7 +194,7 @@ pcps_opencl_acquisition_cc::~pcps_opencl_acquisition_cc()
|
|||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
int pcps_opencl_acquisition_cc::init_opencl_environment(std::string kernel_filename)
|
int pcps_opencl_acquisition_cc::init_opencl_environment(const std::string &kernel_filename)
|
||||||
{
|
{
|
||||||
//get all platforms (drivers)
|
//get all platforms (drivers)
|
||||||
std::vector<cl::Platform> all_platforms;
|
std::vector<cl::Platform> all_platforms;
|
||||||
@ -313,7 +314,7 @@ void pcps_opencl_acquisition_cc::init()
|
|||||||
d_cl_buffer_grid_doppler_wipeoffs = new cl::Buffer *[d_num_doppler_bins];
|
d_cl_buffer_grid_doppler_wipeoffs = new cl::Buffer *[d_num_doppler_bins];
|
||||||
}
|
}
|
||||||
|
|
||||||
for (unsigned int doppler_index = 0; doppler_index < d_num_doppler_bins; doppler_index++)
|
for (uint32_t doppler_index = 0; doppler_index < d_num_doppler_bins; doppler_index++)
|
||||||
{
|
{
|
||||||
d_grid_doppler_wipeoffs[doppler_index] = static_cast<gr_complex *>(volk_gnsssdr_malloc(d_fft_size * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
d_grid_doppler_wipeoffs[doppler_index] = static_cast<gr_complex *>(volk_gnsssdr_malloc(d_fft_size * sizeof(gr_complex), volk_gnsssdr_get_alignment()));
|
||||||
|
|
||||||
@ -359,7 +360,7 @@ void pcps_opencl_acquisition_cc::set_local_code(std::complex<float> *code)
|
|||||||
|
|
||||||
clFFT_ExecuteInterleaved((*d_cl_queue)(), d_cl_fft_plan, d_cl_fft_batch_size,
|
clFFT_ExecuteInterleaved((*d_cl_queue)(), d_cl_fft_plan, d_cl_fft_batch_size,
|
||||||
clFFT_Forward, (*d_cl_buffer_2)(), (*d_cl_buffer_2)(),
|
clFFT_Forward, (*d_cl_buffer_2)(), (*d_cl_buffer_2)(),
|
||||||
0, NULL, NULL);
|
0, nullptr, nullptr);
|
||||||
|
|
||||||
//Conjucate the local code
|
//Conjucate the local code
|
||||||
cl::Kernel kernel = cl::Kernel(d_cl_program, "conj_vector");
|
cl::Kernel kernel = cl::Kernel(d_cl_program, "conj_vector");
|
||||||
@ -406,7 +407,7 @@ void pcps_opencl_acquisition_cc::acquisition_core_volk()
|
|||||||
d_input_power /= static_cast<float>(d_fft_size);
|
d_input_power /= static_cast<float>(d_fft_size);
|
||||||
|
|
||||||
// 2- Doppler frequency search loop
|
// 2- Doppler frequency search loop
|
||||||
for (unsigned int doppler_index = 0; doppler_index < d_num_doppler_bins; doppler_index++)
|
for (uint32_t doppler_index = 0; doppler_index < d_num_doppler_bins; doppler_index++)
|
||||||
{
|
{
|
||||||
// doppler search steps
|
// doppler search steps
|
||||||
doppler = -static_cast<int>(d_doppler_max) + d_doppler_step * doppler_index;
|
doppler = -static_cast<int>(d_doppler_max) + d_doppler_step * doppler_index;
|
||||||
@ -542,7 +543,7 @@ void pcps_opencl_acquisition_cc::acquisition_core_opencl()
|
|||||||
cl::Kernel kernel;
|
cl::Kernel kernel;
|
||||||
|
|
||||||
// 2- Doppler frequency search loop
|
// 2- Doppler frequency search loop
|
||||||
for (unsigned int doppler_index = 0; doppler_index < d_num_doppler_bins; doppler_index++)
|
for (uint32_t doppler_index = 0; doppler_index < d_num_doppler_bins; doppler_index++)
|
||||||
{
|
{
|
||||||
// doppler search steps
|
// doppler search steps
|
||||||
|
|
||||||
@ -562,7 +563,7 @@ void pcps_opencl_acquisition_cc::acquisition_core_opencl()
|
|||||||
|
|
||||||
clFFT_ExecuteInterleaved((*d_cl_queue)(), d_cl_fft_plan, d_cl_fft_batch_size,
|
clFFT_ExecuteInterleaved((*d_cl_queue)(), d_cl_fft_plan, d_cl_fft_batch_size,
|
||||||
clFFT_Forward, (*d_cl_buffer_1)(), (*d_cl_buffer_2)(),
|
clFFT_Forward, (*d_cl_buffer_1)(), (*d_cl_buffer_2)(),
|
||||||
0, NULL, NULL);
|
0, nullptr, nullptr);
|
||||||
|
|
||||||
// Multiply carrier wiped--off, Fourier transformed incoming signal
|
// Multiply carrier wiped--off, Fourier transformed incoming signal
|
||||||
// with the local FFT'd code reference
|
// with the local FFT'd code reference
|
||||||
@ -576,7 +577,7 @@ void pcps_opencl_acquisition_cc::acquisition_core_opencl()
|
|||||||
// compute the inverse FFT
|
// compute the inverse FFT
|
||||||
clFFT_ExecuteInterleaved((*d_cl_queue)(), d_cl_fft_plan, d_cl_fft_batch_size,
|
clFFT_ExecuteInterleaved((*d_cl_queue)(), d_cl_fft_plan, d_cl_fft_batch_size,
|
||||||
clFFT_Inverse, (*d_cl_buffer_2)(), (*d_cl_buffer_2)(),
|
clFFT_Inverse, (*d_cl_buffer_2)(), (*d_cl_buffer_2)(),
|
||||||
0, NULL, NULL);
|
0, nullptr, nullptr);
|
||||||
|
|
||||||
// Compute magnitude
|
// Compute magnitude
|
||||||
kernel = cl::Kernel(d_cl_program, "magnitude_squared");
|
kernel = cl::Kernel(d_cl_program, "magnitude_squared");
|
||||||
@ -735,8 +736,8 @@ int pcps_opencl_acquisition_cc::general_work(int noutput_items,
|
|||||||
// Fill internal buffer with d_max_dwells signal blocks. This step ensures that
|
// Fill internal buffer with d_max_dwells signal blocks. This step ensures that
|
||||||
// consecutive signal blocks will be processed in multi-dwell operation. This is
|
// consecutive signal blocks will be processed in multi-dwell operation. This is
|
||||||
// essential when d_bit_transition_flag = true.
|
// essential when d_bit_transition_flag = true.
|
||||||
unsigned int num_dwells = std::min(static_cast<int>(d_max_dwells - d_in_dwell_count), ninput_items[0]);
|
uint32_t num_dwells = std::min(static_cast<int>(d_max_dwells - d_in_dwell_count), ninput_items[0]);
|
||||||
for (unsigned int i = 0; i < num_dwells; i++)
|
for (uint32_t i = 0; i < num_dwells; i++)
|
||||||
{
|
{
|
||||||
memcpy(d_in_buffer[d_in_dwell_count++], static_cast<const gr_complex *>(input_items[i]),
|
memcpy(d_in_buffer[d_in_dwell_count++], static_cast<const gr_complex *>(input_items[i]),
|
||||||
sizeof(gr_complex) * d_fft_size);
|
sizeof(gr_complex) * d_fft_size);
|
||||||
|
@ -56,6 +56,7 @@
|
|||||||
#include <gnuradio/block.h>
|
#include <gnuradio/block.h>
|
||||||
#include <gnuradio/fft/fft.h>
|
#include <gnuradio/fft/fft.h>
|
||||||
#include <gnuradio/gr_complex.h>
|
#include <gnuradio/gr_complex.h>
|
||||||
|
#include <cstdint>
|
||||||
#include <fstream>
|
#include <fstream>
|
||||||
#include <string>
|
#include <string>
|
||||||
#include <vector>
|
#include <vector>
|
||||||
@ -71,8 +72,8 @@ class pcps_opencl_acquisition_cc;
|
|||||||
typedef boost::shared_ptr<pcps_opencl_acquisition_cc> pcps_opencl_acquisition_cc_sptr;
|
typedef boost::shared_ptr<pcps_opencl_acquisition_cc> pcps_opencl_acquisition_cc_sptr;
|
||||||
|
|
||||||
pcps_opencl_acquisition_cc_sptr
|
pcps_opencl_acquisition_cc_sptr
|
||||||
pcps_make_opencl_acquisition_cc(unsigned int sampled_ms, unsigned int max_dwells,
|
pcps_make_opencl_acquisition_cc(uint32_t sampled_ms, uint32_t max_dwells,
|
||||||
unsigned int doppler_max, long fs_in,
|
uint32_t doppler_max, int64_t fs_in,
|
||||||
int samples_per_ms, int samples_per_code,
|
int samples_per_ms, int samples_per_code,
|
||||||
bool bit_transition_flag,
|
bool bit_transition_flag,
|
||||||
bool dump,
|
bool dump,
|
||||||
@ -88,15 +89,15 @@ class pcps_opencl_acquisition_cc : public gr::block
|
|||||||
{
|
{
|
||||||
private:
|
private:
|
||||||
friend pcps_opencl_acquisition_cc_sptr
|
friend pcps_opencl_acquisition_cc_sptr
|
||||||
pcps_make_opencl_acquisition_cc(unsigned int sampled_ms, unsigned int max_dwells,
|
pcps_make_opencl_acquisition_cc(uint32_t sampled_ms, uint32_t max_dwells,
|
||||||
unsigned int doppler_max, long fs_in,
|
uint32_t doppler_max, int64_t fs_in,
|
||||||
int samples_per_ms, int samples_per_code,
|
int samples_per_ms, int samples_per_code,
|
||||||
bool bit_transition_flag,
|
bool bit_transition_flag,
|
||||||
bool dump,
|
bool dump,
|
||||||
std::string dump_filename);
|
std::string dump_filename);
|
||||||
|
|
||||||
pcps_opencl_acquisition_cc(unsigned int sampled_ms, unsigned int max_dwells,
|
pcps_opencl_acquisition_cc(uint32_t sampled_ms, uint32_t max_dwells,
|
||||||
unsigned int doppler_max, long fs_in,
|
uint32_t doppler_max, int64_t fs_in,
|
||||||
int samples_per_ms, int samples_per_code,
|
int samples_per_ms, int samples_per_code,
|
||||||
bool bit_transition_flag,
|
bool bit_transition_flag,
|
||||||
bool dump,
|
bool dump,
|
||||||
@ -105,30 +106,30 @@ private:
|
|||||||
void calculate_magnitudes(gr_complex* fft_begin, int doppler_shift,
|
void calculate_magnitudes(gr_complex* fft_begin, int doppler_shift,
|
||||||
int doppler_offset);
|
int doppler_offset);
|
||||||
|
|
||||||
int init_opencl_environment(std::string kernel_filename);
|
int init_opencl_environment(const std::string& kernel_filename);
|
||||||
|
|
||||||
long d_fs_in;
|
int64_t d_fs_in;
|
||||||
int d_samples_per_ms;
|
int d_samples_per_ms;
|
||||||
int d_samples_per_code;
|
int d_samples_per_code;
|
||||||
unsigned int d_doppler_resolution;
|
uint32_t d_doppler_resolution;
|
||||||
float d_threshold;
|
float d_threshold;
|
||||||
std::string d_satellite_str;
|
std::string d_satellite_str;
|
||||||
unsigned int d_doppler_max;
|
uint32_t d_doppler_max;
|
||||||
unsigned int d_doppler_step;
|
uint32_t d_doppler_step;
|
||||||
unsigned int d_sampled_ms;
|
uint32_t d_sampled_ms;
|
||||||
unsigned int d_max_dwells;
|
uint32_t d_max_dwells;
|
||||||
unsigned int d_well_count;
|
uint32_t d_well_count;
|
||||||
unsigned int d_fft_size;
|
uint32_t d_fft_size;
|
||||||
unsigned int d_fft_size_pow2;
|
uint32_t d_fft_size_pow2;
|
||||||
int* d_max_doppler_indexs;
|
int* d_max_doppler_indexs;
|
||||||
uint64_t d_sample_counter;
|
uint64_t d_sample_counter;
|
||||||
gr_complex** d_grid_doppler_wipeoffs;
|
gr_complex** d_grid_doppler_wipeoffs;
|
||||||
unsigned int d_num_doppler_bins;
|
uint32_t d_num_doppler_bins;
|
||||||
gr_complex* d_fft_codes;
|
gr_complex* d_fft_codes;
|
||||||
gr::fft::fft_complex* d_fft_if;
|
gr::fft::fft_complex* d_fft_if;
|
||||||
gr::fft::fft_complex* d_ifft;
|
gr::fft::fft_complex* d_ifft;
|
||||||
Gnss_Synchro* d_gnss_synchro;
|
Gnss_Synchro* d_gnss_synchro;
|
||||||
unsigned int d_code_phase;
|
uint32_t d_code_phase;
|
||||||
float d_doppler_freq;
|
float d_doppler_freq;
|
||||||
float d_mag;
|
float d_mag;
|
||||||
float* d_magnitude;
|
float* d_magnitude;
|
||||||
@ -140,12 +141,12 @@ private:
|
|||||||
int d_state;
|
int d_state;
|
||||||
bool d_core_working;
|
bool d_core_working;
|
||||||
bool d_dump;
|
bool d_dump;
|
||||||
unsigned int d_channel;
|
uint32_t d_channel;
|
||||||
std::string d_dump_filename;
|
std::string d_dump_filename;
|
||||||
gr_complex* d_zero_vector;
|
gr_complex* d_zero_vector;
|
||||||
gr_complex** d_in_buffer;
|
gr_complex** d_in_buffer;
|
||||||
std::vector<uint64_t> d_sample_counter_buffer;
|
std::vector<uint64_t> d_sample_counter_buffer;
|
||||||
unsigned int d_in_dwell_count;
|
uint32_t d_in_dwell_count;
|
||||||
|
|
||||||
cl::Platform d_cl_platform;
|
cl::Platform d_cl_platform;
|
||||||
cl::Device d_cl_device;
|
cl::Device d_cl_device;
|
||||||
@ -182,7 +183,7 @@ public:
|
|||||||
/*!
|
/*!
|
||||||
* \brief Returns the maximum peak of grid search.
|
* \brief Returns the maximum peak of grid search.
|
||||||
*/
|
*/
|
||||||
inline unsigned int mag() const
|
inline uint32_t mag() const
|
||||||
{
|
{
|
||||||
return d_mag;
|
return d_mag;
|
||||||
}
|
}
|
||||||
@ -219,7 +220,7 @@ public:
|
|||||||
* \brief Set acquisition channel unique ID
|
* \brief Set acquisition channel unique ID
|
||||||
* \param channel - receiver channel.
|
* \param channel - receiver channel.
|
||||||
*/
|
*/
|
||||||
inline void set_channel(unsigned int channel)
|
inline void set_channel(uint32_t channel)
|
||||||
{
|
{
|
||||||
d_channel = channel;
|
d_channel = channel;
|
||||||
}
|
}
|
||||||
@ -238,7 +239,7 @@ public:
|
|||||||
* \brief Set maximum Doppler grid search
|
* \brief Set maximum Doppler grid search
|
||||||
* \param doppler_max - Maximum Doppler shift considered in the grid search [Hz].
|
* \param doppler_max - Maximum Doppler shift considered in the grid search [Hz].
|
||||||
*/
|
*/
|
||||||
inline void set_doppler_max(unsigned int doppler_max)
|
inline void set_doppler_max(uint32_t doppler_max)
|
||||||
{
|
{
|
||||||
d_doppler_max = doppler_max;
|
d_doppler_max = doppler_max;
|
||||||
}
|
}
|
||||||
@ -247,7 +248,7 @@ public:
|
|||||||
* \brief Set Doppler steps for the grid search
|
* \brief Set Doppler steps for the grid search
|
||||||
* \param doppler_step - Frequency bin of the search grid [Hz].
|
* \param doppler_step - Frequency bin of the search grid [Hz].
|
||||||
*/
|
*/
|
||||||
inline void set_doppler_step(unsigned int doppler_step)
|
inline void set_doppler_step(uint32_t doppler_step)
|
||||||
{
|
{
|
||||||
d_doppler_step = doppler_step;
|
d_doppler_step = doppler_step;
|
||||||
}
|
}
|
||||||
|
@ -86,11 +86,6 @@ if(OPENCL_FOUND)
|
|||||||
opencl/fft_setup.cc # Needs OpenCL
|
opencl/fft_setup.cc # Needs OpenCL
|
||||||
opencl/fft_kernelstring.cc # Needs OpenCL
|
opencl/fft_kernelstring.cc # Needs OpenCL
|
||||||
)
|
)
|
||||||
set(GNSS_SPLIBS_HEADERS ${GNSS_SPLIBS_HEADERS}
|
|
||||||
opencl/fft_execute.h # Needs OpenCL
|
|
||||||
opencl/fft_setup.h # Needs OpenCL
|
|
||||||
opencl/fft_kernelstring.h # Needs OpenCL
|
|
||||||
)
|
|
||||||
endif()
|
endif()
|
||||||
|
|
||||||
include_directories(
|
include_directories(
|
||||||
|
@ -48,9 +48,9 @@
|
|||||||
|
|
||||||
#include "clFFT.h"
|
#include "clFFT.h"
|
||||||
#include "fft_internal.h"
|
#include "fft_internal.h"
|
||||||
#include <math.h>
|
#include <cmath>
|
||||||
#include <stdio.h>
|
#include <cstdio>
|
||||||
#include <stdlib.h>
|
#include <cstdlib>
|
||||||
|
|
||||||
#define max(a, b) (((a) > (b)) ? (a) : (b))
|
#define max(a, b) (((a) > (b)) ? (a) : (b))
|
||||||
#define min(a, b) (((a) < (b)) ? (a) : (b))
|
#define min(a, b) (((a) < (b)) ? (a) : (b))
|
||||||
@ -67,7 +67,7 @@ allocateTemporaryBufferInterleaved(cl_fft_plan *plan, cl_uint batchSize)
|
|||||||
if (plan->tempmemobj)
|
if (plan->tempmemobj)
|
||||||
clReleaseMemObject(plan->tempmemobj);
|
clReleaseMemObject(plan->tempmemobj);
|
||||||
|
|
||||||
plan->tempmemobj = clCreateBuffer(plan->context, CL_MEM_READ_WRITE, tmpLength, NULL, &err);
|
plan->tempmemobj = clCreateBuffer(plan->context, CL_MEM_READ_WRITE, tmpLength, nullptr, &err);
|
||||||
}
|
}
|
||||||
return err;
|
return err;
|
||||||
}
|
}
|
||||||
@ -88,8 +88,8 @@ allocateTemporaryBufferPlannar(cl_fft_plan *plan, cl_uint batchSize)
|
|||||||
if (plan->tempmemobj_imag)
|
if (plan->tempmemobj_imag)
|
||||||
clReleaseMemObject(plan->tempmemobj_imag);
|
clReleaseMemObject(plan->tempmemobj_imag);
|
||||||
|
|
||||||
plan->tempmemobj_real = clCreateBuffer(plan->context, CL_MEM_READ_WRITE, tmpLength, NULL, &err);
|
plan->tempmemobj_real = clCreateBuffer(plan->context, CL_MEM_READ_WRITE, tmpLength, nullptr, &err);
|
||||||
plan->tempmemobj_imag = clCreateBuffer(plan->context, CL_MEM_READ_WRITE, tmpLength, NULL, &terr);
|
plan->tempmemobj_imag = clCreateBuffer(plan->context, CL_MEM_READ_WRITE, tmpLength, nullptr, &terr);
|
||||||
err |= terr;
|
err |= terr;
|
||||||
}
|
}
|
||||||
return err;
|
return err;
|
||||||
@ -126,7 +126,7 @@ clFFT_ExecuteInterleaved(cl_command_queue queue, clFFT_Plan Plan, cl_int batchSi
|
|||||||
cl_int num_events, cl_event *event_list, cl_event *event)
|
cl_int num_events, cl_event *event_list, cl_event *event)
|
||||||
{
|
{
|
||||||
int s;
|
int s;
|
||||||
cl_fft_plan *plan = (cl_fft_plan *)Plan;
|
auto *plan = (cl_fft_plan *)Plan;
|
||||||
if (plan->format != clFFT_InterleavedComplexFormat)
|
if (plan->format != clFFT_InterleavedComplexFormat)
|
||||||
return CL_INVALID_VALUE;
|
return CL_INVALID_VALUE;
|
||||||
|
|
||||||
@ -180,7 +180,7 @@ clFFT_ExecuteInterleaved(cl_command_queue queue, clFFT_Plan Plan, cl_int batchSi
|
|||||||
err |= clSetKernelArg(kernelInfo->kernel, 2, sizeof(cl_int), &dir);
|
err |= clSetKernelArg(kernelInfo->kernel, 2, sizeof(cl_int), &dir);
|
||||||
err |= clSetKernelArg(kernelInfo->kernel, 3, sizeof(cl_int), &s);
|
err |= clSetKernelArg(kernelInfo->kernel, 3, sizeof(cl_int), &s);
|
||||||
|
|
||||||
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, NULL, &gWorkItems, &lWorkItems, 0, NULL, NULL);
|
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, nullptr, &gWorkItems, &lWorkItems, 0, nullptr, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
@ -203,7 +203,7 @@ clFFT_ExecuteInterleaved(cl_command_queue queue, clFFT_Plan Plan, cl_int batchSi
|
|||||||
err |= clSetKernelArg(kernelInfo->kernel, 2, sizeof(cl_int), &dir);
|
err |= clSetKernelArg(kernelInfo->kernel, 2, sizeof(cl_int), &dir);
|
||||||
err |= clSetKernelArg(kernelInfo->kernel, 3, sizeof(cl_int), &s);
|
err |= clSetKernelArg(kernelInfo->kernel, 3, sizeof(cl_int), &s);
|
||||||
|
|
||||||
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, NULL, &gWorkItems, &lWorkItems, 0, NULL, NULL);
|
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, nullptr, &gWorkItems, &lWorkItems, 0, nullptr, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
@ -223,7 +223,7 @@ clFFT_ExecutePlannar(cl_command_queue queue, clFFT_Plan Plan, cl_int batchSize,
|
|||||||
cl_int num_events, cl_event *event_list, cl_event *event)
|
cl_int num_events, cl_event *event_list, cl_event *event)
|
||||||
{
|
{
|
||||||
int s;
|
int s;
|
||||||
cl_fft_plan *plan = (cl_fft_plan *)Plan;
|
auto *plan = (cl_fft_plan *)Plan;
|
||||||
|
|
||||||
if (plan->format != clFFT_SplitComplexFormat)
|
if (plan->format != clFFT_SplitComplexFormat)
|
||||||
return CL_INVALID_VALUE;
|
return CL_INVALID_VALUE;
|
||||||
@ -285,7 +285,7 @@ clFFT_ExecutePlannar(cl_command_queue queue, clFFT_Plan Plan, cl_int batchSize,
|
|||||||
err |= clSetKernelArg(kernelInfo->kernel, 4, sizeof(cl_int), &dir);
|
err |= clSetKernelArg(kernelInfo->kernel, 4, sizeof(cl_int), &dir);
|
||||||
err |= clSetKernelArg(kernelInfo->kernel, 5, sizeof(cl_int), &s);
|
err |= clSetKernelArg(kernelInfo->kernel, 5, sizeof(cl_int), &s);
|
||||||
|
|
||||||
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, NULL, &gWorkItems, &lWorkItems, 0, NULL, NULL);
|
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, nullptr, &gWorkItems, &lWorkItems, 0, nullptr, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
@ -309,7 +309,7 @@ clFFT_ExecutePlannar(cl_command_queue queue, clFFT_Plan Plan, cl_int batchSize,
|
|||||||
err |= clSetKernelArg(kernelInfo->kernel, 4, sizeof(cl_int), &dir);
|
err |= clSetKernelArg(kernelInfo->kernel, 4, sizeof(cl_int), &dir);
|
||||||
err |= clSetKernelArg(kernelInfo->kernel, 5, sizeof(cl_int), &s);
|
err |= clSetKernelArg(kernelInfo->kernel, 5, sizeof(cl_int), &s);
|
||||||
|
|
||||||
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, NULL, &gWorkItems, &lWorkItems, 0, NULL, NULL);
|
err |= clEnqueueNDRangeKernel(queue, kernelInfo->kernel, 1, nullptr, &gWorkItems, &lWorkItems, 0, nullptr, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
@ -327,7 +327,7 @@ cl_int
|
|||||||
clFFT_1DTwistInterleaved(clFFT_Plan Plan, cl_command_queue queue, cl_mem array,
|
clFFT_1DTwistInterleaved(clFFT_Plan Plan, cl_command_queue queue, cl_mem array,
|
||||||
unsigned numRows, unsigned numCols, unsigned startRow, unsigned rowsToProcess, clFFT_Direction dir)
|
unsigned numRows, unsigned numCols, unsigned startRow, unsigned rowsToProcess, clFFT_Direction dir)
|
||||||
{
|
{
|
||||||
cl_fft_plan *plan = (cl_fft_plan *)Plan;
|
auto *plan = (cl_fft_plan *)Plan;
|
||||||
|
|
||||||
unsigned int N = numRows * numCols;
|
unsigned int N = numRows * numCols;
|
||||||
unsigned int nCols = numCols;
|
unsigned int nCols = numCols;
|
||||||
@ -337,12 +337,12 @@ clFFT_1DTwistInterleaved(clFFT_Plan Plan, cl_command_queue queue, cl_mem array,
|
|||||||
int err = 0;
|
int err = 0;
|
||||||
|
|
||||||
cl_device_id device_id;
|
cl_device_id device_id;
|
||||||
err = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(cl_device_id), &device_id, NULL);
|
err = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(cl_device_id), &device_id, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
size_t gSize;
|
size_t gSize;
|
||||||
err = clGetKernelWorkGroupInfo(plan->twist_kernel, device_id, CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &gSize, NULL);
|
err = clGetKernelWorkGroupInfo(plan->twist_kernel, device_id, CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &gSize, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
@ -357,7 +357,7 @@ clFFT_1DTwistInterleaved(clFFT_Plan Plan, cl_command_queue queue, cl_mem array,
|
|||||||
err |= clSetKernelArg(plan->twist_kernel, 4, sizeof(unsigned int), &rToProcess);
|
err |= clSetKernelArg(plan->twist_kernel, 4, sizeof(unsigned int), &rToProcess);
|
||||||
err |= clSetKernelArg(plan->twist_kernel, 5, sizeof(int), &d);
|
err |= clSetKernelArg(plan->twist_kernel, 5, sizeof(int), &d);
|
||||||
|
|
||||||
err |= clEnqueueNDRangeKernel(queue, plan->twist_kernel, 1, NULL, numGlobalThreads, numLocalThreads, 0, NULL, NULL);
|
err |= clEnqueueNDRangeKernel(queue, plan->twist_kernel, 1, nullptr, numGlobalThreads, numLocalThreads, 0, nullptr, nullptr);
|
||||||
|
|
||||||
return err;
|
return err;
|
||||||
}
|
}
|
||||||
@ -366,7 +366,7 @@ cl_int
|
|||||||
clFFT_1DTwistPlannar(clFFT_Plan Plan, cl_command_queue queue, cl_mem array_real, cl_mem array_imag,
|
clFFT_1DTwistPlannar(clFFT_Plan Plan, cl_command_queue queue, cl_mem array_real, cl_mem array_imag,
|
||||||
unsigned numRows, unsigned numCols, unsigned startRow, unsigned rowsToProcess, clFFT_Direction dir)
|
unsigned numRows, unsigned numCols, unsigned startRow, unsigned rowsToProcess, clFFT_Direction dir)
|
||||||
{
|
{
|
||||||
cl_fft_plan *plan = (cl_fft_plan *)Plan;
|
auto *plan = (cl_fft_plan *)Plan;
|
||||||
|
|
||||||
unsigned int N = numRows * numCols;
|
unsigned int N = numRows * numCols;
|
||||||
unsigned int nCols = numCols;
|
unsigned int nCols = numCols;
|
||||||
@ -376,12 +376,12 @@ clFFT_1DTwistPlannar(clFFT_Plan Plan, cl_command_queue queue, cl_mem array_real,
|
|||||||
int err = 0;
|
int err = 0;
|
||||||
|
|
||||||
cl_device_id device_id;
|
cl_device_id device_id;
|
||||||
err = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(cl_device_id), &device_id, NULL);
|
err = clGetCommandQueueInfo(queue, CL_QUEUE_DEVICE, sizeof(cl_device_id), &device_id, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
size_t gSize;
|
size_t gSize;
|
||||||
err = clGetKernelWorkGroupInfo(plan->twist_kernel, device_id, CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &gSize, NULL);
|
err = clGetKernelWorkGroupInfo(plan->twist_kernel, device_id, CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &gSize, nullptr);
|
||||||
if (err)
|
if (err)
|
||||||
return err;
|
return err;
|
||||||
|
|
||||||
@ -397,7 +397,7 @@ clFFT_1DTwistPlannar(clFFT_Plan Plan, cl_command_queue queue, cl_mem array_real,
|
|||||||
err |= clSetKernelArg(plan->twist_kernel, 5, sizeof(unsigned int), &rToProcess);
|
err |= clSetKernelArg(plan->twist_kernel, 5, sizeof(unsigned int), &rToProcess);
|
||||||
err |= clSetKernelArg(plan->twist_kernel, 6, sizeof(int), &d);
|
err |= clSetKernelArg(plan->twist_kernel, 6, sizeof(int), &d);
|
||||||
|
|
||||||
err |= clEnqueueNDRangeKernel(queue, plan->twist_kernel, 1, NULL, numGlobalThreads, numLocalThreads, 0, NULL, NULL);
|
err |= clEnqueueNDRangeKernel(queue, plan->twist_kernel, 1, nullptr, numGlobalThreads, numLocalThreads, 0, nullptr, nullptr);
|
||||||
|
|
||||||
return err;
|
return err;
|
||||||
}
|
}
|
||||||
|
@ -48,13 +48,13 @@
|
|||||||
|
|
||||||
#include "clFFT.h"
|
#include "clFFT.h"
|
||||||
#include "fft_internal.h"
|
#include "fft_internal.h"
|
||||||
#include <assert.h>
|
#include <cassert>
|
||||||
|
#include <cmath>
|
||||||
|
#include <cstdio>
|
||||||
|
#include <cstdlib>
|
||||||
|
#include <cstring>
|
||||||
#include <iostream>
|
#include <iostream>
|
||||||
#include <math.h>
|
|
||||||
#include <sstream>
|
#include <sstream>
|
||||||
#include <stdio.h>
|
|
||||||
#include <stdlib.h>
|
|
||||||
#include <string.h>
|
|
||||||
#include <string>
|
#include <string>
|
||||||
|
|
||||||
using namespace std;
|
using namespace std;
|
||||||
@ -806,13 +806,13 @@ createLocalMemfftKernelString(cl_fft_plan *plan)
|
|||||||
kernelName = string("fft") + num2str(kCount);
|
kernelName = string("fft") + num2str(kCount);
|
||||||
|
|
||||||
*kInfo = (cl_fft_kernel_info *)malloc(sizeof(cl_fft_kernel_info));
|
*kInfo = (cl_fft_kernel_info *)malloc(sizeof(cl_fft_kernel_info));
|
||||||
(*kInfo)->kernel = 0;
|
(*kInfo)->kernel = nullptr;
|
||||||
(*kInfo)->lmem_size = 0;
|
(*kInfo)->lmem_size = 0;
|
||||||
(*kInfo)->num_workgroups = 0;
|
(*kInfo)->num_workgroups = 0;
|
||||||
(*kInfo)->num_workitems_per_workgroup = 0;
|
(*kInfo)->num_workitems_per_workgroup = 0;
|
||||||
(*kInfo)->dir = cl_fft_kernel_x;
|
(*kInfo)->dir = cl_fft_kernel_x;
|
||||||
(*kInfo)->in_place_possible = 1;
|
(*kInfo)->in_place_possible = 1;
|
||||||
(*kInfo)->next = NULL;
|
(*kInfo)->next = nullptr;
|
||||||
(*kInfo)->kernel_name = (char *)malloc(sizeof(char) * (kernelName.size() + 1));
|
(*kInfo)->kernel_name = (char *)malloc(sizeof(char) * (kernelName.size() + 1));
|
||||||
strcpy((*kInfo)->kernel_name, kernelName.c_str());
|
strcpy((*kInfo)->kernel_name, kernelName.c_str());
|
||||||
|
|
||||||
@ -1015,7 +1015,7 @@ createGlobalFFTKernelString(cl_fft_plan *plan, int n, int BS, cl_fft_kernel_dir
|
|||||||
|
|
||||||
kernelName = string("fft") + num2str(kCount);
|
kernelName = string("fft") + num2str(kCount);
|
||||||
*kInfo = (cl_fft_kernel_info *)malloc(sizeof(cl_fft_kernel_info));
|
*kInfo = (cl_fft_kernel_info *)malloc(sizeof(cl_fft_kernel_info));
|
||||||
(*kInfo)->kernel = 0;
|
(*kInfo)->kernel = nullptr;
|
||||||
if (R2 == 1)
|
if (R2 == 1)
|
||||||
(*kInfo)->lmem_size = 0;
|
(*kInfo)->lmem_size = 0;
|
||||||
else
|
else
|
||||||
@ -1033,7 +1033,7 @@ createGlobalFFTKernelString(cl_fft_plan *plan, int n, int BS, cl_fft_kernel_dir
|
|||||||
(*kInfo)->in_place_possible = 1;
|
(*kInfo)->in_place_possible = 1;
|
||||||
else
|
else
|
||||||
(*kInfo)->in_place_possible = 0;
|
(*kInfo)->in_place_possible = 0;
|
||||||
(*kInfo)->next = NULL;
|
(*kInfo)->next = nullptr;
|
||||||
(*kInfo)->kernel_name = (char *)malloc(sizeof(char) * (kernelName.size() + 1));
|
(*kInfo)->kernel_name = (char *)malloc(sizeof(char) * (kernelName.size() + 1));
|
||||||
strcpy((*kInfo)->kernel_name, kernelName.c_str());
|
strcpy((*kInfo)->kernel_name, kernelName.c_str());
|
||||||
|
|
||||||
|
@ -48,11 +48,11 @@
|
|||||||
|
|
||||||
#include "fft_base_kernels.h"
|
#include "fft_base_kernels.h"
|
||||||
#include "fft_internal.h"
|
#include "fft_internal.h"
|
||||||
|
#include <cstdlib>
|
||||||
|
#include <cstring>
|
||||||
#include <iostream>
|
#include <iostream>
|
||||||
#include <limits>
|
#include <limits>
|
||||||
#include <sstream>
|
#include <sstream>
|
||||||
#include <stdlib.h>
|
|
||||||
#include <string.h>
|
|
||||||
#include <string>
|
#include <string>
|
||||||
#include <sys/stat.h>
|
#include <sys/stat.h>
|
||||||
#include <sys/types.h>
|
#include <sys/types.h>
|
||||||
@ -128,37 +128,37 @@ destroy_plan(cl_fft_plan *Plan)
|
|||||||
kernel_info = tmp;
|
kernel_info = tmp;
|
||||||
}
|
}
|
||||||
|
|
||||||
Plan->kernel_info = NULL;
|
Plan->kernel_info = nullptr;
|
||||||
|
|
||||||
if (Plan->kernel_string)
|
if (Plan->kernel_string)
|
||||||
{
|
{
|
||||||
delete Plan->kernel_string;
|
delete Plan->kernel_string;
|
||||||
Plan->kernel_string = NULL;
|
Plan->kernel_string = nullptr;
|
||||||
}
|
}
|
||||||
if (Plan->twist_kernel)
|
if (Plan->twist_kernel)
|
||||||
{
|
{
|
||||||
clReleaseKernel(Plan->twist_kernel);
|
clReleaseKernel(Plan->twist_kernel);
|
||||||
Plan->twist_kernel = NULL;
|
Plan->twist_kernel = nullptr;
|
||||||
}
|
}
|
||||||
if (Plan->program)
|
if (Plan->program)
|
||||||
{
|
{
|
||||||
clReleaseProgram(Plan->program);
|
clReleaseProgram(Plan->program);
|
||||||
Plan->program = NULL;
|
Plan->program = nullptr;
|
||||||
}
|
}
|
||||||
if (Plan->tempmemobj)
|
if (Plan->tempmemobj)
|
||||||
{
|
{
|
||||||
clReleaseMemObject(Plan->tempmemobj);
|
clReleaseMemObject(Plan->tempmemobj);
|
||||||
Plan->tempmemobj = NULL;
|
Plan->tempmemobj = nullptr;
|
||||||
}
|
}
|
||||||
if (Plan->tempmemobj_real)
|
if (Plan->tempmemobj_real)
|
||||||
{
|
{
|
||||||
clReleaseMemObject(Plan->tempmemobj_real);
|
clReleaseMemObject(Plan->tempmemobj_real);
|
||||||
Plan->tempmemobj_real = NULL;
|
Plan->tempmemobj_real = nullptr;
|
||||||
}
|
}
|
||||||
if (Plan->tempmemobj_imag)
|
if (Plan->tempmemobj_imag)
|
||||||
{
|
{
|
||||||
clReleaseMemObject(Plan->tempmemobj_imag);
|
clReleaseMemObject(Plan->tempmemobj_imag);
|
||||||
Plan->tempmemobj_imag = NULL;
|
Plan->tempmemobj_imag = nullptr;
|
||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
@ -201,7 +201,7 @@ int getMaxKernelWorkGroupSize(cl_fft_plan *plan, unsigned int *max_wg_size, unsi
|
|||||||
cl_fft_kernel_info *kInfo = plan->kernel_info;
|
cl_fft_kernel_info *kInfo = plan->kernel_info;
|
||||||
while (kInfo)
|
while (kInfo)
|
||||||
{
|
{
|
||||||
err = clGetKernelWorkGroupInfo(kInfo->kernel, devices[i], CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &wg_size, NULL);
|
err = clGetKernelWorkGroupInfo(kInfo->kernel, devices[i], CL_KERNEL_WORK_GROUP_SIZE, sizeof(size_t), &wg_size, nullptr);
|
||||||
if (err != CL_SUCCESS)
|
if (err != CL_SUCCESS)
|
||||||
return -1;
|
return -1;
|
||||||
|
|
||||||
@ -235,7 +235,7 @@ clFFT_CreatePlan(cl_context context, clFFT_Dim3 n, clFFT_Dimension dim, clFFT_Da
|
|||||||
int i;
|
int i;
|
||||||
cl_int err;
|
cl_int err;
|
||||||
int isPow2 = 1;
|
int isPow2 = 1;
|
||||||
cl_fft_plan *plan = NULL;
|
cl_fft_plan *plan = nullptr;
|
||||||
ostringstream kString;
|
ostringstream kString;
|
||||||
int num_devices;
|
int num_devices;
|
||||||
int gpu_found = 0;
|
int gpu_found = 0;
|
||||||
@ -265,15 +265,15 @@ clFFT_CreatePlan(cl_context context, clFFT_Dim3 n, clFFT_Dimension dim, clFFT_Da
|
|||||||
plan->n = n;
|
plan->n = n;
|
||||||
plan->dim = dim;
|
plan->dim = dim;
|
||||||
plan->format = dataFormat;
|
plan->format = dataFormat;
|
||||||
plan->kernel_info = 0;
|
plan->kernel_info = nullptr;
|
||||||
plan->num_kernels = 0;
|
plan->num_kernels = 0;
|
||||||
plan->twist_kernel = 0;
|
plan->twist_kernel = nullptr;
|
||||||
plan->program = 0;
|
plan->program = nullptr;
|
||||||
plan->temp_buffer_needed = 0;
|
plan->temp_buffer_needed = 0;
|
||||||
plan->last_batch_size = 0;
|
plan->last_batch_size = 0;
|
||||||
plan->tempmemobj = 0;
|
plan->tempmemobj = nullptr;
|
||||||
plan->tempmemobj_real = 0;
|
plan->tempmemobj_real = nullptr;
|
||||||
plan->tempmemobj_imag = 0;
|
plan->tempmemobj_imag = nullptr;
|
||||||
plan->max_localmem_fft_size = 2048;
|
plan->max_localmem_fft_size = 2048;
|
||||||
plan->max_work_item_per_workgroup = 256;
|
plan->max_work_item_per_workgroup = 256;
|
||||||
plan->max_radix = 16;
|
plan->max_radix = 16;
|
||||||
@ -289,7 +289,7 @@ patch_kernel_source:
|
|||||||
getBlockConfigAndKernelString(plan);
|
getBlockConfigAndKernelString(plan);
|
||||||
|
|
||||||
const char *source_str = plan->kernel_string->c_str();
|
const char *source_str = plan->kernel_string->c_str();
|
||||||
plan->program = clCreateProgramWithSource(context, 1, (const char **)&source_str, NULL, &err);
|
plan->program = clCreateProgramWithSource(context, 1, (const char **)&source_str, nullptr, &err);
|
||||||
ERR_MACRO(err);
|
ERR_MACRO(err);
|
||||||
|
|
||||||
err = clGetContextInfo(context, CL_CONTEXT_DEVICES, sizeof(devices), devices, &ret_size);
|
err = clGetContextInfo(context, CL_CONTEXT_DEVICES, sizeof(devices), devices, &ret_size);
|
||||||
@ -299,28 +299,28 @@ patch_kernel_source:
|
|||||||
|
|
||||||
for (i = 0; i < num_devices; i++)
|
for (i = 0; i < num_devices; i++)
|
||||||
{
|
{
|
||||||
err = clGetDeviceInfo(devices[i], CL_DEVICE_TYPE, sizeof(device_type), &device_type, NULL);
|
err = clGetDeviceInfo(devices[i], CL_DEVICE_TYPE, sizeof(device_type), &device_type, nullptr);
|
||||||
ERR_MACRO(err);
|
ERR_MACRO(err);
|
||||||
|
|
||||||
if (device_type == CL_DEVICE_TYPE_GPU)
|
if (device_type == CL_DEVICE_TYPE_GPU)
|
||||||
{
|
{
|
||||||
gpu_found = 1;
|
gpu_found = 1;
|
||||||
err = clBuildProgram(plan->program, 1, &devices[i], "-cl-mad-enable", NULL, NULL);
|
err = clBuildProgram(plan->program, 1, &devices[i], "-cl-mad-enable", nullptr, nullptr);
|
||||||
if (err != CL_SUCCESS)
|
if (err != CL_SUCCESS)
|
||||||
{
|
{
|
||||||
char *build_log;
|
char *build_log;
|
||||||
char devicename[200];
|
char devicename[200];
|
||||||
size_t log_size;
|
size_t log_size;
|
||||||
|
|
||||||
err = clGetProgramBuildInfo(plan->program, devices[i], CL_PROGRAM_BUILD_LOG, 0, NULL, &log_size);
|
err = clGetProgramBuildInfo(plan->program, devices[i], CL_PROGRAM_BUILD_LOG, 0, nullptr, &log_size);
|
||||||
ERR_MACRO(err);
|
ERR_MACRO(err);
|
||||||
|
|
||||||
build_log = (char *)malloc(log_size + 1);
|
build_log = (char *)malloc(log_size + 1);
|
||||||
|
|
||||||
err = clGetProgramBuildInfo(plan->program, devices[i], CL_PROGRAM_BUILD_LOG, log_size, build_log, NULL);
|
err = clGetProgramBuildInfo(plan->program, devices[i], CL_PROGRAM_BUILD_LOG, log_size, build_log, nullptr);
|
||||||
ERR_MACRO(err);
|
ERR_MACRO(err);
|
||||||
|
|
||||||
err = clGetDeviceInfo(devices[i], CL_DEVICE_NAME, sizeof(devicename), devicename, NULL);
|
err = clGetDeviceInfo(devices[i], CL_DEVICE_NAME, sizeof(devicename), devicename, nullptr);
|
||||||
ERR_MACRO(err);
|
ERR_MACRO(err);
|
||||||
|
|
||||||
fprintf(stdout, "FFT program build log on device %s\n", devicename);
|
fprintf(stdout, "FFT program build log on device %s\n", devicename);
|
||||||
@ -370,7 +370,7 @@ patch_kernel_source:
|
|||||||
|
|
||||||
void clFFT_DestroyPlan(clFFT_Plan plan)
|
void clFFT_DestroyPlan(clFFT_Plan plan)
|
||||||
{
|
{
|
||||||
cl_fft_plan *Plan = (cl_fft_plan *)plan;
|
auto *Plan = (cl_fft_plan *)plan;
|
||||||
if (Plan)
|
if (Plan)
|
||||||
{
|
{
|
||||||
destroy_plan(Plan);
|
destroy_plan(Plan);
|
||||||
@ -388,7 +388,7 @@ void clFFT_DumpPlan(clFFT_Plan Plan, FILE *file)
|
|||||||
else
|
else
|
||||||
out = file;
|
out = file;
|
||||||
|
|
||||||
cl_fft_plan *plan = (cl_fft_plan *)Plan;
|
auto *plan = (cl_fft_plan *)Plan;
|
||||||
cl_fft_kernel_info *kInfo = plan->kernel_info;
|
cl_fft_kernel_info *kInfo = plan->kernel_info;
|
||||||
|
|
||||||
while (kInfo)
|
while (kInfo)
|
||||||
|
@ -113,8 +113,7 @@ CustomUDPSignalSource::CustomUDPSignalSource(ConfigurationInterface* configurati
|
|||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
CustomUDPSignalSource::~CustomUDPSignalSource()
|
CustomUDPSignalSource::~CustomUDPSignalSource() = default;
|
||||||
= default;
|
|
||||||
|
|
||||||
|
|
||||||
void CustomUDPSignalSource::connect(gr::top_block_sptr top_block)
|
void CustomUDPSignalSource::connect(gr::top_block_sptr top_block)
|
||||||
|
@ -69,7 +69,7 @@ OsmosdrSignalSource::OsmosdrSignalSource(ConfigurationInterface* configuration,
|
|||||||
|
|
||||||
if (item_type_ == "short")
|
if (item_type_ == "short")
|
||||||
{
|
{
|
||||||
item_size_ = sizeof(short);
|
item_size_ = sizeof(int16_t);
|
||||||
}
|
}
|
||||||
else if (item_type_ == "gr_complex")
|
else if (item_type_ == "gr_complex")
|
||||||
{
|
{
|
||||||
@ -131,7 +131,7 @@ OsmosdrSignalSource::OsmosdrSignalSource(ConfigurationInterface* configuration,
|
|||||||
else
|
else
|
||||||
{
|
{
|
||||||
LOG(WARNING) << item_type_ << " unrecognized item type. Using short.";
|
LOG(WARNING) << item_type_ << " unrecognized item type. Using short.";
|
||||||
item_size_ = sizeof(short);
|
item_size_ = sizeof(int16_t);
|
||||||
}
|
}
|
||||||
|
|
||||||
if (samples_ != 0)
|
if (samples_ != 0)
|
||||||
@ -158,8 +158,7 @@ OsmosdrSignalSource::OsmosdrSignalSource(ConfigurationInterface* configuration,
|
|||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
OsmosdrSignalSource::~OsmosdrSignalSource()
|
OsmosdrSignalSource::~OsmosdrSignalSource() = default;
|
||||||
= default;
|
|
||||||
|
|
||||||
|
|
||||||
void OsmosdrSignalSource::driver_instance()
|
void OsmosdrSignalSource::driver_instance()
|
||||||
|
@ -37,6 +37,7 @@
|
|||||||
#include <boost/shared_ptr.hpp>
|
#include <boost/shared_ptr.hpp>
|
||||||
#include <gnuradio/blocks/file_sink.h>
|
#include <gnuradio/blocks/file_sink.h>
|
||||||
#include <gnuradio/msg_queue.h>
|
#include <gnuradio/msg_queue.h>
|
||||||
|
#include <cstdint>
|
||||||
#include <osmosdr/source.h>
|
#include <osmosdr/source.h>
|
||||||
#include <stdexcept>
|
#include <stdexcept>
|
||||||
#include <string>
|
#include <string>
|
||||||
@ -98,7 +99,7 @@ private:
|
|||||||
|
|
||||||
std::string item_type_;
|
std::string item_type_;
|
||||||
size_t item_size_;
|
size_t item_size_;
|
||||||
long samples_;
|
int64_t samples_;
|
||||||
bool dump_;
|
bool dump_;
|
||||||
std::string dump_filename_;
|
std::string dump_filename_;
|
||||||
|
|
||||||
|
@ -106,8 +106,7 @@ PlutosdrSignalSource::PlutosdrSignalSource(ConfigurationInterface* configuration
|
|||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
PlutosdrSignalSource::~PlutosdrSignalSource()
|
PlutosdrSignalSource::~PlutosdrSignalSource() = default;
|
||||||
= default;
|
|
||||||
|
|
||||||
|
|
||||||
void PlutosdrSignalSource::connect(gr::top_block_sptr top_block)
|
void PlutosdrSignalSource::connect(gr::top_block_sptr top_block)
|
||||||
|
@ -32,9 +32,9 @@
|
|||||||
|
|
||||||
#include "gr_complex_ip_packet_source.h"
|
#include "gr_complex_ip_packet_source.h"
|
||||||
#include <gnuradio/io_signature.h>
|
#include <gnuradio/io_signature.h>
|
||||||
|
#include <cstdint>
|
||||||
#include <utility>
|
#include <utility>
|
||||||
|
|
||||||
|
|
||||||
const int FIFO_SIZE = 1472000;
|
const int FIFO_SIZE = 1472000;
|
||||||
|
|
||||||
|
|
||||||
@ -77,11 +77,11 @@ typedef struct gr_udp_header
|
|||||||
|
|
||||||
gr_complex_ip_packet_source::sptr
|
gr_complex_ip_packet_source::sptr
|
||||||
gr_complex_ip_packet_source::make(std::string src_device,
|
gr_complex_ip_packet_source::make(std::string src_device,
|
||||||
const std::string& origin_address,
|
const std::string &origin_address,
|
||||||
int udp_port,
|
int udp_port,
|
||||||
int udp_packet_size,
|
int udp_packet_size,
|
||||||
int n_baseband_channels,
|
int n_baseband_channels,
|
||||||
const std::string& wire_sample_type,
|
const std::string &wire_sample_type,
|
||||||
size_t item_size,
|
size_t item_size,
|
||||||
bool IQ_swap_)
|
bool IQ_swap_)
|
||||||
{
|
{
|
||||||
@ -100,11 +100,11 @@ gr_complex_ip_packet_source::make(std::string src_device,
|
|||||||
* The private constructor
|
* The private constructor
|
||||||
*/
|
*/
|
||||||
gr_complex_ip_packet_source::gr_complex_ip_packet_source(std::string src_device,
|
gr_complex_ip_packet_source::gr_complex_ip_packet_source(std::string src_device,
|
||||||
__attribute__((unused)) const std::string& origin_address,
|
__attribute__((unused)) const std::string &origin_address,
|
||||||
int udp_port,
|
int udp_port,
|
||||||
int udp_packet_size,
|
int udp_packet_size,
|
||||||
int n_baseband_channels,
|
int n_baseband_channels,
|
||||||
const std::string& wire_sample_type,
|
const std::string &wire_sample_type,
|
||||||
size_t item_size,
|
size_t item_size,
|
||||||
bool IQ_swap_)
|
bool IQ_swap_)
|
||||||
: gr::sync_block("gr_complex_ip_packet_source",
|
: gr::sync_block("gr_complex_ip_packet_source",
|
||||||
@ -328,7 +328,7 @@ void gr_complex_ip_packet_source::demux_samples(gr_vector_void_star output_items
|
|||||||
switch (d_wire_sample_type)
|
switch (d_wire_sample_type)
|
||||||
{
|
{
|
||||||
case 1: // interleaved byte samples
|
case 1: // interleaved byte samples
|
||||||
for (auto & output_item : output_items)
|
for (auto &output_item : output_items)
|
||||||
{
|
{
|
||||||
real = fifo_buff[fifo_read_ptr++];
|
real = fifo_buff[fifo_read_ptr++];
|
||||||
imag = fifo_buff[fifo_read_ptr++];
|
imag = fifo_buff[fifo_read_ptr++];
|
||||||
@ -343,7 +343,7 @@ void gr_complex_ip_packet_source::demux_samples(gr_vector_void_star output_items
|
|||||||
}
|
}
|
||||||
break;
|
break;
|
||||||
case 2: // 4-bit samples
|
case 2: // 4-bit samples
|
||||||
for (auto & output_item : output_items)
|
for (auto &output_item : output_items)
|
||||||
{
|
{
|
||||||
tmp_char2 = fifo_buff[fifo_read_ptr] & 0x0F;
|
tmp_char2 = fifo_buff[fifo_read_ptr] & 0x0F;
|
||||||
if (tmp_char2 >= 8)
|
if (tmp_char2 >= 8)
|
||||||
@ -391,7 +391,7 @@ int gr_complex_ip_packet_source::work(int noutput_items,
|
|||||||
boost::mutex::scoped_lock lock(d_mutex); // hold mutex for duration of this function
|
boost::mutex::scoped_lock lock(d_mutex); // hold mutex for duration of this function
|
||||||
if (fifo_items == 0) return 0;
|
if (fifo_items == 0) return 0;
|
||||||
|
|
||||||
if (output_items.size() > static_cast<long unsigned int>(d_n_baseband_channels))
|
if (output_items.size() > static_cast<uint64_t>(d_n_baseband_channels))
|
||||||
{
|
{
|
||||||
std::cout << "Configuration error: more baseband channels connected than the available in the UDP source\n";
|
std::cout << "Configuration error: more baseband channels connected than the available in the UDP source\n";
|
||||||
exit(0);
|
exit(0);
|
||||||
@ -440,7 +440,7 @@ int gr_complex_ip_packet_source::work(int noutput_items,
|
|||||||
// update fifo items
|
// update fifo items
|
||||||
fifo_items = fifo_items - bytes_requested;
|
fifo_items = fifo_items - bytes_requested;
|
||||||
|
|
||||||
for (long unsigned int n = 0; n < output_items.size(); n++)
|
for (uint64_t n = 0; n < output_items.size(); n++)
|
||||||
{
|
{
|
||||||
produce(static_cast<int>(n), num_samples_readed);
|
produce(static_cast<int>(n), num_samples_readed);
|
||||||
}
|
}
|
||||||
|
@ -84,19 +84,19 @@ private:
|
|||||||
public:
|
public:
|
||||||
typedef boost::shared_ptr<gr_complex_ip_packet_source> sptr;
|
typedef boost::shared_ptr<gr_complex_ip_packet_source> sptr;
|
||||||
static sptr make(std::string src_device,
|
static sptr make(std::string src_device,
|
||||||
const std::string& origin_address,
|
const std::string &origin_address,
|
||||||
int udp_port,
|
int udp_port,
|
||||||
int udp_packet_size,
|
int udp_packet_size,
|
||||||
int n_baseband_channels,
|
int n_baseband_channels,
|
||||||
const std::string& wire_sample_type,
|
const std::string &wire_sample_type,
|
||||||
size_t item_size,
|
size_t item_size,
|
||||||
bool IQ_swap_);
|
bool IQ_swap_);
|
||||||
gr_complex_ip_packet_source(std::string src_device,
|
gr_complex_ip_packet_source(std::string src_device,
|
||||||
const std::string& origin_address,
|
const std::string &origin_address,
|
||||||
int udp_port,
|
int udp_port,
|
||||||
int udp_packet_size,
|
int udp_packet_size,
|
||||||
int n_baseband_channels,
|
int n_baseband_channels,
|
||||||
const std::string& wire_sample_type,
|
const std::string &wire_sample_type,
|
||||||
size_t item_size,
|
size_t item_size,
|
||||||
bool IQ_swap_);
|
bool IQ_swap_);
|
||||||
~gr_complex_ip_packet_source();
|
~gr_complex_ip_packet_source();
|
||||||
|
@ -179,9 +179,9 @@ bool cfg_ad9361_streaming_ch(struct iio_context *ctx, struct stream_cfg *cfg, en
|
|||||||
bool config_ad9361_rx_local(uint64_t bandwidth_,
|
bool config_ad9361_rx_local(uint64_t bandwidth_,
|
||||||
uint64_t sample_rate_,
|
uint64_t sample_rate_,
|
||||||
uint64_t freq_,
|
uint64_t freq_,
|
||||||
const std::string& rf_port_select_,
|
const std::string &rf_port_select_,
|
||||||
const std::string& gain_mode_rx1_,
|
const std::string &gain_mode_rx1_,
|
||||||
const std::string& gain_mode_rx2_,
|
const std::string &gain_mode_rx2_,
|
||||||
double rf_gain_rx1_,
|
double rf_gain_rx1_,
|
||||||
double rf_gain_rx2_)
|
double rf_gain_rx2_)
|
||||||
|
|
||||||
@ -291,13 +291,13 @@ bool config_ad9361_rx_local(uint64_t bandwidth_,
|
|||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
bool config_ad9361_rx_remote(const std::string& remote_host,
|
bool config_ad9361_rx_remote(const std::string &remote_host,
|
||||||
uint64_t bandwidth_,
|
uint64_t bandwidth_,
|
||||||
uint64_t sample_rate_,
|
uint64_t sample_rate_,
|
||||||
uint64_t freq_,
|
uint64_t freq_,
|
||||||
const std::string& rf_port_select_,
|
const std::string &rf_port_select_,
|
||||||
const std::string& gain_mode_rx1_,
|
const std::string &gain_mode_rx1_,
|
||||||
const std::string& gain_mode_rx2_,
|
const std::string &gain_mode_rx2_,
|
||||||
double rf_gain_rx1_,
|
double rf_gain_rx1_,
|
||||||
double rf_gain_rx2_)
|
double rf_gain_rx2_)
|
||||||
{
|
{
|
||||||
@ -543,7 +543,7 @@ bool config_ad9361_lo_local(uint64_t bandwidth_,
|
|||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
bool config_ad9361_lo_remote(const std::string& remote_host,
|
bool config_ad9361_lo_remote(const std::string &remote_host,
|
||||||
uint64_t bandwidth_,
|
uint64_t bandwidth_,
|
||||||
uint64_t sample_rate_,
|
uint64_t sample_rate_,
|
||||||
uint64_t freq_rf_tx_hz_,
|
uint64_t freq_rf_tx_hz_,
|
||||||
@ -680,7 +680,7 @@ bool config_ad9361_lo_remote(const std::string& remote_host,
|
|||||||
}
|
}
|
||||||
|
|
||||||
|
|
||||||
bool ad9361_disable_lo_remote(const std::string& remote_host)
|
bool ad9361_disable_lo_remote(const std::string &remote_host)
|
||||||
{
|
{
|
||||||
std::cout << "AD9361 Acquiring IIO REMOTE context in host " << remote_host << std::endl;
|
std::cout << "AD9361 Acquiring IIO REMOTE context in host " << remote_host << std::endl;
|
||||||
struct iio_context *ctx;
|
struct iio_context *ctx;
|
||||||
|
@ -92,19 +92,19 @@ bool cfg_ad9361_streaming_ch(struct iio_context *ctx, struct stream_cfg *cfg, en
|
|||||||
bool config_ad9361_rx_local(uint64_t bandwidth_,
|
bool config_ad9361_rx_local(uint64_t bandwidth_,
|
||||||
uint64_t sample_rate_,
|
uint64_t sample_rate_,
|
||||||
uint64_t freq_,
|
uint64_t freq_,
|
||||||
const std::string& rf_port_select_,
|
const std::string &rf_port_select_,
|
||||||
const std::string& gain_mode_rx1_,
|
const std::string &gain_mode_rx1_,
|
||||||
const std::string& gain_mode_rx2_,
|
const std::string &gain_mode_rx2_,
|
||||||
double rf_gain_rx1_,
|
double rf_gain_rx1_,
|
||||||
double rf_gain_rx2_);
|
double rf_gain_rx2_);
|
||||||
|
|
||||||
bool config_ad9361_rx_remote(const std::string& remote_host,
|
bool config_ad9361_rx_remote(const std::string &remote_host,
|
||||||
uint64_t bandwidth_,
|
uint64_t bandwidth_,
|
||||||
uint64_t sample_rate_,
|
uint64_t sample_rate_,
|
||||||
uint64_t freq_,
|
uint64_t freq_,
|
||||||
const std::string& rf_port_select_,
|
const std::string &rf_port_select_,
|
||||||
const std::string& gain_mode_rx1_,
|
const std::string &gain_mode_rx1_,
|
||||||
const std::string& gain_mode_rx2_,
|
const std::string &gain_mode_rx2_,
|
||||||
double rf_gain_rx1_,
|
double rf_gain_rx1_,
|
||||||
double rf_gain_rx2_);
|
double rf_gain_rx2_);
|
||||||
|
|
||||||
@ -115,7 +115,7 @@ bool config_ad9361_lo_local(uint64_t bandwidth_,
|
|||||||
int64_t freq_dds_tx_hz_,
|
int64_t freq_dds_tx_hz_,
|
||||||
double scale_dds_dbfs_);
|
double scale_dds_dbfs_);
|
||||||
|
|
||||||
bool config_ad9361_lo_remote(const std::string& remote_host,
|
bool config_ad9361_lo_remote(const std::string &remote_host,
|
||||||
uint64_t bandwidth_,
|
uint64_t bandwidth_,
|
||||||
uint64_t sample_rate_,
|
uint64_t sample_rate_,
|
||||||
uint64_t freq_rf_tx_hz_,
|
uint64_t freq_rf_tx_hz_,
|
||||||
@ -124,7 +124,7 @@ bool config_ad9361_lo_remote(const std::string& remote_host,
|
|||||||
double scale_dds_dbfs_);
|
double scale_dds_dbfs_);
|
||||||
|
|
||||||
|
|
||||||
bool ad9361_disable_lo_remote(const std::string& remote_host);
|
bool ad9361_disable_lo_remote(const std::string &remote_host);
|
||||||
|
|
||||||
bool ad9361_disable_lo_local();
|
bool ad9361_disable_lo_local();
|
||||||
|
|
||||||
|
@ -45,7 +45,7 @@
|
|||||||
const size_t PAGE_SIZE = 0x10000;
|
const size_t PAGE_SIZE = 0x10000;
|
||||||
const unsigned int TEST_REGISTER_TRACK_WRITEVAL = 0x55AA;
|
const unsigned int TEST_REGISTER_TRACK_WRITEVAL = 0x55AA;
|
||||||
|
|
||||||
fpga_switch::fpga_switch(const std::string& device_name)
|
fpga_switch::fpga_switch(const std::string &device_name)
|
||||||
{
|
{
|
||||||
if ((d_device_descriptor = open(device_name.c_str(), O_RDWR | O_SYNC)) == -1)
|
if ((d_device_descriptor = open(device_name.c_str(), O_RDWR | O_SYNC)) == -1)
|
||||||
{
|
{
|
||||||
|
@ -50,7 +50,7 @@ public:
|
|||||||
|
|
||||||
private:
|
private:
|
||||||
int d_device_descriptor; // driver descriptor
|
int d_device_descriptor; // driver descriptor
|
||||||
volatile unsigned *d_map_base; // driver memory map
|
volatile unsigned* d_map_base; // driver memory map
|
||||||
|
|
||||||
// private functions
|
// private functions
|
||||||
unsigned fpga_switch_test_register(unsigned writeval);
|
unsigned fpga_switch_test_register(unsigned writeval);
|
||||||
|
Loading…
x
Reference in New Issue
Block a user