Skip to content

O
OpenXG-RAN
Commit ec286386 authored by hardy's avatar hardy
Browse files
Options

Merge remote-tracking branch 'origin/NR-LogsSanAddrFix' into integration_2022_wk02

parents 60292489 a7b6c4bc
Hide whitespace changes
Inline Side-by-side
Showing with 176 additions and 158 deletions
cmake_targets/CMakeLists.txt
View file @ ec286386
...@@ -1541,6 +1541,10 @@ set(PHY_LDPC_CL_SRC ...@@ -1541,6 +1541,10 @@ set(PHY_LDPC_CL_SRC
${OPENAIR1_DIR}/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_CL.c ${OPENAIR1_DIR}/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_CL.c
${OPENAIR1_DIR}/PHY/CODING/nrLDPC_encoder/ldpc_encoder_optim8segmulti.c ${OPENAIR1_DIR}/PHY/CODING/nrLDPC_encoder/ldpc_encoder_optim8segmulti.c
) )
add_custom_target( nrLDPC_decoder_kernels_CL
COMMAND gcc ${OPENAIR1_DIR}/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_CL.c -dD -DNRLDPC_KERNEL_SOURCE -E -o ${CMAKE_CURRENT_BINARY_DIR}/nrLDPC_decoder_kernels_CL.clc
SOURCES ${OPENAIR1_DIR}/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_CL.c
)
set(PHY_NR_CODINGIF set(PHY_NR_CODINGIF
${OPENAIR1_DIR}/PHY/CODING/nrLDPC_load.c; ${OPENAIR1_DIR}/PHY/CODING/nrLDPC_load.c;
...@@ -1551,6 +1555,7 @@ add_library(ldpc_optim MODULE ${PHY_LDPC_OPTIM_SRC} ) ...@@ -1551,6 +1555,7 @@ add_library(ldpc_optim MODULE ${PHY_LDPC_OPTIM_SRC} )
add_library(ldpc_optim8seg MODULE ${PHY_LDPC_OPTIM8SEG_SRC} ) add_library(ldpc_optim8seg MODULE ${PHY_LDPC_OPTIM8SEG_SRC} )
add_library(ldpc_cl MODULE ${PHY_LDPC_CL_SRC} ) add_library(ldpc_cl MODULE ${PHY_LDPC_CL_SRC} )
target_link_libraries(ldpc_cl OpenCL) target_link_libraries(ldpc_cl OpenCL)
add_dependencies(ldpc_cl nrLDPC_decoder_kernels_CL)
if (CUDA_FOUND) if (CUDA_FOUND)
cuda_add_library(ldpc_cuda MODULE ${PHY_LDPC_CUDA_SRC} ) cuda_add_library(ldpc_cuda MODULE ${PHY_LDPC_CUDA_SRC} )
......
common/config/config_load_configmodule.c
View file @ ec286386
...@@ -224,11 +224,9 @@ configmodule_interface_t *load_configmodule(int argc, ...@@ -224,11 +224,9 @@ configmodule_interface_t *load_configmodule(int argc,
cfgparam = getenv("OAI_CONFIGMODULE"); cfgparam = getenv("OAI_CONFIGMODULE");
} }
/* default different for UE and softmodem because UE doesn't use config file*/ /* default different for UE and softmodem because UE may run without config file */
/* and -O option is not mandatory for UE */ /* and -O option is not mandatory for UE */
/* phy simulators behave as UE */ /* phy simulators behave as UE */
/* test of exec name would better be replaced by a parameter to the l */
/* oad_configmodule function */
if (cfgparam == NULL) { if (cfgparam == NULL) {
tmpflags = tmpflags | CONFIG_NOOOPT; tmpflags = tmpflags | CONFIG_NOOOPT;
...@@ -254,9 +252,14 @@ configmodule_interface_t *load_configmodule(int argc, ...@@ -254,9 +252,14 @@ configmodule_interface_t *load_configmodule(int argc,
} }
cfgptr = calloc(sizeof(configmodule_interface_t),1); cfgptr = calloc(sizeof(configmodule_interface_t),1);
/* argv_info is used to memorize command line options which have been recognized */
/* and to detect unrecognized command line options which might have been specified */
cfgptr->argv_info = calloc(sizeof(int32_t), argc); cfgptr->argv_info = calloc(sizeof(int32_t), argc);
/* argv[0] is the exec name, always Ok */
cfgptr->argv_info[0] |= CONFIG_CMDLINEOPT_PROCESSED; cfgptr->argv_info[0] |= CONFIG_CMDLINEOPT_PROCESSED;
/* when OoptIdx is >0, -O option has been detected at position OoptIdx
* we must memorize arv[OoptIdx is Ok */
if (OoptIdx >= 0) { if (OoptIdx >= 0) {
cfgptr->argv_info[OoptIdx] |= CONFIG_CMDLINEOPT_PROCESSED; cfgptr->argv_info[OoptIdx] |= CONFIG_CMDLINEOPT_PROCESSED;
cfgptr->argv_info[OoptIdx+1] |= CONFIG_CMDLINEOPT_PROCESSED; cfgptr->argv_info[OoptIdx+1] |= CONFIG_CMDLINEOPT_PROCESSED;
......
openair1/PHY/CODING/DOC/LDPCImplementation.md
View file @ ec286386
...@@ -24,12 +24,33 @@ loading `libldpc_cl.so` instead of `libldpc.so`: ...@@ -24,12 +24,33 @@ loading `libldpc_cl.so` instead of `libldpc.so`:
`make ldpc_cl` `make ldpc_cl`
`cp ../../../openair1/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_kernels_CL.cl` This command creates the `libldpc_cl.so` shared library. To perform this build successfully, only the OpenCL header `(/usr/include/CL/opencl.h)` and library `(/usr/lib/x86_64-linux-gnu/libOpenCL.so)`are required, they implement OpenCL API support which is not hardware dependent.
```
Scanning dependencies of target nrLDPC_decoder_kernels_CL
Built target nrLDPC_decoder_kernels_CL
Scanning dependencies of target ldpc_cl
Building C object CMakeFiles/ldpc_cl.dir/usr/local/oai/oai-develop/openairinterface5g/openair1/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_CL.c.o
In file included from /usr/include/CL/cl.h:32,
from /usr/include/CL/opencl.h:38,
from /usr/local/oai/oai-develop/openairinterface5g/openair1/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_CL.c:49:
/usr/include/CL/cl_version.h:34:9: note: #pragma message: cl_version.h: CL_TARGET_OPENCL_VERSION is not defined. Defaulting to 220 (OpenCL 2.2)
#pragma message("cl_version.h: CL_TARGET_OPENCL_VERSION is not defined. Defaulting to 220 (OpenCL 2.2)")
^~~~~~~
Building C object CMakeFiles/ldpc_cl.dir/usr/local/oai/oai-develop/openairinterface5g/openair1/PHY/CODING/nrLDPC_encoder/ldpc_encoder_optim8segmulti.c.o
Linking C shared module libldpc_cl.so
Built target ldpc_cl
```
At runtime, to successfully use hardware acceleration via OpenCL, you need to install vendor specific packages which deliver the required drivers and tools to make use of their GPU (Nvidia, Intel...) , fpga (Xilinx, Intel) or CPU (Intel, AMD, ARM...) through OpenCL.
`./nr-softmodem -O libconfig:gnb.band78.sa.fr1.106PRB.usrpb210.conf:dbgl5 --rfsim --rfsimulator.serveraddr server --sa --log_config.gtpu_log_level info --loader.ldpc.shlibversion _cl` `./nr-softmodem -O libconfig:gnb.band78.sa.fr1.106PRB.usrpb210.conf:dbgl5 --rfsim --rfsimulator.serveraddr server --sa --log_config.gtpu_log_level info --loader.ldpc.shlibversion _cl`
``` [LOADER] library libldpc_cl.so successfully loaded ```
------------------------------------------------ ------------------------------------------------
[LOADER] library libldpc_cl.so successfully loaded
[HW] Platform 0, OpenCL profile FULL_PROFILE [HW] Platform 0, OpenCL profile FULL_PROFILE
[HW] Platform 0, OpenCL version OpenCL 2.1 LINUX [HW] Platform 0, OpenCL version OpenCL 2.1 LINUX
[HW] Device 0 is available [HW] Device 0 is available
...@@ -62,8 +83,9 @@ loading `libldpc_cl.so` instead of `libldpc.so`: ...@@ -62,8 +83,9 @@ loading `libldpc_cl.so` instead of `libldpc.so`:
`./nr-uesoftmodem -r 106 --numerology 1 --band 78 -C 3619200000 --rfsim --sa -O libconfig:/usr/local/oai/conf/nrue_sim.conf:dbgl5 --nokrnmod --loader.ldpc.shlibversion _cl --log_config.hw_log_level info` `./nr-uesoftmodem -r 106 --numerology 1 --band 78 -C 3619200000 --rfsim --sa -O libconfig:/usr/local/oai/conf/nrue_sim.conf:dbgl5 --nokrnmod --loader.ldpc.shlibversion _cl --log_config.hw_log_level info`
```[CONFIG] shlibversion set to _cl from command line ```
............................................................ ............................................................
[CONFIG] shlibversion set to _cl from command line
[CONFIG] loader.ldpc 1 options set from command line [CONFIG] loader.ldpc 1 options set from command line
[LOADER] library libldpc_cl.so successfully loaded [LOADER] library libldpc_cl.so successfully loaded
[HW] Platform 0, OpenCL profile FULL_PROFILE [HW] Platform 0, OpenCL profile FULL_PROFILE
...@@ -94,7 +116,6 @@ loading `libldpc_cl.so` instead of `libldpc.so`: ...@@ -94,7 +116,6 @@ loading `libldpc_cl.so` instead of `libldpc.so`:
[HW] Device 0, max Work Items size for dimension: 1 512 [HW] Device 0, max Work Items size for dimension: 1 512
[HW] Device 0, max Work Items size for dimension: 2 512 [HW] Device 0, max Work Items size for dimension: 2 512
------------------------------------------------------------ ------------------------------------------------------------
​```
``` ```
A mechanism to select ldpc implementation is also available in the `ldpctest` phy simulator via the `-v`option, which can be used to specify the version of the ldpc shared library to be used. A mechanism to select ldpc implementation is also available in the `ldpctest` phy simulator via the `-v`option, which can be used to specify the version of the ldpc shared library to be used.
...@@ -103,7 +124,9 @@ A mechanism to select ldpc implementation is also available in the `ldpctest` ph ...@@ -103,7 +124,9 @@ A mechanism to select ldpc implementation is also available in the `ldpctest` ph
Loading libldpc_cuda.so, the cuda implementation of the ldpc decoder: Loading libldpc_cuda.so, the cuda implementation of the ldpc decoder:
```$ ./ldpctest -v _cuda ```
$ ./ldpctest -v _cuda
ldpctest -v _cuda
Initializing random number generator, seed 0 Initializing random number generator, seed 0
block length 8448: block length 8448:
n_trials 1: n_trials 1:
...@@ -119,18 +142,16 @@ log init done ...@@ -119,18 +142,16 @@ log init done
[CONFIG] loader.ldpc: 1/2 parameters successfully set [CONFIG] loader.ldpc: 1/2 parameters successfully set
[LOADER] library libldpc_cuda.so successfully loaded [LOADER] library libldpc_cuda.so successfully loaded
................................... ...................................
​```
``` ```
Loading libldpc_cl.so, the opencl implementation of the ldpc decoder: Loading libldpc_cl.so, the opencl implementation of the ldpc decoder:
`make ldpc_cl` `make ldpc_cl`
`cp ../../../openair1/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_kernels_CL.cl`
`./ldpctest -v _cl` ```
$ ./ldpctest -v _cl
```$ ./ldpctest -v _cl
Initializing random number generator, seed 0 Initializing random number generator, seed 0
block length 8448: block length 8448:
n_trials 1: n_trials 1:
...@@ -164,7 +185,6 @@ log init done ...@@ -164,7 +185,6 @@ log init done
[HW] Device 0, max Work Items size for dimension: 1 512 [HW] Device 0, max Work Items size for dimension: 1 512
[HW] Device 0, max Work Items size for dimension: 2 512 [HW] Device 0, max Work Items size for dimension: 2 512
................................ ................................
​```
``` ```
...@@ -174,6 +194,6 @@ Libraries implementing the LDPC algorithms must be named `libldpc<_version>.so`, ...@@ -174,6 +194,6 @@ Libraries implementing the LDPC algorithms must be named `libldpc<_version>.so`,
`libldpc_cuda.so`has been tested with the `ldpctest` executable, usage from the softmodem's has to be tested. `libldpc_cuda.so`has been tested with the `ldpctest` executable, usage from the softmodem's has to be tested.
`libldpc_cl`is under development. `libldpc_cl.so`is under development.
[oai Wikis home](https://gitlab.eurecom.fr/oai/openairinterface5g/wikis/home) [oai Wikis home](https://gitlab.eurecom.fr/oai/openairinterface5g/wikis/home)
openair1/PHY/CODING/TESTBENCH/ldpctest.c
View file @ ec286386
...@@ -514,7 +514,7 @@ int main(int argc, char *argv[]) ...@@ -514,7 +514,7 @@ int main(int argc, char *argv[])
unsigned char qbits=8; unsigned char qbits=8;
unsigned int decoded_errors[10000]; // initiate the size of matrix equivalent to size of SNR unsigned int decoded_errors[10000]; // initiate the size of matrix equivalent to size of SNR
int c,i=0, i1 = 0; int c,i=0, i1 = 0;
int loglvl=OAILOG_WARNING;
int n_trials = 1; int n_trials = 1;
double SNR_step = 0.1; double SNR_step = 0.1;
...@@ -525,8 +525,11 @@ int main(int argc, char *argv[]) ...@@ -525,8 +525,11 @@ int main(int argc, char *argv[])
n_iter_stats_t dec_iter[3]; n_iter_stats_t dec_iter[3];
short BG=0,Zc,Kb=0; short BG=0,Zc,Kb=0;
if ( load_configmodule(argc,argv,CONFIG_ENABLECMDLINEONLY) == 0) {
while ((c = getopt (argc, argv, "q:r:s:S:l:G:n:d:i:t:u:hv:")) != -1) exit_fun(" Error, configuration module init failed\n");
} // must be done before specific options parsing to prevent errasing them
while ((c = getopt (argc, argv, "q:r:s:S:l:L:G:n:d:i:t:u:hv:")) != -1)
switch (c) switch (c)
{ {
case 'q': case 'q':
...@@ -544,6 +547,10 @@ int main(int argc, char *argv[]) ...@@ -544,6 +547,10 @@ int main(int argc, char *argv[])
case 'l': case 'l':
block_length = atoi(optarg); block_length = atoi(optarg);
break; break;
case 'L':
loglvl = atoi(optarg);
break;
case 'G': case 'G':
ldpc_version="_cuda"; ldpc_version="_cuda";
...@@ -581,10 +588,11 @@ int main(int argc, char *argv[]) ...@@ -581,10 +588,11 @@ int main(int argc, char *argv[])
printf("BG1 (blocklength > 3840): 1/3, 2/3, 22/25 (8/9) \n"); printf("BG1 (blocklength > 3840): 1/3, 2/3, 22/25 (8/9) \n");
printf("BG2 (blocklength <= 3840): 1/5, 1/3, 2/3 \n\n"); printf("BG2 (blocklength <= 3840): 1/5, 1/3, 2/3 \n\n");
printf("-h This message\n"); printf("-h This message\n");
printf("-L <log level, 0(errors), 1(warning), 2(info) 3(debug) 4 (trace)>\n");
printf("-q Quantization bits, Default: 8\n"); printf("-q Quantization bits, Default: 8\n");
printf("-r Nominator rate, (1, 2, 22), Default: 1\n"); printf("-r Nominator rate, (1, 2, 22), Default: 1\n");
printf("-d Denominator rate, (3, 5, 25), Default: 1\n"); printf("-d Denominator rate, (3, 5, 25), Default: 1\n");
printf("-l Block length (l > 3840 -> BG1, rest BG2 ), Default: 8448\n"); printf("-l Block length (l > 3840 -> BG1, rest BG2 ), Default: 8448\n");
printf("-G give 1 to run cuda for LDPC, Default: 0\n"); printf("-G give 1 to run cuda for LDPC, Default: 0\n");
printf("-n Number of simulation trials, Default: 1\n"); printf("-n Number of simulation trials, Default: 1\n");
//printf("-M MCS2 for TB 2\n"); //printf("-M MCS2 for TB 2\n");
...@@ -603,6 +611,8 @@ int main(int argc, char *argv[]) ...@@ -603,6 +611,8 @@ int main(int argc, char *argv[])
printf("n_trials %d: \n", n_trials); printf("n_trials %d: \n", n_trials);
printf("SNR0 %f: \n", SNR0); printf("SNR0 %f: \n", SNR0);
logInit();
set_glog(loglvl);
if (ldpc_version != NULL) if (ldpc_version != NULL)
load_nrLDPClib(ldpc_version); load_nrLDPClib(ldpc_version);
......
openair1/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_CL.c
View file @ ec286386
...@@ -25,32 +25,37 @@ ...@@ -25,32 +25,37 @@
* \version 1.0 * \version 1.0
* \note initial implem - translation of cuda version * \note initial implem - translation of cuda version
*/ */
#define MAX_ITERATION 2
#define MC 1
#define MAX_OCLDEV 10
#define MAX_OCLRUNTIME 5
typedef struct{
char x;
char y;
short value;
} h_element;
#ifdef NRLDPC_KERNEL_SOURCE
#include "nrLDPC_decoder_kernels_CL.c"
#else
/* uses HW component id for log messages ( --log_config.hw_log_level <warning| info|debug|trace>) */ /* uses HW component id for log messages ( --log_config.hw_log_level <warning| info|debug|trace>) */
#include <stdio.h> #include <stdio.h>
#include <unistd.h> #include <unistd.h>
#include <cuda_runtime.h> #include <sys/stat.h>
#include <CL/opencl.h> #include <CL/opencl.h>
#include "PHY/CODING/nrLDPC_decoder/nrLDPC_types.h" #include "PHY/CODING/nrLDPC_decoder/nrLDPC_types.h"
#include "PHY/CODING/nrLDPC_decoder/nrLDPCdecoder_defs.h" #include "PHY/CODING/nrLDPC_decoder/nrLDPCdecoder_defs.h"
#include "assertions.h" #include "assertions.h"
#include "common/utils/LOG/log.h" #include "common/utils/LOG/log.h"
#define MAX_ITERATION 2
#define MC 1
#define MAX_OCLDEV 10
#define MAX_OCLRUNTIME 5
#define CLSETKERNELARG(A,B,C,D) \ #define CLSETKERNELARG(A,B,C,D) \
rt=clSetKernelArg(A,B,C,D) ;\ rt=clSetKernelArg(A,B,C,D) ;\
AssertFatal(rt == CL_SUCCESS, "Error %d setting kernel argument index %d\n" , (int)rt, B); AssertFatal(rt == CL_SUCCESS, "Error %d setting kernel argument index %d\n" , (int)rt, B);
typedef struct{
char x;
char y;
short value;
} h_element;
#include "../nrLDPC_decoder_LYC/bgs/BG1_compact_in_C.h" #include "../nrLDPC_decoder_LYC/bgs/BG1_compact_in_C.h"
typedef struct{ typedef struct{
...@@ -213,20 +218,31 @@ void get_CompilErr(cl_program program, int pltf) { ...@@ -213,20 +218,31 @@ void get_CompilErr(cl_program program, int pltf) {
} }
size_t load_source(char **source_str) { size_t load_source(char **source_str, char *filename) {
int MAX_SOURCE_SIZE=(500*132);
FILE *fp; FILE *fp;
struct stat st ;
size_t source_size; size_t source_size;
char *src= NULL;
fp = fopen("nrLDPC_decoder_kernels_CL.cl", "r");
AssertFatal(fp,"failed to open cl source: %s\n",strerror(errno));
*source_str = (char*)malloc(MAX_SOURCE_SIZE); if (filename == NULL) {
source_size = fread( *source_str, 1, MAX_SOURCE_SIZE, fp); src = "nrLDPC_decoder_kernels_CL.clc";
fclose( fp ); } else {
return source_size; src = filename;
}
fp = fopen(src, "r");
AssertFatal(fp,"failed to open cl source %s: %s\n",src,strerror(errno));
fstat(fileno(fp), &st);
source_size = st.st_size;
*source_str = (char*)malloc(source_size);
source_size = fread( *source_str, 1, source_size, fp);
fclose( fp );
LOG_I(HW,"Loaded kernel sources from %s %u bytes\n", (filename==NULL)?"embedded cl code":src,(unsigned int)source_size );
return source_size;
} }
/* from here: entry points in decoder shared lib */ /* from here: entry points in decoder shared lib */
int ldpc_autoinit(void) { // called by the library loader int ldpc_autoinit(void) { // called by the library loader
cl_platform_id platforms[10]; cl_platform_id platforms[10];
...@@ -284,11 +300,11 @@ int ldpc_autoinit(void) { // called by the library loader ...@@ -284,11 +300,11 @@ int ldpc_autoinit(void) { // called by the library loader
ocl.runtime[i].dev_tmp = clCreateBuffer(ocl.runtime[i].context, CL_MEM_READ_ONLY|CL_MEM_HOST_WRITE_ONLY, 68*384, NULL, (cl_int *)&rt); ocl.runtime[i].dev_tmp = clCreateBuffer(ocl.runtime[i].context, CL_MEM_READ_ONLY|CL_MEM_HOST_WRITE_ONLY, 68*384, NULL, (cl_int *)&rt);
AssertFatal(rt == CL_SUCCESS, "Error %d creating buffer dev_tmp for platform %i \n" , (int)rt, i); AssertFatal(rt == CL_SUCCESS, "Error %d creating buffer dev_tmp for platform %i \n" , (int)rt, i);
char *source_str; char *source_str;
size_t source_size=load_source(&source_str); size_t source_size=load_source(&source_str,"nrLDPC_decoder_kernels_CL.clc");
cl_program program = clCreateProgramWithSource(ocl.runtime[i].context, 1, cl_program program = clCreateProgramWithSource(ocl.runtime[i].context, 1,
(const char **)&source_str, (const size_t *)&source_size, (cl_int *)&rt); (const char **)&source_str, (const size_t *)&source_size, (cl_int *)&rt);
AssertFatal(rt == CL_SUCCESS, "Error %d creating program for platform %i \n" , (int)rt, i); AssertFatal(rt == CL_SUCCESS, "Error %d creating program for platform %i \n" , (int)rt, i);
rt = clBuildProgram(program, ocl.runtime[i].num_devices,ocl.runtime[i].devices, NULL, NULL, NULL); rt = clBuildProgram(program, ocl.runtime[i].num_devices,ocl.runtime[i].devices, NULL /* compile options */, NULL, NULL);
if (rt == CL_BUILD_PROGRAM_FAILURE) { if (rt == CL_BUILD_PROGRAM_FAILURE) {
get_CompilErr(program,i); get_CompilErr(program,i);
} }
...@@ -434,3 +450,4 @@ int32_t nrLDPC_decod(t_nrLDPC_dec_params* p_decParams, int8_t* p_llr, int8_t* p_ ...@@ -434,3 +450,4 @@ int32_t nrLDPC_decod(t_nrLDPC_dec_params* p_decParams, int8_t* p_llr, int8_t* p_
return MAX_ITERATION; return MAX_ITERATION;
} }
#endif //NRLDPC_KERNEL_SOURCE
openair1/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_kernels_CL.cl → openair1/PHY/CODING/nrLDPC_decoder/nrLDPC_decoder_kernels_CL.c
View file @ ec286386
...@@ -29,16 +29,7 @@ ...@@ -29,16 +29,7 @@
* \note initial implem - translation of cuda version * \note initial implem - translation of cuda version
* \warning * \warning
*/ */
#define define MAX_ITERATION 2
#define MC 1
#define INT32_MAX 2147483647
typedef struct{
char x;
char y;
short value;
} h_element;
//__global char dev_dt [46*68*384]; //__global char dev_dt [46*68*384];
//__local char *dev_t; //__local char *dev_t;
...@@ -46,7 +37,7 @@ typedef struct{ ...@@ -46,7 +37,7 @@ typedef struct{
//__global unsigned char dev_tmp[68*384]; //__global unsigned char dev_tmp[68*384];
#define INT32_MAX 2147483647
//__constant h_element dev_h_compact1[46*19] = {}; // used in kernel 1 //__constant h_element dev_h_compact1[46*19] = {}; // used in kernel 1
//__constant h_element dev_h_compact2[68*30] = {}; // used in kernel 2 //__constant h_element dev_h_compact2[68*30] = {}; // used in kernel 2
...@@ -299,4 +290,3 @@ __kernel void pack_decoded_bit(__global unsigned char * dev_llr, __global unsign ...@@ -299,4 +290,3 @@ __kernel void pack_decoded_bit(__global unsigned char * dev_llr, __global unsign
} }
} }
} }
openair1/PHY/CODING/nrLDPC_load.c
View file @ ec286386
...@@ -46,14 +46,14 @@ static loader_shlibfunc_t shlib_fdesc[3]; ...@@ -46,14 +46,14 @@ static loader_shlibfunc_t shlib_fdesc[3];
/* arguments used when called from phy simulators exec's which do not use the config module */ /* arguments used when called from phy simulators exec's which do not use the config module */
/* arg is used to initialize the config module so that the loader works as expected */ /* arg is used to initialize the config module so that the loader works as expected */
char *arg[64]={"ldpctest","-O","cmdlineonly::dbgl0",NULL,NULL}; char *arg[64]={"ldpctest",NULL};
int load_nrLDPClib(char *version) { int load_nrLDPClib(char *version) {
char *ptr = (char*)config_get_if(); char *ptr = (char*)config_get_if();
char libname[64]="ldpc"; char libname[64]="ldpc";
if ( ptr==NULL ) {// phy simulators, config module possibly not loaded if ( ptr==NULL ) {// phy simulators, config module possibly not loaded
load_configmodule(0,(char **)NULL,CONFIG_ENABLECMDLINEONLY) ; load_configmodule(1,arg,CONFIG_ENABLECMDLINEONLY) ;
logInit(); logInit();
} }
shlib_fdesc[0].fname = "nrLDPC_decod"; shlib_fdesc[0].fname = "nrLDPC_decod";
......
openair1/PHY/INIT/nr_parms.c
View file @ ec286386
...@@ -250,11 +250,9 @@ int nr_init_frame_parms(nfapi_nr_config_request_scf_t* cfg, ...@@ -250,11 +250,9 @@ int nr_init_frame_parms(nfapi_nr_config_request_scf_t* cfg,
int Ncp = NFAPI_CP_NORMAL; int Ncp = NFAPI_CP_NORMAL;
int mu = cfg->ssb_config.scs_common.value; int mu = cfg->ssb_config.scs_common.value;
#if DISABLE_LOG_X
printf("Initializing frame parms for mu %d, N_RB %d, Ncp %d\n",mu, fp->N_RB_DL, Ncp);
#else
LOG_I(PHY,"Initializing frame parms for mu %d, N_RB %d, Ncp %d\n",mu, fp->N_RB_DL, Ncp); LOG_I(PHY,"Initializing frame parms for mu %d, N_RB %d, Ncp %d\n",mu, fp->N_RB_DL, Ncp);
#endif
if (Ncp == NFAPI_CP_EXTENDED) if (Ncp == NFAPI_CP_EXTENDED)
AssertFatal(mu == NR_MU_2,"Invalid cyclic prefix %d for numerology index %d\n", Ncp, mu); AssertFatal(mu == NR_MU_2,"Invalid cyclic prefix %d for numerology index %d\n", Ncp, mu);
...@@ -343,11 +341,8 @@ int nr_init_frame_parms_ue(NR_DL_FRAME_PARMS *fp, ...@@ -343,11 +341,8 @@ int nr_init_frame_parms_ue(NR_DL_FRAME_PARMS *fp,
AssertFatal(fp->ul_CarrierFreq == (fp->dl_CarrierFreq + uplink_frequency_offset), "Disagreement in uplink frequency for band %d: ul_CarrierFreq = %lu Hz vs expected %lu Hz\n", fp->nr_band, fp->ul_CarrierFreq, fp->dl_CarrierFreq + uplink_frequency_offset); AssertFatal(fp->ul_CarrierFreq == (fp->dl_CarrierFreq + uplink_frequency_offset), "Disagreement in uplink frequency for band %d: ul_CarrierFreq = %lu Hz vs expected %lu Hz\n", fp->nr_band, fp->ul_CarrierFreq, fp->dl_CarrierFreq + uplink_frequency_offset);
#if DISABLE_LOG_X
printf("Initializing UE frame parms for mu %d, N_RB %d, Ncp %d\n",fp->numerology_index, fp->N_RB_DL, Ncp);
#else
LOG_I(PHY,"Initializing frame parms for mu %d, N_RB %d, Ncp %d\n",fp->numerology_index, fp->N_RB_DL, Ncp); LOG_I(PHY,"Initializing frame parms for mu %d, N_RB %d, Ncp %d\n",fp->numerology_index, fp->N_RB_DL, Ncp);
#endif
if (Ncp == NFAPI_CP_EXTENDED) if (Ncp == NFAPI_CP_EXTENDED)
AssertFatal(fp->numerology_index == NR_MU_2,"Invalid cyclic prefix %d for numerology index %d\n", Ncp, fp->numerology_index); AssertFatal(fp->numerology_index == NR_MU_2,"Invalid cyclic prefix %d for numerology index %d\n", Ncp, fp->numerology_index);
......
openair1/PHY/LTE_UE_TRANSPORT/pcfich_ue.c
View file @ ec286386
...@@ -46,10 +46,9 @@ void pcfich_unscrambling(LTE_DL_FRAME_PARMS *frame_parms, ...@@ -46,10 +46,9 @@ void pcfich_unscrambling(LTE_DL_FRAME_PARMS *frame_parms,
uint32_t i; uint32_t i;
uint8_t reset; uint8_t reset;
uint32_t x1, x2, s=0; uint32_t x1=0, x2, s=0; // x1 is set in lte_gold_generic
reset = 1; reset = 1;
// x1 is set in lte_gold_generic
x2 = ((((2*frame_parms->Nid_cell)+1)*(1+subframe))<<9) + frame_parms->Nid_cell; //this is c_init in 36.211 Sec 6.7.1 x2 = ((((2*frame_parms->Nid_cell)+1)*(1+subframe))<<9) + frame_parms->Nid_cell; //this is c_init in 36.211 Sec 6.7.1
for (i=0; i<32; i++) { for (i=0; i<32; i++) {
......
openair1/PHY/NR_UE_TRANSPORT/nr_dlsch_decoding.c
View file @ ec286386
...@@ -642,7 +642,6 @@ uint32_t nr_dlsch_decoding(PHY_VARS_NR_UE *phy_vars_ue, ...@@ -642,7 +642,6 @@ uint32_t nr_dlsch_decoding(PHY_VARS_NR_UE *phy_vars_ue,
if (LOG_DEBUGFLAG(DEBUG_DLSCH_DECOD)) if (LOG_DEBUGFLAG(DEBUG_DLSCH_DECOD))
LOG_I(PHY,"Segmentation: C %d, K %d\n",harq_process->C,harq_process->K); LOG_I(PHY,"Segmentation: C %d, K %d\n",harq_process->C,harq_process->K);
opp_enabled=1;
Kr = harq_process->K; // [hna] overwrites this line "Kr = p_decParams->Z*kb" Kr = harq_process->K; // [hna] overwrites this line "Kr = p_decParams->Z*kb"
Kr_bytes = Kr>>3; Kr_bytes = Kr>>3;
offset = 0; offset = 0;
......
openair1/PHY/NR_UE_TRANSPORT/nr_dlsch_demodulation.c
View file @ ec286386
...@@ -379,13 +379,10 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue, ...@@ -379,13 +379,10 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue,
len = (pilots==1)? ((config_type==NFAPI_NR_DMRS_TYPE1)?nb_rb*(12-6*dlsch0_harq->n_dmrs_cdm_groups): nb_rb*(12-4*dlsch0_harq->n_dmrs_cdm_groups)):(nb_rb*12); len = (pilots==1)? ((config_type==NFAPI_NR_DMRS_TYPE1)?nb_rb*(12-6*dlsch0_harq->n_dmrs_cdm_groups): nb_rb*(12-4*dlsch0_harq->n_dmrs_cdm_groups)):(nb_rb*12);
stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]); stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
#if DISABLE_LOG_X if (cpumeas(CPUMEAS_GETSTATE))
printf("[AbsSFN %u.%d] Slot%d Symbol %d type %d: Pilot/Data extraction %5.2f \n", LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d type %d: Pilot/Data extraction %5.2f \n",
frame,nr_slot_rx,slot,symbol,type,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0)); frame,nr_slot_rx,slot,symbol,type,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#else
LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d type %d: Pilot/Data extraction %5.2f \n",
frame,nr_slot_rx,slot,symbol,type,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#endif
start_meas(&ue->generic_stat_bis[proc->thread_id][slot]); start_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
n_tx = dlsch0_harq->Nl; n_tx = dlsch0_harq->Nl;
...@@ -402,11 +399,9 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue, ...@@ -402,11 +399,9 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue,
nb_rb_pdsch); nb_rb_pdsch);
stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]); stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
#if DISABLE_LOG_X if (cpumeas(CPUMEAS_GETSTATE))
printf("[AbsSFN %u.%d] Slot%d Symbol %d: Channel Scale %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0)); LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d: Channel Scale %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#else
LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d: Channel Scale %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#endif
start_meas(&ue->generic_stat_bis[proc->thread_id][slot]); start_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
if (first_symbol_flag==1) { if (first_symbol_flag==1) {
if (beamforming_mode==0){ if (beamforming_mode==0){
...@@ -463,11 +458,8 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue, ...@@ -463,11 +458,8 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue,
#endif #endif
stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]); stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
#if DISABLE_LOG_X if (cpumeas(CPUMEAS_GETSTATE))
printf("[AbsSFN %u.%d] Slot%d Symbol %d first_symbol_flag %d: Channel Level %5.2f \n",frame,nr_slot_rx,slot,symbol,first_symbol_flag,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0)); LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d first_symbol_flag %d: Channel Level %5.2f \n",frame,nr_slot_rx,slot,symbol,first_symbol_flag,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#else
LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d first_symbol_flag %d: Channel Level %5.2f \n",frame,nr_slot_rx,slot,symbol,first_symbol_flag,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#endif
start_meas(&ue->generic_stat_bis[proc->thread_id][slot]); start_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
// Now channel compensation // Now channel compensation
...@@ -515,11 +507,9 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue, ...@@ -515,11 +507,9 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue,
} }
stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]); stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
#if DISABLE_LOG_X if (cpumeas(CPUMEAS_GETSTATE))
printf("[AbsSFN %u.%d] Slot%d Symbol %d log2_maxh %d channel_level %d: Channel Comp %5.2f \n", frame, nr_slot_rx, slot, symbol, pdsch_vars[gNB_id]->log2_maxh, proc->channel_level, ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0)); LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d log2_maxh %d channel_level %d: Channel Comp %5.2f \n", frame, nr_slot_rx, slot, symbol, pdsch_vars[gNB_id]->log2_maxh, proc->channel_level, ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#else
LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d log2_maxh %d channel_level %d: Channel Comp %5.2f \n", frame, nr_slot_rx, slot, symbol, pdsch_vars[gNB_id]->log2_maxh, proc->channel_level, ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#endif
start_meas(&ue->generic_stat_bis[proc->thread_id][slot]); start_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
if (frame_parms->nb_antennas_rx > 1) { if (frame_parms->nb_antennas_rx > 1) {
...@@ -574,11 +564,9 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue, ...@@ -574,11 +564,9 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue,
} }
stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]); stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
#if DISABLE_LOG_X if (cpumeas(CPUMEAS_GETSTATE))
printf("[AbsSFN %u.%d] Slot%d Symbol %d: Channel Combine %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0)); LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d: Channel Combine %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#else
LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d: Channel Combine %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#endif
start_meas(&ue->generic_stat_bis[proc->thread_id][slot]); start_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
/* Store the valid DL RE's */ /* Store the valid DL RE's */
pdsch_vars[gNB_id]->dl_valid_re[symbol-1] = len; pdsch_vars[gNB_id]->dl_valid_re[symbol-1] = len;
...@@ -657,11 +645,8 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue, ...@@ -657,11 +645,8 @@ int nr_rx_pdsch(PHY_VARS_NR_UE *ue,
} }
stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]); stop_meas(&ue->generic_stat_bis[proc->thread_id][slot]);
#if DISABLE_LOG_X if (cpumeas(CPUMEAS_GETSTATE))
printf("[AbsSFN %u.%d] Slot%d Symbol %d: LLR Computation %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0)); LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d: LLR Computation %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#else
LOG_D(PHY, "[AbsSFN %u.%d] Slot%d Symbol %d: LLR Computation %5.2f \n",frame,nr_slot_rx,slot,symbol,ue->generic_stat_bis[proc->thread_id][slot].p_time/(cpuf*1000.0));
#endif
// Please keep it: useful for debugging // Please keep it: useful for debugging
#ifdef DEBUG_PDSCH_RX #ifdef DEBUG_PDSCH_RX
......
openair1/PHY/NR_UE_TRANSPORT/nr_initial_sync.c
View file @ ec286386
...@@ -353,21 +353,15 @@ int nr_initial_sync(UE_nr_rxtx_proc_t *proc, ...@@ -353,21 +353,15 @@ int nr_initial_sync(UE_nr_rxtx_proc_t *proc,
if( (abs(ue->common_vars.freq_offset) > 150) && (ret == 0) ) if( (abs(ue->common_vars.freq_offset) > 150) && (ret == 0) )
{ {
ret=-1; ret=-1;
#if DISABLE_LOG_X
printf("Ignore MIB with high freq offset [%d Hz] estimation \n",ue->common_vars.freq_offset);
#else
LOG_E(HW, "Ignore MIB with high freq offset [%d Hz] estimation \n",ue->common_vars.freq_offset); LOG_E(HW, "Ignore MIB with high freq offset [%d Hz] estimation \n",ue->common_vars.freq_offset);
#endif
}*/ }*/
if (ret==0) { // PBCH found so indicate sync to higher layers and configure frame parameters if (ret==0) { // PBCH found so indicate sync to higher layers and configure frame parameters
//#ifdef DEBUG_INITIAL_SYNCH //#ifdef DEBUG_INITIAL_SYNCH
#if DISABLE_LOG_X
printf("[UE%d] In synch, rx_offset %d samples\n",ue->Mod_id, ue->rx_offset);
#else
LOG_I(PHY, "[UE%d] In synch, rx_offset %d samples\n",ue->Mod_id, ue->rx_offset); LOG_I(PHY, "[UE%d] In synch, rx_offset %d samples\n",ue->Mod_id, ue->rx_offset);
#endif
//#endif //#endif
if (ue->UE_scan_carrier == 0) { if (ue->UE_scan_carrier == 0) {
......
openair1/PHY/TOOLS/time_meas.c
View file @ ec286386
...@@ -51,24 +51,7 @@ double get_cpu_freq_GHz(void) ...@@ -51,24 +51,7 @@ double get_cpu_freq_GHz(void)
return cpu_freq_GHz; return cpu_freq_GHz;
} }
int cpumeas(int action)
{
switch (action) {
case CPUMEAS_ENABLE:
opp_enabled = 1;
break;
case CPUMEAS_DISABLE:
opp_enabled = 0;
break;
case CPUMEAS_GETSTATE:
default:
break;
}
return opp_enabled;
}
void print_meas_now(time_stats_t *ts, void print_meas_now(time_stats_t *ts,
const char *name, const char *name,
......
openair1/PHY/TOOLS/time_meas.h
View file @ ec286386
...@@ -115,7 +115,24 @@ static inline uint32_t rdtsc_oai(void) { ...@@ -115,7 +115,24 @@ static inline uint32_t rdtsc_oai(void) {
#define CPUMEAS_DISABLE 0 #define CPUMEAS_DISABLE 0
#define CPUMEAS_ENABLE 1 #define CPUMEAS_ENABLE 1
#define CPUMEAS_GETSTATE 2 #define CPUMEAS_GETSTATE 2
int cpumeas(int action); static inline int cpumeas(int action) {
switch (action) {
case CPUMEAS_ENABLE:
opp_enabled = 1;
break;
case CPUMEAS_DISABLE:
opp_enabled = 0;
break;
case CPUMEAS_GETSTATE:
default:
break;
}
return opp_enabled;
}
static inline void start_meas(time_stats_t *ts) { static inline void start_meas(time_stats_t *ts) {
if (opp_enabled) { if (opp_enabled) {
if (ts->meas_flag==0) { if (ts->meas_flag==0) {
......
openair1/SCHED_NR_UE/phy_procedures_nr_ue.c
View file @ ec286386
...@@ -855,8 +855,8 @@ int nr_ue_pdsch_procedures(PHY_VARS_NR_UE *ue, UE_nr_rxtx_proc_t *proc, int gNB_ ...@@ -855,8 +855,8 @@ int nr_ue_pdsch_procedures(PHY_VARS_NR_UE *ue, UE_nr_rxtx_proc_t *proc, int gNB_
} else AssertFatal(1==0,"Not RA_PDSCH, SI_PDSCH or PDSCH\n"); } else AssertFatal(1==0,"Not RA_PDSCH, SI_PDSCH or PDSCH\n");
stop_meas(&ue->dlsch_llr_stats_parallelization[proc->thread_id][slot]); stop_meas(&ue->dlsch_llr_stats_parallelization[proc->thread_id][slot]);
LOG_D(PHY, "[AbsSFN %d.%d] LLR Computation Symbol %d %5.2f \n",frame_rx,nr_slot_rx,m,ue->dlsch_llr_stats_parallelization[proc->thread_id][slot].p_time/(cpuf*1000.0)); if (cpumeas(CPUMEAS_GETSTATE))
LOG_D(PHY, "[AbsSFN %d.%d] LLR Computation Symbol %d %5.2f \n",frame_rx,nr_slot_rx,m,ue->dlsch_llr_stats_parallelization[proc->thread_id][slot].p_time/(cpuf*1000.0));
if(first_symbol_flag) { if(first_symbol_flag) {
proc->first_symbol_available = 1; proc->first_symbol_available = 1;
} }
...@@ -1031,14 +1031,13 @@ bool nr_ue_dlsch_procedures(PHY_VARS_NR_UE *ue, ...@@ -1031,14 +1031,13 @@ bool nr_ue_dlsch_procedures(PHY_VARS_NR_UE *ue,
LOG_D(PHY, "In %s DL PDU length in bits: %d, in bytes: %d \n", __FUNCTION__, dlsch0->harq_processes[harq_pid]->TBS, dlsch0->harq_processes[harq_pid]->TBS / 8); LOG_D(PHY, "In %s DL PDU length in bits: %d, in bytes: %d \n", __FUNCTION__, dlsch0->harq_processes[harq_pid]->TBS, dlsch0->harq_processes[harq_pid]->TBS / 8);
stop_meas(&ue->dlsch_decoding_stats[proc->thread_id]);
if (cpumeas(CPUMEAS_GETSTATE)) {
LOG_D(PHY, " --> Unscrambling for CW0 %5.3f\n",
stop_meas(&ue->dlsch_decoding_stats[proc->thread_id]); (ue->dlsch_unscrambling_stats.p_time)/(cpuf*1000.0));
LOG_D(PHY, " --> Unscrambling for CW0 %5.3f\n", LOG_D(PHY, "AbsSubframe %d.%d --> LDPC Decoding for CW0 %5.3f\n",
(ue->dlsch_unscrambling_stats.p_time)/(cpuf*1000.0)); frame_rx%1024, nr_slot_rx,(ue->dlsch_decoding_stats[proc->thread_id].p_time)/(cpuf*1000.0));
LOG_D(PHY, "AbsSubframe %d.%d --> LDPC Decoding for CW0 %5.3f\n", }
frame_rx%1024, nr_slot_rx,(ue->dlsch_decoding_stats[proc->thread_id].p_time)/(cpuf*1000.0));
if(is_cw1_active) { if(is_cw1_active) {
// start ldpc decode for CW 1 // start ldpc decode for CW 1
...@@ -1090,21 +1089,15 @@ bool nr_ue_dlsch_procedures(PHY_VARS_NR_UE *ue, ...@@ -1090,21 +1089,15 @@ bool nr_ue_dlsch_procedures(PHY_VARS_NR_UE *ue,
LOG_T(PHY,"CWW sequential dlsch decoding, ret1 = %d\n", ret1); LOG_T(PHY,"CWW sequential dlsch decoding, ret1 = %d\n", ret1);
} }
stop_meas(&ue->dlsch_decoding_stats[proc->thread_id]); stop_meas(&ue->dlsch_decoding_stats[proc->thread_id]);
if (cpumeas(CPUMEAS_GETSTATE)) {
LOG_D(PHY, " --> Unscrambling for CW1 %5.3f\n", LOG_D(PHY, " --> Unscrambling for CW1 %5.3f\n",
(ue->dlsch_unscrambling_stats.p_time)/(cpuf*1000.0)); (ue->dlsch_unscrambling_stats.p_time)/(cpuf*1000.0));
LOG_D(PHY, "AbsSubframe %d.%d --> ldpc Decoding for CW1 %5.3f\n", LOG_D(PHY, "AbsSubframe %d.%d --> ldpc Decoding for CW1 %5.3f\n",
frame_rx%1024, nr_slot_rx,(ue->dlsch_decoding_stats[proc->thread_id].p_time)/(cpuf*1000.0)); frame_rx%1024, nr_slot_rx,(ue->dlsch_decoding_stats[proc->thread_id].p_time)/(cpuf*1000.0));
LOG_D(PHY,"AbsSubframe %d.%d --> ldpc Decoding for CW1 %5.3f\n", }
frame_rx%1024, nr_slot_rx,(ue->dlsch_decoding_stats[proc->thread_id].p_time)/(cpuf*1000.0)); LOG_D(PHY, "harq_pid: %d, TBS expected dlsch1: %d \n", harq_pid, dlsch1->harq_processes[harq_pid]->TBS);
LOG_D(PHY, "harq_pid: %d, TBS expected dlsch1: %d \n", harq_pid, dlsch1->harq_processes[harq_pid]->TBS);
} }
LOG_D(PHY," ------ end ldpc decoder for AbsSubframe %d.%d ------ decoded in %d \n", frame_rx, nr_slot_rx, ret);
// send to mac // send to mac
if (ue->if_inst && ue->if_inst->dl_indication) { if (ue->if_inst && ue->if_inst->dl_indication) {
ue->if_inst->dl_indication(&dl_indication, ul_time_alignment); ue->if_inst->dl_indication(&dl_indication, ul_time_alignment);
...@@ -1405,8 +1398,8 @@ void *UE_thread_slot1_dl_processing(void *arg) { ...@@ -1405,8 +1398,8 @@ void *UE_thread_slot1_dl_processing(void *arg) {
stop_meas(&ue->ue_front_end_per_slot_stat[proc->thread_id][1]); stop_meas(&ue->ue_front_end_per_slot_stat[proc->thread_id][1]);
LOG_D(PHY, "[AbsSFN %d.%d] Slot1: FFT + Channel Estimate + Pdsch Proc Slot0 %5.2f \n",frame_rx,nr_slot_rx,ue->ue_front_end_per_slot_stat[proc->thread_id][1].p_time/(cpuf*1000.0)); if (cpumeas(CPUMEAS_GETSTATE))
LOG_D(PHY, "[AbsSFN %d.%d] Slot1: FFT + Channel Estimate + Pdsch Proc Slot0 %5.2f \n",frame_rx,nr_slot_rx,ue->ue_front_end_per_slot_stat[proc->thread_id][1].p_time/(cpuf*1000.0));
//wait until pdcch is decoded //wait until pdcch is decoded
uint32_t wait = 0; uint32_t wait = 0;
...@@ -1495,7 +1488,8 @@ void *UE_thread_slot1_dl_processing(void *arg) { ...@@ -1495,7 +1488,8 @@ void *UE_thread_slot1_dl_processing(void *arg) {
//printf("Set available LLR slot1 to 1 AbsSubframe %d.%d \n",frame_rx,nr_slot_rx); //printf("Set available LLR slot1 to 1 AbsSubframe %d.%d \n",frame_rx,nr_slot_rx);
stop_meas(&ue->pdsch_procedures_per_slot_stat[proc->thread_id][1]); stop_meas(&ue->pdsch_procedures_per_slot_stat[proc->thread_id][1]);
LOG_D(PHY, "[AbsSFN %d.%d] Slot1: LLR Computation %5.2f \n",frame_rx,nr_slot_rx,ue->pdsch_procedures_per_slot_stat[proc->thread_id][1].p_time/(cpuf*1000.0)); if (cpumeas(CPUMEAS_GETSTATE))
LOG_D(PHY, "[AbsSFN %d.%d] Slot1: LLR Computation %5.2f \n",frame_rx,nr_slot_rx,ue->pdsch_procedures_per_slot_stat[proc->thread_id][1].p_time/(cpuf*1000.0));
if (pthread_mutex_lock(&proc->mutex_slot1_dl_processing) != 0) { if (pthread_mutex_lock(&proc->mutex_slot1_dl_processing) != 0) {
LOG_E( PHY, "[SCHED][UE] error locking mutex for UE RXTX\n" ); LOG_E( PHY, "[SCHED][UE] error locking mutex for UE RXTX\n" );
...@@ -1864,9 +1858,10 @@ int phy_procedures_nrUE_RX(PHY_VARS_NR_UE *ue, ...@@ -1864,9 +1858,10 @@ int phy_procedures_nrUE_RX(PHY_VARS_NR_UE *ue,
dlsch_parallel); dlsch_parallel);
stop_meas(&ue->dlsch_procedures_stat[proc->thread_id]); stop_meas(&ue->dlsch_procedures_stat[proc->thread_id]);
LOG_D(PHY, "[SFN %d] Slot1: Pdsch Proc %5.2f\n",nr_slot_rx,ue->pdsch_procedures_stat[proc->thread_id].p_time/(cpuf*1000.0)); if (cpumeas(CPUMEAS_GETSTATE)) {
LOG_D(PHY, "[SFN %d] Slot0 Slot1: Dlsch Proc %5.2f\n",nr_slot_rx,ue->dlsch_procedures_stat[proc->thread_id].p_time/(cpuf*1000.0)); LOG_D(PHY, "[SFN %d] Slot1: Pdsch Proc %5.2f\n",nr_slot_rx,ue->pdsch_procedures_stat[proc->thread_id].p_time/(cpuf*1000.0));
LOG_D(PHY, "[SFN %d] Slot0 Slot1: Dlsch Proc %5.2f\n",nr_slot_rx,ue->dlsch_procedures_stat[proc->thread_id].p_time/(cpuf*1000.0));
}
// deactivate dlsch once dlsch proc is done // deactivate dlsch once dlsch proc is done
ue->dlsch[proc->thread_id][gNB_id][0]->active = 0; ue->dlsch[proc->thread_id][gNB_id][0]->active = 0;
...@@ -1941,6 +1936,8 @@ if (nr_slot_rx==9) { ...@@ -1941,6 +1936,8 @@ if (nr_slot_rx==9) {
} }
stop_meas(&ue->generic_stat); stop_meas(&ue->generic_stat);
if (cpumeas(CPUMEAS_GETSTATE))
LOG_D(PHY,"after tubo until end of Rx %5.2f \n",ue->generic_stat.p_time/(cpuf*1000.0));
#ifdef EMOS #ifdef EMOS
phy_procedures_emos_UE_RX(ue,slot,gNB_id); phy_procedures_emos_UE_RX(ue,slot,gNB_id);
...@@ -1950,7 +1947,9 @@ phy_procedures_emos_UE_RX(ue,slot,gNB_id); ...@@ -1950,7 +1947,9 @@ phy_procedures_emos_UE_RX(ue,slot,gNB_id);
VCD_SIGNAL_DUMPER_DUMP_FUNCTION_BY_NAME(VCD_SIGNAL_DUMPER_FUNCTIONS_PHY_PROCEDURES_UE_RX, VCD_FUNCTION_OUT); VCD_SIGNAL_DUMPER_DUMP_FUNCTION_BY_NAME(VCD_SIGNAL_DUMPER_FUNCTIONS_PHY_PROCEDURES_UE_RX, VCD_FUNCTION_OUT);
stop_meas(&ue->phy_proc_rx[proc->thread_id]); stop_meas(&ue->phy_proc_rx[proc->thread_id]);
LOG_D(PHY, "------FULL RX PROC [SFN %d]: %5.2f ------\n",nr_slot_rx,ue->phy_proc_rx[proc->thread_id].p_time/(cpuf*1000.0)); if (cpumeas(CPUMEAS_GETSTATE))
LOG_D(PHY, "------FULL RX PROC [SFN %d]: %5.2f ------\n",nr_slot_rx,ue->phy_proc_rx[proc->thread_id].p_time/(cpuf*1000.0));
//#endif //pdsch //#endif //pdsch
......
openair1/SIMULATION/NR_PHY/dlsim.c
View file @ ec286386
...@@ -402,7 +402,7 @@ int main(int argc, char **argv) ...@@ -402,7 +402,7 @@ int main(int argc, char **argv)
NR_UE_MAC_INST_t *UE_mac; NR_UE_MAC_INST_t *UE_mac;
int cyclic_prefix_type = NFAPI_CP_NORMAL; int cyclic_prefix_type = NFAPI_CP_NORMAL;
int run_initial_sync=0; int run_initial_sync=0;
int loglvl=OAILOG_INFO; int loglvl=OAILOG_WARNING;
//float target_error_rate = 0.01; //float target_error_rate = 0.01;
int css_flag=0; int css_flag=0;
...@@ -647,6 +647,7 @@ int main(int argc, char **argv) ...@@ -647,6 +647,7 @@ int main(int argc, char **argv)
printf("%s -h(elp) -p(extended_prefix) -N cell_id -f output_filename -F input_filename -g channel_model -n n_frames -t Delayspread -s snr0 -S snr1 -x transmission_mode -y TXant -z RXant -i Intefrence0 -j Interference1 -A interpolation_file -C(alibration offset dB) -N CellId\n", printf("%s -h(elp) -p(extended_prefix) -N cell_id -f output_filename -F input_filename -g channel_model -n n_frames -t Delayspread -s snr0 -S snr1 -x transmission_mode -y TXant -z RXant -i Intefrence0 -j Interference1 -A interpolation_file -C(alibration offset dB) -N CellId\n",
argv[0]); argv[0]);
printf("-h This message\n"); printf("-h This message\n");
printf("-L <log level, 0(errors), 1(warning), 2(info) 3(debug) 4 (trace)>\n");
//printf("-p Use extended prefix mode\n"); //printf("-p Use extended prefix mode\n");
//printf("-d Use TDD\n"); //printf("-d Use TDD\n");
printf("-n Number of frames to simulate\n"); printf("-n Number of frames to simulate\n");
......
openair1/SIMULATION/NR_PHY/ulsim.c
View file @ ec286386
...@@ -275,7 +275,7 @@ int main(int argc, char **argv) ...@@ -275,7 +275,7 @@ int main(int argc, char **argv)
//unsigned char frame_type = 0; //unsigned char frame_type = 0;
NR_DL_FRAME_PARMS *frame_parms; NR_DL_FRAME_PARMS *frame_parms;
int loglvl = OAILOG_INFO; int loglvl = OAILOG_WARNING;
//uint64_t SSB_positions=0x01; //uint64_t SSB_positions=0x01;
uint16_t nb_symb_sch = 12; uint16_t nb_symb_sch = 12;
int start_symbol = 0; int start_symbol = 0;
...@@ -603,6 +603,7 @@ int main(int argc, char **argv) ...@@ -603,6 +603,7 @@ int main(int argc, char **argv)
//printf("-C Generate Calibration information for Abstraction (effective SNR adjustment to remove Pe bias w.r.t. AWGN)\n"); //printf("-C Generate Calibration information for Abstraction (effective SNR adjustment to remove Pe bias w.r.t. AWGN)\n");
printf("-F Input filename (.txt format) for RX conformance testing\n"); printf("-F Input filename (.txt format) for RX conformance testing\n");
printf("-G Offset of samples to read from file (0 default)\n"); printf("-G Offset of samples to read from file (0 default)\n");
printf("-L <log level, 0(errors), 1(warning), 2(info) 3(debug) 4 (trace)>\n");
printf("-M Multiple SSB positions in burst\n"); printf("-M Multiple SSB positions in burst\n");
printf("-N Nid_cell\n"); printf("-N Nid_cell\n");
printf("-O oversampling factor (1,2,4,8,16)\n"); printf("-O oversampling factor (1,2,4,8,16)\n");
...@@ -643,7 +644,7 @@ int main(int argc, char **argv) ...@@ -643,7 +644,7 @@ int main(int argc, char **argv)
else if (N_RB_UL == 106) bandwidth = 40; else if (N_RB_UL == 106) bandwidth = 40;
else if (N_RB_UL == 32) bandwidth = 50; else if (N_RB_UL == 32) bandwidth = 50;
else { printf("Add N_RB_UL %d\n",N_RB_UL); exit(-1); } else { printf("Add N_RB_UL %d\n",N_RB_UL); exit(-1); }
LOG_I( PHY,"++++++++++++++++++++++++++++++++++++++++++++++%i+++++++++++++++++++++++++++++++++++++++++",loglvl);
if (openair0_cfg[0].threequarter_fs == 1) sampling_frequency*=.75; if (openair0_cfg[0].threequarter_fs == 1) sampling_frequency*=.75;
UE2gNB = new_channel_desc_scm(n_tx, n_rx, channel_model, UE2gNB = new_channel_desc_scm(n_tx, n_rx, channel_model,
......
Markdown is supported
0% or
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment