material_eval.cu

// CUDA Chess Material Evaluation
// Ported from PTX by Helmsman (JetsonClaw1)
// Optimized for RTX 4050 (sm_89) by Forgemaster
//
// Board encoding: 64 uint32 values
//   0 = empty, 100 = pawn, 320 = knight, 330 = bishop,
//   500 = rook, 900 = queen, 20000 = king
//
// Batch evaluation: process multiple boards in parallel

#include <cstdio>
#include <cstdint>

#define BOARD_SIZE 64
#define WARP_SIZE 32

// Piece values for evaluation
__device__ __constant__ int32_t PIECE_VALUES[7] = {
    0,      // empty
    100,    // pawn
    320,    // knight
    330,    // bishop
    500,    // rook
    900,    // queen
    20000   // king
};

// Piece-square tables for positional evaluation (optional extension)
// For now: material only

__global__ void batch_material_eval(
    const int32_t* __restrict__ boards,  // [num_boards * 64]
    int32_t* __restrict__ scores,        // [num_boards]
    int num_boards
) {
    int tid = threadIdx.x + blockIdx.x * blockDim.x;
    int boards_per_block = blockDim.x / BOARD_SIZE;
    int board_idx = tid / BOARD_SIZE;
    int square_idx = tid % BOARD_SIZE;
    
    if (board_idx >= num_boards) return;
    
    int32_t piece = boards[board_idx * BOARD_SIZE + square_idx];
    
    // Atomic add to this board's score
    atomicAdd(&scores[board_idx], piece);
}

// Host wrapper
void evaluate_boards(const int32_t* h_boards, int32_t* h_scores, int num_boards) {
    int32_t* d_boards;
    int32_t* d_scores;
    
    cudaMalloc(&d_boards, num_boards * BOARD_SIZE * sizeof(int32_t));
    cudaMalloc(&d_scores, num_boards * sizeof(int32_t));
    cudaMemset(d_scores, 0, num_boards * sizeof(int32_t));
    
    cudaMemcpy(d_boards, h_boards, num_boards * BOARD_SIZE * sizeof(int32_t), cudaMemcpyHostToDevice);
    
    int threads = BOARD_SIZE * 8;  // 512 threads
    int blocks = (num_boards * BOARD_SIZE + threads - 1) / threads;
    
    batch_material_eval<<<blocks, threads>>>(d_boards, d_scores, num_boards);
    
    cudaMemcpy(h_scores, d_scores, num_boards * sizeof(int32_t), cudaMemcpyDeviceToHost);
    
    cudaFree(d_boards);
    cudaFree(d_scores);
}

int main() {
    // Test: evaluate starting position
    int32_t board[BOARD_SIZE] = {0};
    // Standard starting position piece values
    // Row 0: black pieces (rook, knight, bishop, queen, king, bishop, knight, rook)
    board[0] = 500; board[1] = 320; board[2] = 330; board[3] = 900;
    board[4] = 20000; board[5] = 330; board[6] = 320; board[7] = 500;
    // Row 1: black pawns
    for (int i = 8; i < 16; i++) board[i] = 100;
    // Row 6: white pawns
    for (int i = 48; i < 56; i++) board[i] = 100;
    // Row 7: white pieces
    board[56] = 500; board[57] = 320; board[58] = 330; board[59] = 900;
    board[60] = 20000; board[61] = 330; board[62] = 320; board[63] = 500;
    
    int32_t score = 0;
    evaluate_boards(board, &score, 1);
    
    printf("Starting position material score: %d\n", score);
    printf("Expected: %d (mirror positions cancel out)\n", 40000); // 2 kings
    
    return 0;
}
EOF

git init
git add -A
git commit -m "Forgemaster handoff: CUDA chess eval from Jetson PTX

- material_eval.cu: CUDA kernel with batch evaluation
- Ported from Helmsman's PTX kernel (8 regs, 512 threads)
- Extended with host wrapper, constant piece values
- Ready for sm_89 optimization and minimax wrapper
- Bridge Report context in README"

gh repo create Lucineer/forgemaster-chess-eval --public --description "CUDA chess evaluation — Forgemaster handoff from Jetson Bridge" --source=. --push 2>&1 | tail -3

echo ""
echo "=== All repos this session ==="
echo "ptx-room:     github.com/Lucineer/ptx-room (v0.2.1, all gates working)"
echo "chess-dojo-v2: github.com/Lucineer/chess-dojo-v2 (v0.2, python-chess)"
echo "plato-chess:  github.com/Lucineer/plato-chess-dojo (v0.2)"
echo "zeroclaws:    github.com/Lucineer/zeroclaws (Bridge Report #001)"
echo "cudaclaw:     github.com/Lucineer/cudaclaw (PTX submodule)"
echo "forge-eval:   github.com/Lucineer/forgemaster-chess-eval (CUDA handoff)"