CLAUDE: GpuBuf<T> RAII device buffer; test_avg_td converted (C++17 style ruling for new host code)

Co-Authored-By: Claude Fable 5 <noreply@anthropic.com>

eclipse_setup/nsight_per_kernel_tests.md

@@ -58,6 +58,12 @@ One debug configuration per test binary — create once, reuse forever:
 2. Copy `src/tests/test_avg_td.cu` → `test_<name>.cu`, adapt buffers/launch.
 3. Java side: export a test case with `com.elphel.imagej.gpu.testdata`
    (`TestDataWriter`; see `AvgTdExport` as the model — export ALL inputs
-   plus `expected_*` from the CPU oracle).
+   plus `expected_*` from the CPU oracle). To eyeball any buffer or C++
+   result in ImageJ without the raw importer:
+   `TestDataView.viewCase(dir, name).show()` / `viewResults(dir, name)` /
+   `showAll(dir)` — dims come from the manifest, TD buffers render as tile
+   grids (presentCltData quadrant layout).
+   Host-code style for new tests: C++17 + RAII (`tp_gpu_buf.h` GpuBuf<T>),
+   no malloc/free or manual cudaFree.
 4. `build_tests.sh test_<name>`, run, then create the debug configuration
    (steps above).

src/tests/test_avg_td.cu

@@ -27,6 +27,7 @@
 
 #include "tp_consolidate.h"
 #include "tests/tp_test_data.h"
+#include "tests/tp_gpu_buf.h"
 
 int main(int argc, char **argv)
 {
@@ -60,10 +61,8 @@ int main(int argc, char **argv)
 	printf("num_sensors=%d num_tiles=%d\n", num_sensors, num_tiles);
 
 	float * gpu_td_sens = data.gpuFloat("td_sens");
-	float * gpu_td_avg  {};
-	float * gpu_counts  {};
-	checkCudaErrors(cudaMalloc(&gpu_td_avg, (size_t) num_tiles * TD_TILE_FLOATS * sizeof(float)));
-	checkCudaErrors(cudaMalloc(&gpu_counts, (size_t) num_tiles * sizeof(float)));
+	GpuBuf<float> gpu_td_avg((size_t) num_tiles * TD_TILE_FLOATS); // RAII, freed on scope exit
+	GpuBuf<float> gpu_counts((size_t) num_tiles);
 
 	dim3 blocks(num_tiles);
 	dim3 threads(TD_TILE_FLOATS);
@@ -72,24 +71,22 @@ int main(int argc, char **argv)
 				num_sensors,
 				gpu_td_sens,
 				num_tiles,
-				gpu_td_avg,
-				gpu_counts);
+				gpu_td_avg.dev(),
+				gpu_counts.dev());
 		checkCudaErrors(cudaGetLastError());
 	}
 	checkCudaErrors(cudaDeviceSynchronize());
 	printf("clt_average_sensors: %d run(s) done\n", num_runs);
 
-	data.saveResult("td_avg", gpu_td_avg, {num_tiles, TD_TILE_FLOATS});
-	data.saveResult("counts", gpu_counts, {num_tiles});
+	data.saveResult("td_avg", gpu_td_avg.dev(), {num_tiles, TD_TILE_FLOATS});
+	data.saveResult("counts", gpu_counts.dev(), {num_tiles});
 
 	int fail = 0;
-	int r = data.compare("td_avg", gpu_td_avg, tol);
+	int r = data.compare("td_avg", gpu_td_avg.dev(), tol);
 	if (r > 0) fail = 1;
-	r = data.compare("counts", gpu_counts, 0.0f); // counts must match exactly
+	r = data.compare("counts", gpu_counts.dev(), 0.0f); // counts must match exactly
 	if (r > 0) fail = 1;
 
-	checkCudaErrors(cudaFree(gpu_td_avg));
-	checkCudaErrors(cudaFree(gpu_counts));
 	printf("%s: %s\n", argv[0], fail ? "FAIL" : "PASS");
 	return fail ? EXIT_FAILURE : EXIT_SUCCESS;
 }

src/tests/tp_gpu_buf.h

+/*
+ * tp_gpu_buf.h
+ *
+ * Minimal RAII device buffer for per-kernel test mains: allocation and
+ * cudaFree tied to scope, so a test cannot leak GPU memory and stays short.
+ * Style ruling for NEW host code in this repo: C++17, RAII, std::vector —
+ * no malloc/free, no manual cudaFree bookkeeping (old harness left as-is).
+ *
+ *  Created on: Jul 10, 2026
+ *      Author: Claude (Anthropic), for Elphel
+ */
+
+#ifndef SRC_TESTS_TP_GPU_BUF_H_
+#define SRC_TESTS_TP_GPU_BUF_H_
+
+#include <vector>
+#include <cuda_runtime.h>
+#include <helper_cuda.h>   // checkCudaErrors
+
+template <typename T> class GpuBuf {
+public:
+	explicit GpuBuf(size_t elems) : m_elems(elems) {
+		checkCudaErrors(cudaMalloc(&m_dev, elems * sizeof(T)));
+	}
+	~GpuBuf() {
+		if (m_dev) cudaFree(m_dev);
+	}
+	GpuBuf(const GpuBuf &) = delete;             // no accidental double-free
+	GpuBuf & operator=(const GpuBuf &) = delete;
+	GpuBuf(GpuBuf && other) noexcept : m_dev(other.m_dev), m_elems(other.m_elems) {
+		other.m_dev = nullptr;
+	}
+
+	T *      dev()   const { return m_dev; }
+	size_t   elems() const { return m_elems; }
+
+	std::vector<T> download() const {
+		std::vector<T> host(m_elems);
+		checkCudaErrors(cudaMemcpy(host.data(), m_dev, m_elems * sizeof(T),
+				cudaMemcpyDeviceToHost));
+		return host;
+	}
+	void upload(const std::vector<T> & host) {
+		checkCudaErrors(cudaMemcpy(m_dev, host.data(), m_elems * sizeof(T),
+				cudaMemcpyHostToDevice));
+	}
+
+private:
+	T *    m_dev {};
+	size_t m_elems {};
+};
+
+#endif /* SRC_TESTS_TP_GPU_BUF_H_ */