Repository: rapidsai/kvikio (language: python)

Issue:
Forward-merge branch-24.08 into branch-24.10
Forward-merge triggered by push to branch-24.08 that creates a PR to keep branch-24.10 up-to-date. If this PR is unable to be immediately merged due to conflicts, it will remain open for the team to manually merge. See [forward-merger docs](https://docs.rapids.ai/maintainers/forward-merger/) for more info.
Verify behavior of `KVIKIO_BOUNCE_BUFFER_SIZE`
When attempting to analyze large task sizes, it appears that the size of a copy is always capped at 16 MiB. 

Even with `KVIKIO_BOUNCE_BUFFER_SIZE` of 40 MB and `KVIKIO_TASK_SIZE` of 40 MB, I observe copies capped at ~16 MiB in the profile. Is there a way to enable larger task sizes?

![Image](https://github.com/user-attachments/assets/e5760f8c-7eeb-457b-9c79-878108f96ab7)


`nvtx3/nvtx3.hpp` file not found for CPU-only build.
```sh
path/to/kvikio/cpp/include/kvikio/utils.hpp:28:10: fatal error: 'nvtx3/nvtx3.hpp' file not found
   28 | #include <nvtx3/nvtx3.hpp>
      |          ^~~~~~~~~~~~~~~~~
```


Discussion / hints:

Fix patch:
diff --git a/.devcontainer/cuda11.8-conda/devcontainer.json b/.devcontainer/cuda11.8-conda/devcontainer.json
index 1f9d029417..a9d1b897bc 100644
--- a/.devcontainer/cuda11.8-conda/devcontainer.json
+++ b/.devcontainer/cuda11.8-conda/devcontainer.json
@@ -5,17 +5,17 @@
     "args": {
       "CUDA": "11.8",
       "PYTHON_PACKAGE_MANAGER": "conda",
-      "BASE": "rapidsai/devcontainers:24.08-cpp-cuda11.8-mambaforge-ubuntu22.04"
+      "BASE": "rapidsai/devcontainers:24.10-cpp-cuda11.8-mambaforge-ubuntu22.04"
     }
   },
   "runArgs": [
     "--rm",
     "--name",
-    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.08-cuda11.8-conda"
+    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.10-cuda11.8-conda"
   ],
   "hostRequirements": {"gpu": "optional"},
   "features": {
-    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.8": {}
+    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.10": {}
   },
   "overrideFeatureInstallOrder": [
     "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils"
diff --git a/.devcontainer/cuda11.8-pip/devcontainer.json b/.devcontainer/cuda11.8-pip/devcontainer.json
index 1472bd95bb..3130ab8687 100644
--- a/.devcontainer/cuda11.8-pip/devcontainer.json
+++ b/.devcontainer/cuda11.8-pip/devcontainer.json
@@ -5,17 +5,17 @@
     "args": {
       "CUDA": "11.8",
       "PYTHON_PACKAGE_MANAGER": "pip",
-      "BASE": "rapidsai/devcontainers:24.08-cpp-cuda11.8-ubuntu22.04"
+      "BASE": "rapidsai/devcontainers:24.10-cpp-cuda11.8-ubuntu22.04"
     }
   },
   "runArgs": [
     "--rm",
     "--name",
-    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.08-cuda11.8-pip"
+    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.10-cuda11.8-pip"
   ],
   "hostRequirements": {"gpu": "optional"},
   "features": {
-    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.8": {}
+    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.10": {}
   },
   "overrideFeatureInstallOrder": [
     "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils"
diff --git a/.devcontainer/cuda12.5-conda/devcontainer.json b/.devcontainer/cuda12.5-conda/devcontainer.json
index b815b954b9..70a8a37430 100644
--- a/.devcontainer/cuda12.5-conda/devcontainer.json
+++ b/.devcontainer/cuda12.5-conda/devcontainer.json
@@ -5,17 +5,17 @@
     "args": {
       "CUDA": "12.5",
       "PYTHON_PACKAGE_MANAGER": "conda",
-      "BASE": "rapidsai/devcontainers:24.08-cpp-mambaforge-ubuntu22.04"
+      "BASE": "rapidsai/devcontainers:24.10-cpp-mambaforge-ubuntu22.04"
     }
   },
   "runArgs": [
     "--rm",
     "--name",
-    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.08-cuda12.5-conda"
+    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.10-cuda12.5-conda"
   ],
   "hostRequirements": {"gpu": "optional"},
   "features": {
-    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.8": {}
+    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.10": {}
   },
   "overrideFeatureInstallOrder": [
     "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils"
diff --git a/.devcontainer/cuda12.5-pip/devcontainer.json b/.devcontainer/cuda12.5-pip/devcontainer.json
index b4996728db..6b14f93e2d 100644
--- a/.devcontainer/cuda12.5-pip/devcontainer.json
+++ b/.devcontainer/cuda12.5-pip/devcontainer.json
@@ -5,17 +5,17 @@
     "args": {
       "CUDA": "12.5",
       "PYTHON_PACKAGE_MANAGER": "pip",
-      "BASE": "rapidsai/devcontainers:24.08-cpp-cuda12.5-ubuntu22.04"
+      "BASE": "rapidsai/devcontainers:24.10-cpp-cuda12.5-ubuntu22.04"
     }
   },
   "runArgs": [
     "--rm",
     "--name",
-    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.08-cuda12.5-pip"
+    "${localEnv:USER:anon}-rapids-${localWorkspaceFolderBasename}-24.10-cuda12.5-pip"
   ],
   "hostRequirements": {"gpu": "optional"},
   "features": {
-    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.8": {}
+    "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils:24.10": {}
   },
   "overrideFeatureInstallOrder": [
     "ghcr.io/rapidsai/devcontainers/features/rapids-build-utils"
diff --git a/.github/workflows/build.yaml b/.github/workflows/build.yaml
index 7b826f0cb4..1f8796e67f 100644
--- a/.github/workflows/build.yaml
+++ b/.github/workflows/build.yaml
@@ -28,7 +28,7 @@ concurrency:
 jobs:
   cpp-build:
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-build.yaml@branch-24.10
     with:
       build_type: ${{ inputs.build_type || 'branch' }}
       branch: ${{ inputs.branch }}
@@ -37,7 +37,7 @@ jobs:
   python-build:
     needs: [cpp-build]
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-python-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-python-build.yaml@branch-24.10
     with:
       build_type: ${{ inputs.build_type || 'branch' }}
       branch: ${{ inputs.branch }}
@@ -46,7 +46,7 @@ jobs:
   upload-conda:
     needs: [cpp-build, python-build]
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-upload-packages.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-upload-packages.yaml@branch-24.10
     with:
       build_type: ${{ inputs.build_type || 'branch' }}
       branch: ${{ inputs.branch }}
@@ -56,7 +56,7 @@ jobs:
     if: github.ref_type == 'branch'
     needs: python-build
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/custom-job.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/custom-job.yaml@branch-24.10
     with:
       arch: "amd64"
       branch: ${{ inputs.branch }}
@@ -68,7 +68,7 @@ jobs:
       sha: ${{ inputs.sha }}
   wheel-build-cpp:
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.10
     with:
       matrix_filter: group_by([.ARCH, (.CUDA_VER|split(".")|map(tonumber)|.[0])]) | map(max_by(.PY_VER|split(".")|map(tonumber)))
       build_type: ${{ inputs.build_type || 'branch' }}
@@ -79,7 +79,7 @@ jobs:
   wheel-build-python:
     needs: wheel-build-cpp
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.10
     with:
       build_type: ${{ inputs.build_type || 'branch' }}
       branch: ${{ inputs.branch }}
@@ -89,7 +89,7 @@ jobs:
   wheel-publish-cpp:
     needs: wheel-build-cpp
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/wheels-publish.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/wheels-publish.yaml@branch-24.10
     with:
       build_type: ${{ inputs.build_type || 'branch' }}
       branch: ${{ inputs.branch }}
@@ -100,7 +100,7 @@ jobs:
   wheel-publish-python:
     needs: wheel-build-python
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/wheels-publish.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/wheels-publish.yaml@branch-24.10
     with:
       build_type: ${{ inputs.build_type || 'branch' }}
       branch: ${{ inputs.branch }}
diff --git a/.github/workflows/pr.yaml b/.github/workflows/pr.yaml
index f4dc586470..4499514060 100644
--- a/.github/workflows/pr.yaml
+++ b/.github/workflows/pr.yaml
@@ -23,38 +23,38 @@ jobs:
       - wheel-python-build
       - wheel-python-tests
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/pr-builder.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/pr-builder.yaml@branch-24.10
   checks:
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/checks.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/checks.yaml@branch-24.10
   conda-cpp-build:
     needs: checks
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-build.yaml@branch-24.10
     with:
       build_type: pull-request
   conda-cpp-tests:
     needs: conda-cpp-build
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-tests.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-tests.yaml@branch-24.10
     with:
       build_type: pull-request
   conda-python-build:
     needs: conda-cpp-build
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-python-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-python-build.yaml@branch-24.10
     with:
       build_type: pull-request
   conda-python-tests:
     needs: conda-python-build
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-python-tests.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-python-tests.yaml@branch-24.10
     with:
       build_type: pull-request
   docs-build:
     needs: conda-python-build
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/custom-job.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/custom-job.yaml@branch-24.10
     with:
       build_type: pull-request
       node_type: "gpu-v100-latest-1"
@@ -63,7 +63,7 @@ jobs:
       run_script: "ci/build_docs.sh"
   devcontainer:
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/build-in-devcontainer.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/build-in-devcontainer.yaml@branch-24.10
     with:
       arch: '["amd64"]'
       cuda: '["12.5"]'
@@ -73,7 +73,7 @@ jobs:
         sccache -s;
   wheel-cpp-build:
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.10
     with:
       matrix_filter: group_by([.ARCH, (.CUDA_VER|split(".")|map(tonumber)|.[0])]) | map(max_by(.PY_VER|split(".")|map(tonumber)))
       build_type: pull-request
@@ -81,14 +81,14 @@ jobs:
   wheel-python-build:
     needs: wheel-cpp-build
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/wheels-build.yaml@branch-24.10
     with:
       build_type: pull-request
       script: ci/build_wheel_python.sh
   wheel-python-tests:
     needs: wheel-python-build
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/wheels-test.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/wheels-test.yaml@branch-24.10
     with:
       build_type: pull-request
       script: ci/test_wheel.sh
diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml
index f794394103..d89e3ab71c 100644
--- a/.pre-commit-config.yaml
+++ b/.pre-commit-config.yaml
@@ -20,7 +20,7 @@ repos:
         files: python/.*
         args: ["--config", "python/kvikio/pyproject.toml"]
   - repo: https://github.com/PyCQA/flake8
-    rev: 6.0.0
+    rev: 7.1.1
     hooks:
       - id: flake8
         args: ["--config=.flake8"]
@@ -82,7 +82,7 @@ repos:
             ^CHANGELOG.md$
           )
   - repo: https://github.com/rapidsai/pre-commit-hooks
-    rev: v0.2.0
+    rev: v0.4.0
     hooks:
       - id: verify-copyright
         files: |
diff --git a/CHANGELOG.md b/CHANGELOG.md
index e86398d239..ef921bf18a 100644
--- a/CHANGELOG.md
+++ b/CHANGELOG.md
@@ -1,3 +1,53 @@
+# kvikio 24.10.00 (9 Oct 2024)
+
+## 🚨 Breaking Changes
+
+- Remove now unnecessary variable ([#466](https://github.com/rapidsai/kvikio/pull/466)) [@vyasr](https://github.com/vyasr)
+
+## 🐛 Bug Fixes
+
+- Fix no-cuda build ([#484](https://github.com/rapidsai/kvikio/pull/484)) [@madsbk](https://github.com/madsbk)
+- Add license to kvikio wheel ([#481](https://github.com/rapidsai/kvikio/pull/481)) [@raydouglass](https://github.com/raydouglass)
+- Fix release-after-main bug in `get_primary_cuda_context` ([#472](https://github.com/rapidsai/kvikio/pull/472)) [@madsbk](https://github.com/madsbk)
+- Intentionally leak static CUDA resources to avoid crash (part 2) ([#462](https://github.com/rapidsai/kvikio/pull/462)) [@kingcrimsontianyu](https://github.com/kingcrimsontianyu)
+- Fix several issues with wheel builds ([#439](https://github.com/rapidsai/kvikio/pull/439)) [@KyleFromNVIDIA](https://github.com/KyleFromNVIDIA)
+- Fix missing 24.08 =&gt; 24.10 renames ([#423](https://github.com/rapidsai/kvikio/pull/423)) [@madsbk](https://github.com/madsbk)
+
+## 📖 Documentation
+
+- Docs cleanup ([#480](https://github.com/rapidsai/kvikio/pull/480)) [@madsbk](https://github.com/madsbk)
+
+## 🚀 New Features
+
+- Update nvcomp to 4.0.1 ([#449](https://github.com/rapidsai/kvikio/pull/449)) [@vuule](https://github.com/vuule)
+
+## 🛠️ Improvements
+
+- nvtx support macos ([#474](https://github.com/rapidsai/kvikio/pull/474)) [@madsbk](https://github.com/madsbk)
+- Update update-version.sh to use packaging lib ([#469](https://github.com/rapidsai/kvikio/pull/469)) [@AyodeAwe](https://github.com/AyodeAwe)
+- Remove now unnecessary variable ([#466](https://github.com/rapidsai/kvikio/pull/466)) [@vyasr](https://github.com/vyasr)
+- Use CI workflow branch &#39;branch-24.10&#39; again ([#465](https://github.com/rapidsai/kvikio/pull/465)) [@jameslamb](https://github.com/jameslamb)
+- Update to flake8 7.1.1. ([#463](https://github.com/rapidsai/kvikio/pull/463)) [@bdice](https://github.com/bdice)
+- remove pytest dependency in kvikio conda recipe ([#461](https://github.com/rapidsai/kvikio/pull/461)) [@jameslamb](https://github.com/jameslamb)
+- Add support for Python 3.12, relax numcodecs pin ([#458](https://github.com/rapidsai/kvikio/pull/458)) [@jameslamb](https://github.com/jameslamb)
+- Skip cufile and nvtx3 when cuda isn&#39;t found ([#455](https://github.com/rapidsai/kvikio/pull/455)) [@madsbk](https://github.com/madsbk)
+- Non-static bounce buffer option ([#454](https://github.com/rapidsai/kvikio/pull/454)) [@madsbk](https://github.com/madsbk)
+- Use nvcomp defaults for algo options. ([#450](https://github.com/rapidsai/kvikio/pull/450)) [@bdice](https://github.com/bdice)
+- Update rapidsai/pre-commit-hooks ([#448](https://github.com/rapidsai/kvikio/pull/448)) [@KyleFromNVIDIA](https://github.com/KyleFromNVIDIA)
+- Make the bounce buffer size configurable ([#447](https://github.com/rapidsai/kvikio/pull/447)) [@madsbk](https://github.com/madsbk)
+- Update `build.sh` ([#445](https://github.com/rapidsai/kvikio/pull/445)) [@madsbk](https://github.com/madsbk)
+- Drop Python 3.9 support ([#443](https://github.com/rapidsai/kvikio/pull/443)) [@jameslamb](https://github.com/jameslamb)
+- Remove NumPy &lt;2 pin ([#441](https://github.com/rapidsai/kvikio/pull/441)) [@seberg](https://github.com/seberg)
+- C++ Test ([#440](https://github.com/rapidsai/kvikio/pull/440)) [@madsbk](https://github.com/madsbk)
+- `config_path()`: return a reference ([#438](https://github.com/rapidsai/kvikio/pull/438)) [@madsbk](https://github.com/madsbk)
+- Stream sync before calling cufile ([#437](https://github.com/rapidsai/kvikio/pull/437)) [@madsbk](https://github.com/madsbk)
+- Speeding up Python tests ([#435](https://github.com/rapidsai/kvikio/pull/435)) [@madsbk](https://github.com/madsbk)
+- Cython refactoring ([#434](https://github.com/rapidsai/kvikio/pull/434)) [@madsbk](https://github.com/madsbk)
+- Update pre-commit hooks ([#433](https://github.com/rapidsai/kvikio/pull/433)) [@KyleFromNVIDIA](https://github.com/KyleFromNVIDIA)
+- Improve update-version.sh ([#432](https://github.com/rapidsai/kvikio/pull/432)) [@bdice](https://github.com/bdice)
+- Use tool.scikit-build.cmake.version, set scikit-build-core minimum-version ([#430](https://github.com/rapidsai/kvikio/pull/430)) [@jameslamb](https://github.com/jameslamb)
+- build.sh: --pydevelop ([#429](https://github.com/rapidsai/kvikio/pull/429)) [@madsbk](https://github.com/madsbk)
+
 # kvikio 24.08.00 (7 Aug 2024)
 
 ## 🐛 Bug Fixes
diff --git a/README.md b/README.md
index 4df538aa2d..e787f8a4e0 100644
--- a/README.md
+++ b/README.md
@@ -2,7 +2,7 @@
 
 ## Summary
 
-KvikIO is a Python and C++ library for high performance file IO. It provides C++ and Python
+KvikIO (pronounced "kuh-VICK-eye-oh", see [here](https://ordnet.dk/ddo_en/dict?query=kvik) for pronunciation of kvik) is a Python and C++ library for high performance file IO. It provides C++ and Python
 bindings to [cuFile](https://docs.nvidia.com/gpudirect-storage/api-reference-guide/index.html),
 which enables [GPUDirect Storage (GDS)](https://developer.nvidia.com/blog/gpudirect-storage/).
 KvikIO also works efficiently when GDS isn't available and can read/write both host and device data seamlessly.
@@ -15,9 +15,92 @@ The C++ library is header-only making it easy to include in [existing projects](
 * A Python [Zarr](https://zarr.readthedocs.io/en/stable/) backend for reading and writing GPU data to file seamlessly.
 * Concurrent reads and writes using an internal thread pool.
 * Non-blocking API.
-* Handle both host and device IO seamlessly.
+* Transparently handles reads and writes to/from memory on both host and device.
 * Provides Python bindings to [nvCOMP](https://github.com/NVIDIA/nvcomp).
 
+
 ### Documentation
  * Python: <https://docs.rapids.ai/api/kvikio/nightly/>
  * C++: <https://docs.rapids.ai/api/libkvikio/nightly/>
+
+
+### Examples
+
+#### Python
+```python
+import cupy
+import kvikio
+
+def main(path):
+    a = cupy.arange(100)
+    f = kvikio.CuFile(path, "w")
+    # Write whole array to file
+    f.write(a)
+    f.close()
+
+    b = cupy.empty_like(a)
+    f = kvikio.CuFile(path, "r")
+    # Read whole array from file
+    f.read(b)
+    assert all(a == b)
+    f.close()
+
+    # Use contexmanager
+    c = cupy.empty_like(a)
+    with kvikio.CuFile(path, "r") as f:
+        f.read(c)
+    assert all(a == c)
+
+    # Non-blocking read
+    d = cupy.empty_like(a)
+    with kvikio.CuFile(path, "r") as f:
+        future1 = f.pread(d[:50])
+        future2 = f.pread(d[50:], file_offset=d[:50].nbytes)
+        # Note: must wait for futures before exiting block
+        # at which point the file is closed.
+        future1.get()  # Wait for first read
+        future2.get()  # Wait for second read
+    assert all(a == d)
+
+
+if __name__ == "__main__":
+    main("/tmp/kvikio-hello-world-file")
+```
+
+#### C++
+```c++
+#include <cstddef>
+#include <cuda_runtime.h>
+#include <kvikio/file_handle.hpp>
+using namespace std;
+
+int main()
+{
+  // Create two arrays `a` and `b`
+  constexpr std::size_t size = 100;
+  void *a = nullptr;
+  void *b = nullptr;
+  cudaMalloc(&a, size);
+  cudaMalloc(&b, size);
+
+  // Write `a` to file
+  kvikio::FileHandle fw("test-file", "w");
+  size_t written = fw.write(a, size);
+  fw.close();
+
+  // Read file into `b`
+  kvikio::FileHandle fr("test-file", "r");
+  size_t read = fr.read(b, size);
+  fr.close();
+
+  // Read file into `b` in parallel using 16 threads
+  kvikio::default_thread_pool::reset(16);
+  {
+    // FileHandles have RAII semantics
+    kvikio::FileHandle f("test-file", "r");
+    future<size_t> future = f.pread(b_dev, sizeof(a), 0);  // Non-blocking
+    size_t read = future.get(); // Blocking
+    // Notice, `f` closes automatically on destruction.
+  }
+}
+```
diff --git a/VERSION b/VERSION
index ec8489fda9..7c7ba04436 100644
--- a/VERSION
+++ b/VERSION
@@ -1,1 +1,1 @@
-24.08.00
+24.10.00
diff --git a/build.sh b/build.sh
index ca9de72a49..c02d0c876a 100755
--- a/build.sh
+++ b/build.sh
@@ -18,20 +18,21 @@ ARGS=$*
 # script, and that this script resides in the repo dir!
 REPODIR=$(cd $(dirname $0); pwd)
 
-VALIDARGS="clean libkvikio kvikio -v -g -n -s --ptds -h"
-HELP="$0 [clean] [libkvikio] [kvikio] [-v] [-g] [-n] [-s] [--ptds] [--cmake-args=\"<args>\"] [-h]
+VALIDARGS="clean libkvikio kvikio -v -g -n --pydevelop -h"
+HELP="$0 [clean] [libkvikio] [kvikio] [-v] [-g] [-n] [--cmake-args=\"<args>\"] [-h]
    clean                       - remove all existing build artifacts and configuration (start over)
    libkvikio                   - build and install the libkvikio C++ code
-   kvikio                      - build and install the kvikio Python package
+   kvikio                      - build and install the kvikio Python package (requires libkvikio)
    -v                          - verbose build mode
    -g                          - build for debug
    -n                          - no install step
+   --pydevelop                 - Install Python packages in editable mode
    --cmake-args=\\\"<args>\\\" - pass arbitrary list of CMake configuration options (escape all quotes in argument)
    -h                          - print this text
    default action (no args) is to build and install 'libkvikio' and 'kvikio' targets
 "
 LIBKVIKIO_BUILD_DIR=${LIBKVIKIO_BUILD_DIR:=${REPODIR}/cpp/build}
-KVIKIO_BUILD_DIR="${REPODIR}/python/build ${REPODIR}/python/_skbuild"
+KVIKIO_BUILD_DIR="${REPODIR}/python/kvikio/build/"
 BUILD_DIRS="${LIBKVIKIO_BUILD_DIR} ${KVIKIO_BUILD_DIR}"
 
 # Set defaults for vars modified by flags to this script
@@ -39,6 +40,8 @@ VERBOSE_FLAG=""
 BUILD_TYPE=Release
 INSTALL_TARGET=install
 RAN_CMAKE=0
+PYTHON_ARGS_FOR_INSTALL="-v --no-build-isolation --no-deps --config-settings rapidsai.disable-cuda=true"
+
 
 # Set defaults for vars that may not have been defined externally
 # If INSTALL_PREFIX is not set, check PREFIX, then check
@@ -108,18 +111,20 @@ fi
 # Process flags
 if hasArg -v; then
     VERBOSE_FLAG=-v
+    export SKBUILD_BUILD_VERBOSE=true
+    export SKBUILD_LOGGING_LEVEL=INFO
     set -x
 fi
 if hasArg -g; then
     BUILD_TYPE=Debug
+    export SKBUILD_INSTALL_STRIP=false
+    export SKBUILD_CMAKE_BUILD_TYPE=Debug
 fi
 if hasArg -n; then
     INSTALL_TARGET=""
 fi
-
-# Append `-DFIND_KVIKIO_CPP=ON` to EXTRA_CMAKE_ARGS unless a user specified the option.
-if [[ "${EXTRA_CMAKE_ARGS}" != *"DFIND_KVIKIO_CPP"* ]]; then
-    EXTRA_CMAKE_ARGS="${EXTRA_CMAKE_ARGS} -DFIND_KVIKIO_CPP=ON"
+if hasArg --pydevelop; then
+    PYTHON_ARGS_FOR_INSTALL="${PYTHON_ARGS_FOR_INSTALL} -e"
 fi
 
 # If clean given, run it prior to any other steps
@@ -144,7 +149,7 @@ if (( NUMARGS == 0 )) || hasArg libkvikio; then
     cmake --build "${LIBKVIKIO_BUILD_DIR}" -j${PARALLEL_LEVEL} ${VERBOSE_FLAG}
     if [[ ${INSTALL_TARGET} != "" ]]; then
         echo "installing libkvikio..."
-        cmake --build "${LIBKVIKIO_BUILD_DIR}" --target install -v ${VERBOSE_FLAG}
+        cmake --build "${LIBKVIKIO_BUILD_DIR}" --target install ${VERBOSE_FLAG}
     fi
 fi
 
@@ -153,5 +158,5 @@ if (( NUMARGS == 0 )) || hasArg kvikio; then
     echo "building kvikio..."
     cd ${REPODIR}/python/kvikio
     SKBUILD_CMAKE_ARGS="-DCMAKE_PREFIX_PATH=${INSTALL_PREFIX};-DCMAKE_LIBRARY_PATH=${LIBKVIKIO_BUILD_DIR};${EXTRA_CMAKE_ARGS}" \
-        python -m pip install --no-build-isolation --no-deps --config-settings rapidsai.disable-cuda=true .
+        python -m pip install ${PYTHON_ARGS_FOR_INSTALL} .
 fi
diff --git a/ci/release/update-version.sh b/ci/release/update-version.sh
index 585c9d7697..6b9458b273 100755
--- a/ci/release/update-version.sh
+++ b/ci/release/update-version.sh
@@ -17,14 +17,13 @@ CURRENT_MINOR=$(echo $CURRENT_TAG | awk '{split($0, a, "."); print a[2]}')
 CURRENT_PATCH=$(echo $CURRENT_TAG | awk '{split($0, a, "."); print a[3]}')
 CURRENT_SHORT_TAG=${CURRENT_MAJOR}.${CURRENT_MINOR}
 
-#Get <major>.<minor> for next version
+# Get <major>.<minor> for next version
 NEXT_MAJOR=$(echo $NEXT_FULL_TAG | awk '{split($0, a, "."); print a[1]}')
 NEXT_MINOR=$(echo $NEXT_FULL_TAG | awk '{split($0, a, "."); print a[2]}')
 NEXT_SHORT_TAG=${NEXT_MAJOR}.${NEXT_MINOR}
-NEXT_UCX_PY_VERSION="$(curl -sL https://version.gpuci.io/rapids/${NEXT_SHORT_TAG}).*"
 
 # Need to distutils-normalize the original version
-NEXT_SHORT_TAG_PEP440=$(python -c "from setuptools.extern import packaging; print(packaging.version.Version('${NEXT_SHORT_TAG}'))")
+NEXT_SHORT_TAG_PEP440=$(python -c "from packaging.version import Version; print(Version('${NEXT_SHORT_TAG}'))")
 
 echo "Preparing release $CURRENT_TAG => $NEXT_FULL_TAG"
 
@@ -44,7 +43,7 @@ for DEP in "${DEPENDENCIES[@]}"; do
     sed_runner "/-.* ${DEP}\(-cu[[:digit:]]\{2\}\)\{0,1\}==/ s/==.*/==${NEXT_SHORT_TAG_PEP440}.*,>=0.0.0a0/g" "${FILE}"
   done
   for FILE in python/*/pyproject.toml; do
-    sed_runner "/\"${DEP}==/ s/==.*\"/==${NEXT_SHORT_TAG_PEP440}.*,>=0.0.0a0\"/g" ${FILE}
+    sed_runner "/\"${DEP}==/ s/==.*\"/==${NEXT_SHORT_TAG_PEP440}.*,>=0.0.0a0\"/g" "${FILE}"
   done
 done
 
diff --git a/conda/environments/all_cuda-118_arch-aarch64.yaml b/conda/environments/all_cuda-118_arch-aarch64.yaml
index 6e2e085874..65ca39fa34 100644
--- a/conda/environments/all_cuda-118_arch-aarch64.yaml
+++ b/conda/environments/all_cuda-118_arch-aarch64.yaml
@@ -18,18 +18,18 @@ dependencies:
 - doxygen=1.9.1
 - gcc_linux-aarch64=11.*
 - ninja
-- numcodecs <0.12.0
-- numpy>=1.23,<2.0a0
+- numcodecs !=0.12.0
+- numpy>=1.23,<3.0a0
 - numpydoc
 - nvcc_linux-aarch64=11.8
-- nvcomp==3.0.6
+- nvcomp==4.0.1
 - packaging
 - pre-commit
 - pytest
 - pytest-cov
-- python>=3.9,<3.12
+- python>=3.10,<3.13
 - rapids-build-backend>=0.3.0,<0.4.0.dev0
-- scikit-build-core>=0.7.0
+- scikit-build-core>=0.10.0
 - sphinx
 - sphinx-click
 - sphinx_rtd_theme
diff --git a/conda/environments/all_cuda-118_arch-x86_64.yaml b/conda/environments/all_cuda-118_arch-x86_64.yaml
index 1cbc7c0a80..a020690e64 100644
--- a/conda/environments/all_cuda-118_arch-x86_64.yaml
+++ b/conda/environments/all_cuda-118_arch-x86_64.yaml
@@ -20,18 +20,18 @@ dependencies:
 - libcufile-dev=1.4.0.31
 - libcufile=1.4.0.31
 - ninja
-- numcodecs <0.12.0
-- numpy>=1.23,<2.0a0
+- numcodecs !=0.12.0
+- numpy>=1.23,<3.0a0
 - numpydoc
 - nvcc_linux-64=11.8
-- nvcomp==3.0.6
+- nvcomp==4.0.1
 - packaging
 - pre-commit
 - pytest
 - pytest-cov
-- python>=3.9,<3.12
+- python>=3.10,<3.13
 - rapids-build-backend>=0.3.0,<0.4.0.dev0
-- scikit-build-core>=0.7.0
+- scikit-build-core>=0.10.0
 - sphinx
 - sphinx-click
 - sphinx_rtd_theme
diff --git a/conda/environments/all_cuda-125_arch-aarch64.yaml b/conda/environments/all_cuda-125_arch-aarch64.yaml
index 4fcd7aff02..31145241d7 100644
--- a/conda/environments/all_cuda-125_arch-aarch64.yaml
+++ b/conda/environments/all_cuda-125_arch-aarch64.yaml
@@ -19,17 +19,17 @@ dependencies:
 - gcc_linux-aarch64=11.*
 - libcufile-dev
 - ninja
-- numcodecs <0.12.0
-- numpy>=1.23,<2.0a0
+- numcodecs !=0.12.0
+- numpy>=1.23,<3.0a0
 - numpydoc
-- nvcomp==3.0.6
+- nvcomp==4.0.1
 - packaging
 - pre-commit
 - pytest
 - pytest-cov
-- python>=3.9,<3.12
+- python>=3.10,<3.13
 - rapids-build-backend>=0.3.0,<0.4.0.dev0
-- scikit-build-core>=0.7.0
+- scikit-build-core>=0.10.0
 - sphinx
 - sphinx-click
 - sphinx_rtd_theme
diff --git a/conda/environments/all_cuda-125_arch-x86_64.yaml b/conda/environments/all_cuda-125_arch-x86_64.yaml
index d19547dc34..4d7d0be7c6 100644
--- a/conda/environments/all_cuda-125_arch-x86_64.yaml
+++ b/conda/environments/all_cuda-125_arch-x86_64.yaml
@@ -19,17 +19,17 @@ dependencies:
 - gcc_linux-64=11.*
 - libcufile-dev
 - ninja
-- numcodecs <0.12.0
-- numpy>=1.23,<2.0a0
+- numcodecs !=0.12.0
+- numpy>=1.23,<3.0a0
 - numpydoc
-- nvcomp==3.0.6
+- nvcomp==4.0.1
 - packaging
 - pre-commit
 - pytest
 - pytest-cov
-- python>=3.9,<3.12
+- python>=3.10,<3.13
 - rapids-build-backend>=0.3.0,<0.4.0.dev0
-- scikit-build-core>=0.7.0
+- scikit-build-core>=0.10.0
 - sphinx
 - sphinx-click
 - sphinx_rtd_theme
diff --git a/conda/recipes/kvikio/build.sh b/conda/recipes/kvikio/build.sh
index 462f06826b..af4b40291f 100644
--- a/conda/recipes/kvikio/build.sh
+++ b/conda/recipes/kvikio/build.sh
@@ -1,4 +1,4 @@
 #!/bin/bash
 # Copyright (c) 2018-2024, NVIDIA CORPORATION.
 
-./build.sh kvikio
+./build.sh -v kvikio
diff --git a/conda/recipes/kvikio/conda_build_config.yaml b/conda/recipes/kvikio/conda_build_config.yaml
index 4055133583..35cb51e2f9 100644
--- a/conda/recipes/kvikio/conda_build_config.yaml
+++ b/conda/recipes/kvikio/conda_build_config.yaml
@@ -20,4 +20,4 @@ cmake_version:
   - ">=3.26.4,!=3.30.0"
 
 nvcomp_version:
-  - "=3.0.6"
+  - "=4.0.1"
diff --git a/conda/recipes/kvikio/meta.yaml b/conda/recipes/kvikio/meta.yaml
index 2f2e308ae7..4a352012e3 100644
--- a/conda/recipes/kvikio/meta.yaml
+++ b/conda/recipes/kvikio/meta.yaml
@@ -62,15 +62,15 @@ requirements:
     - cuda-version ={{ cuda_version }}
     - nvcomp {{ nvcomp_version }}
     - rapids-build-backend >=0.3.0,<0.4.0.dev0
-    - scikit-build-core >=0.7.0
+    - scikit-build-core >=0.10.0
     - libkvikio ={{ version }}
   run:
     - python
-    - numpy >=1.23,<2.0a0
+    - numpy >=1.23,<3.0a0
     - cupy >=12.0.0
     - zarr
     # See https://github.com/zarr-developers/numcodecs/pull/475
-    - numcodecs <0.12.0
+    - numcodecs !=0.12.0
     - packaging
     - {{ pin_compatible('cuda-version', max_pin='x', min_pin='x') }}
     {% if cuda_major == "11" %}
@@ -78,8 +78,6 @@ requirements:
     {% endif %}
 
 test:
-  requires:
-    - pytest
   imports:
     - kvikio
 
diff --git a/conda/recipes/libkvikio/build.sh b/conda/recipes/libkvikio/build.sh
index 81f7c4fa30..82117d66a4 100644
--- a/conda/recipes/libkvikio/build.sh
+++ b/conda/recipes/libkvikio/build.sh
@@ -1,4 +1,4 @@
 #!/bin/bash
-# Copyright (c) 2023, NVIDIA CORPORATION.
+# Copyright (c) 2023-2024, NVIDIA CORPORATION.
 
-./build.sh -n libkvikio
+./build.sh -v -n libkvikio
diff --git a/cpp/CMakeLists.txt b/cpp/CMakeLists.txt
index f4f3f13109..5990405b1c 100644
--- a/cpp/CMakeLists.txt
+++ b/cpp/CMakeLists.txt
@@ -37,6 +37,7 @@ rapids_cmake_build_type(Release)
 
 # build options
 option(KvikIO_BUILD_EXAMPLES "Configure CMake to build examples" ON)
+option(KvikIO_BUILD_TESTS "Configure CMake to build tests" ON)
 
 rapids_cmake_support_conda_env(conda_env MODIFY_PREFIX_PATH)
 
@@ -55,32 +56,37 @@ rapids_find_package(
   INSTALL_EXPORT_SET kvikio-exports
 )
 
-rapids_find_package(
-  cuFile
-  BUILD_EXPORT_SET kvikio-exports
-  INSTALL_EXPORT_SET kvikio-exports
-)
-if(NOT cuFile_FOUND)
-  message(WARNING "Building KvikIO without cuFile")
-else()
-  file(READ "${cuFile_INCLUDE_DIRS}/cufile.h" CUFILE_H_STR)
-  string(FIND "${CUFILE_H_STR}" "cuFileBatchIOSetUp" cuFileBatchIOSetUp_location)
-  if(cuFileBatchIOSetUp_location EQUAL "-1")
-    set(cuFile_BATCH_API_FOUND FALSE)
-  else()
-    set(cuFile_BATCH_API_FOUND TRUE)
-  endif()
-  message(STATUS "Found cuFile Batch API: ${cuFile_BATCH_API_FOUND}")
-  string(FIND "${CUFILE_H_STR}" "cuFileReadAsync" cuFileReadAsync_location)
-  if(cuFileReadAsync_location EQUAL "-1")
-    set(cuFile_STREAM_API_FOUND FALSE)
+if(CUDAToolkit_FOUND)
+  rapids_find_package(
+    cuFile
+    BUILD_EXPORT_SET kvikio-exports
+    INSTALL_EXPORT_SET kvikio-exports
+  )
+  if(NOT cuFile_FOUND)
+    message(
+      WARNING "Cannot find cuFile - KvikIO will still work but won't use GPUDirect Storage (GDS)"
+    )
   else()
-    set(cuFile_STREAM_API_FOUND TRUE)
+    file(READ "${cuFile_INCLUDE_DIRS}/cufile.h" CUFILE_H_STR)
+    string(FIND "${CUFILE_H_STR}" "cuFileBatchIOSetUp" cuFileBatchIOSetUp_location)
+    if(cuFileBatchIOSetUp_location EQUAL "-1")
+      set(cuFile_BATCH_API_FOUND FALSE)
+    else()
+      set(cuFile_BATCH_API_FOUND TRUE)
+    endif()
+    message(STATUS "Found cuFile Batch API: ${cuFile_BATCH_API_FOUND}")
+    string(FIND "${CUFILE_H_STR}" "cuFileReadAsync" cuFileReadAsync_location)
+    if(cuFileReadAsync_location EQUAL "-1")
+      set(cuFile_STREAM_API_FOUND FALSE)
+    else()
+      set(cuFile_STREAM_API_FOUND TRUE)
+    endif()
+    message(STATUS "Found cuFile Stream API: ${cuFile_STREAM_API_FOUND}")
   endif()
-  message(STATUS "Found cuFile Stream API: ${cuFile_STREAM_API_FOUND}")
+
+  include(cmake/thirdparty/get_nvtx.cmake)
 endif()
 
-include(cmake/thirdparty/get_nvtx.cmake)
 include(cmake/thirdparty/get_thread_pool.cmake)
 
 # library targets
@@ -129,7 +135,8 @@ target_include_directories(
                    "$<INSTALL_INTERFACE:include>"
 )
 target_link_libraries(
-  kvikio INTERFACE Threads::Threads ${CMAKE_DL_LIBS} nvtx3::nvtx3-cpp BS::thread_pool
+  kvikio INTERFACE Threads::Threads BS::thread_pool ${CMAKE_DL_LIBS}
+                   $<TARGET_NAME_IF_EXISTS:nvtx3::nvtx3-cpp>
 )
 target_compile_features(kvikio INTERFACE cxx_std_17)
 
@@ -138,10 +145,17 @@ if(KvikIO_BUILD_EXAMPLES)
   add_subdirectory(examples)
 endif()
 
-# optionally build tests
-if(KvikIO_BUILD_TESTS AND CMAKE_PROJECT_NAME STREQUAL PROJECT_NAME)
+if(CUDAToolkit_FOUND
+   AND KvikIO_BUILD_TESTS
+   AND CMAKE_PROJECT_NAME STREQUAL PROJECT_NAME
+)
   include(cmake/thirdparty/get_gtest.cmake)
-  include(CTest) # calls enable_testing()
+
+  # include CTest module -- automatically calls enable_testing()
+  include(CTest)
+
+  # Always print verbose output when tests fail if run using `make test`.
+  list(APPEND CMAKE_CTEST_ARGUMENTS "--output-on-failure")
   add_subdirectory(tests)
 endif()
 
diff --git a/cpp/doxygen/main_page.md b/cpp/doxygen/main_page.md
index 2b404f835e..21a33b1d45 100644
--- a/cpp/doxygen/main_page.md
+++ b/cpp/doxygen/main_page.md
@@ -47,9 +47,9 @@ Install the **nightly release** from the ``rapidsai-nightly`` channel with the f
 # Install in existing environment
 mamba install -c rapidsai-nightly -c conda-forge libkvikio
 # Create new environment (CUDA 11.8)
-mamba create -n libkvikio-env -c rapidsai-nightly -c conda-forge python=3.11 cuda-version=11.8 libkvikio
+mamba create -n libkvikio-env -c rapidsai-nightly -c conda-forge python=3.12 cuda-version=11.8 libkvikio
 # Create new environment (CUDA 12.5)
-mamba create -n libkvikio-env -c rapidsai-nightly -c conda-forge python=3.11 cuda-version=12.5 libkvikio
+mamba create -n libkvikio-env -c rapidsai-nightly -c conda-forge python=3.12 cuda-version=12.5 libkvikio
 ```
 
 ---
@@ -85,14 +85,28 @@ Set the environment variable `KVIKIO_COMPAT_MODE` to enable/disable compatibilit
   - when running in Windows Subsystem for Linux (WSL).
   - when `/run/udev` isn't readable, which typically happens when running inside a docker image not launched with `--volume /run/udev:/run/udev:ro`.
 
+This setting can also be controlled by `defaults::compat_mode()` and `defaults::compat_mode_reset()`.
+
+
 #### Thread Pool (KVIKIO_NTHREADS)
 KvikIO can use multiple threads for IO automatically. Set the environment variable `KVIKIO_NTHREADS` to the number of threads in the thread pool. If not set, the default value is 1.
 
+This setting can also be controlled by `defaults::thread_pool_nthreads()` and `defaults::thread_pool_nthreads_reset()`.
+
 #### Task Size (KVIKIO_TASK_SIZE)
 KvikIO splits parallel IO operations into multiple tasks. Set the environment variable `KVIKIO_TASK_SIZE` to the maximum task size (in bytes). If not set, the default value is 4194304 (4 MiB).
 
+This setting can also be controlled by `defaults::task_size()` and `defaults::task_size_reset()`.
+
 #### GDS Threshold (KVIKIO_GDS_THRESHOLD)
-In order to improve performance of small IO, `.pread()` and `.pwrite()` implement a shortcut that circumvent the threadpool and use the POSIX backend directly. Set the environment variable `KVIKIO_GDS_THRESHOLD` to the minimum size (in bytes) to use GDS. If not set, the default value is 1048576 (1 MiB).
+To improve performance of small IO requests, `.pread()` and `.pwrite()` implement a shortcut that circumvents the threadpool and uses the POSIX backend directly. Set the environment variable `KVIKIO_GDS_THRESHOLD` to the minimum size (in bytes) to use GDS. If not set, the default value is 1048576 (1 MiB).
+
+This setting can also be controlled by `defaults::gds_threshold()` and `defaults::gds_threshold_reset()`.
+
+#### Size of the Bounce Buffer (KVIKIO_GDS_THRESHOLD)
+KvikIO might have to use intermediate host buffers (one per thread) when copying between files and device memory. Set the environment variable ``KVIKIO_BOUNCE_BUFFER_SIZE`` to the size (in bytes) of these "bounce" buffers. If not set, the default value is 16777216 (16 MiB).
+
+This setting can also be controlled by `defaults::bounce_buffer_size()` and `defaults::bounce_buffer_size_reset()`.
 
 
 ## Example
diff --git a/cpp/examples/CMakeLists.txt b/cpp/examples/CMakeLists.txt
index c12ddb2e52..5124c9ef71 100644
--- a/cpp/examples/CMakeLists.txt
+++ b/cpp/examples/CMakeLists.txt
@@ -14,19 +14,23 @@
 
 set(TEST_INSTALL_PATH bin/tests/libkvikio)
 
+# Example: basic_io
+
 if(CUDAToolkit_FOUND)
-  add_executable(BASIC_IO_TEST basic_io.cpp)
-  set_target_properties(BASIC_IO_TEST PROPERTIES INSTALL_RPATH "\$ORIGIN/../../lib")
-  target_include_directories(BASIC_IO_TEST PRIVATE ../include ${cuFile_INCLUDE_DIRS})
-  target_link_libraries(BASIC_IO_TEST PRIVATE kvikio CUDA::cudart)
+  add_executable(BASIC_IO_EXAMPLE basic_io.cpp)
+  set_target_properties(BASIC_IO_EXAMPLE PROPERTIES INSTALL_RPATH "\$ORIGIN/../../lib")
+  target_include_directories(BASIC_IO_EXAMPLE PRIVATE ../include ${cuFile_INCLUDE_DIRS})
+  target_link_libraries(BASIC_IO_EXAMPLE PRIVATE kvikio CUDA::cudart)
 
   if(CMAKE_COMPILER_IS_GNUCXX)
     set(KVIKIO_CXX_FLAGS "-Wall;-Werror;-Wno-unknown-pragmas")
-    target_compile_options(BASIC_IO_TEST PRIVATE "$<$<COMPILE_LANGUAGE:CXX>:${KVIKIO_CXX_FLAGS}>")
+    target_compile_options(
+      BASIC_IO_EXAMPLE PRIVATE "$<$<COMPILE_LANGUAGE:CXX>:${KVIKIO_CXX_FLAGS}>"
+    )
   endif()
 
   install(
-    TARGETS BASIC_IO_TEST
+    TARGETS BASIC_IO_EXAMPLE
     COMPONENT testing
     DESTINATION ${TEST_INSTALL_PATH}
     EXCLUDE_FROM_ALL
@@ -35,20 +39,22 @@ else()
   message(STATUS "Cannot build the basic_io example when CUDA is not found")
 endif()
 
-add_executable(BASIC_NO_CUDA_TEST basic_no_cuda.cpp)
-set_target_properties(BASIC_NO_CUDA_TEST PROPERTIES INSTALL_RPATH "\$ORIGIN/../../lib")
-target_include_directories(BASIC_NO_CUDA_TEST PRIVATE ../include)
-target_link_libraries(BASIC_NO_CUDA_TEST PRIVATE kvikio)
+# Example: basic_no_cuda
+
+add_executable(BASIC_NO_CUDA_EXAMPLE basic_no_cuda.cpp)
+set_target_properties(BASIC_NO_CUDA_EXAMPLE PROPERTIES INSTALL_RPATH "\$ORIGIN/../../lib")
+target_include_directories(BASIC_NO_CUDA_EXAMPLE PRIVATE ../include)
+target_link_libraries(BASIC_NO_CUDA_EXAMPLE PRIVATE kvikio)
 
 if(CMAKE_COMPILER_IS_GNUCXX)
   set(KVIKIO_CXX_FLAGS "-Wall;-Werror;-Wno-unknown-pragmas")
   target_compile_options(
-    BASIC_NO_CUDA_TEST PRIVATE "$<$<COMPILE_LANGUAGE:CXX>:${KVIKIO_CXX_FLAGS}>"
+    BASIC_NO_CUDA_EXAMPLE PRIVATE "$<$<COMPILE_LANGUAGE:CXX>:${KVIKIO_CXX_FLAGS}>"
   )
 endif()
 
 install(
-  TARGETS BASIC_NO_CUDA_TEST
+  TARGETS BASIC_NO_CUDA_EXAMPLE
   COMPONENT testing
   DESTINATION ${TEST_INSTALL_PATH}
   EXCLUDE_FROM_ALL
diff --git a/cpp/examples/downstream/cmake/get_kvikio.cmake b/cpp/examples/downstream/cmake/get_kvikio.cmake
index 8516bc332f..13f5bf50c2 100644
--- a/cpp/examples/downstream/cmake/get_kvikio.cmake
+++ b/cpp/examples/downstream/cmake/get_kvikio.cmake
@@ -30,4 +30,4 @@ function(find_and_configure_kvikio MIN_VERSION)
 
 endfunction()
 
-find_and_configure_kvikio("24.06")
+find_and_configure_kvikio("24.10")
diff --git a/cpp/include/kvikio/bounce_buffer.hpp b/cpp/include/kvikio/bounce_buffer.hpp
new file mode 100644
index 0000000000..8160be3c5c
--- /dev/null
+++ b/cpp/include/kvikio/bounce_buffer.hpp
@@ -0,0 +1,158 @@
+/*
+ * Copyright (c) 2024, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+#pragma once
+
+#include <mutex>
+#include <stack>
+
+#include <kvikio/defaults.hpp>
+
+namespace kvikio {
+
+/**
+ * @brief Singleton class to retain host memory allocations
+ *
+ * Call `AllocRetain::get` to get an allocation that will be retained when it
+ * goes out of scope (RAII). The size of all retained allocations are the same.
+ */
+class AllocRetain {
+ private:
+  std::mutex _mutex{};
+  // Stack of free allocations
+  std::stack<void*> _free_allocs{};
+  // The size of each allocation in `_free_allocs`
+  std::size_t _size{defaults::bounce_buffer_size()};
+
+  /**
+   * @brief An host memory allocation
+   */
+  class Alloc {
+   private:
+    AllocRetain* _manager;
+    void* _alloc;
+    std::size_t const _size;
+
+   public:
+    Alloc(AllocRetain* manager, void* alloc, std::size_t size)
+      : _manager(manager), _alloc{alloc}, _size{size}
+    {
+    }
+    Alloc(Alloc const&)            = delete;
+    Alloc& operator=(Alloc const&) = delete;
+    Alloc(Alloc&& o)               = delete;
+    Alloc& operator=(Alloc&& o)    = delete;
+    ~Alloc() noexcept { _manager->put(_alloc, _size); }
+    void* get() noexcept { return _alloc; }
+    std::size_t size() noexcept { return _size; }
+  };
+
+  AllocRetain() = default;
+
+  // Notice, we do not clear the allocations at destruction thus the allocations leaks
+  // at exit. We do this because `AllocRetain::instance()` stores the allocations in a
+  // static stack that are destructed below main, which is not allowed in CUDA:
+  // <https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#initialization>
+  ~AllocRetain() noexcept = default;
+
+  /**
+   * @brief Free all retained allocations
+   *
+   * NB: The `_mutex` must be taken prior to calling this function.
+   *
+   * @return The number of bytes cleared
+   */
+  std::size_t _clear()
+  {
+    std::size_t ret = _free_allocs.size() * _size;
+    while (!_free_allocs.empty()) {
+      CUDA_DRIVER_TRY(cudaAPI::instance().MemFreeHost(_free_allocs.top()));
+      _free_allocs.pop();
+    }
+    return ret;
+  }
+
+  /**
+   * @brief Ensure the sizes of the retained allocations match `defaults::bounce_buffer_size()`
+   *
+   * NB: `_mutex` must be taken prior to calling this function.
+   */
+  void _ensure_alloc_size()
+  {
+    auto const bounce_buffer_size = defaults::bounce_buffer_size();
+    if (_size != bounce_buffer_size) {
+      _clear();
+      _size = bounce_buffer_size;
+    }
+  }
+
+ public:
+  [[nodiscard]] Alloc get()
+  {
+    std::lock_guard const lock(_mutex);
+    _ensure_alloc_size();
+
+    // Check if we have an allocation available
+    if (!_free_allocs.empty()) {
+      void* ret = _free_allocs.top();
+      _free_allocs.pop();
+      return Alloc(this, ret, _size);
+    }
+
+    // If no available allocation, allocate and register a new one
+    void* alloc{};
+    // Allocate page-locked host memory
+    CUDA_DRIVER_TRY(cudaAPI::instance().MemHostAlloc(&alloc, _size, CU_MEMHOSTREGISTER_PORTABLE));
+    return Alloc(this, alloc, _size);
+  }
+
+  void put(void* alloc, std::size_t size)
+  {
+    std::lock_guard const lock(_mutex);
+    _ensure_alloc_size();
+
+    // If the size of `alloc` matches the sizes of the retained allocations,
+    // it is added to the set of free allocation otherwise it is freed.
+    if (size == _size) {
+      _free_allocs.push(alloc);
+    } else {
+      CUDA_DRIVER_TRY(cudaAPI::instance().MemFreeHost(alloc));
+    }
+  }
+
+  /**
+   * @brief Free all retained allocations
+   *
+   * @return The number of bytes cleared
+   */
+  std::size_t clear()
+  {
+    std::lock_guard const lock(_mutex);
+    return _clear();
+  }
+
+  static AllocRetain& instance()
+  {
+    static AllocRetain _instance;
+    return _instance;
+  }
+
+  AllocRetain(AllocRetain const&)            = delete;
+  AllocRetain& operator=(AllocRetain const&) = delete;
+  AllocRetain(AllocRetain&& o)               = delete;
+  AllocRetain& operator=(AllocRetain&& o)    = delete;
+};
+
+}  // namespace kvikio
diff --git a/cpp/include/kvikio/cufile_config.hpp b/cpp/include/kvikio/cufile_config.hpp
index bfffb1eb05..b1457880bb 100644
--- a/cpp/include/kvikio/cufile_config.hpp
+++ b/cpp/include/kvikio/cufile_config.hpp
@@ -19,15 +19,17 @@
 #include <filesystem>
 #include <string>
 
+#include <kvikio/utils.hpp>
+
 namespace kvikio {
 namespace detail {
 
-[[nodiscard]] inline const std::string lookup_config_path()
+[[nodiscard]] inline const char* lookup_config_path()
 {
   const char* env = std::getenv("CUFILE_ENV_PATH_JSON");
   if (env != nullptr && std::filesystem::exists(env)) { return env; }
   if (std::filesystem::exists("/etc/cufile.json")) { return "/etc/cufile.json"; }
-  return std::string();
+  return "";
 }
 
 }  // namespace detail
@@ -39,7 +41,7 @@ namespace detail {
  *
  * @return The filepath to the cufile.json file or the empty string if it isn't found.
  */
-[[nodiscard]] inline const std::string config_path()
+[[nodiscard]] KVIKIO_EXPORT inline const std::string& config_path()
 {
   static const std::string ret = detail::lookup_config_path();
   return ret;
diff --git a/cpp/include/kvikio/defaults.hpp b/cpp/include/kvikio/defaults.hpp
index c192763ccd..c812c6e251 100644
--- a/cpp/include/kvikio/defaults.hpp
+++ b/cpp/include/kvikio/defaults.hpp
@@ -21,7 +21,6 @@
 #include <sstream>
 #include <stdexcept>
 #include <string>
-#include <utility>
 
 #include <BS_thread_pool.hpp>
 
@@ -83,11 +82,14 @@ class defaults {
   bool _compat_mode;
   std::size_t _task_size;
   std::size_t _gds_threshold;
+  std::size_t _bounce_buffer_size;
 
   static unsigned int get_num_threads_from_env()
   {
     const int ret = detail::getenv_or("KVIKIO_NTHREADS", 1);
-    if (ret <= 0) { throw std::invalid_argument("KVIKIO_NTHREADS has to be a positive integer"); }
+    if (ret <= 0) {
+      throw std::invalid_argument("KVIKIO_NTHREADS has to be a positive integer greater than zero");
+    }
     return ret;
   }
 
@@ -107,18 +109,28 @@ class defaults {
     {
       const ssize_t env = detail::getenv_or("KVIKIO_TASK_SIZE", 4 * 1024 * 1024);
       if (env <= 0) {
-        throw std::invalid_argument("KVIKIO_TASK_SIZE has to be a positive integer");
+        throw std::invalid_argument(
+          "KVIKIO_TASK_SIZE has to be a positive integer greater than zero");
       }
       _task_size = env;
     }
     // Determine the default value of `gds_threshold`
     {
       const ssize_t env = detail::getenv_or("KVIKIO_GDS_THRESHOLD", 1024 * 1024);
-      if (env <= 0) {
+      if (env < 0) {
         throw std::invalid_argument("KVIKIO_GDS_THRESHOLD has to be a positive integer");
       }
       _gds_threshold = env;
     }
+    // Determine the default value of `bounce_buffer_size`
+    {
+      const ssize_t env = detail::getenv_or("KVIKIO_BOUNCE_BUFFER_SIZE", 16 * 1024 * 1024);
+      if (env <= 0) {
+        throw std::invalid_argument(
+          "KVIKIO_BOUNCE_BUFFER_SIZE has to be a positive integer greater than zero");
+      }
+      _bounce_buffer_size = env;
+    }
   }
 
   static defaults* instance()
@@ -191,7 +203,13 @@ class defaults {
    *
    * @param nthreads The number of threads to use.
    */
-  static void thread_pool_nthreads_reset(unsigned int nthreads) { thread_pool().reset(nthreads); }
+  static void thread_pool_nthreads_reset(unsigned int nthreads)
+  {
+    if (nthreads == 0) {
+      throw std::invalid_argument("number of threads must be a positive integer greater than zero");
+    }
+    thread_pool().reset(nthreads);
+  }
 
   /**
    * @brief Get the default task size used for parallel IO operations.
@@ -208,7 +226,13 @@ class defaults {
    *
    * @param nbytes The default task size in bytes.
    */
-  static void task_size_reset(std::size_t nbytes) { instance()->_task_size = nbytes; }
+  static void task_size_reset(std::size_t nbytes)
+  {
+    if (nbytes == 0) {
+      throw std::invalid_argument("task size must be a positive integer greater than zero");
+    }
+    instance()->_task_size = nbytes;
+  }
 
   /**
    * @brief Get the default GDS threshold, which is the minimum size to use GDS (in bytes).
@@ -228,6 +252,30 @@ class defaults {
    * @param nbytes The default GDS threshold size in bytes.
    */
   static void gds_threshold_reset(std::size_t nbytes) { instance()->_gds_threshold = nbytes; }
+
+  /**
+   * @brief Get the size of the bounce buffer used to stage data in host memory.
+   *
+   * Set the value using `kvikio::default::bounce_buffer_size_reset()` or by setting the
+   * `KVIKIO_BOUNCE_BUFFER_SIZE` environment variable. If not set, the value is 16 MiB.
+   *
+   * @return The bounce buffer size in bytes.
+   */
+  [[nodiscard]] static std::size_t bounce_buffer_size() { return instance()->_bounce_buffer_size; }
+
+  /**
+   * @brief Reset the size of the bounce buffer used to stage data in host memory.
+   *
+   * @param nbytes The bounce buffer size in bytes.
+   */
+  static void bounce_buffer_size_reset(std::size_t nbytes)
+  {
+    if (nbytes == 0) {
+      throw std::invalid_argument(
+        "size of the bounce buffer must be a positive integer greater than zero");
+    }
+    instance()->_bounce_buffer_size = nbytes;
+  }
 };
 
 }  // namespace kvikio
diff --git a/cpp/include/kvikio/file_handle.hpp b/cpp/include/kvikio/file_handle.hpp
index 80779c5282..f84e792489 100644
--- a/cpp/include/kvikio/file_handle.hpp
+++ b/cpp/include/kvikio/file_handle.hpp
@@ -286,7 +286,7 @@ class FileHandle {
    *
    * @return The number of bytes
    */
-  [[nodiscard]] inline std::size_t nbytes() const
+  [[nodiscard]] std::size_t nbytes() const
   {
     if (closed()) { return 0; }
     if (_nbytes == 0) { _nbytes = detail::get_file_size(_fd_direct_off); }
@@ -314,16 +314,26 @@ class FileHandle {
    * @param file_offset Offset in the file to read from.
    * @param devPtr_offset Offset relative to the `devPtr_base` pointer to read into.
    * This parameter should be used only with registered buffers.
+   * @param sync_default_stream Synchronize the CUDA default (null) stream prior to calling cuFile.
+   * Contrary to most of the non-async CUDA API, cuFile does not have the semantic of being ordered
+   * with respect to other non-cuFile work in the default stream. By enabling `sync_default_stream`,
+   * KvikIO will synchronize the default stream and order the operation with respect to other work
+   * in the null stream. When in KvikIO's compatibility mode or when accessing host memory, the
+   * operation is always default stream ordered like the rest of the non-async CUDA API. In this
+   * case, the value of `sync_default_stream` is ignored.
    * @return Size of bytes that were successfully read.
    */
   std::size_t read(void* devPtr_base,
                    std::size_t size,
                    std::size_t file_offset,
-                   std::size_t devPtr_offset)
+                   std::size_t devPtr_offset,
+                   bool sync_default_stream = true)
   {
     if (_compat_mode) {
       return posix_device_read(_fd_direct_off, devPtr_base, size, file_offset, devPtr_offset);
     }
+    if (sync_default_stream) { CUDA_DRIVER_TRY(cudaAPI::instance().StreamSynchronize(nullptr)); }
+
     KVIKIO_NVTX_FUNC_RANGE("cufileRead()", size);
     ssize_t ret = cuFileAPI::instance().Read(
       _handle, devPtr_base, size, convert_size2off(file_offset), convert_size2off(devPtr_offset));
@@ -351,20 +361,30 @@ class FileHandle {
    * `devPtr_base` must remain set to the base address used in the `buffer_register` call.
    * @param size Size in bytes to write.
    * @param file_offset Offset in the file to write from.
-   * @param devPtr_offset Offset relative to the `devPtr_base` pointer to write from.
-   * This parameter should be used only with registered buffers.
+   * @param devPtr_offset Offset relative to the `devPtr_base` pointer to write from. This parameter
+   * should be used only with registered buffers.
+   * @param sync_default_stream Synchronize the CUDA default (null) stream prior to calling cuFile.
+   * Contrary to most of the non-async CUDA API, cuFile does not have the semantic of being ordered
+   * with respect to other non-cuFile work in the default stream. By enabling `sync_default_stream`,
+   * KvikIO will synchronize the default stream and order the operation with respect to other work
+   * in the null stream. When in KvikIO's compatibility mode or when accessing host memory, the
+   * operation is always default stream ordered like the rest of the non-async CUDA API. In this
+   * case, the value of `sync_default_stream` is ignored.
    * @return Size of bytes that were successfully written.
    */
   std::size_t write(const void* devPtr_base,
                     std::size_t size,
                     std::size_t file_offset,
-                    std::size_t devPtr_offset)
+                    std::size_t devPtr_offset,
+                    bool sync_default_stream = true)
   {
     _nbytes = 0;  // Invalidate the computed file size
 
     if (_compat_mode) {
       return posix_device_write(_fd_direct_off, devPtr_base, size, file_offset, devPtr_offset);
     }
+    if (sync_default_stream) { CUDA_DRIVER_TRY(cudaAPI::instance().StreamSynchronize(nullptr)); }
+
     KVIKIO_NVTX_FUNC_RANGE("cufileWrite()", size);
     ssize_t ret = cuFileAPI::instance().Write(
       _handle, devPtr_base, size, convert_size2off(file_offset), convert_size2off(devPtr_offset));
@@ -396,13 +416,21 @@ class FileHandle {
    * @param file_offset Offset in the file to read from.
    * @param task_size Size of each task in bytes.
    * @param gds_threshold Minimum buffer size to use GDS and the thread pool.
+   * @param sync_default_stream Synchronize the CUDA default (null) stream prior to calling cuFile.
+   * Contrary to most of the non-async CUDA API, cuFile does not have the semantic of being ordered
+   * with respect to other non-cuFile work in the default stream. By enabling `sync_default_stream`,
+   * KvikIO will synchronize the default stream and order the operation with respect to other work
+   * in the null stream. When in KvikIO's compatibility mode or when accessing host memory, the
+   * operation is always default stream ordered like the rest of the non-async CUDA API. In this
+   * case, the value of `sync_default_stream` is ignored.
    * @return Future that on completion returns the size of bytes that were successfully read.
    */
   std::future<std::size_t> pread(void* buf,
                                  std::size_t size,
                                  std::size_t file_offset   = 0,
                                  std::size_t task_size     = defaults::task_size(),
-                                 std::size_t gds_threshold = defaults::gds_threshold())
+                                 std::size_t gds_threshold = defaults::gds_threshold(),
+                                 bool sync_default_stream  = true)
   {
     if (is_host_memory(buf)) {
       auto op = [this](void* hostPtr_base,
@@ -427,13 +455,19 @@ class FileHandle {
       return std::async(std::launch::deferred, task);
     }
 
+    // Let's synchronize once instead of in each task.
+    if (sync_default_stream && !_compat_mode) {
+      PushAndPopContext c(ctx);
+      CUDA_DRIVER_TRY(cudaAPI::instance().StreamSynchronize(nullptr));
+    }
+
     // Regular case that use the threadpool and run the tasks in parallel
     auto task = [this, ctx](void* devPtr_base,
                             std::size_t size,
                             std::size_t file_offset,
                             std::size_t devPtr_offset) -> std::size_t {
       PushAndPopContext c(ctx);
-      return read(devPtr_base, size, file_offset, devPtr_offset);
+      return read(devPtr_base, size, file_offset, devPtr_offset, /* sync_default_stream = */ false);
     };
     auto [devPtr_base, base_size, devPtr_offset] = get_alloc_info(buf, &ctx);
     return parallel_io(task, devPtr_base, size, file_offset, task_size, devPtr_offset);
@@ -457,13 +491,21 @@ class FileHandle {
    * @param file_offset Offset in the file to write from.
    * @param task_size Size of each task in bytes.
    * @param gds_threshold Minimum buffer size to use GDS and the thread pool.
+   * @param sync_default_stream Synchronize the CUDA default (null) stream prior to calling cuFile.
+   * Contrary to most of the non-async CUDA API, cuFile does not have the semantic of being ordered
+   * with respect to other non-cuFile work in the default stream. By enabling `sync_default_stream`,
+   * KvikIO will synchronize the default stream and order the operation with respect to other work
+   * in the null stream. When in KvikIO's compatibility mode or when accessing host memory, the
+   * operation is always default stream ordered like the rest of the non-async CUDA API. In this
+   * case, the value of `sync_default_stream` is ignored.
    * @return Future that on completion returns the size of bytes that were successfully written.
    */
   std::future<std::size_t> pwrite(const void* buf,
                                   std::size_t size,
                                   std::size_t file_offset   = 0,
                                   std::size_t task_size     = defaults::task_size(),
-                                  std::size_t gds_threshold = defaults::gds_threshold())
+                                  std::size_t gds_threshold = defaults::gds_threshold(),
+                                  bool sync_default_stream  = true)
   {
     if (is_host_memory(buf)) {
       auto op = [this](const void* hostPtr_base,
@@ -488,13 +530,20 @@ class FileHandle {
       return std::async(std::launch::deferred, task);
     }
 
+    // Let's synchronize once instead of in each task.
+    if (sync_default_stream && !_compat_mode) {
+      PushAndPopContext c(ctx);
+      CUDA_DRIVER_TRY(cudaAPI::instance().StreamSynchronize(nullptr));
+    }
+
     // Regular case that use the threadpool and run the tasks in parallel
     auto op = [this, ctx](const void* devPtr_base,
                           std::size_t size,
                           std::size_t file_offset,
                           std::size_t devPtr_offset) -> std::size_t {
       PushAndPopContext c(ctx);
-      return write(devPtr_base, size, file_offset, devPtr_offset);
+      return write(
+        devPtr_base, size, file_offset, devPtr_offset, /* sync_default_stream = */ false);
     };
     auto [devPtr_base, base_size, devPtr_offset] = get_alloc_info(buf, &ctx);
     return parallel_io(op, devPtr_base, size, file_offset, task_size, devPtr_offset);
@@ -571,8 +620,8 @@ class FileHandle {
    * `devPtr_base` must remain set to the base address used in the `buffer_register` call.
    * @param size Size in bytes to read.
    * @param file_offset Offset in the file to read from.
-   * @param devPtr_offset Offset relative to the `devPtr_base` pointer to read into.
-   * This parameter should be used only with registered buffers.
+   * @param devPtr_offset Offset relative to the `devPtr_base` pointer to read into. This parameter
+   * should be used only with registered buffers.
    * @param stream CUDA stream in which to enqueue the operation. If NULL, make this operation
    * synchronous.
    * @return A future object that must be kept alive until all data has been read to disk e.g.
@@ -663,8 +712,8 @@ class FileHandle {
    * `devPtr_base` must remain set to the base address used in the `buffer_register` call.
    * @param size Size in bytes to write.
    * @param file_offset Offset in the file to write from.
-   * @param devPtr_offset Offset relative to the `devPtr_base` pointer to write from.
-   * This parameter should be used only with registered buffers.
+   * @param devPtr_offset Offset relative to the `devPtr_base` pointer to write from. This parameter
+   * should be used only with registered buffers.
    * @param stream CUDA stream in which to enqueue the operation. If NULL, make this operation
    * synchronous.
    * @return A future object that must be kept alive until all data has been written to disk e.g.
diff --git a/cpp/include/kvikio/posix_io.hpp b/cpp/include/kvikio/posix_io.hpp
index cba399a449..9a28e06eec 100644
--- a/cpp/include/kvikio/posix_io.hpp
+++ b/cpp/include/kvikio/posix_io.hpp
@@ -19,19 +19,15 @@
 #include <cstddef>
 #include <cstdlib>
 #include <map>
-#include <mutex>
-#include <stack>
 #include <thread>
 
-#include <cstring>
+#include <kvikio/bounce_buffer.hpp>
 #include <kvikio/error.hpp>
 #include <kvikio/shim/cuda.hpp>
 #include <kvikio/utils.hpp>
 
 namespace kvikio {
 
-inline constexpr std::size_t posix_bounce_buffer_size = 2 << 23;  // 16 MiB
-
 namespace detail {
 
 /**
@@ -46,16 +42,14 @@ class StreamsByThread {
 
  public:
   StreamsByThread() = default;
-  ~StreamsByThread() noexcept
-  {
-    for (auto& [_, stream] : _streams) {
-      try {
-        CUDA_DRIVER_TRY(cudaAPI::instance().StreamDestroy(stream));
-      } catch (const CUfileException& e) {
-        std::cerr << e.what() << std::endl;
-      }
-    }
-  }
+
+  // Here we intentionally do not destroy in the destructor the CUDA resources
+  // (e.g. CUstream) with static storage duration, but instead let them leak
+  // on program termination. This is to prevent undefined behavior in CUDA. See
+  // <https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#initialization>
+  // This also prevents crash (segmentation fault) if clients call
+  // cuDevicePrimaryCtxReset() or cudaDeviceReset() before program termination.
+  ~StreamsByThread() = default;
 
   static CUstream get(CUcontext ctx, std::thread::id thd_id)
   {
@@ -87,80 +81,6 @@ class StreamsByThread {
   StreamsByThread& operator=(StreamsByThread&& o)    = delete;
 };
 
-/**
- * @brief Singleton class to retain host memory allocations
- *
- * Call `AllocRetain::get` to get an allocation that will be retained when it
- * goes out of scope (RAII). The size of all allocations are `posix_bounce_buffer_size`.
- */
-class AllocRetain {
- private:
-  std::stack<void*> _free_allocs;
-  std::mutex _mutex;
-
- public:
-  class Alloc {
-   private:
-    AllocRetain* _manager;
-    void* _alloc;
-
-   public:
-    Alloc(AllocRetain* manager, void* alloc) : _manager(manager), _alloc{alloc} {}
-    Alloc(const Alloc&)            = delete;
-    Alloc& operator=(Alloc const&) = delete;
-    Alloc(Alloc&& o)               = delete;
-    Alloc& operator=(Alloc&& o)    = delete;
-    ~Alloc() noexcept { _manager->put(_alloc); }
-    void* get() noexcept { return _alloc; }
-  };
-
-  AllocRetain() = default;
-  [[nodiscard]] Alloc get()
-  {
-    const std::lock_guard lock(_mutex);
-    // Check if we have an allocation available
-    if (!_free_allocs.empty()) {
-      void* ret = _free_allocs.top();
-      _free_allocs.pop();
-      return Alloc(this, ret);
-    }
-
-    // If no available allocation, allocate and register a new one
-    void* alloc{};
-    // Allocate page-locked host memory
-    CUDA_DRIVER_TRY(cudaAPI::instance().MemHostAlloc(
-      &alloc, posix_bounce_buffer_size, CU_MEMHOSTREGISTER_PORTABLE));
-    return Alloc(this, alloc);
-  }
-
-  void put(void* alloc)
-  {
-    const std::lock_guard lock(_mutex);
-    _free_allocs.push(alloc);
-  }
-
-  void clear()
-  {
-    const std::lock_guard lock(_mutex);
-    while (!_free_allocs.empty()) {
-      CUDA_DRIVER_TRY(cudaAPI::instance().MemFreeHost(_free_allocs.top()));
-      _free_allocs.pop();
-    }
-  }
-
-  static AllocRetain& instance()
-  {
-    static AllocRetain _instance;
-    return _instance;
-  }
-
-  AllocRetain(const AllocRetain&)            = delete;
-  AllocRetain& operator=(AllocRetain const&) = delete;
-  AllocRetain(AllocRetain&& o)               = delete;
-  AllocRetain& operator=(AllocRetain&& o)    = delete;
-  ~AllocRetain() noexcept                    = default;
-};
-
 /**
  * @brief Read or write host memory to or from disk using POSIX
  *
@@ -230,7 +150,7 @@ std::size_t posix_device_io(int fd,
   CUdeviceptr devPtr      = convert_void2deviceptr(devPtr_base) + devPtr_offset;
   off_t cur_file_offset   = convert_size2off(file_offset);
   off_t byte_remaining    = convert_size2off(size);
-  const off_t chunk_size2 = convert_size2off(posix_bounce_buffer_size);
+  const off_t chunk_size2 = convert_size2off(alloc.size());
 
   // Get a stream for the current CUDA context and thread
   CUstream stream = StreamsByThread::get();
diff --git a/cpp/include/kvikio/utils.hpp b/cpp/include/kvikio/utils.hpp
index 918983a919..3c77c1c853 100644
--- a/cpp/include/kvikio/utils.hpp
+++ b/cpp/include/kvikio/utils.hpp
@@ -21,13 +21,31 @@
 #include <iostream>
 #include <map>
 #include <optional>
+#include <stdexcept>
 #include <tuple>
+#include <type_traits>
 
+#ifdef KVIKIO_CUDA_FOUND
 #include <nvtx3/nvtx3.hpp>
+#endif
 
 #include <kvikio/error.hpp>
 #include <kvikio/shim/cuda.hpp>
 
+// Macros used for defining symbol visibility, only GLIBC is supported.
+// Since KvikIO is header-only, we rely on the linker to disambiguate inline functions
+// that have (or return) static references. To do this, the relevant function must have
+// `__attribute__((visibility("default")))`. If not, then if KvikIO is used in two
+// different DSOs, the function will appear twice, and there will be two static objects.
+// See <https://github.com/rapidsai/kvikio/issues/442>.
+#if (defined(__GNUC__) || defined(__clang__)) && !defined(__MINGW32__) && !defined(__MINGW64__)
+#define KVIKIO_EXPORT __attribute__((visibility("default")))
+#define KVIKIO_HIDDEN __attribute__((visibility("hidden")))
+#else
+#define KVIKIO_EXPORT
+#define KVIKIO_HIDDEN
+#endif
+
 namespace kvikio {
 
 // cuFile defines a page size to 4 KiB
@@ -55,6 +73,29 @@ inline constexpr std::size_t page_size = 4096;
   return reinterpret_cast<CUdeviceptr>(devPtr);
 }
 
+/**
+ * @brief Help function to convert value to 64 bit signed integer
+ */
+template <typename T, std::enable_if_t<std::is_integral_v<T>>* = nullptr>
+[[nodiscard]] std::int64_t convert_to_64bit(T value)
+{
+  if constexpr (std::numeric_limits<T>::max() > std::numeric_limits<std::int64_t>::max()) {
+    if (value > std::numeric_limits<std::int64_t>::max()) {
+      throw std::overflow_error("convert_to_64bit(x): x too large to fit std::int64_t");
+    }
+  }
+  return std::int64_t(value);
+}
+
+/**
+ * @brief Help function to convert value to 64 bit float
+ */
+template <typename T, std::enable_if_t<std::is_floating_point_v<T>>* = nullptr>
+[[nodiscard]] double convert_to_64bit(T value)
+{
+  return double(value);
+}
+
 /**
  * @brief Check if `ptr` points to host memory (as opposed to device memory)
  *
@@ -101,33 +142,6 @@ constexpr bool is_host_memory(const void* ptr) { return true; }
   return ret;
 }
 
-/**
- * @brief RAII wrapper for a CUDA primary context
- */
-class CudaPrimaryContext {
- public:
-  CUdevice dev{};
-  CUcontext ctx{};
-
-  CudaPrimaryContext(int device_ordinal)
-  {
-    CUDA_DRIVER_TRY(cudaAPI::instance().DeviceGet(&dev, device_ordinal));
-    CUDA_DRIVER_TRY(cudaAPI::instance().DevicePrimaryCtxRetain(&ctx, dev));
-  }
-  CudaPrimaryContext(const CudaPrimaryContext&)            = delete;
-  CudaPrimaryContext& operator=(CudaPrimaryContext const&) = delete;
-  CudaPrimaryContext(CudaPrimaryContext&&)                 = delete;
-  CudaPrimaryContext&& operator=(CudaPrimaryContext&&)     = delete;
-  ~CudaPrimaryContext()
-  {
-    try {
-      CUDA_DRIVER_TRY(cudaAPI::instance().DevicePrimaryCtxRelease(dev), CUfileException);
-    } catch (const CUfileException& e) {
-      std::cerr << e.what() << std::endl;
-    }
-  }
-};
-
 /**
  * @brief Given a device ordinal, return the primary context of the device.
  *
@@ -136,11 +150,24 @@ class CudaPrimaryContext {
  * @param ordinal Device ordinal - an integer between 0 and the number of CUDA devices
  * @return Primary CUDA context
  */
-[[nodiscard]] inline CUcontext get_primary_cuda_context(int ordinal)
+[[nodiscard]] KVIKIO_EXPORT inline CUcontext get_primary_cuda_context(int ordinal)
 {
-  static std::map<int, CudaPrimaryContext> _primary_contexts;
-  _primary_contexts.try_emplace(ordinal, ordinal);
-  return _primary_contexts.at(ordinal).ctx;
+  static std::map<int, CUcontext> _cache;
+  static std::mutex _mutex;
+  std::lock_guard const lock(_mutex);
+
+  if (_cache.find(ordinal) == _cache.end()) {
+    CUdevice dev{};
+    CUcontext ctx{};
+    CUDA_DRIVER_TRY(cudaAPI::instance().DeviceGet(&dev, ordinal));
+
+    // Notice, we let the primary context leak at program exit. We do this because `_cache`
+    // is static and we are not allowed to call `cuDevicePrimaryCtxRelease()` after main:
+    // <https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#initialization>
+    CUDA_DRIVER_TRY(cudaAPI::instance().DevicePrimaryCtxRetain(&ctx, dev));
+    _cache.emplace(ordinal, ctx);
+  }
+  return _cache.at(ordinal);
 }
 
 /**
@@ -266,6 +293,7 @@ inline bool is_future_done(const T& future)
   return future.wait_for(std::chrono::seconds(0)) != std::future_status::timeout;
 }
 
+#ifdef KVIKIO_CUDA_FOUND
 /**
  * @brief Tag type for libkvikio's NVTX domain.
  */
@@ -280,10 +308,11 @@ struct libkvikio_domain {
   {                                                           \
     nvtx3::event_attributes                                   \
     {                                                         \
-      msg, nvtx3::payload { val }                             \
+      msg, nvtx3::payload { convert_to_64bit(val) }           \
     }                                                         \
   }
 #define GET_KVIKIO_NVTX_FUNC_RANGE_MACRO(_1, _2, NAME, ...) NAME
+#endif
 
 /**
  * @brief Convenience macro for generating an NVTX range in the `libkvikio` domain
@@ -304,9 +333,15 @@ struct libkvikio_domain {
  * }
  * ```
  */
+#ifdef KVIKIO_CUDA_FOUND
 #define KVIKIO_NVTX_FUNC_RANGE(...)                                  \
   GET_KVIKIO_NVTX_FUNC_RANGE_MACRO(                                  \
     __VA_ARGS__, KVIKIO_NVTX_FUNC_RANGE_2, KVIKIO_NVTX_FUNC_RANGE_1) \
   (__VA_ARGS__)
+#else
+#define KVIKIO_NVTX_FUNC_RANGE(...) \
+  do {                              \
+  } while (0)
+#endif
 
 }  // namespace kvikio
diff --git a/dependencies.yaml b/dependencies.yaml
index 16f059de19..6c6919a250 100644
--- a/dependencies.yaml
+++ b/dependencies.yaml
@@ -17,11 +17,13 @@ files:
       - py_version
       - rapids_build_skbuild
       - run
+      - test_cpp
       - test_python
   test_cpp:
     output: none
     includes:
       - cuda_version
+      - test_cpp
   test_python:
     output: none
     includes:
@@ -98,7 +100,7 @@ dependencies:
     common:
       - output_types: [conda, requirements, pyproject]
         packages:
-          - cmake>=3.26.4,!=3.30.0
+          - &cmake_ver cmake>=3.26.4,!=3.30.0
           - ninja
   build-cpp:
     common:
@@ -139,7 +141,7 @@ dependencies:
     common:
       - output_types: conda
         packages: &libkvikio_packages
-          - libkvikio==24.8.*,>=0.0.0a0
+          - libkvikio==24.10.*,>=0.0.0a0
     specific:
       - output_types: [requirements, pyproject]
         matrices:
@@ -147,12 +149,12 @@ dependencies:
               cuda: "12.*"
               cuda_suffixed: "true"
             packages:
-              - libkvikio-cu12==24.8.*,>=0.0.0a0
+              - libkvikio-cu12==24.10.*,>=0.0.0a0
           - matrix:
               cuda: "11.*"
               cuda_suffixed: "true"
             packages:
-              - libkvikio-cu11==24.8.*,>=0.0.0a0
+              - libkvikio-cu11==24.10.*,>=0.0.0a0
           - {matrix: null, packages: *libkvikio_packages}
   build-py-wrapper:
     common:
@@ -200,7 +202,7 @@ dependencies:
     common:
       - output_types: conda
         packages:
-          - nvcomp==3.0.6
+          - nvcomp==4.0.1
     specific:
       - output_types: conda
         matrices:
@@ -277,10 +279,6 @@ dependencies:
     specific:
       - output_types: conda
         matrices:
-          - matrix:
-              py: "3.9"
-            packages:
-              - python=3.9
           - matrix:
               py: "3.10"
             packages:
@@ -289,9 +287,13 @@ dependencies:
               py: "3.11"
             packages:
               - python=3.11
+          - matrix:
+              py: "3.12"
+            packages:
+              - python=3.12
           - matrix:
             packages:
-              - python>=3.9,<3.12
+              - python>=3.10,<3.13
   rapids_build_skbuild:
     common:
       - output_types: [conda, requirements, pyproject]
@@ -299,19 +301,24 @@ dependencies:
           - rapids-build-backend>=0.3.0,<0.4.0.dev0
       - output_types: conda
         packages:
-          - scikit-build-core>=0.7.0
+          - scikit-build-core>=0.10.0
       - output_types: [requirements, pyproject]
         packages:
-          - scikit-build-core[pyproject]>=0.7.0
+          - scikit-build-core[pyproject]>=0.10.0
   run:
     common:
       - output_types: [conda, requirements, pyproject]
         packages:
-          - numpy>=1.23,<2.0a0
+          - numpy>=1.23,<3.0a0
           - zarr
           # See https://github.com/zarr-developers/numcodecs/pull/475
-          - numcodecs <0.12.0
+          - numcodecs !=0.12.0
           - packaging
+  test_cpp:
+    common:
+      - output_types: conda
+        packages:
+          - *cmake_ver
   test_python:
     common:
       - output_types: [conda, requirements, pyproject]
@@ -329,9 +336,3 @@ dependencies:
           - matrix: # All CUDA 11 versions
             packages:
               - cuda-python>=11.7.1,<12.0a0
-  test_python_legate:
-    common:
-      - output_types: [conda, requirements, pyproject]
-        packages:
-          - *dask
-          - distributed>=2022.05.2
diff --git a/docs/source/conf.py b/docs/source/conf.py
index 089a8033f6..0ac87ceae7 100644
--- a/docs/source/conf.py
+++ b/docs/source/conf.py
@@ -81,6 +81,13 @@
 # The name of the Pygments (syntax highlighting) style to use.
 pygments_style = None
 
+autodoc_default_options = {
+    'members': True,
+    'member-order': 'bysource',
+    'special-members': '__init__',
+    'undoc-members': True,
+    'exclude-members': '__weakref__'
+}
 
 # -- Options for HTML output -------------------------------------------------
 
diff --git a/docs/source/install.rst b/docs/source/install.rst
index 3e7985401c..c286aca2b9 100644
--- a/docs/source/install.rst
+++ b/docs/source/install.rst
@@ -16,9 +16,9 @@ Install the **stable release** from the ``rapidsai`` channel like:
   # Install in existing environment
   mamba install -c rapidsai -c conda-forge kvikio
   # Create new environment (CUDA 11.8)
-  mamba create -n kvikio-env -c rapidsai -c conda-forge python=3.11 cuda-version=11.8 kvikio
+  mamba create -n kvikio-env -c rapidsai -c conda-forge python=3.12 cuda-version=11.8 kvikio
   # Create new environment (CUDA 12.5)
-  mamba create -n kvikio-env -c rapidsai -c conda-forge python=3.11 cuda-version=12.5 kvikio
+  mamba create -n kvikio-env -c rapidsai -c conda-forge python=3.12 cuda-version=12.5 kvikio
 
 Install the **nightly release** from the ``rapidsai-nightly`` channel like:
 
@@ -27,9 +27,9 @@ Install the **nightly release** from the ``rapidsai-nightly`` channel like:
   # Install in existing environment
   mamba install -c rapidsai-nightly -c conda-forge kvikio
   # Create new environment (CUDA 11.8)
-  mamba create -n kvikio-env -c rapidsai-nightly -c conda-forge python=3.11 cuda-version=11.8 kvikio
+  mamba create -n kvikio-env -c rapidsai-nightly -c conda-forge python=3.12 cuda-version=11.8 kvikio
   # Create new environment (CUDA 12.5)
-  mamba create -n kvikio-env -c rapidsai-nightly -c conda-forge python=3.11 cuda-version=12.5 kvikio
+  mamba create -n kvikio-env -c rapidsai-nightly -c conda-forge python=3.12 cuda-version=12.5 kvikio
 
 
 .. note::
@@ -39,7 +39,7 @@ Install the **nightly release** from the ``rapidsai-nightly`` channel like:
 Build from source
 -----------------
 
-In order to setup a development environment run:
+In order to setup a development environment, we recommend Conda:
 
 .. code-block::
 
@@ -48,11 +48,11 @@ In order to setup a development environment run:
   # CUDA 12.5
   mamba env create --name kvikio-dev --file conda/environments/all_cuda-125_arch-x86_64.yaml
 
-To build and install the extension run:
+The Python library depends on the C++ library, thus we build and install both:
 
 .. code-block::
 
-  ./build.sh kvikio
+  ./build.sh libkvikio kvikio
 
 
 One might have to define ``CUDA_HOME`` to the path to the CUDA installation.
diff --git a/docs/source/runtime_settings.rst b/docs/source/runtime_settings.rst
index 2d03eb2f87..631ba0c937 100644
--- a/docs/source/runtime_settings.rst
+++ b/docs/source/runtime_settings.rst
@@ -10,17 +10,29 @@ Set the environment variable ``KVIKIO_COMPAT_MODE`` to enable/disable compatibil
   * when running in Windows Subsystem for Linux (WSL).
   * when ``/run/udev`` isn't readable, which typically happens when running inside a docker image not launched with ``--volume /run/udev:/run/udev:ro``.
 
+This setting can also be controlled by :py:func:`kvikio.defaults.compat_mode`, :py:func:`kvikio.defaults.compat_mode_reset`, and :py:func:`kvikio.defaults.set_compat_mode`.
+
 
 Thread Pool ``KVIKIO_NTHREADS``
 -------------------------------
 KvikIO can use multiple threads for IO automatically. Set the environment variable ``KVIKIO_NTHREADS`` to the number of threads in the thread pool. If not set, the default value is 1.
 
+This setting can also be controlled by :py:func:`kvikio.defaults.get_num_threads`, :py:func:`kvikio.defaults.num_threads_reset`, and :py:func:`kvikio.defaults.set_num_threads`.
 
 Task Size ``KVIKIO_TASK_SIZE``
 ------------------------------
 KvikIO splits parallel IO operations into multiple tasks. Set the environment variable ``KVIKIO_TASK_SIZE`` to the maximum task size (in bytes). If not set, the default value is 4194304 (4 MiB).
 
+This setting can also be controlled by :py:func:`kvikio.defaults.task_size`, :py:func:`kvikio.defaults.task_size_reset`, and :py:func:`kvikio.defaults.set_task_size`.
 
 GDS Threshold ``KVIKIO_GDS_THRESHOLD``
 --------------------------------------
 In order to improve performance of small IO, ``.pread()`` and ``.pwrite()`` implement a shortcut that circumvent the threadpool and use the POSIX backend directly. Set the environment variable ``KVIKIO_GDS_THRESHOLD`` to the minimum size (in bytes) to use GDS. If not set, the default value is 1048576 (1 MiB).
+
+This setting can also be controlled by :py:func:`kvikio.defaults.gds_threshold`, :py:func:`kvikio.defaults.gds_threshold_reset`, and :py:func:`kvikio.defaults.set_gds_threshold`.
+
+Size of the Bounce Buffer ``KVIKIO_BOUNCE_BUFFER_SIZE``
+-------------------------------------------------------
+KvikIO might have to use intermediate host buffers (one per thread) when copying between files and device memory. Set the environment variable ``KVIKIO_BOUNCE_BUFFER_SIZE`` to the size (in bytes) of these "bounce" buffers. If not set, the default value is 16777216 (16 MiB).
+
+This setting can also be controlled by :py:func:`kvikio.defaults.bounce_buffer_size`, :py:func:`kvikio.defaults.bounce_buffer_size_reset`, and :py:func:`kvikio.defaults.set_bounce_buffer_size`.
diff --git a/python/kvikio/CMakeLists.txt b/python/kvikio/CMakeLists.txt
index 9217645637..2b9278fcd1 100644
--- a/python/kvikio/CMakeLists.txt
+++ b/python/kvikio/CMakeLists.txt
@@ -26,33 +26,20 @@ project(
   LANGUAGES CXX CUDA
 )
 
-option(FIND_KVIKIO_CPP
-       "Search for existing KVIKIO C++ installations before defaulting to local files" OFF
-)
-
 # TODO: Should we symlink FindcuFile.cmake into python/cmake? find cuFile
 include(../../cpp/cmake/Modules/FindcuFile.cmake)
 
-if(FIND_KVIKIO_CPP)
-  find_package(KvikIO "${RAPIDS_VERSION}")
-else()
-  set(KvikIO_FOUND OFF)
-endif()
+find_package(KvikIO REQUIRED "${RAPIDS_VERSION}")
 
 find_package(CUDAToolkit REQUIRED)
 
-set(cython_lib_dir kvikio)
-
-if(NOT KvikIO_FOUND)
-  add_subdirectory(../../cpp kvikio-cpp)
-  install(TARGETS kvikio DESTINATION ${cython_lib_dir})
-endif()
-
 include(rapids-cython-core)
 rapids_cython_init()
 
 add_subdirectory(cmake)
 
+set(cython_lib_dir kvikio)
+
 # It would be better to factor nvcomp out into its own wheel. Until that is available, we vendor it
 # here.
 install_aliased_imported_targets(
diff --git a/python/kvikio/LICENSE b/python/kvikio/LICENSE
new file mode 120000
index 0000000000..30cff7403d
--- /dev/null
+++ b/python/kvikio/LICENSE
@@ -0,0 +1,1 @@
+../../LICENSE
\ No newline at end of file
diff --git a/python/kvikio/cmake/thirdparty/get_nvcomp.cmake b/python/kvikio/cmake/thirdparty/get_nvcomp.cmake
index ff39366bff..9361624c07 100644
--- a/python/kvikio/cmake/thirdparty/get_nvcomp.cmake
+++ b/python/kvikio/cmake/thirdparty/get_nvcomp.cmake
@@ -1,5 +1,5 @@
 # =============================================================================
-# Copyright (c) 2021-2023, NVIDIA CORPORATION.
+# Copyright (c) 2021-2024, NVIDIA CORPORATION.
 #
 # Licensed under the Apache License, Version 2.0 (the "License"); you may not use this file except
 # in compliance with the License. You may obtain a copy of the License at
@@ -18,11 +18,7 @@ set(KVIKIO_USE_PROPRIETARY_BINARY ON)
 function(find_and_configure_nvcomp)
 
   include(${rapids-cmake-dir}/cpm/nvcomp.cmake)
-  rapids_cpm_nvcomp(
-    BUILD_EXPORT_SET kvikio-exports
-    INSTALL_EXPORT_SET kvikio-exports
-    USE_PROPRIETARY_BINARY ${KVIKIO_USE_PROPRIETARY_BINARY}
-  )
+  rapids_cpm_nvcomp(USE_PROPRIETARY_BINARY ${KVIKIO_USE_PROPRIETARY_BINARY})
 
   # Per-thread default stream
   if(TARGET nvcomp AND PER_THREAD_DEFAULT_STREAM)
diff --git a/python/kvikio/kvikio/__init__.py b/python/kvikio/kvikio/__init__.py
index 5e2704adfa..883ac9e784 100644
--- a/python/kvikio/kvikio/__init__.py
+++ b/python/kvikio/kvikio/__init__.py
@@ -1,18 +1,12 @@
 # Copyright (c) 2021-2024, NVIDIA CORPORATION. All rights reserved.
 # See file LICENSE for terms.
 
-from ._lib import libkvikio  # type: ignore
-from ._version import __git_commit__, __version__  # noqa: F401
-from .cufile import CuFile  # noqa: F401
+from kvikio._lib import driver_properties  # type: ignore
+from kvikio._version import __git_commit__, __version__
+from kvikio.cufile import CuFile
 
-
-def memory_register(buf) -> None:
-    return libkvikio.memory_register(buf)
-
-
-def memory_deregister(buf) -> None:
-    libkvikio.memory_deregister(buf)
+# TODO: Wrap nicely, maybe as a dataclass?
+DriverProperties = driver_properties.DriverProperties
 
 
-# TODO: Wrap nicely, maybe as a dataclass?
-DriverProperties = libkvikio.DriverProperties
+__all__ = ["__git_commit__", "__version__", "CuFile"]
diff --git a/python/kvikio/kvikio/_lib/CMakeLists.txt b/python/kvikio/kvikio/_lib/CMakeLists.txt
index 04cff87c08..c77d8e3df1 100644
--- a/python/kvikio/kvikio/_lib/CMakeLists.txt
+++ b/python/kvikio/kvikio/_lib/CMakeLists.txt
@@ -1,5 +1,5 @@
 # =============================================================================
-# Copyright (c) 2022-2023, NVIDIA CORPORATION.
+# Copyright (c) 2022-2024, NVIDIA CORPORATION.
 #
 # Licensed under the Apache License, Version 2.0 (the "License"); you may not use this file except
 # in compliance with the License. You may obtain a copy of the License at
@@ -12,8 +12,10 @@
 # the License.
 # =============================================================================
 
-# Set the list of Cython files to build
-set(cython_modules arr.pyx libnvcomp.pyx libnvcomp_ll.pyx libkvikio.pyx)
+# Set the list of Cython files to build, one .so per file
+set(cython_modules arr.pyx buffer.pyx defaults.pyx driver_properties.pyx file_handle.pyx future.pyx
+                   libnvcomp.pyx libnvcomp_ll.pyx
+)
 
 rapids_cython_create_modules(
   CXX
diff --git a/python/kvikio/kvikio/_lib/arr.pxd b/python/kvikio/kvikio/_lib/arr.pxd
index 7b79987c0b..a9d2b7e7f1 100644
--- a/python/kvikio/kvikio/_lib/arr.pxd
+++ b/python/kvikio/kvikio/_lib/arr.pxd
@@ -1,4 +1,4 @@
-# Copyright (c) 2020-2021, NVIDIA CORPORATION. All rights reserved.
+# Copyright (c) 2020-2024, NVIDIA CORPORATION. All rights reserved.
 # See file LICENSE for terms.
 
 # distutils: language = c++
@@ -6,6 +6,7 @@
 
 
 from libc.stdint cimport uintptr_t
+from libcpp.utility cimport pair
 
 
 cdef class Array:
@@ -25,3 +26,8 @@ cdef class Array:
     cpdef bint _f_contiguous(self)
     cpdef bint _contiguous(self)
     cpdef Py_ssize_t _nbytes(self)
+
+
+cdef pair[uintptr_t, size_t] parse_buffer_argument(
+    buf, size, bint accept_host_buffer
+) except *
diff --git a/python/kvikio/kvikio/_lib/arr.pyx b/python/kvikio/kvikio/_lib/arr.pyx
index d8a7b7b5a3..793a414b17 100644
--- a/python/kvikio/kvikio/_lib/arr.pyx
+++ b/python/kvikio/kvikio/_lib/arr.pyx
@@ -1,4 +1,4 @@
-# Copyright (c) 2020-2021, NVIDIA CORPORATION. All rights reserved.
+# Copyright (c) 2020-2024, NVIDIA CORPORATION. All rights reserved.
 # See file LICENSE for terms.
 
 # cython: language_level=3
@@ -294,3 +294,24 @@ cpdef asarray(obj):
         return obj
     else:
         return Array(obj)
+
+
+cdef pair[uintptr_t, size_t] parse_buffer_argument(
+    buf, size, bint accept_host_buffer
+) except *:
+    """Parse `buf` and `size` argument and return a pointer and nbytes"""
+    if not isinstance(buf, Array):
+        buf = Array(buf)
+    cdef Array arr = buf
+    if not arr._contiguous():
+        raise ValueError("Array must be contiguous")
+    if not accept_host_buffer and not arr.cuda:
+        raise ValueError("Non-CUDA buffers not supported")
+    cdef size_t nbytes
+    if size is None:
+        nbytes = arr.nbytes
+    elif size > arr.nbytes:
+        raise ValueError("Size is greater than the size of the buffer")
+    else:
+        nbytes = size
+    return pair[uintptr_t, size_t](arr.ptr, nbytes)
diff --git a/python/kvikio/kvikio/_lib/buffer.pyx b/python/kvikio/kvikio/_lib/buffer.pyx
new file mode 100644
index 0000000000..3b90f09816
--- /dev/null
+++ b/python/kvikio/kvikio/_lib/buffer.pyx
@@ -0,0 +1,35 @@
+# Copyright (c) 2024, NVIDIA CORPORATION. All rights reserved.
+# See file LICENSE for terms.
+
+# distutils: language = c++
+# cython: language_level=3
+
+
+from kvikio._lib.arr cimport Array
+
+
+cdef extern from "<kvikio/buffer.hpp>" nogil:
+    void cpp_memory_register "kvikio::memory_register"(const void* devPtr) except +
+    void cpp_memory_deregister "kvikio::memory_deregister"(const void* devPtr) except +
+
+
+def memory_register(buf) -> None:
+    if not isinstance(buf, Array):
+        buf = Array(buf)
+    cdef Array arr = buf
+    cpp_memory_register(<void*>arr.ptr)
+
+
+def memory_deregister(buf) -> None:
+    if not isinstance(buf, Array):
+        buf = Array(buf)
+    cdef Array arr = buf
+    cpp_memory_deregister(<void*>arr.ptr)
+
+
+cdef extern from "<kvikio/bounce_buffer.hpp>" nogil:
+    size_t cpp_alloc_retain_clear "kvikio::AllocRetain::instance().clear"() except +
+
+
+def bounce_buffer_free() -> int:
+    return cpp_alloc_retain_clear()
diff --git a/python/kvikio/kvikio/_lib/defaults.pyx b/python/kvikio/kvikio/_lib/defaults.pyx
new file mode 100644
index 0000000000..f59cad5cb4
--- /dev/null
+++ b/python/kvikio/kvikio/_lib/defaults.pyx
@@ -0,0 +1,65 @@
+# Copyright (c) 2024, NVIDIA CORPORATION. All rights reserved.
+# See file LICENSE for terms.
+
+# distutils: language = c++
+# cython: language_level=3
+
+from libcpp cimport bool
+
+
+cdef extern from "<kvikio/defaults.hpp>" nogil:
+    bool cpp_compat_mode "kvikio::defaults::compat_mode"() except +
+    void cpp_compat_mode_reset \
+        "kvikio::defaults::compat_mode_reset"(bool enable) except +
+    unsigned int cpp_thread_pool_nthreads \
+        "kvikio::defaults::thread_pool_nthreads"() except +
+    void cpp_thread_pool_nthreads_reset \
+        "kvikio::defaults::thread_pool_nthreads_reset" (unsigned int nthreads) except +
+    size_t cpp_task_size "kvikio::defaults::task_size"() except +
+    void cpp_task_size_reset "kvikio::defaults::task_size_reset"(size_t nbytes) except +
+    size_t cpp_gds_threshold "kvikio::defaults::gds_threshold"() except +
+    void cpp_gds_threshold_reset \
+        "kvikio::defaults::gds_threshold_reset"(size_t nbytes) except +
+    size_t cpp_bounce_buffer_size "kvikio::defaults::bounce_buffer_size"() except +
+    void cpp_bounce_buffer_size_reset \
+        "kvikio::defaults::bounce_buffer_size_reset"(size_t nbytes) except +
+
+
+def compat_mode() -> bool:
+    return cpp_compat_mode()
+
+
+def compat_mode_reset(enable: bool) -> None:
+    cpp_compat_mode_reset(enable)
+
+
+def thread_pool_nthreads() -> int:
+    return cpp_thread_pool_nthreads()
+
+
+def thread_pool_nthreads_reset(nthreads: int) -> None:
+    cpp_thread_pool_nthreads_reset(nthreads)
+
+
+def task_size() -> int:
+    return cpp_task_size()
+
+
+def task_size_reset(nbytes: int) -> None:
+    cpp_task_size_reset(nbytes)
+
+
+def gds_threshold() -> int:
+    return cpp_gds_threshold()
+
+
+def gds_threshold_reset(nbytes: int) -> None:
+    cpp_gds_threshold_reset(nbytes)
+
+
+def bounce_buffer_size() -> int:
+    return cpp_bounce_buffer_size()
+
+
+def bounce_buffer_size_reset(nbytes: int) -> None:
+    cpp_bounce_buffer_size_reset(nbytes)
diff --git a/python/kvikio/kvikio/_lib/driver_properties.pyx b/python/kvikio/kvikio/_lib/driver_properties.pyx
new file mode 100644
index 0000000000..674ef14cde
--- /dev/null
+++ b/python/kvikio/kvikio/_lib/driver_properties.pyx
@@ -0,0 +1,85 @@
+# Copyright (c) 2024, NVIDIA CORPORATION. All rights reserved.
+# See file LICENSE for terms.
+
+# distutils: language = c++
+# cython: language_level=3
+
+
+from libcpp cimport bool
+
+
+cdef extern from "<kvikio/driver.hpp>" nogil:
+    cdef cppclass cpp_DriverProperties "kvikio::DriverProperties":
+        cpp_DriverProperties() except +
+        bool is_gds_available() except +
+        unsigned int get_nvfs_major_version() except +
+        unsigned int get_nvfs_minor_version() except +
+        bool get_nvfs_allow_compat_mode() except +
+        bool get_nvfs_poll_mode() except +
+        size_t get_nvfs_poll_thresh_size() except +
+        void set_nvfs_poll_mode(bool enable) except +
+        void set_nvfs_poll_thresh_size(size_t size_in_kb) except +
+        size_t get_max_device_cache_size() except +
+        void set_max_device_cache_size(size_t size_in_kb) except +
+        size_t get_per_buffer_cache_size() except +
+        size_t get_max_pinned_memory_size() except +
+        void set_max_pinned_memory_size(size_t size_in_kb) except +
+
+
+cdef class DriverProperties:
+    cdef cpp_DriverProperties _handle
+
+    @property
+    def is_gds_available(self) -> bool:
+        try:
+            return self._handle.is_gds_available()
+        except RuntimeError:
+            return False
+
+    @property
+    def major_version(self) -> bool:
+        return self._handle.get_nvfs_major_version()
+
+    @property
+    def minor_version(self) -> bool:
+        return self._handle.get_nvfs_minor_version()
+
+    @property
+    def allow_compat_mode(self) -> bool:
+        return self._handle.get_nvfs_allow_compat_mode()
+
+    @property
+    def poll_mode(self) -> bool:
+        return self._handle.get_nvfs_poll_mode()
+
+    @poll_mode.setter
+    def poll_mode(self, enable: bool) -> None:
+        self._handle.set_nvfs_poll_mode(enable)
+
+    @property
+    def poll_thresh_size(self) -> int:
+        return self._handle.get_nvfs_poll_thresh_size()
+
+    @poll_thresh_size.setter
+    def poll_thresh_size(self, size_in_kb: int) -> None:
+        self._handle.set_nvfs_poll_thresh_size(size_in_kb)
+
+    @property
+    def max_device_cache_size(self) -> int:
+        return self._handle.get_max_device_cache_size()
+
+    @max_device_cache_size.setter
+    def max_device_cache_size(self, size_in_kb: int) -> None:
+        self._handle.set_max_device_cache_size(size_in_kb)
+
+    @property
+    def per_buffer_cache_size(self) -> int:
+        return self._handle.get_per_buffer_cache_size()
+
+    @property
+    def max_pinned_memory_size(self) -> int:
+        return self._handle.get_max_pinned_memory_size()
+
+    @max_pinned_memory_size.setter
+    def max_pinned_memory_size(self, size_in_kb: int) -> None:
+        self._handle.set_max_pinned_memory_size(size_in_kb)
diff --git a/python/kvikio/kvikio/_lib/file_handle.pyx b/python/kvikio/kvikio/_lib/file_handle.pyx
new file mode 100644
index 0000000000..7a8de368ef
--- /dev/null
+++ b/python/kvikio/kvikio/_lib/file_handle.pyx
@@ -0,0 +1,177 @@
+# Copyright (c) 2021-2024, NVIDIA CORPORATION. All rights reserved.
+# See file LICENSE for terms.
+
+# distutils: language = c++
+# cython: language_level=3
+
+import pathlib
+from typing import Optional
+
+from posix cimport fcntl
+
+from libc.stdint cimport uintptr_t
+from libcpp cimport bool
+from libcpp.string cimport string
+from libcpp.utility cimport move, pair
+
+from kvikio._lib.arr cimport parse_buffer_argument
+from kvikio._lib.future cimport (
+    IOFuture,
+    IOFutureStream,
+    _wrap_io_future,
+    _wrap_stream_future,
+    cpp_StreamFuture,
+    future,
+)
+
+from kvikio._lib import defaults
+
+
+cdef extern from "cuda.h":
+    ctypedef void* CUstream
+
+
+cdef extern from "<kvikio/file_handle.hpp>" namespace "kvikio" nogil:
+    cdef cppclass FileHandle:
+        FileHandle() except +
+        FileHandle(int fd) except +
+        FileHandle(
+            string file_path,
+            string flags,
+        ) except +
+        FileHandle(
+            string file_path,
+            string flags,
+            fcntl.mode_t mode
+        ) except +
+        void close()
+        bool closed()
+        int fd()
+        int fd_open_flags() except +
+        future[size_t] pread(
+            void* devPtr,
+            size_t size,
+            size_t file_offset,
+            size_t task_size
+        ) except +
+        future[size_t] pwrite(
+            void* devPtr,
+            size_t size,
+            size_t file_offset,
+            size_t task_size
+        ) except +
+        size_t read(
+            void* devPtr_base,
+            size_t size,
+            size_t file_offset,
+            size_t devPtr_offset
+        ) except +
+        size_t write(
+            void* devPtr_base,
+            size_t size,
+            size_t file_offset,
+            size_t devPtr_offset
+        ) except +
+        cpp_StreamFuture read_async(
+            void* devPtr_base,
+            size_t size,
+            size_t file_offset,
+            size_t devPtr_offset,
+            CUstream stream
+        ) except +
+        cpp_StreamFuture write_async(
+            void* devPtr_base,
+            size_t size,
+            size_t file_offset,
+            size_t devPtr_offset,
+            CUstream stream
+        ) except +
+
+
+cdef class CuFile:
+    """File handle for GPUDirect Storage (GDS)"""
+    cdef FileHandle _handle
+
+    def __init__(self, file_path, flags="r"):
+        self._handle = move(
+            FileHandle(
+                str.encode(str(pathlib.Path(file_path))),
+                str.encode(str(flags))
+            )
+        )
+
+    def close(self) -> None:
+        self._handle.close()
+
+    def closed(self) -> bool:
+        return self._handle.closed()
+
+    def fileno(self) -> int:
+        return self._handle.fd()
+
+    def open_flags(self) -> int:
+        return self._handle.fd_open_flags()
+
+    def pread(self, buf, size: Optional[int], file_offset: int, task_size) -> IOFuture:
+        cdef pair[uintptr_t, size_t] info = parse_buffer_argument(buf, size, True)
+        return _wrap_io_future(
+            self._handle.pread(
+                <void*>info.first,
+                info.second,
+                file_offset,
+                task_size if task_size else defaults.task_size()
+            )
+        )
+
+    def pwrite(self, buf, size: Optional[int], file_offset: int, task_size) -> IOFuture:
+        cdef pair[uintptr_t, size_t] info = parse_buffer_argument(buf, size, True)
+        return _wrap_io_future(
+            self._handle.pwrite(
+                <void*>info.first,
+                info.second,
+                file_offset,
+                task_size if task_size else defaults.task_size()
+            )
+        )
+
+    def read(self, buf, size: Optional[int], file_offset: int, dev_offset: int) -> int:
+        cdef pair[uintptr_t, size_t] info = parse_buffer_argument(buf, size, False)
+        return self._handle.read(
+            <void*>info.first,
+            info.second,
+            file_offset,
+            dev_offset,
+        )
+
+    def write(self, buf, size: Optional[int], file_offset: int, dev_offset: int) -> int:
+        cdef pair[uintptr_t, size_t] info = parse_buffer_argument(buf, size, False)
+        return self._handle.write(
+            <void*>info.first,
+            info.second,
+            file_offset,
+            dev_offset,
+        )
+
+    def read_async(self, buf, size: Optional[int], file_offset: int, dev_offset: int,
+                   st: uintptr_t) -> IOFutureStream:
+        stream = <CUstream>st
+        cdef pair[uintptr_t, size_t] info = parse_buffer_argument(buf, size, False)
+        return _wrap_stream_future(self._handle.read_async(
+            <void*>info.first,
+            info.second,
+            file_offset,
+            dev_offset,
+            stream,
+        ))
+
+    def write_async(self, buf, size: Optional[int], file_offset: int, dev_offset: int,
+                    st: uintptr_t) -> IOFutureStream:
+        stream = <CUstream>st
+        cdef pair[uintptr_t, size_t] info = parse_buffer_argument(buf, size, False)
+        return _wrap_stream_future(self._handle.write_async(
+            <void*>info.first,
+            info.second,
+            file_offset,
+            dev_offset,
+            stream,
+        ))
diff --git a/python/kvikio/kvikio/_lib/future.pxd b/python/kvikio/kvikio/_lib/future.pxd
new file mode 100644
index 0000000000..4d564b37a9
--- /dev/null
+++ b/python/kvikio/kvikio/_lib/future.pxd
@@ -0,0 +1,30 @@
+# Copyright (c) 2024, NVIDIA CORPORATION. All rights reserved.
+# See file LICENSE for terms.
+
+# distutils: language = c++
+# cython: language_level=3
+
+
+# Cython doesn't have bindings to std::future, so we make them ourself.
+# See <https://github.com/cython/cython/issues/6336>
+cdef extern from "<future>" namespace "std" nogil:
+    cdef cppclass future[T]:
+        future() except +
+        T get() except +
+
+
+cdef extern from "<kvikio/stream.hpp>" nogil:
+    cdef cppclass cpp_StreamFuture "kvikio::StreamFuture":
+        cpp_StreamFuture() except +
+        size_t check_bytes_done() except +
+
+
+cdef class IOFutureStream:
+    cdef cpp_StreamFuture _handle
+
+cdef IOFutureStream _wrap_stream_future(cpp_StreamFuture &fut)
+
+cdef class IOFuture:
+    cdef future[size_t] _handle
+
+cdef IOFuture _wrap_io_future(future[size_t] &fut)
diff --git a/python/kvikio/kvikio/_lib/future.pyx b/python/kvikio/kvikio/_lib/future.pyx
new file mode 100644
index 0000000000..da6ab308dc
--- /dev/null
+++ b/python/kvikio/kvikio/_lib/future.pyx
@@ -0,0 +1,44 @@
+# Copyright (c) 2024, NVIDIA CORPORATION. All rights reserved.
+# See file LICENSE for terms.
+
+# distutils: language = c++
+# cython: language_level=3
+
+from libcpp cimport bool
+from libcpp.utility cimport move
+
+
+cdef extern from "<kvikio/utils.hpp>" namespace "kvikio" nogil:
+    bool is_future_done[T](const T& future) except +
+
+
+cdef class IOFutureStream:
+    """Wrap a C++ StreamFuture in a Python object"""
+    def check_bytes_done(self) -> int:
+        return self._handle.check_bytes_done()
+
+
+cdef IOFutureStream _wrap_stream_future(cpp_StreamFuture &fut):
+    """Wrap a C++ future (of a `size_t`) in a `IOFuture` instance"""
+    ret = IOFutureStream()
+    ret._handle = move(fut)
+    return ret
+
+
+cdef class IOFuture:
+    """C++ future for CuFile reads and writes"""
+    def get(self) -> int:
+        cdef size_t ret
+        with nogil:
+            ret = self._handle.get()
+        return ret
+
+    def done(self) -> bool:
+        return is_future_done(self._handle)
+
+
+cdef IOFuture _wrap_io_future(future[size_t] &fut):
+    """Wrap a C++ future (of a `size_t`) in a `IOFuture` instance"""
+    ret = IOFuture()
+    ret._handle = move(fut)
+    return ret
diff --git a/python/kvikio/kvikio/_lib/kvikio_cxx_api.pxd b/python/kvikio/kvikio/_lib/kvikio_cxx_api.pxd
deleted file mode 100644
index afb644f7e6..0000000000
--- a/python/kvikio/kvikio/_lib/kvikio_cxx_api.pxd
+++ /dev/null
@@ -1,124 +0,0 @@
-# Copyright (c) 2021-2024, NVIDIA CORPORATION. All rights reserved.
-# See file LICENSE for terms.
-
-# distutils: language = c++
-# cython: language_level=3
-
-from posix cimport fcntl
-
-from libcpp cimport bool
-from libcpp.string cimport string
-from libcpp.utility cimport pair
-from libcpp.vector cimport vector
-
-
-cdef extern from "cuda.h":
-    ctypedef void* CUstream
-
-
-cdef extern from "<future>" namespace "std" nogil:
-    cdef cppclass future[T]:
-        future() except +
-        T get() except +
-
-
-cdef extern from "<kvikio/stream.hpp>" namespace "kvikio" nogil:
-    cdef cppclass StreamFuture:
-        StreamFuture() except +
-        StreamFuture(StreamFuture&&) except +
-        size_t check_bytes_done() except +
-
-
-cdef extern from "<kvikio/utils.hpp>" namespace "kvikio" nogil:
-    bool is_future_done[T](const T& future) except +
-
-
-cdef extern from "<kvikio/driver.hpp>" namespace "kvikio" nogil:
-    cdef cppclass DriverProperties:
-        DriverProperties() except +
-        bool is_gds_available() except +
-        unsigned int get_nvfs_major_version() except +
-        unsigned int get_nvfs_minor_version() except +
-        bool get_nvfs_allow_compat_mode() except +
-        bool get_nvfs_poll_mode() except +
-        size_t get_nvfs_poll_thresh_size() except +
-        void set_nvfs_poll_mode(bool enable) except +
-        void set_nvfs_poll_thresh_size(size_t size_in_kb) except +
-        size_t get_max_device_cache_size() except +
-        void set_max_device_cache_size(size_t size_in_kb) except +
-        size_t get_per_buffer_cache_size() except +
-        size_t get_max_pinned_memory_size() except +
-        void set_max_pinned_memory_size(size_t size_in_kb) except +
-
-
-cdef extern from "<kvikio/buffer.hpp>" namespace "kvikio" nogil:
-    void memory_register(const void* devPtr) except +
-    void memory_deregister(const void* devPtr) except +
-
-
-cdef extern from "<kvikio/defaults.hpp>" namespace "kvikio::defaults" nogil:
-    bool compat_mode() except +
-    void compat_mode_reset(bool enable) except +
-    unsigned int thread_pool_nthreads() except +
-    void thread_pool_nthreads_reset(unsigned int nthreads) except +
-    size_t task_size() except +
-    void task_size_reset(size_t nbytes) except +
-    size_t gds_threshold() except +
-    void gds_threshold_reset(size_t nbytes) except +
-
-
-cdef extern from "<kvikio/file_handle.hpp>" namespace "kvikio" nogil:
-    cdef cppclass FileHandle:
-        FileHandle() except +
-        FileHandle(int fd) except +
-        FileHandle(
-            string file_path,
-            string flags,
-        ) except +
-        FileHandle(
-            string file_path,
-            string flags,
-            fcntl.mode_t mode
-        ) except +
-        void close()
-        bool closed()
-        int fd()
-        int fd_open_flags() except +
-        future[size_t] pread(
-            void* devPtr,
-            size_t size,
-            size_t file_offset,
-            size_t task_size
-        ) except +
-        future[size_t] pwrite(
-            void* devPtr,
-            size_t size,
-            size_t file_offset,
-            size_t task_size
-        ) except +
-        size_t read(
-            void* devPtr_base,
-            size_t size,
-            size_t file_offset,
-            size_t devPtr_offset
-        ) except +
-        size_t write(
-            void* devPtr_base,
-            size_t size,
-            size_t file_offset,
-            size_t devPtr_offset
-        ) except +
-        StreamFuture read_async(
-            void* devPtr_base,
-            size_t size,
-            size_t file_offset,
-            size_t devPtr_offset,
-            CUstream stream
-        ) except +
-        StreamFuture write_async(
-            void* devPtr_base,
-            size_t size,
-            size_t file_offset,
-            size_t devPtr_offset,
-            CUstream stream
-        ) except +
diff --git a/python/kvikio/kvikio/_lib/libkvikio.pyx b/python/kvikio/kvikio/_lib/libkvikio.pyx
deleted file mode 100644
index c9c539877f..0000000000
--- a/python/kvikio/kvikio/_lib/libkvikio.pyx
+++ /dev/null
@@ -1,264 +0,0 @@
-# Copyright (c) 2021-2024, NVIDIA CORPORATION. All rights reserved.
-# See file LICENSE for terms.
-
-# distutils: language = c++
-# cython: language_level=3
-
-import pathlib
-from typing import Optional
-
-from libc.stdint cimport uintptr_t
-from libcpp.utility cimport move, pair
-
-from . cimport kvikio_cxx_api
-from .arr cimport Array
-from .kvikio_cxx_api cimport CUstream, FileHandle, StreamFuture, future, is_future_done
-
-
-cdef class IOFutureStream:
-    """Wrap a C++ StreamFuture in a Python object"""
-    cdef StreamFuture _handle
-
-    def check_bytes_done(self) -> int:
-        return self._handle.check_bytes_done()
-
-
-cdef IOFutureStream _wrap_stream_future(StreamFuture &fut):
-    """Wrap a C++ future (of a `size_t`) in a `IOFuture` instance"""
-    ret = IOFutureStream()
-    ret._handle = move(fut)
-    return ret
-
-
-cdef class IOFuture:
-    """C++ future for CuFile reads and writes"""
-    cdef future[size_t] _handle
-
-    def get(self) -> int:
-        cdef size_t ret
-        with nogil:
-            ret = self._handle.get()
-        return ret
-
-    def done(self) -> bool:
-        return is_future_done(self._handle)
-
-
-cdef IOFuture _wrap_io_future(future[size_t] &fut):
-    """Wrap a C++ future (of a `size_t`) in a `IOFuture` instance"""
-    ret = IOFuture()
-    ret._handle = move(fut)
-    return ret
-
-
-def memory_register(buf) -> None:
-    if not isinstance(buf, Array):
-        buf = Array(buf)
-    cdef Array arr = buf
-    kvikio_cxx_api.memory_register(<void*>arr.ptr)
-
-
-def memory_deregister(buf) -> None:
-    if not isinstance(buf, Array):
-        buf = Array(buf)
-    cdef Array arr = buf
-    kvikio_cxx_api.memory_deregister(<void*>arr.ptr)
-
-
-def compat_mode() -> bool:
-    return kvikio_cxx_api.compat_mode()
-
-
-def compat_mode_reset(enable: bool) -> None:
-    kvikio_cxx_api.compat_mode_reset(enable)
-
-
-def thread_pool_nthreads() -> int:
-    return kvikio_cxx_api.thread_pool_nthreads()
-
-
-def thread_pool_nthreads_reset(nthreads: int) -> None:
-    kvikio_cxx_api.thread_pool_nthreads_reset(nthreads)
-
-
-def task_size() -> int:
-    return kvikio_cxx_api.task_size()
-
-
-def task_size_reset(nbytes: int) -> None:
-    kvikio_cxx_api.task_size_reset(nbytes)
-
-
-def gds_threshold() -> int:
-    return kvikio_cxx_api.gds_threshold()
-
-
-def gds_threshold_reset(nbytes: int) -> None:
-    kvikio_cxx_api.gds_threshold_reset(nbytes)
-
-
-cdef pair[uintptr_t, size_t] _parse_buffer(buf, size, bint accept_host_buffer) except *:
-    """Parse `buf` and `size` argument and return a pointer and nbytes"""
-    if not isinstance(buf, Array):
-        buf = Array(buf)
-    cdef Array arr = buf
-    if not arr._contiguous():
-        raise ValueError("Array must be contiguous")
-    if not accept_host_buffer and not arr.cuda:
-        raise ValueError("Non-CUDA buffers not supported")
-    cdef size_t nbytes
-    if size is None:
-        nbytes = arr.nbytes
-    elif size > arr.nbytes:
-        raise ValueError("Size is greater than the size of the buffer")
-    else:
-        nbytes = size
-    return pair[uintptr_t, size_t](arr.ptr, nbytes)
-
-
-cdef class CuFile:
-    """ File handle for GPUDirect Storage (GDS) """
-    cdef FileHandle _handle
-
-    def __init__(self, file_path, flags="r"):
-        self._handle = move(
-            FileHandle(
-                str.encode(str(pathlib.Path(file_path))),
-                str.encode(str(flags))
-            )
-        )
-
-    def close(self) -> None:
-        self._handle.close()
-
-    def closed(self) -> bool:
-        return self._handle.closed()
-
-    def fileno(self) -> int:
-        return self._handle.fd()
-
-    def open_flags(self) -> int:
-        return self._handle.fd_open_flags()
-
-    def pread(self, buf, size: Optional[int], file_offset: int, task_size) -> IOFuture:
-        cdef pair[uintptr_t, size_t] info = _parse_buffer(buf, size, True)
-        return _wrap_io_future(
-            self._handle.pread(
-                <void*>info.first,
-                info.second,
-                file_offset,
-                task_size if task_size else kvikio_cxx_api.task_size()
-            )
-        )
-
-    def pwrite(self, buf, size: Optional[int], file_offset: int, task_size) -> IOFuture:
-        cdef pair[uintptr_t, size_t] info = _parse_buffer(buf, size, True)
-        return _wrap_io_future(
-            self._handle.pwrite(
-                <void*>info.first,
-                info.second,
-                file_offset,
-                task_size if task_size else kvikio_cxx_api.task_size()
-            )
-        )
-
-    def read(self, buf, size: Optional[int], file_offset: int, dev_offset: int) -> int:
-        cdef pair[uintptr_t, size_t] info = _parse_buffer(buf, size, False)
-        return self._handle.read(
-            <void*>info.first,
-            info.second,
-            file_offset,
-            dev_offset,
-        )
-
-    def write(self, buf, size: Optional[int], file_offset: int, dev_offset: int) -> int:
-        cdef pair[uintptr_t, size_t] info = _parse_buffer(buf, size, False)
-        return self._handle.write(
-            <void*>info.first,
-            info.second,
-            file_offset,
-            dev_offset,
-        )
-
-    def read_async(self, buf, size: Optional[int], file_offset: int, dev_offset: int,
-                   st: uintptr_t) -> IOFutureStream:
-        stream = <CUstream>st
-        cdef pair[uintptr_t, size_t] info = _parse_buffer(buf, size, False)
-        return _wrap_stream_future(self._handle.read_async(
-            <void*>info.first,
-            info.second,
-            file_offset,
-            dev_offset,
-            stream,
-        ))
-
-    def write_async(self, buf, size: Optional[int], file_offset: int, dev_offset: int,
-                    st: uintptr_t) -> IOFutureStream:
-        stream = <CUstream>st
-        cdef pair[uintptr_t, size_t] info = _parse_buffer(buf, size, False)
-        return _wrap_stream_future(self._handle.write_async(
-            <void*>info.first,
-            info.second,
-            file_offset,
-            dev_offset,
-            stream,
-        ))
-
-
-cdef class DriverProperties:
-    cdef kvikio_cxx_api.DriverProperties _handle
-
-    @property
-    def is_gds_available(self) -> bool:
-        try:
-            return self._handle.is_gds_available()
-        except RuntimeError:
-            return False
-
-    @property
-    def major_version(self) -> bool:
-        return self._handle.get_nvfs_major_version()
-
-    @property
-    def minor_version(self) -> bool:
-        return self._handle.get_nvfs_minor_version()
-
-    @property
-    def allow_compat_mode(self) -> bool:
-        return self._handle.get_nvfs_allow_compat_mode()
-
-    @property
-    def poll_mode(self) -> bool:
-        return self._handle.get_nvfs_poll_mode()
-
-    @poll_mode.setter
-    def poll_mode(self, enable: bool) -> None:
-        self._handle.set_nvfs_poll_mode(enable)
-
-    @property
-    def poll_thresh_size(self) -> int:
-        return self._handle.get_nvfs_poll_thresh_size()
-
-    @poll_thresh_size.setter
-    def poll_thresh_size(self, size_in_kb: int) -> None:
-        self._handle.set_nvfs_poll_thresh_size(size_in_kb)
-
-    @property
-    def max_device_cache_size(self) -> int:
-        return self._handle.get_max_device_cache_size()
-
-    @max_device_cache_size.setter
-    def max_device_cache_size(self, size_in_kb: int) -> None:
-        self._handle.set_max_device_cache_size(size_in_kb)
-
-    @property
-    def per_buffer_cache_size(self) -> int:
-        return self._handle.get_per_buffer_cache_size()
-
-    @property
-    def max_pinned_memory_size(self) -> int:
-        return self._handle.get_max_pinned_memory_size()
-
-    @max_pinned_memory_size.setter
-    def max_pinned_memory_size(self, size_in_kb: int) -> None:
-        self._handle.set_max_pinned_memory_size(size_in_kb)
diff --git a/python/kvikio/kvikio/_lib/libnvcomp_ll.pyx b/python/kvikio/kvikio/_lib/libnvcomp_ll.pyx
index 0e3fdabe4e..46c7b399a9 100644
--- a/python/kvikio/kvikio/_lib/libnvcomp_ll.pyx
+++ b/python/kvikio/kvikio/_lib/libnvcomp_ll.pyx
@@ -357,6 +357,7 @@ from kvikio._lib.nvcomp_ll_cxx_api cimport (
     nvcompBatchedLZ4CompressGetTempSize,
     nvcompBatchedLZ4DecompressAsync,
     nvcompBatchedLZ4DecompressGetTempSize,
+    nvcompBatchedLZ4DefaultOpts,
     nvcompBatchedLZ4GetDecompressSizeAsync,
     nvcompBatchedLZ4Opts_t,
 )
@@ -371,20 +372,24 @@ class nvCompBatchAlgorithmLZ4(nvCompBatchAlgorithm):
 
     HEADER_SIZE_BYTES: size_t = sizeof(uint32_t)
 
-    def __init__(self, data_type: int = 0, has_header: bool = True):
+    def __init__(self, data_type: int = None, has_header: bool = True):
         """Initialize the codec.
 
         Parameters
         ----------
-        data_type: int
-            Source data type.
+        data_type: int or None
+            Source data type. If None, uses nvcomp default options.
         has_header: bool
             Whether the compressed data has a header.
             This enables data compatibility between numcodecs LZ4 codec,
             which has the header and nvCOMP LZ4 codec which does not
             require the header.
         """
-        self.options = nvcompBatchedLZ4Opts_t(data_type)
+        if data_type is None:
+            self.options = nvcompBatchedLZ4DefaultOpts
+        else:
+            self.options = nvcompBatchedLZ4Opts_t(data_type)
+
         self.has_header = has_header
 
         # Note on LZ4 header structure: numcodecs LZ4 codec prepends
@@ -621,6 +626,7 @@ from kvikio._lib.nvcomp_ll_cxx_api cimport (
     nvcompBatchedGdeflateCompressGetTempSize,
     nvcompBatchedGdeflateDecompressAsync,
     nvcompBatchedGdeflateDecompressGetTempSize,
+    nvcompBatchedGdeflateDefaultOpts,
     nvcompBatchedGdeflateGetDecompressSizeAsync,
     nvcompBatchedGdeflateOpts_t,
 )
@@ -633,8 +639,11 @@ class nvCompBatchAlgorithmGdeflate(nvCompBatchAlgorithm):
 
     options: nvcompBatchedGdeflateOpts_t
 
-    def __init__(self, algo: int = 0):
-        self.options = nvcompBatchedGdeflateOpts_t(algo)
+    def __init__(self, algo: int = None):
+        if algo is None:
+            self.options = nvcompBatchedGdeflateDefaultOpts
+        else:
+            self.options = nvcompBatchedGdeflateOpts_t(algo)
 
     def _get_comp_temp_size(
         self,
@@ -756,6 +765,7 @@ from kvikio._lib.nvcomp_ll_cxx_api cimport (
     nvcompBatchedZstdCompressGetTempSize,
     nvcompBatchedZstdDecompressAsync,
     nvcompBatchedZstdDecompressGetTempSize,
+    nvcompBatchedZstdDefaultOpts,
     nvcompBatchedZstdGetDecompressSizeAsync,
     nvcompBatchedZstdOpts_t,
 )
@@ -769,7 +779,7 @@ class nvCompBatchAlgorithmZstd(nvCompBatchAlgorithm):
     options: nvcompBatchedZstdOpts_t
 
     def __init__(self):
-        self.options = nvcompBatchedZstdOpts_t(0)
+        self.options = nvcompBatchedZstdDefaultOpts
 
     def _get_comp_temp_size(
         self,
@@ -891,6 +901,7 @@ from kvikio._lib.nvcomp_ll_cxx_api cimport (
     nvcompBatchedSnappyCompressGetTempSize,
     nvcompBatchedSnappyDecompressAsync,
     nvcompBatchedSnappyDecompressGetTempSize,
+    nvcompBatchedSnappyDefaultOpts,
     nvcompBatchedSnappyGetDecompressSizeAsync,
     nvcompBatchedSnappyOpts_t,
 )
@@ -904,7 +915,7 @@ class nvCompBatchAlgorithmSnappy(nvCompBatchAlgorithm):
     options: nvcompBatchedSnappyOpts_t
 
     def __init__(self):
-        self.options = nvcompBatchedSnappyOpts_t(0)
+        self.options = nvcompBatchedSnappyDefaultOpts
 
     def _get_comp_temp_size(
         self,
@@ -1026,6 +1037,7 @@ from kvikio._lib.nvcomp_ll_cxx_api cimport (
     nvcompBatchedDeflateCompressGetTempSize,
     nvcompBatchedDeflateDecompressAsync,
     nvcompBatchedDeflateDecompressGetTempSize,
+    nvcompBatchedDeflateDefaultOpts,
     nvcompBatchedDeflateGetDecompressSizeAsync,
     nvcompBatchedDeflateOpts_t,
 )
@@ -1038,14 +1050,17 @@ class nvCompBatchAlgorithmDeflate(nvCompBatchAlgorithm):
 
     options: nvcompBatchedDeflateOpts_t
 
-    def __init__(self, algo: int = 0):
-        self.options = nvcompBatchedDeflateOpts_t(algo)
+    def __init__(self, algo: int = None):
+        if algo is None:
+            self.options = nvcompBatchedDeflateDefaultOpts
+        else:
+            self.options = nvcompBatchedDeflateOpts_t(algo)
 
     def _get_comp_temp_size(
         self,
         size_t batch_size,
         size_t max_uncompressed_chunk_bytes,
-    ) -> (nvcompStatus_t, size_t):
+    ) -> tuple[nvcompStatus_t, size_t]:
         cdef size_t temp_bytes = 0
 
         err = nvcompBatchedDeflateCompressGetTempSize(
diff --git a/python/kvikio/kvikio/_lib/nvcomp_ll_cxx_api.pxd b/python/kvikio/kvikio/_lib/nvcomp_ll_cxx_api.pxd
index a6cbb6bdd5..6a23eb5cd1 100644
--- a/python/kvikio/kvikio/_lib/nvcomp_ll_cxx_api.pxd
+++ b/python/kvikio/kvikio/_lib/nvcomp_ll_cxx_api.pxd
@@ -47,6 +47,8 @@ cdef extern from "nvcomp/lz4.h" nogil:
     ctypedef struct nvcompBatchedLZ4Opts_t:
         nvcompType_t data_type
 
+    cdef nvcompBatchedLZ4Opts_t nvcompBatchedLZ4DefaultOpts
+
     # Compression API.
     cdef nvcompStatus_t nvcompBatchedLZ4CompressGetTempSize(
         size_t batch_size,
@@ -109,6 +111,8 @@ cdef extern from "nvcomp/gdeflate.h" nogil:
     ctypedef struct nvcompBatchedGdeflateOpts_t:
         int algo
 
+    cdef nvcompBatchedGdeflateOpts_t nvcompBatchedGdeflateDefaultOpts
+
     # Compression API.
     cdef nvcompStatus_t nvcompBatchedGdeflateCompressGetTempSize(
         size_t batch_size,
@@ -171,6 +175,8 @@ cdef extern from "nvcomp/zstd.h" nogil:
     ctypedef struct nvcompBatchedZstdOpts_t:
         int reserved
 
+    cdef nvcompBatchedZstdOpts_t nvcompBatchedZstdDefaultOpts
+
     # Compression API.
     cdef nvcompStatus_t nvcompBatchedZstdCompressGetTempSize(
         size_t batch_size,
@@ -233,6 +239,8 @@ cdef extern from "nvcomp/snappy.h" nogil:
     ctypedef struct nvcompBatchedSnappyOpts_t:
         int reserved
 
+    cdef nvcompBatchedSnappyOpts_t nvcompBatchedSnappyDefaultOpts
+
     # Compression API.
     cdef nvcompStatus_t nvcompBatchedSnappyCompressGetTempSize(
         size_t batch_size,
@@ -296,6 +304,8 @@ cdef extern from "nvcomp/deflate.h" nogil:
     ctypedef struct nvcompBatchedDeflateOpts_t:
         int algo
 
+    cdef nvcompBatchedDeflateOpts_t nvcompBatchedDeflateDefaultOpts
+
     # Compression API.
     cdef nvcompStatus_t nvcompBatchedDeflateCompressGetTempSize(
         size_t batch_size,
diff --git a/python/kvikio/kvikio/benchmarks/single_node_io.py b/python/kvikio/kvikio/benchmarks/single_node_io.py
index 4d47a80791..bca29ef90d 100644
--- a/python/kvikio/kvikio/benchmarks/single_node_io.py
+++ b/python/kvikio/kvikio/benchmarks/single_node_io.py
@@ -14,6 +14,7 @@
 from dask.utils import format_bytes, parse_bytes
 
 import kvikio
+import kvikio.buffer
 import kvikio.defaults
 from kvikio.benchmarks.utils import parse_directory, pprint_sys_info
 
@@ -38,7 +39,7 @@ def run_cufile(args):
     file_path = args.dir / "kvikio-single-file"
     data = create_data(args.nbytes)
     if args.pre_register_buffer:
-        kvikio.memory_register(data)
+        kvikio.buffer.memory_register(data)
 
     # Write
     f = kvikio.CuFile(file_path, flags="w")
@@ -57,7 +58,7 @@ def run_cufile(args):
     assert res == args.nbytes, f"IO mismatch, expected {args.nbytes} got {res}"
 
     if args.pre_register_buffer:
-        kvikio.memory_deregister(data)
+        kvikio.buffer.memory_deregister(data)
 
     return read_time, write_time
 
@@ -73,7 +74,7 @@ def run_cufile_multiple_files_multiple_arrays(args):
     arrays = [create_data(chunksize) for _ in range(args.nthreads)]
     if args.pre_register_buffer:
         for array in arrays:
-            kvikio.memory_register(array)
+            kvikio.buffer.memory_register(array)
 
     # Write
     files = [kvikio.CuFile(file_path % i, flags="w") for i in range(args.nthreads)]
@@ -95,7 +96,7 @@ def run_cufile_multiple_files_multiple_arrays(args):
 
     if args.pre_register_buffer:
         for array in arrays:
-            kvikio.memory_deregister(array)
+            kvikio.buffer.memory_deregister(array)
 
     return read_time, write_time
 
@@ -108,7 +109,7 @@ def run_cufile_multiple_files(args):
     file_path = str(args.dir / "cufile-p-%03d")
     data = create_data(args.nbytes)
     if args.pre_register_buffer:
-        kvikio.memory_register(data)
+        kvikio.buffer.memory_register(data)
 
     # Write
     files = [kvikio.CuFile(file_path % i, flags="w") for i in range(args.nthreads)]
@@ -133,7 +134,7 @@ def run_cufile_multiple_files(args):
     assert res == args.nbytes, f"IO mismatch, expected {args.nbytes} got {res}"
 
     if args.pre_register_buffer:
-        kvikio.memory_deregister(data)
+        kvikio.buffer.memory_deregister(data)
 
     return read_time, write_time
 
@@ -149,7 +150,7 @@ def run_cufile_multiple_arrays(args):
     arrays = [create_data(chunksize) for _ in range(args.nthreads)]
     if args.pre_register_buffer:
         for array in arrays:
-            kvikio.memory_register(array)
+            kvikio.buffer.memory_register(array)
 
     # Write
     f = kvikio.CuFile(file_path, flags="w")
@@ -174,7 +175,7 @@ def run_cufile_multiple_arrays(args):
 
     if args.pre_register_buffer:
         for array in arrays:
-            kvikio.memory_deregister(array)
+            kvikio.buffer.memory_deregister(array)
 
     return read_time, write_time
 
diff --git a/python/kvikio/kvikio/buffer.py b/python/kvikio/kvikio/buffer.py
new file mode 100644
index 0000000000..62bbc754b4
--- /dev/null
+++ b/python/kvikio/kvikio/buffer.py
@@ -0,0 +1,41 @@
+# Copyright (c) 2024, NVIDIA CORPORATION. All rights reserved.
+# See file LICENSE for terms.
+
+from kvikio._lib import buffer  # type: ignore
+
+
+def memory_register(buf) -> None:
+    """Register a device memory allocation with cuFile.
+
+    Warning
+    -------
+    This API is intended for usecases where the memory is used as a streaming
+    buffer that is reused across multiple cuFile IO operations.
+
+    Parameters
+    ----------
+    buf: buffer-like or array-like
+        Device buffer to register .
+    """
+    return buffer.memory_register(buf)
+
+
+def memory_deregister(buf) -> None:
+    """Deregister an already registered device memory from cuFile.
+
+    Parameters
+    ----------
+    buf: buffer-like or array-like
+        Device buffer to deregister .
+    """
+    buffer.memory_deregister(buf)
+
+
+def bounce_buffer_free() -> int:
+    """Free the host allocations used as bounce buffers.
+
+    Returns
+    -------
+    Number of bytes freed.
+    """
+    return buffer.bounce_buffer_free()
diff --git a/python/kvikio/kvikio/cufile.py b/python/kvikio/kvikio/cufile.py
index 7bba7bc566..ead7bc5f7a 100644
--- a/python/kvikio/kvikio/cufile.py
+++ b/python/kvikio/kvikio/cufile.py
@@ -4,7 +4,7 @@
 import pathlib
 from typing import Optional, Union
 
-from ._lib import libkvikio  # type: ignore
+from kvikio._lib import file_handle  # type: ignore
 
 
 class IOFutureStream:
@@ -85,7 +85,7 @@ def __init__(self, file: Union[pathlib.Path, str], flags: str = "r"):
             "a" -> "open for writing, appending to the end of file if it exists"
             "+" -> "open for updating (reading and writing)"
         """
-        self._handle = libkvikio.CuFile(file, flags)
+        self._handle = file_handle.CuFile(file, flags)
 
     def close(self) -> None:
         """Deregister the file and close the file"""
diff --git a/python/kvikio/kvikio/defaults.py b/python/kvikio/kvikio/defaults.py
index c300272ef4..a0ff265873 100644
--- a/python/kvikio/kvikio/defaults.py
+++ b/python/kvikio/kvikio/defaults.py
@@ -1,10 +1,10 @@
-# Copyright (c) 2021-2023, NVIDIA CORPORATION. All rights reserved.
+# Copyright (c) 2021-2024, NVIDIA CORPORATION. All rights reserved.
 # See file LICENSE for terms.
 
 
 import contextlib
 
-from ._lib import libkvikio  # type: ignore
+import kvikio._lib.defaults
 
 
 def compat_mode() -> bool:
@@ -23,12 +23,12 @@ def compat_mode() -> bool:
     - when `/run/udev` isn't readable, which typically happens when running inside
     a docker image not launched with `--volume /run/udev:/run/udev:ro`
 
-    Return
-    ------
+    Returns
+    -------
     bool
         Whether KvikIO is running in compatibility mode or not.
     """
-    return libkvikio.compat_mode()
+    return kvikio._lib.defaults.compat_mode()
 
 
 def compat_mode_reset(enable: bool) -> None:
@@ -41,7 +41,7 @@ def compat_mode_reset(enable: bool) -> None:
     enable : bool
         Set to True to enable and False to disable compatibility mode
     """
-    libkvikio.compat_mode_reset(enable)
+    kvikio._lib.defaults.compat_mode_reset(enable)
 
 
 @contextlib.contextmanager
@@ -68,12 +68,12 @@ def get_num_threads() -> int:
     Set the default value using `num_threads_reset()` or by setting the
     `KVIKIO_NTHREADS` environment variable. If not set, the default value is 1.
 
-    Return
-    ------
+    Returns
+    -------
     nthreads: int
         The number of threads in the current thread pool.
     """
-    return libkvikio.thread_pool_nthreads()
+    return kvikio._lib.defaults.thread_pool_nthreads()
 
 
 def num_threads_reset(nthreads: int) -> None:
@@ -92,7 +92,7 @@ def num_threads_reset(nthreads: int) -> None:
         the `KVIKIO_NTHREADS` environment variable. If not set, the default value
         is 1.
     """
-    libkvikio.thread_pool_nthreads_reset(nthreads)
+    kvikio._lib.defaults.thread_pool_nthreads_reset(nthreads)
 
 
 @contextlib.contextmanager
@@ -119,12 +119,12 @@ def task_size() -> int:
     the `KVIKIO_TASK_SIZE` environment variable. If not set,
     the default value is 4 MiB.
 
-    Return
-    ------
+    Returns
+    -------
     nbytes: int
         The default task size in bytes.
     """
-    return libkvikio.task_size()
+    return kvikio._lib.defaults.task_size()
 
 
 def task_size_reset(nbytes: int) -> None:
@@ -135,7 +135,7 @@ def task_size_reset(nbytes: int) -> None:
     nbytes : int
         The default task size in bytes.
     """
-    libkvikio.task_size_reset(nbytes)
+    kvikio._lib.defaults.task_size_reset(nbytes)
 
 
 @contextlib.contextmanager
@@ -163,15 +163,15 @@ def gds_threshold() -> int:
     backend directly.
 
     Set the default value using `gds_threshold_reset()` or by setting the
-    `KVIKIO_TASK_SIZE` environment variable. If not set, the default value
-    is 1 MiB.
+    `KVIKIO_GDS_THRESHOLD` environment variable. If not set, the default
+    value is 1 MiB.
 
-    Return
-    ------
+    Returns
+    -------
     nbytes : int
         The default GDS threshold size in bytes.
     """
-    return libkvikio.gds_threshold()
+    return kvikio._lib.defaults.gds_threshold()
 
 
 def gds_threshold_reset(nbytes: int) -> None:
@@ -182,7 +182,7 @@ def gds_threshold_reset(nbytes: int) -> None:
     nbytes : int
         The default GDS threshold size in bytes.
     """
-    libkvikio.gds_threshold_reset(nbytes)
+    kvikio._lib.defaults.gds_threshold_reset(nbytes)
 
 
 @contextlib.contextmanager
@@ -200,3 +200,46 @@ def set_gds_threshold(nbytes: int):
         yield
     finally:
         gds_threshold_reset(old_value)
+
+
+def bounce_buffer_size() -> int:
+    """Get the size of the bounce buffer used to stage data in host memory.
+
+    Set the value using `bounce_buffer_size_reset()` or by setting the
+    `KVIKIO_BOUNCE_BUFFER_SIZE` environment variable. If not set, the
+    value is 16 MiB.
+
+    Return
+    ------
+    nbytes : int
+        The bounce buffer size in bytes.
+    """
+    return kvikio._lib.defaults.bounce_buffer_size()
+
+
+def bounce_buffer_size_reset(nbytes: int) -> None:
+    """Reset the size of the bounce buffer used to stage data in host memory.
+
+    Parameters
+    ----------
+    nbytes : int
+        The bounce buffer size in bytes.
+    """
+    kvikio._lib.defaults.bounce_buffer_size_reset(nbytes)
+
+
+@contextlib.contextmanager
+def set_bounce_buffer_size(nbytes: int):
+    """Context for resetting the the size of the bounce buffer.
+
+    Parameters
+    ----------
+    nbytes : int
+        The bounce buffer size in bytes.
+    """
+    old_value = bounce_buffer_size()
+    try:
+        bounce_buffer_size_reset(nbytes)
+        yield
+    finally:
+        bounce_buffer_size_reset(old_value)
diff --git a/python/kvikio/pyproject.toml b/python/kvikio/pyproject.toml
index 824641964d..046e157a11 100644
--- a/python/kvikio/pyproject.toml
+++ b/python/kvikio/pyproject.toml
@@ -5,7 +5,7 @@
 build-backend = "rapids_build_backend.build"
 requires = [
     "rapids-build-backend>=0.3.0,<0.4.0.dev0",
-    "scikit-build-core[pyproject]>=0.7.0",
+    "scikit-build-core[pyproject]>=0.10.0",
 ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`.
 
 [project]
@@ -17,11 +17,11 @@ authors = [
     { name = "NVIDIA Corporation" },
 ]
 license = { text = "Apache 2.0" }
-requires-python = ">=3.9"
+requires-python = ">=3.10"
 dependencies = [
     "cupy-cuda11x>=12.0.0",
-    "numcodecs <0.12.0",
-    "numpy>=1.23,<2.0a0",
+    "numcodecs !=0.12.0",
+    "numpy>=1.23,<3.0a0",
     "packaging",
     "zarr",
 ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`.
@@ -31,9 +31,9 @@ classifiers = [
     "Topic :: Scientific/Engineering",
     "License :: OSI Approved :: Apache Software License",
     "Programming Language :: Python",
-    "Programming Language :: Python :: 3.9",
     "Programming Language :: Python :: 3.10",
     "Programming Language :: Python :: 3.11",
+    "Programming Language :: Python :: 3.12",
 ]
 
 [project.optional-dependencies]
@@ -49,7 +49,7 @@ Homepage = "https://github.com/rapidsai/kvikio"
 
 [tool.black]
 line-length = 88
-target-version = ["py39"]
+target-version = ["py310"]
 include = '\.py?$'
 exclude = '''
 /(
@@ -114,18 +114,20 @@ matrix-entry = "cuda_suffixed=true"
 requires = [
     "cmake>=3.26.4,!=3.30.0",
     "cython>=3.0.0",
-    "libkvikio==24.8.*,>=0.0.0a0",
+    "libkvikio==24.10.*,>=0.0.0a0",
     "ninja",
 ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`.
 
 [tool.scikit-build]
 build-dir = "build/{wheel_tag}"
 cmake.build-type = "Release"
-cmake.minimum-version = "3.26.4"
+cmake.version = "CMakeLists.txt"
+minimum-version = "build-system.requires"
 ninja.make-fallback = true
 sdist.exclude = ["*tests*"]
 sdist.reproducible = true
 wheel.packages = ["kvikio"]
+wheel.exclude = ["*.pyx", "CMakeLists.txt"]
 
 [tool.scikit-build.metadata.version]
 provider = "scikit_build_core.metadata.regex"
diff --git a/python/libkvikio/pyproject.toml b/python/libkvikio/pyproject.toml
index b40b42dbbd..3be7cbc0ae 100644
--- a/python/libkvikio/pyproject.toml
+++ b/python/libkvikio/pyproject.toml
@@ -5,7 +5,7 @@
 build-backend = "rapids_build_backend.build"
 requires = [
     "rapids-build-backend>=0.3.0,<0.4.0.dev0",
-    "scikit-build-core[pyproject]>=0.7.0",
+    "scikit-build-core[pyproject]>=0.10.0",
 ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`.
 
 [project]
@@ -23,9 +23,9 @@ classifiers = [
     "Topic :: Scientific/Engineering",
     "License :: OSI Approved :: Apache Software License",
     "Programming Language :: Python",
-    "Programming Language :: Python :: 3.9",
     "Programming Language :: Python :: 3.10",
     "Programming Language :: Python :: 3.11",
+    "Programming Language :: Python :: 3.12",
 ]
 
 [project.urls]
@@ -34,12 +34,14 @@ Homepage = "https://github.com/rapidsai/kvikio"
 [tool.scikit-build]
 build-dir = "build/{wheel_tag}"
 cmake.build-type = "Release"
-cmake.minimum-version = "3.26.4"
+cmake.version = "CMakeLists.txt"
+minimum-version = "build-system.requires"
 ninja.make-fallback = true
 sdist.exclude = ["*tests*"]
 sdist.reproducible = true
 wheel.packages = ["libkvikio"]
 wheel.py-api = "py3"
+wheel.install-dir = "libkvikio"
 
 [tool.scikit-build.metadata.version]
 provider = "scikit_build_core.metadata.regex"
@@ -54,3 +56,6 @@ requires = [
     "cmake>=3.26.4,!=3.30.0",
     "ninja",
 ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`.
+
+[project.entry-points."cmake.prefix"]
+libkvikio = "libkvikio"


Test patch:
diff --git a/.github/workflows/test.yaml b/.github/workflows/test.yaml
index 959c63a39f..ef093ee79d 100644
--- a/.github/workflows/test.yaml
+++ b/.github/workflows/test.yaml
@@ -16,7 +16,7 @@ on:
 jobs:
   cpp-tests:
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-tests.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-cpp-tests.yaml@branch-24.10
     with:
       build_type: nightly
       branch: ${{ inputs.branch }}
@@ -24,7 +24,7 @@ jobs:
       sha: ${{ inputs.sha }}
   python-tests:
     secrets: inherit
-    uses: rapidsai/shared-workflows/.github/workflows/conda-python-tests.yaml@branch-24.08
+    uses: rapidsai/shared-workflows/.github/workflows/conda-python-tests.yaml@branch-24.10
     with:
       build_type: nightly
       branch: ${{ inputs.branch }}
diff --git a/ci/run_ctests.sh b/ci/run_ctests.sh
index 069490facf..dcb938cbdd 100755
--- a/ci/run_ctests.sh
+++ b/ci/run_ctests.sh
@@ -6,5 +6,9 @@ set -euo pipefail
 # Support customizing the ctests' install location
 cd "${INSTALL_PREFIX:-${CONDA_PREFIX:-/usr}}/bin/tests/libkvikio/"
 
-# Run BASIC_IO_TEST
-./BASIC_IO_TEST
+# Run basic tests
+./BASIC_IO_EXAMPLE
+./BASIC_NO_CUDA_EXAMPLE
+
+# Run gtests
+ctest --no-tests=error --output-on-failure "$@"
diff --git a/cpp/cmake/thirdparty/get_gtest.cmake b/cpp/cmake/thirdparty/get_gtest.cmake
index 1133c846c3..10e6b026d9 100644
--- a/cpp/cmake/thirdparty/get_gtest.cmake
+++ b/cpp/cmake/thirdparty/get_gtest.cmake
@@ -1,5 +1,5 @@
 # =============================================================================
-# Copyright (c) 2021-2023, NVIDIA CORPORATION.
+# Copyright (c) 2021-2024, NVIDIA CORPORATION.
 #
 # Licensed under the Apache License, Version 2.0 (the "License"); you may not use this file except
 # in compliance with the License. You may obtain a copy of the License at
@@ -17,24 +17,7 @@ function(find_and_configure_gtest)
   include(${rapids-cmake-dir}/cpm/gtest.cmake)
 
   # Find or install GoogleTest
-  rapids_cpm_gtest(
-    BUILD_EXPORT_SET kvikio-testing-exports INSTALL_EXPORT_SET kvikio-testing-exports
-  )
-
-  if(GTest_ADDED)
-    rapids_export(
-      BUILD GTest
-      VERSION ${GTest_VERSION}
-      EXPORT_SET GTestTargets
-      GLOBAL_TARGETS gtest gmock gtest_main gmock_main
-      NAMESPACE GTest::
-    )
-
-    include("${rapids-cmake-dir}/export/find_package_root.cmake")
-    rapids_export_find_package_root(
-      BUILD GTest [=[${CMAKE_CURRENT_LIST_DIR}]=] EXPORT_SET kvikio-testing-exports
-    )
-  endif()
+  rapids_cpm_gtest(BUILD_STATIC)
 
 endfunction()
 
diff --git a/cpp/tests/CMakeLists.txt b/cpp/tests/CMakeLists.txt
new file mode 100644
index 0000000000..429bd8b722
--- /dev/null
+++ b/cpp/tests/CMakeLists.txt
@@ -0,0 +1,43 @@
+# =============================================================================
+# Copyright (c) 2024, NVIDIA CORPORATION.
+#
+# Licensed under the Apache License, Version 2.0 (the "License"); you may not use this file except
+# in compliance with the License. You may obtain a copy of the License at
+#
+# http://www.apache.org/licenses/LICENSE-2.0
+#
+# Unless required by applicable law or agreed to in writing, software distributed under the License
+# is distributed on an "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express
+# or implied. See the License for the specific language governing permissions and limitations under
+# the License.
+# =============================================================================
+
+# ##################################################################################################
+# enable testing -----------------------------------------------------------------------------------
+# ##################################################################################################
+enable_testing()
+
+include(rapids-test)
+rapids_test_init()
+
+file(GLOB SOURCES "*.cpp")
+add_executable(cpp_tests ${SOURCES})
+set_target_properties(
+  cpp_tests
+  PROPERTIES RUNTIME_OUTPUT_DIRECTORY "$<BUILD_INTERFACE:${KvikIO_BINARY_DIR}/gtests>"
+             CXX_STANDARD 17
+             CXX_STANDARD_REQUIRED ON
+             # For std:: support of __int128_t. Can be removed once using cuda::std
+             CXX_EXTENSIONS ON
+             CUDA_STANDARD 17
+             CUDA_STANDARD_REQUIRED ON
+)
+target_link_libraries(cpp_tests PRIVATE kvikio::kvikio GTest::gmock GTest::gtest)
+rapids_test_add(
+  NAME cpp_tests
+  COMMAND cpp_tests
+  GPUS 1
+  INSTALL_COMPONENT_SET testing
+)
+
+rapids_test_install_relocatable(INSTALL_COMPONENT_SET testing DESTINATION bin/tests/libkvikio)
diff --git a/cpp/tests/main.cpp b/cpp/tests/main.cpp
new file mode 100644
index 0000000000..73acda862a
--- /dev/null
+++ b/cpp/tests/main.cpp
@@ -0,0 +1,23 @@
+/*
+ * Copyright (c) 2024, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#include <gtest/gtest.h>
+
+int main(int argc, char** argv)
+{
+  ::testing::InitGoogleTest(&argc, argv);
+  return RUN_ALL_TESTS();
+}
diff --git a/cpp/tests/test_basic_io.cpp b/cpp/tests/test_basic_io.cpp
new file mode 100644
index 0000000000..12ccb6d428
--- /dev/null
+++ b/cpp/tests/test_basic_io.cpp
@@ -0,0 +1,43 @@
+/*
+ * Copyright (c) 2024, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#include <kvikio/file_handle.hpp>
+
+#include "utils.hpp"
+
+using namespace kvikio::test;
+
+TEST(BasicIO, write_read)
+{
+  TempDir tmp_dir{false};
+  auto filepath = tmp_dir.path() / "test";
+
+  auto dev_a = DevBuffer::arange(100);
+  auto dev_b = DevBuffer::zero_like(dev_a);
+
+  {
+    kvikio::FileHandle f(filepath, "w");
+    auto nbytes = f.write(dev_a.ptr, dev_a.nbytes, 0, 0);
+    EXPECT_EQ(nbytes, dev_a.nbytes);
+  }
+
+  {
+    kvikio::FileHandle f(filepath, "r");
+    auto nbytes = f.read(dev_b.ptr, dev_b.nbytes, 0, 0);
+    EXPECT_EQ(nbytes, dev_b.nbytes);
+    expect_equal(dev_a, dev_b);
+  }
+}
diff --git a/cpp/tests/utils.hpp b/cpp/tests/utils.hpp
new file mode 100644
index 0000000000..56a2cd5c45
--- /dev/null
+++ b/cpp/tests/utils.hpp
@@ -0,0 +1,172 @@
+/*
+ * Copyright (c) 2024, NVIDIA CORPORATION.
+ *
+ * Licensed under the Apache License, Version 2.0 (the "License");
+ * you may not use this file except in compliance with the License.
+ * You may obtain a copy of the License at
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+#pragma once
+
+#include <cstdint>
+#include <filesystem>
+#include <iostream>
+#include <numeric>
+#include <string>
+#include <vector>
+
+#include <cuda_runtime_api.h>
+#include <gtest/gtest.h>
+
+#include <kvikio/error.hpp>
+
+namespace kvikio::test {
+
+/**                                                                                  \
+ * @brief Error checking macro for CUDA runtime API functions.                       \
+ *                                                                                   \
+ * Invokes a CUDA runtime API function call. If the call does not return             \
+ * `cudaSuccess`, invokes cudaGetLastError() to clear the error and throws an        \
+ * exception detailing the CUDA error that occurred                                  \
+ *                                                                                   \
+ * Defaults to throwing std::runtime_error, but a custom exception may also be       \
+ * specified.                                                                        \
+ *                                                                                   \
+ * Example:                                                                          \
+ * ```c++                                                                            \
+ *                                                                                   \
+ * // Throws std::runtime_error if `cudaMalloc` fails                                \
+ * KVIKIO_CHECK_CUDA(cudaMalloc(&p, 100));                                           \
+ *                                                                                   \
+ * // Throws std::runtime_error if `cudaMalloc` fails                                \
+ * KVIKIO_CHECK_CUDA(cudaMalloc(&p, 100), std::runtime_error);                       \
+ * ```                                                                               \
+ *                                                                                   \
+ */                                                                                  \
+#define KVIKIO_CHECK_CUDA(...)                                                       \
+  GET_KVIKIO_CHECK_CUDA_MACRO(__VA_ARGS__, KVIKIO_CHECK_CUDA_2, KVIKIO_CHECK_CUDA_1) \
+  (__VA_ARGS__)
+#define GET_KVIKIO_CHECK_CUDA_MACRO(_1, _2, NAME, ...) NAME
+#define KVIKIO_CHECK_CUDA_2(_call, _exception_type)                                             \
+  do {                                                                                          \
+    cudaError_t const error = (_call);                                                          \
+    if (cudaSuccess != error) {                                                                 \
+      cudaGetLastError();                                                                       \
+      /*NOLINTNEXTLINE(bugprone-macro-parentheses)*/                                            \
+      throw _exception_type{std::string{"CUDA error at: "} + __FILE__ + ":" +                   \
+                            KVIKIO_STRINGIFY(__LINE__) + ": " + cudaGetErrorName(error) + " " + \
+                            cudaGetErrorString(error)};                                         \
+    }                                                                                           \
+  } while (0)
+#define KVIKIO_CHECK_CUDA_1(_call) KVIKIO_CHECK_CUDA_2(_call, std::runtime_error)
+
+/**
+ * @brief Help class to create a temporary directory.
+ */
+class TempDir {
+ public:
+  TempDir(const bool cleanup = true) : _cleanup{cleanup}
+  {
+    std::string tpl{std::filesystem::temp_directory_path() / "kvikio.XXXXXX"};
+    if (mkdtemp(tpl.data()) == nullptr) {
+      throw std::runtime_error("TempDir: cannot make tmpdir: " + tpl);
+    }
+    _dir_path = tpl;
+  }
+  ~TempDir() noexcept
+  {
+    if (_cleanup) {
+      try {
+        std::filesystem::remove_all(_dir_path);
+      } catch (...) {
+        std::cout << "error while trying to remove " << _dir_path.string() << std::endl;
+      }
+    }
+  }
+
+  TempDir(const TempDir&)              = delete;
+  TempDir& operator=(TempDir const&)   = delete;
+  TempDir(const TempDir&&)             = delete;
+  TempDir&& operator=(TempDir const&&) = delete;
+
+  const std::filesystem::path& path() { return _dir_path; }
+
+  operator std::string() { return path(); }
+
+ private:
+  const bool _cleanup;
+  std::filesystem::path _dir_path{};
+};
+
+/**
+ * @brief Help class for creating and comparing buffers.
+ */
+class DevBuffer {
+ public:
+  const std::size_t nelem;
+  const std::size_t nbytes;
+  void* ptr{nullptr};
+
+  DevBuffer(std::size_t nelem) : nelem{nelem}, nbytes{nelem * sizeof(std::int64_t)}
+  {
+    KVIKIO_CHECK_CUDA(cudaMalloc(&ptr, nbytes));
+  }
+  DevBuffer(const std::vector<std::int64_t>& host_buffer) : DevBuffer{host_buffer.size()}
+  {
+    KVIKIO_CHECK_CUDA(cudaMemcpy(ptr, host_buffer.data(), nbytes, cudaMemcpyHostToDevice));
+  }
+
+  ~DevBuffer() noexcept { cudaFree(ptr); }
+
+  [[nodiscard]] static DevBuffer arange(std::size_t nelem, std::int64_t start = 0)
+  {
+    std::vector<std::int64_t> host_buffer(nelem);
+    std::iota(host_buffer.begin(), host_buffer.end(), start);
+    return DevBuffer{host_buffer};
+  }
+
+  [[nodiscard]] static DevBuffer zero_like(const DevBuffer& prototype)
+  {
+    DevBuffer ret{prototype.nelem};
+    KVIKIO_CHECK_CUDA(cudaMemset(ret.ptr, 0, ret.nbytes));
+    return ret;
+  }
+
+  [[nodiscard]] std::vector<std::int64_t> to_vector() const
+  {
+    std::vector<std::int64_t> ret(nelem);
+    KVIKIO_CHECK_CUDA(cudaMemcpy(ret.data(), this->ptr, nbytes, cudaMemcpyDeviceToHost));
+    return ret;
+  }
+
+  void pprint() const
+  {
+    std::cout << "DevBuffer(";
+    for (auto item : to_vector()) {
+      std::cout << static_cast<int64_t>(item) << ", ";
+    }
+    std::cout << ")" << std::endl;
+  }
+};
+
+/**
+ * @brief Check that two buffers are equal
+ */
+inline void expect_equal(const DevBuffer& a, const DevBuffer& b)
+{
+  EXPECT_EQ(a.nbytes, b.nbytes);
+  auto a_vec = a.to_vector();
+  auto b_vec = b.to_vector();
+  for (std::size_t i = 0; i < a.nelem; ++i) {
+    EXPECT_EQ(a_vec[i], b_vec[i]) << "Mismatch at index " << i;
+  }
+}
+
+}  // namespace kvikio::test
diff --git a/python/kvikio/tests/test_async_io.py b/python/kvikio/tests/test_async_io.py
index 194a921ec5..5ddc7536d8 100644
--- a/python/kvikio/tests/test_async_io.py
+++ b/python/kvikio/tests/test_async_io.py
@@ -54,5 +54,5 @@ def test_read_write(tmp_path, size):
         == future_stream3.check_bytes_done()
         == b.nbytes
     )
-    assert all(a == b)
-    assert all(a == c)
+    cupy.testing.assert_array_equal(a, b)
+    cupy.testing.assert_array_equal(a, c)
diff --git a/python/kvikio/tests/test_basic_io.py b/python/kvikio/tests/test_basic_io.py
index 49c3bb544d..e1e9932e23 100644
--- a/python/kvikio/tests/test_basic_io.py
+++ b/python/kvikio/tests/test_basic_io.py
@@ -8,6 +8,7 @@
 import pytest
 
 import kvikio
+import kvikio.buffer
 import kvikio.defaults
 
 cupy = pytest.importorskip("cupy")
@@ -22,33 +23,42 @@ def check_bit_flags(x: int, y: int) -> bool:
 @pytest.mark.parametrize("size", [1, 10, 100, 1000, 1024, 4096, 4096 * 10])
 @pytest.mark.parametrize("nthreads", [1, 3, 4, 16])
 @pytest.mark.parametrize("tasksize", [199, 1024])
-def test_read_write(tmp_path, xp, gds_threshold, size, nthreads, tasksize):
+def test_write(tmp_path, xp, gds_threshold, size, nthreads, tasksize):
     """Test basic read/write"""
     filename = tmp_path / "test-file"
 
     with kvikio.defaults.set_num_threads(nthreads):
         with kvikio.defaults.set_task_size(tasksize):
-            # Write file
             a = xp.arange(size)
             f = kvikio.CuFile(filename, "w")
             assert not f.closed
             assert check_bit_flags(f.open_flags(), os.O_WRONLY)
             assert f.write(a) == a.nbytes
-
-            # Try to read file opened in write-only mode
-            with pytest.raises(RuntimeError, match="Operation not permitted"):
-                f.read(a)
-
-            # Close file
             f.close()
             assert f.closed
 
-            # Read file into a new array and compare
+            b = numpy.fromfile(filename, dtype=a.dtype)
+            xp.testing.assert_array_equal(a, b)
+
+
+@pytest.mark.parametrize("size", [1, 10, 100, 1000, 1024, 4096, 4096 * 10])
+@pytest.mark.parametrize("nthreads", [1, 3, 4, 16])
+@pytest.mark.parametrize("tasksize", [199, 1024])
+def test_read(tmp_path, xp, gds_threshold, size, nthreads, tasksize):
+    """Test basic read/write"""
+    filename = tmp_path / "test-file"
+
+    with kvikio.defaults.set_num_threads(nthreads):
+        with kvikio.defaults.set_task_size(tasksize):
+            a = numpy.arange(size)
+            a.tofile(filename)
+            os.sync()
+
             b = xp.empty_like(a)
             f = kvikio.CuFile(filename, "r")
             assert check_bit_flags(f.open_flags(), os.O_RDONLY)
             assert f.read(b) == b.nbytes
-            assert all(a == b)
+            xp.testing.assert_array_equal(a, b)
 
 
 def test_file_handle_context(tmp_path):
@@ -61,10 +71,34 @@ def test_file_handle_context(tmp_path):
         assert check_bit_flags(f.open_flags(), os.O_RDWR)
         assert f.write(a) == a.nbytes
         assert f.read(b) == b.nbytes
-        assert all(a == b)
+        cupy.testing.assert_array_equal(a, b)
     assert f.closed
 
 
+def test_no_file_error(tmp_path):
+    """Test "No such file" error"""
+
+    filename = tmp_path / "test-file"
+    with pytest.raises(RuntimeError, match="Unable to open file: No such file"):
+        kvikio.CuFile(filename, "r")
+
+
+def test_incorrect_open_mode_error(tmp_path, xp):
+    """Test incorrect mode errors"""
+    filename = tmp_path / "test-file"
+    a = numpy.arange(10)
+    a.tofile(filename)
+    os.sync()
+
+    with kvikio.CuFile(filename, "r") as f:
+        with pytest.raises(RuntimeError, match="Operation not permitted"):
+            f.write(xp.arange(10))
+
+    with kvikio.CuFile(filename, "w") as f:
+        with pytest.raises(RuntimeError, match="Operation not permitted"):
+            f.read(xp.arange(10))
+
+
 @pytest.mark.skipif(
     kvikio.defaults.compat_mode(),
     reason="cannot test `set_compat_mode` when already running in compatibility mode",
@@ -109,7 +143,7 @@ def test_write_to_files_in_chunks(tmp_path, xp, gds_threshold):
     b = xp.empty_like(a)
     f = kvikio.CuFile(filename, "r")
     assert f.read(b) == b.nbytes
-    assert all(a == b)
+    xp.testing.assert_array_equal(a, b)
 
 
 @pytest.mark.parametrize("nthreads", [1, 3, 16])
@@ -131,7 +165,7 @@ def test_read_write_slices(tmp_path, xp, gds_threshold, nthreads, tasksize, star
                 assert f.write(a[start:end]) == a[start:end].nbytes
             with kvikio.CuFile(filename, "r") as f:
                 assert f.read(b[start:end]) == b[start:end].nbytes
-            assert all(a == b)
+            xp.testing.assert_array_equal(a, b)
 
 
 @pytest.mark.parametrize("size", [1, 10, 100, 1000, 1024, 4096, 4096 * 10])
@@ -188,7 +222,7 @@ def test_no_current_cuda_context(tmp_path, xp, gds_threshold):
         with with_no_cuda_context():
             f.write(a)
         f.read(b)
-    assert all(a == b)
+    xp.testing.assert_array_equal(a, b)
 
 
 @pytest.mark.skipif(
@@ -223,3 +257,39 @@ def test_multiple_gpus(tmp_path, xp, gds_threshold):
                 with cupy.cuda.Device(0):
                     assert f.read(a1) == a1.nbytes
             assert bytes(a0) == bytes(a1)
+
+
+@pytest.mark.parametrize("size", [1, 10, 100, 1000])
+@pytest.mark.parametrize("tasksize", [1, 10, 100, 1000])
+@pytest.mark.parametrize("buffer_size", [1, 10, 100, 1000])
+def test_different_bounce_buffer_sizes(tmp_path, size, tasksize, buffer_size):
+    """Test different bounce buffer sizes"""
+    filename = tmp_path / "test-file"
+    with kvikio.defaults.set_compat_mode(True), kvikio.defaults.set_num_threads(10):
+        with kvikio.defaults.set_task_size(tasksize):
+            with kvikio.defaults.set_bounce_buffer_size(buffer_size):
+                with kvikio.CuFile(filename, "w+") as f:
+                    a = cupy.arange(size)
+                    b = cupy.empty_like(a)
+                    f.write(a)
+                    assert f.read(b) == b.nbytes
+                    cupy.testing.assert_array_equal(a, b)
+
+
+def test_bounce_buffer_free(tmp_path):
+    """Test freeing the bounce buffer allocations"""
+    filename = tmp_path / "test-file"
+    kvikio.buffer.bounce_buffer_free()
+    with kvikio.defaults.set_compat_mode(True), kvikio.defaults.set_num_threads(1):
+        with kvikio.CuFile(filename, "w") as f:
+            with kvikio.defaults.set_bounce_buffer_size(1024):
+                # Notice, since the bounce buffer size is only checked when the buffer
+                # is used, we populate the bounce buffer in between we clear it.
+                f.write(cupy.arange(10))
+                assert kvikio.buffer.bounce_buffer_free() == 1024
+                assert kvikio.buffer.bounce_buffer_free() == 0
+                f.write(cupy.arange(10))
+            with kvikio.defaults.set_bounce_buffer_size(2048):
+                f.write(cupy.arange(10))
+                assert kvikio.buffer.bounce_buffer_free() == 2048
+                assert kvikio.buffer.bounce_buffer_free() == 0
diff --git a/python/kvikio/tests/test_defaults.py b/python/kvikio/tests/test_defaults.py
index 27b2d6187a..39892a784d 100644
--- a/python/kvikio/tests/test_defaults.py
+++ b/python/kvikio/tests/test_defaults.py
@@ -1,4 +1,4 @@
-# Copyright (c) 2022, NVIDIA CORPORATION. All rights reserved.
+# Copyright (c) 2022-2024, NVIDIA CORPORATION. All rights reserved.
 # See file LICENSE for terms.
 
 
@@ -32,6 +32,11 @@ def test_num_threads():
         assert kvikio.defaults.get_num_threads() == 4
     assert before == kvikio.defaults.get_num_threads()
 
+    with pytest.raises(ValueError, match="positive integer greater than zero"):
+        kvikio.defaults.num_threads_reset(0)
+    with pytest.raises(OverflowError, match="negative value"):
+        kvikio.defaults.num_threads_reset(-1)
+
 
 def test_task_size():
     """Test changing `task_size`"""
@@ -42,3 +47,38 @@ def test_task_size():
         kvikio.defaults.task_size_reset(4)
         assert kvikio.defaults.task_size() == 4
     assert before == kvikio.defaults.task_size()
+
+    with pytest.raises(ValueError, match="positive integer greater than zero"):
+        kvikio.defaults.task_size_reset(0)
+    with pytest.raises(OverflowError, match="negative value"):
+        kvikio.defaults.task_size_reset(-1)
+
+
+def test_gds_threshold():
+    """Test changing `gds_threshold`"""
+
+    before = kvikio.defaults.gds_threshold()
+    with kvikio.defaults.set_gds_threshold(3):
+        assert kvikio.defaults.gds_threshold() == 3
+        kvikio.defaults.gds_threshold_reset(4)
+        assert kvikio.defaults.gds_threshold() == 4
+    assert before == kvikio.defaults.gds_threshold()
+
+    with pytest.raises(OverflowError, match="negative value"):
+        kvikio.defaults.gds_threshold_reset(-1)
+
+
+def test_bounce_buffer_size():
+    """Test changing `bounce_buffer_size`"""
+
+    before = kvikio.defaults.bounce_buffer_size()
+    with kvikio.defaults.set_bounce_buffer_size(3):
+        assert kvikio.defaults.bounce_buffer_size() == 3
+        kvikio.defaults.bounce_buffer_size_reset(4)
+        assert kvikio.defaults.bounce_buffer_size() == 4
+    assert before == kvikio.defaults.bounce_buffer_size()
+
+    with pytest.raises(ValueError, match="positive integer greater than zero"):
+        kvikio.defaults.bounce_buffer_size_reset(0)
+    with pytest.raises(OverflowError, match="negative value"):
+        kvikio.defaults.bounce_buffer_size_reset(-1)