Hip poc/api buffer size

HIP_API.md

@@ -64,6 +64,7 @@ All functions are declared in [`hip/hipCompress.h`](hip/hipCompress.h).
 | Function | Description |
 |----------|-------------|
 | `hipCopyToWaveletLayout` | Copy from strided grid → contiguous wavelet buffer (zero-pad + optional RMS) |
+| `hipComputeRMS` | Convenience wrapper: RMS over extraction window only (no copy) |
 | `hipCopyFromWaveletLayout` | Copy from wavelet buffer → strided grid (extraction window only) |
 
 ### Compression / Decompression

hip/hipCompress.cpp

@@ -292,6 +292,26 @@ hipError_t hipCopyToWaveletLayout(
     return launch_err;
 }
 
+hipError_t hipComputeRMS(
+    const float* d_src,
+    int ldimx, int ldimxy,
+    int x0, int y0, int z0,
+    int ex, int ey, int ez,
+    double* d_rms_out,
+    hipCompressPlan* plan,
+    hipStream_t user_stream)
+{
+    if (!plan) return hipErrorInvalidValue;
+    plan->last_error = HIP_COMPRESS_SUCCESS;
+    if (!d_rms_out)
+        PLAN_ERROR(plan, HIP_COMPRESS_ERROR_NULL_OUTPUT, hipErrorInvalidValue);
+    return hipCopyToWaveletLayout(
+        d_src, ldimx, ldimxy,
+        x0, y0, z0, ex, ey, ez,
+        /*d_dst=*/nullptr, d_rms_out,
+        plan, user_stream);
+}
+
 hipError_t hipCopyFromWaveletLayout(
     const float* d_src,
     float* d_dst,
@@ -363,6 +383,18 @@ hipError_t hipDecompress(
     return launch_err;
 }
 
+hipError_t hipCompressBufferSize(const hipCompressPlan* plan, size_t* size)
+{
+    if (!plan) return hipErrorInvalidValue;
+    if (!size) {
+        plan->last_error = HIP_COMPRESS_ERROR_NULL_OUTPUT;
+        return hipErrorInvalidValue;
+    }
+    plan->last_error = HIP_COMPRESS_SUCCESS;
+    *size = (size_t)plan->nx * plan->ny * plan->nz * sizeof(float);
+    return hipSuccess;
+}
+
 hipError_t hipCompressMaxOutputSize(const hipCompressPlan* plan, size_t* size)
 {
     if (!plan) return hipErrorInvalidValue;

hip/hipCompress.h

@@ -105,6 +105,19 @@ hipError_t hipCopyToWaveletLayout(
     hipCompressPlan* plan,
     hipStream_t user_stream);
 
+// Convenience wrapper: RMS-only mode of hipCopyToWaveletLayout.
+// Computes RMS over the ex*ey*ez extraction window; writes no wavelet buffer.
+// Equivalent to hipCopyToWaveletLayout(..., d_dst=NULL, d_rms_out, ...).
+// Returns HIP_COMPRESS_ERROR_NULL_OUTPUT if d_rms_out is NULL.
+hipError_t hipComputeRMS(
+    const float* d_src,
+    int ldimx, int ldimxy,
+    int x0, int y0, int z0,
+    int ex, int ey, int ez,
+    double* d_rms_out,
+    hipCompressPlan* plan,
+    hipStream_t user_stream);
+
 // Copy from a wavelet-layout buffer back to a strided destination volume.
 // Only the ex*ey*ez extraction samples are written; padding is skipped.
 // Wavelet dims are derived from the extraction window.
@@ -158,6 +171,9 @@ hipError_t hipDecompress(
     hipCompressPlan* plan,
     hipStream_t user_stream);
 
+// Uncompressed wavelet-layout buffer size in bytes (for allocation).
+hipError_t hipCompressBufferSize(const hipCompressPlan* plan, size_t* size);
+
 // Upper bound on compressed output size in bytes (for allocation).
 hipError_t hipCompressMaxOutputSize(const hipCompressPlan* plan, size_t* size);
 

tests/test_compress_api_hip.cpp

@@ -106,6 +106,23 @@ static bool test_plan_lifecycle()
     size_t expected = (size_t)(8 + 8 * nb + 4) + (size_t)nb * 4 * 32768;
     if (max_sz != expected) { printf("  FAIL: MaxOutputSize=%zu expected=%zu\n", max_sz, expected); hipCompressDestroyPlan(plan); return false; }
     printf("  MaxOutputSize=%zu: PASS\n", max_sz);
+
+    // BufferSize
+    size_t buf_sz = 0;
+    HIPCHECK(hipCompressBufferSize(plan, &buf_sz));
+    size_t expected_buf = (size_t)64 * 64 * 64 * sizeof(float);
+    if (buf_sz != expected_buf) { printf("  FAIL: BufferSize=%zu expected=%zu\n", buf_sz, expected_buf); hipCompressDestroyPlan(plan); return false; }
+    printf("  BufferSize=%zu: PASS\n", buf_sz);
+    hipCompressDestroyPlan(plan);
+
+    // BufferSize with non-cubic dimensions
+    err = hipCompressCreatePlan(&plan, 128, 64, 32, 0);
+    if (err != hipSuccess) { printf("  FAIL: create 128x64x32 plan\n"); return false; }
+    buf_sz = 0;
+    HIPCHECK(hipCompressBufferSize(plan, &buf_sz));
+    expected_buf = (size_t)128 * 64 * 32 * sizeof(float);
+    if (buf_sz != expected_buf) { printf("  FAIL: BufferSize 128x64x32=%zu expected=%zu\n", buf_sz, expected_buf); hipCompressDestroyPlan(plan); return false; }
+    printf("  BufferSize 128x64x32=%zu: PASS\n", buf_sz);
     hipCompressDestroyPlan(plan);
 
     return true;
@@ -1461,6 +1478,20 @@ static bool test_copy_modes()
            rms_match ? "ok" : "FAIL");
     if (!rms_match) pass = false;
 
+    // hipComputeRMS wrapper: should match RMS-only path.
+    double h_rms3 = 0;
+    HIPCHECK(hipComputeRMS(
+        d_src, N, N * N,
+        0, 0, 0, N, N, N,
+        plan->d_rms, plan, 0));
+    HIPCHECK(hipDeviceSynchronize());
+    HIPCHECK(hipMemcpy(&h_rms3, plan->d_rms, sizeof(double), hipMemcpyDeviceToHost));
+    bool wrapper_match = (fabsf((float)h_rms3 - (float)h_rms1) < 1e-10f);
+    printf("  hipComputeRMS wrapper: rms=%.4e (match=%s): %s\n",
+           h_rms3, wrapper_match ? "yes" : "no",
+           wrapper_match ? "ok" : "FAIL");
+    if (!wrapper_match) pass = false;
+
     // Both NULL should error
     hipError_t err = hipCopyToWaveletLayout(
         d_src, N, N * N,
@@ -1470,6 +1501,20 @@ static bool test_copy_modes()
     printf("  both-null: error=%s: %s\n", err_ok ? "yes" : "no", err_ok ? "ok" : "FAIL");
     if (!err_ok) pass = false;
 
+    // hipComputeRMS with null d_rms_out -> NULL_OUTPUT
+    hipError_t err_rms = hipComputeRMS(
+        d_src, N, N * N,
+        0, 0, 0, N, N, N,
+        nullptr, plan, 0);
+    hipCompressError_t lib_err = hipCompressGetLastError(plan);
+    bool err_rms_ok = (err_rms != hipSuccess) &&
+                      (lib_err == HIP_COMPRESS_ERROR_NULL_OUTPUT);
+    printf("  hipComputeRMS null-output: error=%s code=%s: %s\n",
+           err_rms != hipSuccess ? "yes" : "no",
+           hipCompressErrorString(lib_err),
+           err_rms_ok ? "ok" : "FAIL");
+    if (!err_rms_ok) pass = false;
+
     printf("  mode dispatch: %s\n", pass ? "PASS" : "FAIL");
     hipFree(d_src); hipFree(d_dst);
     hipCompressDestroyPlan(plan);
@@ -1834,7 +1879,7 @@ static void bench_radial_sinc_sweep()
         int best_fwd_threads = 1, best_inv_threads = 1;
 
         std::vector<int> thread_sweep;
-        for (int t = 1; t <= max_threads; t *= 2) thread_sweep.push_back(t);
+        for (int t = 8; t <= max_threads; t *= 2) thread_sweep.push_back(t);
         if (thread_sweep.back() != max_threads) thread_sweep.push_back(max_threads);
 
         for (int nt : thread_sweep) {
@@ -2367,6 +2412,12 @@ static bool test_error_codes()
         if (err != hipErrorInvalidValue) { printf("  FAIL [MaxOutputSize null plan]\n"); pass = false; }
         else printf("  ok [MaxOutputSize null plan]\n");
     }
+    {
+        size_t sz = 0;
+        hipError_t err = hipCompressBufferSize(nullptr, &sz);
+        if (err != hipErrorInvalidValue) { printf("  FAIL [BufferSize null plan]\n"); pass = false; }
+        else printf("  ok [BufferSize null plan]\n");
+    }
 
     // Allocate a valid plan for the remaining tests
     hipCompressPlan* plan = nullptr;
@@ -2571,6 +2622,14 @@ static bool test_error_codes()
             pass = false;
     }
 
+    // --- BufferSize errors ---
+    {
+        hipError_t err = hipCompressBufferSize(plan, nullptr);
+        if (!check_error("BufferSize null size", plan, err,
+                          hipErrorInvalidValue, HIP_COMPRESS_ERROR_NULL_OUTPUT))
+            pass = false;
+    }
+
     // --- GetLastError / ErrorString ---
     {
         hipCompressError_t e = hipCompressGetLastError(nullptr);