Projeto final: pequeno caminho de inferência e relatório

Objetivo e critérios de aceitação

Mantenha model.mjs do laboratório anterior,crie capstone-lab.mjs ao lado dele e execute node capstone-lab.mjs. Conecte uma camada de convolução,viés,requantização,ReLU e max pooling. Informe valores separadamente do custo. Isso não é um classificador completo nem um mecanismo de treinamento. Você pode inspecionar manualmente os arquivos por um fluxo suportado de arquivos no Digital Design Studio instalado; a execução dentro do DDS não está estabelecida.

Conclua quando os acumuladores de convolução direta e produtos escalares de recortes concordarem,os casos de teste intermediários/finais coincidirem,e seu relatório incluir dimensões,MACs,ciclos presumidos,utilização de posições e bytes únicos dos tensores. Acrescente um caso de entrada alterada calculado manualmente e de forma independente. Mantenha comando,versão dos arquivos e premissas com o log. Investigue uma comparação falha em vez de apenas editar seu valor esperado.

Referência resolvida

A entrada é 4×4 com valores de 1 a 16. O kernel [1,0,−1;1,0,−1;1,0,−1] subtrai a amostra direita da esquerda em cada linha. Cada linha dá −2, portanto toda janela soma −6. O viés oito dá quatro acumuladores iguais a dois. O multiplicador 0.5 e o ponto zero igual a zero produzem o código um. ReLU o mantém e o pooling 2×2 dá [1]. Preserve também casos de borda assimétricos; saídas iguais podem ocultar erros de ordenação.

A dimensão [M,Ho,Wo] é [1,2,2]. Conte 36 MACs, incluindo termos de pesos zero. Com dois PEs de um MAC compartilhando a redução de cada saída, suponha cinco ciclos MAC mais um ciclo de combinação/escrita por saída, sem sobreposição nem pausas. O total modelado é 24 ciclos. A utilização das posições apenas de MACs é 36/(4×5×2)=90%. Isso não é latência medida de dispositivo.

Os tensores int8 únicos usam 16 bytes de entrada, 9 de pesos e 4 de saídas da convolução, mais um viés de quatro bytes. A saída de pooling usa um byte. O tráfego real pode incluir recargas e somas parciais. As escalas sx=sw=1, sy=2 explicam o multiplicador 0.5.

Mudar o viés para zero prevê acumulador −6, código −3, valor ativado zero e pooling zero. Alterar apenas a primeira entrada de um para dois dá primeiro acumulador três, empate 1.5 arredondando para dois, três saídas restantes um e pooling dois. Teste as mudanças separadamente. Mais PEs ou omitir pesos zero exige um novo cronograma explícito preservando os valores.

Arquivo executável

import assert from 'node:assert/strict';
import {conv2d, im2col, dot, requantize, maxPool2x2} from './model.mjs';
const x = [Array.from({length: 4}, (_, r) => Array.from({length: 4}, (_, c) => r * 4 + c + 1))];
const kernel = [[1, 0, -1], [1, 0, -1], [1, 0, -1]];
const weights = [[kernel]], bias = [8];
const reference = conv2d(x, weights, {bias});
const flattenedWeights = kernel.flat();
const rows = im2col(x, 3, 3);
const accumulators = rows.map(row => dot(row, flattenedWeights) + bias[0]);
assert.deepEqual(accumulators, reference[0].flat());
assert.deepEqual(accumulators, [2, 2, 2, 2]);
const activated = reference[0].map(row => row.map(v => Math.max(0, requantize(v, 0.5))));
const pooled = maxPool2x2(activated);
assert.deepEqual(activated, [[1, 1], [1, 1]]);
assert.deepEqual(pooled, [[1]]);
// Cycle estimate with explicit assumptions: all MAC operands local, no stalls,
// one output at a time, K=9, two single-MAC PEs, no overlap with output write.
const macs = rows.length * flattenedWeights.length;
const peCount = 2, macCyclesPerOutput = Math.ceil(flattenedWeights.length / peCount);
const modelCycles = rows.length * (macCyclesPerOutput + 1);
const usefulSlots = macs, scheduledSlots = rows.length * macCyclesPerOutput * peCount;
assert.equal(macs, 36);
assert.equal(modelCycles, 24);
assert.equal(usefulSlots / scheduledSlots, 0.9);
const report = {shape:[1, 2, 2], accumulators, activated, pooled, macs, peCount,
  modelCycles, macSlotUtilization:usefulSlots / scheduledSlots,
  uniqueInputBytes:16, weightBytes:9, outputBytes:4, biasBytes:4,
  scope:'Software reference and schedule estimate only; no synthesis or DDS run'};
console.log(JSON.stringify(report, null, 2));

Projeto final RTL: convolução inteira

Os anexos cnn_window3x3.sv e cnn_conv3x3.sv conectam janelas a nove produtos 8×8 com sinal, uma soma/viés de 24 bits com sinal, divisão por dois com arredondamento para o par mais próximo, saturação int8, ReLU e um máximo 2×2. A geometria é o contrato anterior de 4×4, canal único e janelas 3×3 sem preenchimento. As escalas de entrada/pesos são um, a escala de saída é dois e todos os pontos zero são zero. O kernel padrão é [1,0,−1;1,0,−1;1,0,−1], com viés oito. Os parâmetros do kernel podem mudar; verifique novamente o limite da soma de 24 bits.

Compare out_acc, out_code e out_activation separadamente, com out_row/out_col e transferências de saída aceitas. Done sobe após o quarto resultado ser aceito; pooled então é válido. Dezesseis entradas aceitas não implicam conclusão. Este RTL usa nove produtos paralelos e dois estágios registrados, em vez da estimativa anterior de dois PEs e 24 ciclos. Frequência/potência exigem evidências de síntese ou de dispositivo.

Execute o testbench compartilhado com autoverificação:

iverilog -g2012 -s tb_cnn_stream -o cnn-sim cnn_window3x3.sv cnn_conv3x3.sv tb_cnn_stream.sv
vvp cnn-sim

Os acumuladores padrão são dois, códigos/ativações são um e a saída de pooling é um. As verificações incluem limites com sinal, saturação, empates positivos/negativos, um kernel denso com sinal, pausas e reset. Verifique PASS frames=6 windows=24 outputs=24 kernels=2 em seu log real. Mude a primeira entrada para dois: a primeira soma três arredonda para o código dois e pooled torna-se dois. Depois, escolha outro kernel com sinal, calcule manualmente sua primeira janela, atualize a função independente de coeficientes e execute novamente todos os casos. Não copie resultados do DUT para o modelo de referência.