wu-arch/vortex
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

fix: isolate unused Blackwell writeback lanes

Browse Source
This commit is contained in:
Zhongdi LUO
2026-07-13 10:27:08 +00:00
parent 9560f9cab6
commit a0b08f4c81
1 changed files with 11 additions and 9 deletions
Download Patch File Download Diff File Expand all files Collapse all files

20
hw/rtl/core/VX_tensor_blackwell_core.sv
View File

@@ -78,7 +78,9 @@ module VX_tensor_blackwell_core_block import VX_gpu_pkg::*; #(
wire [`NW_WIDTH-1:0] writeback_wid;
wire [4:0] writeback_rd;
logic writeback_ready;
wire [`NUM_THREADS-1:0][`XLEN-1:0] writeback_data;
wire [7:0][`XLEN-1:0] tensor_writeback_data;
wire [`NUM_THREADS-1:0][`XLEN-1:0] writeback_data =
tensor_writeback_data[`NUM_THREADS-1:0];
wire metadata_valid = !metadata_queue_empty;
wire bwgmma = metadata_valid &&
@@ -135,14 +137,14 @@ module VX_tensor_blackwell_core_block import VX_gpu_pkg::*; #(
.io_writeback_bits_last(writeback_last),
.io_writeback_bits_wid(writeback_wid),
.io_writeback_bits_rd(writeback_rd),
.io_writeback_bits_data_0(writeback_data[0]),
.io_writeback_bits_data_1(writeback_data[1]),
.io_writeback_bits_data_2(writeback_data[2]),
.io_writeback_bits_data_3(writeback_data[3]),
.io_writeback_bits_data_4(writeback_data[4]),
.io_writeback_bits_data_5(writeback_data[5]),
.io_writeback_bits_data_6(writeback_data[6]),
.io_writeback_bits_data_7(writeback_data[7]),
.io_writeback_bits_data_0(tensor_writeback_data[0]),
.io_writeback_bits_data_1(tensor_writeback_data[1]),
.io_writeback_bits_data_2(tensor_writeback_data[2]),
.io_writeback_bits_data_3(tensor_writeback_data[3]),
.io_writeback_bits_data_4(tensor_writeback_data[4]),
.io_writeback_bits_data_5(tensor_writeback_data[5]),
.io_writeback_bits_data_6(tensor_writeback_data[6]),
.io_writeback_bits_data_7(tensor_writeback_data[7]),
.io_respA_ready(tmem_if.rsp_ready),
.io_respA_valid(tmem_if.rsp_valid),