Skip to content
Snippets Groups Projects
Commit 82b80e51 authored by Reinier van der Walle's avatar Reinier van der Walle
Browse files

processed review comments

parent 083a2b94
No related branches found
No related tags found
2 merge requests!100Removed text for XSub that is now written in Confluence Subband correlator...,!93resolves L2SDP-191
Expand all Show whitespace changes
Inline Side-by-side
applications/lofar2/designs/lofar2_unb2b_ring/lofar2_unb2b_ring.cl
+ 42
− 35
View file @ 82b80e51
......@@ -37,7 +37,7 @@
#define LANE_DIRECTION 1
// Nof lanes = 1 - 8
#define NOF_LANES 8
#define USE_DP_LAYER true
#define USE_DP_LAYER
#define ETH_HEADER_SIZE 2
#define DP_HEADER_SIZE 3
#ifdef EMULATOR
......@@ -45,7 +45,7 @@
#else
#define PAYLOAD_SIZE 750 // = 750*8 bytes = 6000 bytes
#endif
#if USE_DP_LAYER
#ifdef USE_DP_LAYER
#define BLOCK_LENGTH (PAYLOAD_SIZE+DP_HEADER_SIZE)
#else
#define BLOCK_LENGTH (PAYLOAD_SIZE+ETH_HEADER_SIZE)
......@@ -532,7 +532,7 @@ __kernel void block_validate_decode()
for (int x = 0; x < NOF_ERR_COUNTS; x++)
reg.parameters.err_cnt[x] = 0;
#if USE_DP_LAYER
#ifdef USE_DP_LAYER
union dp_packet packets[2]; //one to read and one to write
#else
union eth_packet packets[2]; //one to read and one to write
......@@ -561,6 +561,7 @@ __kernel void block_validate_decode()
else {
if (i != BLOCK_LENGTH-1)
reg.parameters.err_cnt[ERR_BI] += 1;
#pragma unroll
for (int err = 0; err < NOF_ERR_COUNTS; err++){
if (err != ERR_BI)
reg.parameters.err_cnt[err] += ((input_10GbE.err & (1 << err)) >> err);
......@@ -592,7 +593,7 @@ __kernel void block_validate_decode()
line_out.dp_channel = 0;
if (j == 0) {
line_out.flags |= FLAG_FIRST;
#if USE_DP_LAYER
#ifdef USE_DP_LAYER
line_out.dp_bsn = (packets[writeIndex].packet.dp_header.dp_sync_and_bsn & MASK_BSN); //62:0 = bsn
line_out.dp_channel = packets[writeIndex].packet.dp_header.dp_channel;
......@@ -644,7 +645,7 @@ __kernel void rx_split()
}
}
#if USE_DP_LAYER
#ifdef USE_DP_LAYER
__attribute__((max_global_work_dim(0)))
#ifndef EMULATOR
__attribute__((autorun))
......@@ -720,7 +721,7 @@ __kernel void no_validate_bsn_at_sync()
/* ----- End of ring_rx ----- */
/* ----- ring_tx ----- */
#if USE_DP_LAYER
#ifdef USE_DP_LAYER
__attribute__((max_global_work_dim(0)))
#ifndef EMULATOR
__attribute__((autorun))
......@@ -776,7 +777,7 @@ __attribute__((autorun))
#endif
__kernel void no_validate_channel()
{
uint no_param_arr[8] = [-1, -1, -1, -1, -1, -1, -1, -1];
uint no_param_arr[8] = [~0, ~0, ~0, ~0, ~0, ~0, ~0, ~0];
while(1){
// handle MM read/write requests
handle_ro_mm_request(CH_VALIDATE_CHANNEL, no_param_arr);
......@@ -826,7 +827,7 @@ __kernel void tx_encode()
ushort dp_channel = 0;
for (int j = 0; j < BLOCK_LENGTH; j++){
#if USE_DP_LAYER
#ifdef USE_DP_LAYER
if(j == 0 || (j > DP_HEADER_SIZE)){
#else
if(j == 0 || (j > ETH_HEADER_SIZE)){
......@@ -836,8 +837,10 @@ __kernel void tx_encode()
output_10GbE.flags = 0;
output_10GbE.err = 0;
if(j == 0){
#if USE_DP_LAYER
switch(j)
{
case 0:
#ifdef USE_DP_LAYER
dp_channel = input_dp.dp_channel + 1; //Add 1 hop.
if ((input_dp.flags & FLAG_SYNC)==FLAG_SYNC){
dp_sync_and_bsn = (input_dp.dp_bsn | MASK_SYNC); // set bsn and sync
......@@ -848,24 +851,28 @@ __kernel void tx_encode()
#endif
output_10GbE.flags = FLAG_FIRST;
output_10GbE.data = c_header_out[0];
}
else if (j == 1){
#if USE_DP_LAYER
break;
case 1:
#ifdef USE_DP_LAYER
output_10GbE.data = (c_header_out[1] | ((uint64_t) dp_channel));
#else
output_10GbE.data = c_header_out[1];
#endif
}
#if USE_DP_LAYER
else if (j == 2){
break;
#ifdef USE_DP_LAYER
case 2:
output_10GbE.data = dp_sync_and_bsn;
}
break;
#endif
else {
output_10GbE.data = input_dp.data;
}
if (j == BLOCK_LENGTH-1){
case (BLOCK_LENGTH-1):
output_10GbE.flags = FLAG_LAST;
// no break, we also want to execute the default case.
default:
output_10GbE.data = input_dp.data;
}
write_channel_intel(tx_sosi_channels[laneIndex], output_10GbE);
......
applications/ta2/bsp/hardware/lofar2_unb2b_ring_bsp/top.vhd
+ 145
− 142
View file @ 82b80e51
This diff is collapsed.
applications/ta2/ip/ta2_channel_cross/ta2_channel_cross.vhd
+ 27
− 20
View file @ 82b80e51
......@@ -55,7 +55,9 @@
-- | | | generics |
-- +-----------+---------+--------------------------------------------------------+
-- Remark:
-- . It may be nice to be able to configure a larger empty field to support g_nof_bytes > 32. Keep
-- . This IP should be configured according to the corresponding IO channel in the OpenCL code.
-- . It may be nice to be able to configure a larger empty field to support g_nof_bytes > 32
-- but that would mean that the data structure in the OpenCL code must be adapted. Keep
-- in mind that IO channels must be a multiple of 8 bits (bytes).
LIBRARY IEEE, common_lib, dp_lib, technology_lib;
......@@ -67,7 +69,7 @@ USE technology_lib.technology_pkg.ALL;
ENTITY ta2_channel_cross IS
GENERIC (
g_nof_streams : NATURAL;
g_nof_bytes : POSITIVE; -- Max = 32
g_nof_bytes : POSITIVE; -- nof bytes in payload field, Max = 32
g_reverse_bytes : BOOLEAN := TRUE;
g_fifo_size : NATURAL := 8;
g_use_err : BOOLEAN := FALSE;
......@@ -104,13 +106,18 @@ ARCHITECTURE str OF ta2_channel_cross IS
CONSTANT c_data_w : NATURAL := c_byte_w * g_nof_bytes;
CONSTANT c_empty_w : NATURAL := ceil_log2(g_nof_bytes);
CONSTANT c_err_w : NATURAL := sel_a_b(g_use_err, g_err_w, 0);
CONSTANT c_bsn_w : NATURAL := sel_a_b(g_use_err, g_err_w, 0);
CONSTANT c_channel_w : NATURAL := sel_a_b(g_use_err, g_err_w, 0);
CONSTANT c_bsn_w : NATURAL := sel_a_b(g_use_bsn, g_bsn_w, 0);
CONSTANT c_channel_w : NATURAL := sel_a_b(g_use_channel, g_channel_w, 0);
CONSTANT c_err_offset : NATURAL := (g_nof_bytes+1 * c_byte_w);
CONSTANT c_bsn_offset : NATURAL := c_err_offset+c_err_w;
CONSTANT c_channel_offset : NATURAL := c_bsn_offset+c_bsn_w;
CONSTANT c_sop_offset : NATURAL := g_nof_bytes*c_byte_w;
CONSTANT c_eop_offset : NATURAL := g_nof_bytes*c_byte_w+1;
CONSTANT c_sync_offset : NATURAL := g_nof_bytes*c_byte_w+2;
CONSTANT c_empty_offset : NATURAL := c_byte_w*(g_nof_bytes+1);
SIGNAL dp_latency_adapter_tx_snk_in_arr : t_dp_sosi_arr(g_nof_streams-1 DOWNTO 0);
SIGNAL dp_latency_adapter_tx_snk_out_arr : t_dp_siso_arr(g_nof_streams-1 DOWNTO 0);
SIGNAL dp_latency_adapter_tx_src_out_arr : t_dp_sosi_arr(g_nof_streams-1 DOWNTO 0);
......@@ -189,6 +196,16 @@ BEGIN
kernel_src_out_arr(stream).data(c_data_w-1 DOWNTO 0) <= dp_latency_adapter_tx_src_out_arr(stream).data(c_data_w-1 DOWNTO 0);
END GENERATE;
-- Assign control signals to correct data fields.
kernel_src_out_arr(stream).data(c_sop_offset) <= dp_latency_adapter_tx_src_out_arr(stream).sop;
kernel_src_out_arr(stream).data(c_eop_offset) <= dp_latency_adapter_tx_src_out_arr(stream).eop;
kernel_src_out_arr(stream).data(c_sync_offset) <= dp_latency_adapter_tx_src_out_arr(stream).sync WHEN g_use_sync ELSE '0';
kernel_src_out_arr(stream).data(c_empty_offset-1 DOWNTO c_empty_offset-c_empty_w) <= dp_latency_adapter_tx_src_out_arr(stream).empty(c_empty_w-1 DOWNTO 0);
kernel_src_out_arr(stream).valid <= dp_latency_adapter_tx_src_out_arr(stream).valid;
dp_latency_adapter_tx_src_in_arr(stream).ready <= kernel_src_in_arr(stream).ready;
dp_latency_adapter_tx_src_in_arr(stream).xon <= '1';
-- Assign optional meta data signals to correct data fields.
gen_err_out : IF g_use_err GENERATE
kernel_src_out_arr(stream).data(c_err_offset+c_err_w-1 DOWNTO c_err_offset) <= dp_latency_adapter_tx_src_out_arr(stream).err(c_err_w-1 DOWNTO 0);
END GENERATE;
......@@ -201,16 +218,6 @@ BEGIN
kernel_src_out_arr(stream).data(c_channel_offset+c_channel_w-1 DOWNTO c_channel_offset) <= dp_latency_adapter_tx_src_out_arr(stream).channel(c_channel_w-1 DOWNTO 0);
END GENERATE;
-- Assign control signals to correct data fields.
kernel_src_out_arr(stream).data(g_nof_bytes*c_byte_w+0) <= dp_latency_adapter_tx_src_out_arr(stream).sop;
kernel_src_out_arr(stream).data(g_nof_bytes*c_byte_w+1) <= dp_latency_adapter_tx_src_out_arr(stream).eop;
kernel_src_out_arr(stream).data(g_nof_bytes*c_byte_w+2) <= dp_latency_adapter_tx_src_out_arr(stream).sync WHEN g_use_sync ELSE '0';
kernel_src_out_arr(stream).data(c_byte_w*(g_nof_bytes+1)-1 DOWNTO c_byte_w*(g_nof_bytes+1)-c_empty_w) <= dp_latency_adapter_tx_src_out_arr(stream).empty(c_empty_w-1 DOWNTO 0);
kernel_src_out_arr(stream).valid <= dp_latency_adapter_tx_src_out_arr(stream).valid;
dp_latency_adapter_tx_src_in_arr(stream).ready <= kernel_src_in_arr(stream).ready;
dp_latency_adapter_tx_src_in_arr(stream).xon <= '1';
-- kernel_snk_in -> dp_src_out
----------------------------------------------------------------------------
-- Data mapping
......
0% or .
You are about to add 0 people to the discussion. Proceed with caution.
Please register or to comment