diff --git a/.beads/.gitignore b/.beads/.gitignore new file mode 100644 index 0000000..374adb8 --- /dev/null +++ b/.beads/.gitignore @@ -0,0 +1,32 @@ +# SQLite databases +*.db +*.db?* +*.db-journal +*.db-wal +*.db-shm + +# Daemon runtime files +daemon.lock +daemon.log +daemon.pid +bd.sock + +# Local version tracking (prevents upgrade notification spam after git ops) +.local_version + +# Legacy database files +db.sqlite +bd.db + +# Merge artifacts (temporary files from 3-way merge) +beads.base.jsonl +beads.base.meta.json +beads.left.jsonl +beads.left.meta.json +beads.right.jsonl +beads.right.meta.json + +# Keep JSONL exports and config (source of truth for git) +!issues.jsonl +!metadata.json +!config.json diff --git a/.beads/README.md b/.beads/README.md new file mode 100644 index 0000000..50f281f --- /dev/null +++ b/.beads/README.md @@ -0,0 +1,81 @@ +# Beads - AI-Native Issue Tracking + +Welcome to Beads! This repository uses **Beads** for issue tracking - a modern, AI-native tool designed to live directly in your codebase alongside your code. + +## What is Beads? + +Beads is issue tracking that lives in your repo, making it perfect for AI coding agents and developers who want their issues close to their code. No web UI required - everything works through the CLI and integrates seamlessly with git. + +**Learn more:** [github.com/steveyegge/beads](https://github.com/steveyegge/beads) + +## Quick Start + +### Essential Commands + +```bash +# Create new issues +bd create "Add user authentication" + +# View all issues +bd list + +# View issue details +bd show + +# Update issue status +bd update --status in_progress +bd update --status done + +# Sync with git remote +bd sync +``` + +### Working with Issues + +Issues in Beads are: +- **Git-native**: Stored in `.beads/issues.jsonl` and synced like code +- **AI-friendly**: CLI-first design works perfectly with AI coding agents +- **Branch-aware**: Issues can follow your branch workflow +- **Always in sync**: Auto-syncs with your commits + +## Why Beads? + +✨ **AI-Native Design** +- Built specifically for AI-assisted development workflows +- CLI-first interface works seamlessly with AI coding agents +- No context switching to web UIs + +🚀 **Developer Focused** +- Issues live in your repo, right next to your code +- Works offline, syncs when you push +- Fast, lightweight, and stays out of your way + +🔧 **Git Integration** +- Automatic sync with git commits +- Branch-aware issue tracking +- Intelligent JSONL merge resolution + +## Get Started with Beads + +Try Beads in your own projects: + +```bash +# Install Beads +curl -sSL https://raw.githubusercontent.com/steveyegge/beads/main/scripts/install.sh | bash + +# Initialize in your repo +bd init + +# Create your first issue +bd create "Try out Beads" +``` + +## Learn More + +- **Documentation**: [github.com/steveyegge/beads/docs](https://github.com/steveyegge/beads/tree/main/docs) +- **Quick Start Guide**: Run `bd quickstart` +- **Examples**: [github.com/steveyegge/beads/examples](https://github.com/steveyegge/beads/tree/main/examples) + +--- + +*Beads: Issue tracking that moves at the speed of thought* ⚡ diff --git a/.beads/config.yaml b/.beads/config.yaml new file mode 100644 index 0000000..f242785 --- /dev/null +++ b/.beads/config.yaml @@ -0,0 +1,62 @@ +# Beads Configuration File +# This file configures default behavior for all bd commands in this repository +# All settings can also be set via environment variables (BD_* prefix) +# or overridden with command-line flags + +# Issue prefix for this repository (used by bd init) +# If not set, bd init will auto-detect from directory name +# Example: issue-prefix: "myproject" creates issues like "myproject-1", "myproject-2", etc. +# issue-prefix: "" + +# Use no-db mode: load from JSONL, no SQLite, write back after each command +# When true, bd will use .beads/issues.jsonl as the source of truth +# instead of SQLite database +# no-db: false + +# Disable daemon for RPC communication (forces direct database access) +# no-daemon: false + +# Disable auto-flush of database to JSONL after mutations +# no-auto-flush: false + +# Disable auto-import from JSONL when it's newer than database +# no-auto-import: false + +# Enable JSON output by default +# json: false + +# Default actor for audit trails (overridden by BD_ACTOR or --actor) +# actor: "" + +# Path to database (overridden by BEADS_DB or --db) +# db: "" + +# Auto-start daemon if not running (can also use BEADS_AUTO_START_DAEMON) +# auto-start-daemon: true + +# Debounce interval for auto-flush (can also use BEADS_FLUSH_DEBOUNCE) +# flush-debounce: "5s" + +# Git branch for beads commits (bd sync will commit to this branch) +# IMPORTANT: Set this for team projects so all clones use the same sync branch. +# This setting persists across clones (unlike database config which is gitignored). +# Can also use BEADS_SYNC_BRANCH env var for local override. +# If not set, bd sync will require you to run 'bd config set sync.branch '. +# sync-branch: "beads-sync" + +# Multi-repo configuration (experimental - bd-307) +# Allows hydrating from multiple repositories and routing writes to the correct JSONL +# repos: +# primary: "." # Primary repo (where this database lives) +# additional: # Additional repos to hydrate from (read-only) +# - ~/beads-planning # Personal planning repo +# - ~/work-planning # Work planning repo + +# Integration settings (access with 'bd config get/set') +# These are stored in the database, not in this file: +# - jira.url +# - jira.project +# - linear.url +# - linear.api-key +# - github.org +# - github.repo diff --git a/.beads/issues.jsonl b/.beads/issues.jsonl new file mode 100644 index 0000000..a39f34e --- /dev/null +++ b/.beads/issues.jsonl @@ -0,0 +1,15 @@ +{"id":"gpu-support-0y1","title":"Phase 3: Kernel Loading Utilities","description":"Consolidate kernel loading utilities.\n\n## Analysis\n- ART has KernelLoader (kernels/metal/, kernels/opencl/ conventions)\n- gpu-test-framework has KernelResourceLoader (generic, cached)\n\n## Decision\nKeep KernelResourceLoader in gpu-test-framework as-is.\nAdd OpenCL-specific convenience methods if needed.\n\n## Tasks\n- Review if KernelLoader conventions needed in gpu-support\n- Add any missing functionality to KernelResourceLoader\n- Document kernel resource path conventions\n\n## Acceptance Criteria\n- [ ] Kernel loading works for extracted compute infrastructure\n- [ ] Convention documented\n\nContext: Depends on gpu-support-5wc (Phase 2)","status":"closed","priority":3,"issue_type":"feature","created_at":"2025-12-28T17:01:52.69214-08:00","updated_at":"2025-12-29T00:14:21.149884-08:00","closed_at":"2025-12-29T00:14:21.149884-08:00","close_reason":"KernelLoader added with conventions, caching, 12 tests pass. Package documentation complete.","dependencies":[{"issue_id":"gpu-support-0y1","depends_on_id":"gpu-support-5wc","type":"blocks","created_at":"2025-12-28T17:02:43.572244-08:00","created_by":"daemon"}]} +{"id":"gpu-support-5wc","title":"Phase 2: OpenCL Implementation","description":"Extract OpenCL implementation classes from ART.\n\n## Components\n1. OpenCLContext - Singleton context manager with reference counting\n2. OpenCLKernel - Kernel compilation, async execution, events\n3. OpenCLBuffer - Float buffer wrapper using OpenCL API\n\n## Package\n`com.hellblazer.luciferase.resource.compute` (context, kernel)\n`com.hellblazer.luciferase.resource.compute.memory` (buffer)\n\n## Key Patterns\n- Singleton context persists until JVM shutdown (macOS OpenCL cleanup crashes)\n- Out-of-order queue execution when supported\n- Event-based async kernel execution\n\n## Integration\n- OpenCLBuffer uses existing CLBufferHandle.translateError()\n- Tests extend CICompatibleGPUTest for CI compatibility\n\n## Acceptance Criteria\n- [ ] OpenCL context initializes correctly\n- [ ] Kernels compile and execute\n- [ ] Integration tests pass on local machine\n- [ ] Tests skip gracefully in CI without OpenCL\n\nContext: Depends on gpu-support-6e9 (Phase 1)","status":"closed","priority":2,"issue_type":"feature","created_at":"2025-12-28T17:01:42.204995-08:00","updated_at":"2025-12-29T00:05:42.348939-08:00","closed_at":"2025-12-29T00:05:42.348939-08:00","close_reason":"Phase 2 complete: OpenCLContext singleton, OpenCLBuffer, OpenCLKernel, BackendSelector - 251 tests pass","dependencies":[{"issue_id":"gpu-support-5wc","depends_on_id":"gpu-support-6e9","type":"blocks","created_at":"2025-12-28T17:02:43.493285-08:00","created_by":"daemon"}]} +{"id":"gpu-support-6e9","title":"Phase 1: Core Compute Interfaces","description":"Extract foundational interfaces and enums from ART.\n\n## Components\n1. GPUBuffer interface - Common buffer abstraction\n2. ComputeKernel interface - Unified kernel API with BufferAccess enum, exceptions\n3. GPUBackend enum - Backend selection (Metal initially disabled)\n4. BackendSelector - Auto-selection with CI detection\n5. GPUErrorClassifier - Programming vs recoverable error classification\n\n## Package\n`com.hellblazer.luciferase.resource.compute`\n`com.hellblazer.luciferase.resource.compute.memory`\n\n## Notes\n- Metal/BGFX detection disabled initially (no BGFX dependency)\n- All interfaces are generic, not ART-specific\n- Unit tests for each component\n\n## Acceptance Criteria\n- [ ] All interfaces compile in gpu-support\n- [ ] Unit tests pass\n- [ ] No ART-specific imports remain\n\nContext: Parent epic gpu-support-bsy","status":"closed","priority":2,"issue_type":"feature","created_at":"2025-12-28T17:01:30.296533-08:00","updated_at":"2025-12-28T18:11:02.279248-08:00","closed_at":"2025-12-28T18:11:02.279248-08:00","close_reason":"Closed"} +{"id":"gpu-support-6pw","title":"Extract OpenCLKernel implementation","description":"Extract OpenCLKernel from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/compute/OpenCLKernel.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/OpenCLKernel.java`\n\n## Changes Required\n- Update package declaration\n- Update GPUBuffer import\n- Update OpenCLBuffer import\n- Update ComputeKernel import\n- Update GPUBackend import\n\n## Key Features\n- Kernel compilation with build log on failure\n- Buffer argument binding\n- Scalar (float, int) argument setting\n- Local memory argument support\n- 1D/2D/3D execution with optional local work size\n- Async execution with events\n- clFinish() synchronization\n\n## Dependencies\n- Requires OpenCLContext, OpenCLBuffer, ComputeKernel interface\n\nContext: Parent feature gpu-support-5wc","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:06:08.709736-08:00","updated_at":"2025-12-28T19:59:07.163136-08:00","closed_at":"2025-12-28T19:59:07.163136-08:00","close_reason":"OpenCLKernel extracted - 16 tests pass, all kernel operations verified","dependencies":[{"issue_id":"gpu-support-6pw","depends_on_id":"gpu-support-gij","type":"blocks","created_at":"2025-12-28T17:06:31.104057-08:00","created_by":"daemon"},{"issue_id":"gpu-support-6pw","depends_on_id":"gpu-support-ilr","type":"blocks","created_at":"2025-12-28T17:06:31.184783-08:00","created_by":"daemon"}]} +{"id":"gpu-support-97u","title":"Phase 2 Integration Tests","description":"Create integration tests for Phase 2 OpenCL components.\n\n## Tests Required\n1. OpenCLContextTest\n - Singleton behavior\n - Acquire/release reference counting\n - Context/queue/device handle validity\n\n2. OpenCLKernelTest\n - Kernel compilation (simple vector_add kernel)\n - Argument setting\n - Execution with various work sizes\n - Error handling for invalid kernels\n\n3. OpenCLBufferTest\n - Buffer allocation\n - Upload/download float data\n - Size validation\n\n4. ComputeIntegrationTest\n - Full workflow: context -\u003e buffer -\u003e kernel -\u003e execute -\u003e read\n\n## Test Base Class\nExtend CICompatibleGPUTest for automatic OpenCL detection and CI skip\n\n## Test Kernel\nUse simple vector_add.cl kernel for validation\n\n## Acceptance Criteria\n- [ ] All tests pass on local machine with OpenCL\n- [ ] Tests skip gracefully in CI without OpenCL\n- [ ] No resource leaks (use @AfterEach cleanup)\n\nContext: Parent feature gpu-support-5wc","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:06:26.00741-08:00","updated_at":"2025-12-28T20:04:38.861567-08:00","closed_at":"2025-12-28T20:04:38.861567-08:00","close_reason":"Phase 2 integration tests complete - 9 tests covering full compute workflow, SAXPY, 2D execution, resource cleanup","dependencies":[{"issue_id":"gpu-support-97u","depends_on_id":"gpu-support-6pw","type":"blocks","created_at":"2025-12-28T17:06:31.343886-08:00","created_by":"daemon"}]} +{"id":"gpu-support-9go","title":"Extract GPUBackend enum","description":"Extract GPUBackend enum from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/compute/GPUBackend.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUBackend.java`\n\n## Changes Required\n- Update package declaration\n- DISABLE Metal detection initially (remove BGFX dependency)\n- isMetalAvailable() should return false unconditionally for now\n- Keep METAL enum value but mark as unavailable\n- Update OpenCLContext import\n\n## Notes\nMetal support can be added later when/if BGFX is added to gpu-support\n\nContext: Parent feature gpu-support-6e9","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:03:07.645662-08:00","updated_at":"2025-12-28T19:12:04.193389-08:00","closed_at":"2025-12-28T19:12:04.193389-08:00","close_reason":"Added isAvailable() with cached Metal/OpenCL detection and testResetAvailability()","dependencies":[{"issue_id":"gpu-support-9go","depends_on_id":"gpu-support-ad2","type":"blocks","created_at":"2025-12-28T17:05:48.848777-08:00","created_by":"daemon"}]} +{"id":"gpu-support-ad2","title":"Extract ComputeKernel interface","description":"Extract ComputeKernel interface from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/compute/ComputeKernel.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/ComputeKernel.java`\n\n## Changes Required\n- Update package declaration\n- Update GPUBuffer import to new location\n- Includes: BufferAccess enum, KernelCompilationException, KernelExecutionException\n\n## Dependencies\n- Requires GPUBuffer interface to exist first\n\nContext: Parent feature gpu-support-6e9","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:02:59.802566-08:00","updated_at":"2025-12-28T18:06:48.978099-08:00","closed_at":"2025-12-28T18:06:48.978099-08:00","close_reason":"Closed","dependencies":[{"issue_id":"gpu-support-ad2","depends_on_id":"gpu-support-e63","type":"blocks","created_at":"2025-12-28T17:05:48.692692-08:00","created_by":"daemon"}]} +{"id":"gpu-support-bsy","title":"Extract ART OpenCL Compute Infrastructure to gpu-support","description":"## Epic: Extract ART OpenCL Compute Infrastructure\n\n### Goal\nExtract the mature, production-ready GPU compute infrastructure from ART repository into gpu-support framework for reuse by ART, Luciferase, and future projects.\n\n### Source Location\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/`\n\n### Components to Extract\n- **compute/**: GPUBackend, BackendSelector, GPUErrorClassifier, OpenCLContext, OpenCLKernel, ComputeKernel\n- **memory/**: GPUBuffer, OpenCLBuffer \n- **kernels/**: KernelLoader (consolidate with existing KernelResourceLoader)\n\n### Target Package Structure\n```\ncom.hellblazer.luciferase.resource.compute/\n GPUBackend.java, BackendSelector.java, GPUErrorClassifier.java\n ComputeKernel.java, OpenCLContext.java, OpenCLKernel.java\ncom.hellblazer.luciferase.resource.compute.memory/\n GPUBuffer.java, OpenCLBuffer.java\n```\n\n### Key Patterns to Preserve\n1. Singleton OpenCL context with reference counting\n2. Threshold-based GPU/CPU execution selection\n3. Programming vs recoverable error classification\n4. Graceful CI environment handling\n5. Integration with existing CLBufferHandle\n\n### Success Criteria\n- ART can switch to using gpu-support's compute infrastructure\n- Luciferase can use same infrastructure for ESVO\n- All extracted code has comprehensive tests\n- CI runs tests with graceful skip when OpenCL unavailable\n\nContext: .pm/CONTEXT_PROTOCOL.md (when established)","status":"closed","priority":1,"issue_type":"epic","created_at":"2025-12-28T17:00:22.064372-08:00","updated_at":"2025-12-29T00:15:05.568204-08:00","closed_at":"2025-12-29T00:15:05.568204-08:00","close_reason":"OpenCL compute infrastructure extracted to gpu-support. Phase 4 (ART migration) is a separate task in ART repo."} +{"id":"gpu-support-cbr","title":"Extract BackendSelector","description":"Extract BackendSelector from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/compute/BackendSelector.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/BackendSelector.java`\n\n## Changes Required\n- Update package declaration\n- Update GPUBackend import\n- Rename ART_GPU_BACKEND env var to GPU_BACKEND (generic)\n- Rename ART_GPU_DISABLE env var to GPU_DISABLE (generic)\n\n## Key Features to Preserve\n- CI environment detection\n- Priority-based backend selection\n- Forced backend via environment variable\n\nContext: Parent feature gpu-support-6e9","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:03:14.299639-08:00","updated_at":"2025-12-28T20:01:27.530397-08:00","closed_at":"2025-12-28T20:01:27.530397-08:00","close_reason":"BackendSelector extracted with dual env var support - 17 tests pass","dependencies":[{"issue_id":"gpu-support-cbr","depends_on_id":"gpu-support-9go","type":"blocks","created_at":"2025-12-28T17:05:48.92771-08:00","created_by":"daemon"}]} +{"id":"gpu-support-e63","title":"Extract GPUBuffer interface","description":"Extract GPUBuffer interface from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/memory/GPUBuffer.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/memory/GPUBuffer.java`\n\n## Changes Required\n- Update package declaration\n- Remove any ART-specific imports (none expected)\n\n## Test\nCreate unit test verifying interface compilation\n\nContext: Parent feature gpu-support-6e9","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:02:52.795707-08:00","updated_at":"2025-12-28T18:04:16.430635-08:00","closed_at":"2025-12-28T18:04:16.430635-08:00","close_reason":"Closed"} +{"id":"gpu-support-gij","title":"Extract OpenCLContext singleton","description":"Extract OpenCLContext from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/compute/OpenCLContext.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/OpenCLContext.java`\n\n## Changes Required\n- Update package declaration\n- Rename art.gpu.disable property to luciferase.gpu.disable\n\n## Key Patterns to Preserve\n- Singleton with reference counting (acquire/release)\n- Out-of-order queue when supported\n- NO cleanup on release (macOS OpenCL crash prevention)\n- GPU/CPU device fallback\n\n## Integration Notes\n- This is the core context that OpenCLKernel and OpenCLBuffer depend on\n\nContext: Parent feature gpu-support-5wc","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:05:58.594406-08:00","updated_at":"2025-12-28T19:07:44.456012-08:00","closed_at":"2025-12-28T19:07:44.456012-08:00","close_reason":"OpenCLContext extracted with testReset(), dual property support, reference counting tests","dependencies":[{"issue_id":"gpu-support-gij","depends_on_id":"gpu-support-ipz","type":"blocks","created_at":"2025-12-28T17:06:35.807026-08:00","created_by":"daemon"}]} +{"id":"gpu-support-ilr","title":"Extract OpenCLBuffer implementation","description":"Extract OpenCLBuffer from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/memory/OpenCLBuffer.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/memory/OpenCLBuffer.java`\n\n## Changes Required\n- Update package declaration\n- Update GPUBuffer import\n- CLBufferHandle.translateError() import unchanged (already in resource module)\n\n## Key Features\n- Float buffer abstraction\n- upload(FloatBuffer) / upload(float[])\n- download(FloatBuffer) / download(float[])\n- Protected constructor for subclasses (CLBufferAdapter pattern)\n- Size validation on transfers\n\n## Future Enhancement (Phase 5)\nConsider refactoring to wrap CLBufferHandle internally for RAII benefits\n\nContext: Parent feature gpu-support-5wc","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:06:16.180484-08:00","updated_at":"2025-12-28T19:52:42.541562-08:00","closed_at":"2025-12-28T19:52:42.541562-08:00","close_reason":"OpenCLBuffer tests fixed - 10 tests pass with simplified test pattern","dependencies":[{"issue_id":"gpu-support-ilr","depends_on_id":"gpu-support-gij","type":"blocks","created_at":"2025-12-28T17:06:31.264386-08:00","created_by":"daemon"}]} +{"id":"gpu-support-ipz","title":"Phase 1 Unit Tests","description":"Create unit tests for Phase 1 components.\n\n## Tests Required\n1. GPUBufferTest - Interface contract verification (mock implementation)\n2. ComputeKernelTest - Interface contract, exception types\n3. GPUBackendTest - Enum values, availability checks\n4. BackendSelectorTest - Selection logic, CI detection, env vars\n5. GPUErrorClassifierTest - Error classification logic\n\n## Test Patterns\n- Use Mockito for interface testing\n- Test error classification with sample exception messages\n- Test CI environment detection with env var mocking\n\n## Acceptance Criteria\n- [ ] All tests pass\n- [ ] Coverage \u003e 80% for classifier logic\n\nContext: Parent feature gpu-support-6e9","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:05:42.229259-08:00","updated_at":"2025-12-28T18:11:02.203471-08:00","closed_at":"2025-12-28T18:11:02.203471-08:00","close_reason":"Closed","dependencies":[{"issue_id":"gpu-support-ipz","depends_on_id":"gpu-support-kdp","type":"blocks","created_at":"2025-12-28T17:05:49.00115-08:00","created_by":"daemon"},{"issue_id":"gpu-support-ipz","depends_on_id":"gpu-support-cbr","type":"blocks","created_at":"2025-12-28T17:05:49.077511-08:00","created_by":"daemon"},{"issue_id":"gpu-support-ipz","depends_on_id":"gpu-support-e63","type":"blocks","created_at":"2025-12-28T17:10:14.155949-08:00","created_by":"daemon"},{"issue_id":"gpu-support-ipz","depends_on_id":"gpu-support-ad2","type":"blocks","created_at":"2025-12-28T17:10:14.229753-08:00","created_by":"daemon"},{"issue_id":"gpu-support-ipz","depends_on_id":"gpu-support-9go","type":"blocks","created_at":"2025-12-28T17:10:14.30489-08:00","created_by":"daemon"}]} +{"id":"gpu-support-kdp","title":"Extract GPUErrorClassifier","description":"Extract GPUErrorClassifier from ART.\n\n## Source\n`/Users/hal.hildebrand/git/ART/art-modules/art-cortical/src/main/java/com/hellblazer/art/cortical/gpu/compute/GPUErrorClassifier.java`\n\n## Target\n`resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUErrorClassifier.java`\n\n## Changes Required\n- Update package declaration\n- Update ComputeKernel.KernelCompilationException import\n- Update ComputeKernel.KernelExecutionException import\n\n## Key Features to Preserve\n- Programming error detection (fail fast)\n- Recoverable error detection (allow CPU fallback)\n- OpenCL error code extraction from messages\n- Error code to name translation\n\nContext: Parent feature gpu-support-6e9","status":"closed","priority":2,"issue_type":"task","created_at":"2025-12-28T17:03:23.717012-08:00","updated_at":"2025-12-28T18:10:23.076314-08:00","closed_at":"2025-12-28T18:10:23.076314-08:00","close_reason":"Closed","dependencies":[{"issue_id":"gpu-support-kdp","depends_on_id":"gpu-support-ad2","type":"blocks","created_at":"2025-12-28T17:05:48.769925-08:00","created_by":"daemon"}]} +{"id":"gpu-support-trf","title":"Phase 4: ART Migration","description":"Migrate ART to use gpu-support compute infrastructure.\n\n## Tasks\n1. Update ART pom.xml to depend on gpu-support 1.0.5+\n2. Update ART imports from art.cortical.gpu to luciferase.resource.compute\n3. Remove extracted code from ART (compute/, memory/, kernels/ packages)\n4. Run full ART test suite to validate\n\n## Risk Mitigation\n- Keep ART working on separate branch until validated\n- Run performance comparison before/after\n\n## Acceptance Criteria\n- [ ] ART builds successfully with new dependency\n- [ ] All ART GPU tests pass\n- [ ] No duplicate code remains in ART\n- [ ] Performance within 5% of original\n\nContext: Depends on gpu-support-0y1 (Phase 3)","status":"open","priority":2,"issue_type":"feature","created_at":"2025-12-28T17:02:02.780414-08:00","updated_at":"2025-12-28T17:02:02.780414-08:00","dependencies":[{"issue_id":"gpu-support-trf","depends_on_id":"gpu-support-0y1","type":"blocks","created_at":"2025-12-28T17:02:43.652183-08:00","created_by":"daemon"}]} diff --git a/.beads/metadata.json b/.beads/metadata.json new file mode 100644 index 0000000..c787975 --- /dev/null +++ b/.beads/metadata.json @@ -0,0 +1,4 @@ +{ + "database": "beads.db", + "jsonl_export": "issues.jsonl" +} \ No newline at end of file diff --git a/.gitattributes b/.gitattributes new file mode 100644 index 0000000..807d598 --- /dev/null +++ b/.gitattributes @@ -0,0 +1,3 @@ + +# Use bd merge for beads JSONL files +.beads/issues.jsonl merge=beads diff --git a/.gitignore b/.gitignore index f2eb62c..7441b3f 100644 --- a/.gitignore +++ b/.gitignore @@ -36,3 +36,4 @@ build/ ### Mac OS ### .DS_Store +.pm/ diff --git a/AGENTS.md b/AGENTS.md new file mode 100644 index 0000000..a657a7f --- /dev/null +++ b/AGENTS.md @@ -0,0 +1,194 @@ +# Agent Instructions + +## Workflow + +This project uses **bd** (beads) for issue tracking. Run `bd onboard` to get started. + +```bash +bd ready # Find available work +bd show # View issue details +bd update --status in_progress # Claim work +bd close # Complete work +bd sync # Sync with git +``` + +### Session Completion + +Work is NOT complete until `git push` succeeds. + +```bash +git pull --rebase +bd sync +git push +git status # MUST show "up to date with origin" +``` + +--- + +## GPU Compute API Reference + +### Quick Start + +```java +var compute = ComputeService.getInstance(); +float[] result = compute.vectorAdd(a, b); +``` + +### Package Structure + +``` +com.hellblazer.luciferase.resource.compute +├── ComputeService # High-level API (start here) +├── GPUBackend # METAL | OPENCL | CPU_FALLBACK +├── BackendSelector # Auto-selects best backend +├── KernelLoader # Load .cl files from resources +└── opencl/ + ├── OpenCLContext # Singleton, manages device/queue + ├── OpenCLBuffer # GPU memory, use try-with-resources + └── OpenCLKernel # Compile and execute kernels +``` + +### ComputeService Methods + +| Method | Signature | Description | +|--------|-----------|-------------| +| `getInstance` | `() → ComputeService` | Singleton | +| `isGPUAvailable` | `() → boolean` | Check before custom ops | +| `getBackend` | `() → GPUBackend` | METAL, OPENCL, or CPU_FALLBACK | +| `vectorAdd` | `(float[], float[]) → float[]` | a + b | +| `saxpy` | `(float, float[], float[]) → float[]` | alpha*x + y | +| `scale` | `(float[], float) → float[]` | data * scalar | +| `sum` | `(float[]) → float` | Sum all | +| `min` | `(float[]) → float` | Minimum | +| `max` | `(float[]) → float` | Maximum | +| `createOperation` | `(String, String, String) → ComputeOperation` | Custom kernel | + +### Patterns + +**Built-in operations:** +```java +var compute = ComputeService.getInstance(); +float[] sum = compute.vectorAdd(a, b); +float[] result = compute.saxpy(2.0f, x, y); +float[] scaled = compute.scale(data, 0.5f); +float total = compute.sum(data); +``` + +**Custom kernel:** +```java +String kernel = """ + __kernel void op(__global const float* in, + __global float* out, + const int size) { + int i = get_global_id(0); + if (i < size) out[i] = in[i] * 2.0f; + } + """; + +try (var op = compute.createOperation("name", kernel, "op")) { + op.setInput(0, inputArray); + op.setOutput(1, size); + op.setArg(2, size); + float[] result = op.execute(size); +} +``` + +**Low-level control:** +```java +try (var kernel = OpenCLKernel.create("name"); + var bufIn = OpenCLBuffer.createWithData(data, READ_ONLY); + var bufOut = OpenCLBuffer.create(size, WRITE_ONLY)) { + + kernel.compile(source, "entryPoint"); + kernel.setBufferArg(0, bufIn, READ); + kernel.setBufferArg(1, bufOut, WRITE); + kernel.setIntArg(2, size); + kernel.execute(size); + kernel.finish(); + bufOut.download(result); +} +``` + +### Kernel Template + +```c +__kernel void name( + __global const float* input, // arg 0 + __global float* output, // arg 1 + const float scalar, // arg 2 + const int size) // arg 3 +{ + int gid = get_global_id(0); + if (gid < size) { // bounds check required + output[gid] = input[gid] * scalar; + } +} +``` + +### Built-in Kernels + +`resources/kernels/opencl/`: + +| File | Functions | +|------|-----------| +| `vector_add.cl` | `vectorAdd` | +| `saxpy.cl` | `saxpy`, `saxpyInPlace` | +| `reduce.cl` | `reduceSum`, `reduceMax`, `reduceMin` | +| `transform.cl` | `scale`, `addScalar`, `clampValues`, `absolute`, `square`, `squareRoot` | + +Load: `KernelLoader.loadOpenCLKernel("vector_add")` + +### Configuration + +| Env Variable | Values | Effect | +|--------------|--------|--------| +| `GPU_BACKEND` | `metal`, `opencl`, `cpu` | Force backend | +| `GPU_DISABLE` | `true` | Disable GPU | + +### Error Handling + +```java +// GPU unavailable - built-ins auto-fallback to CPU +// Custom ops throw IllegalStateException +if (!compute.isGPUAvailable()) { /* handle */ } + +// Compilation +catch (ComputeKernel.KernelCompilationException e) + +// Execution +catch (ComputeKernel.KernelExecutionException e) +``` + +### Common Mistakes + +| Wrong | Right | +|-------|-------| +| `stack.ints(N)` for arg | `clSetKernelArg1i(k, i, N)` | +| No bounds check in kernel | `if (gid < size)` | +| Forgetting `close()` | Use try-with-resources | + +### Key Files + +| Purpose | Path | +|---------|------| +| High-level API | `resource/.../compute/ComputeService.java` | +| Usage guide | `resource/COMPUTE.md` | +| Examples | `resource/.../compute/examples/*.java` | +| Tests | `resource/.../compute/ComputeServiceTest.java` | + +### Testing + +```java +// Skip if no GPU +if (!compute.isGPUAvailable()) return; + +// CI annotation +@DisabledIfEnvironmentVariable(named = "CI", matches = "true") +``` + +### Build + +```bash +./mvnw test -pl resource # Test compute module +./mvnw test -Dtest=ComputeServiceTest # Specific test +``` diff --git a/README.md b/README.md index e6a6a1a..1cbdc32 100644 --- a/README.md +++ b/README.md @@ -14,6 +14,34 @@ The GPU Support Framework consists of two primary modules designed for the [Luci Built on LWJGL 3.3.6, targeting Java 25, and designed for cross-platform GPU compute applications. +## GPU Compute + +Vector operations with automatic GPU acceleration and CPU fallback: + +```java +var compute = ComputeService.getInstance(); + +// Built-in operations +float[] sum = compute.vectorAdd(a, b); +float[] result = compute.saxpy(2.0f, x, y); // 2*x + y +float[] scaled = compute.scale(data, 0.5f); + +// Reductions +float total = compute.sum(data); +float max = compute.max(data); + +// Custom kernels +try (var op = compute.createOperation("multiply", source, "multiply")) { + op.setInput(0, a); + op.setInput(1, b); + op.setOutput(2, size); + op.setArg(3, size); + float[] product = op.execute(size); +} +``` + +See [resource/COMPUTE.md](resource/COMPUTE.md) for the full usage guide. + ## Key Features ### Resource Management @@ -181,8 +209,20 @@ gpu-support/ │ │ ├── ResourceHandle.java │ │ ├── MemoryPool.java │ │ ├── ResourceTracker.java +│ │ ├── compute/ # GPU compute infrastructure +│ │ │ ├── ComputeService.java # High-level API +│ │ │ ├── GPUBackend.java +│ │ │ ├── BackendSelector.java +│ │ │ ├── KernelLoader.java +│ │ │ └── opencl/ # OpenCL implementation +│ │ │ ├── OpenCLContext.java +│ │ │ ├── OpenCLBuffer.java +│ │ │ └── OpenCLKernel.java │ │ ├── opencl/ # OpenCL resource handles │ │ └── opengl/ # OpenGL resource handles +│ ├── src/main/resources/ +│ │ └── kernels/opencl/ # Built-in OpenCL kernels +│ ├── COMPUTE.md # Compute usage guide │ └── README.md │ ├── gpu-test-framework/ # GPU testing infrastructure @@ -346,6 +386,7 @@ The framework includes: ## Documentation ### Primary Documentation +- **[GPU Compute Guide](resource/COMPUTE.md)** - ComputeService API and kernel writing - **[Resource Module README](resource/README.md)** - Detailed resource management guide - **[GPU Test Framework README](gpu-test-framework/README.md)** - Testing framework overview - **[Usage Guide](gpu-test-framework/USAGE_GUIDE.md)** - Comprehensive usage examples diff --git a/gpu-test-framework/src/test/java/com/hellblazer/gpu/test/examples/SimpleMatrixMultiplyTest.java b/gpu-test-framework/src/test/java/com/hellblazer/gpu/test/examples/SimpleMatrixMultiplyTest.java index 3ffaa97..d01d5ef 100644 --- a/gpu-test-framework/src/test/java/com/hellblazer/gpu/test/examples/SimpleMatrixMultiplyTest.java +++ b/gpu-test-framework/src/test/java/com/hellblazer/gpu/test/examples/SimpleMatrixMultiplyTest.java @@ -201,10 +201,10 @@ private float[] multiplyGPU(float[] A, float[] B, int N) throws Exception { checkError(errcode.get(0), "clCreateKernel"); // Set kernel arguments - CL10.clSetKernelArg(kernel, 0, memA); - CL10.clSetKernelArg(kernel, 1, memB); - CL10.clSetKernelArg(kernel, 2, memC); - CL10.clSetKernelArg(kernel, 3, stack.ints(N)); + CL10.clSetKernelArg1p(kernel, 0, memA); + CL10.clSetKernelArg1p(kernel, 1, memB); + CL10.clSetKernelArg1p(kernel, 2, memC); + CL10.clSetKernelArg1i(kernel, 3, N); // Execute kernel var globalWorkSize = stack.pointers(N, N); diff --git a/resource/COMPUTE.md b/resource/COMPUTE.md new file mode 100644 index 0000000..9960c54 --- /dev/null +++ b/resource/COMPUTE.md @@ -0,0 +1,303 @@ +# GPU Compute Guide + +GPU-accelerated vector operations with automatic CPU fallback. + +## Setup + +Add the dependency: + +```xml + + com.hellblazer + luciferase-resource + 1.0.5-SNAPSHOT + +``` + +## Basic Usage + +```java +var compute = ComputeService.getInstance(); + +// Vector addition +float[] a = {1, 2, 3, 4}; +float[] b = {5, 6, 7, 8}; +float[] sum = compute.vectorAdd(a, b); // [6, 8, 10, 12] + +// SAXPY: result = alpha * x + y +float[] result = compute.saxpy(2.0f, a, b); // [7, 10, 13, 16] + +// Scale +float[] scaled = compute.scale(a, 0.5f); // [0.5, 1, 1.5, 2] + +// Reductions +float total = compute.sum(a); // 10 +float max = compute.max(a); // 4 +float min = compute.min(a); // 1 +``` + +The service automatically uses GPU when available, falls back to CPU otherwise. + +## Check Backend + +```java +var compute = ComputeService.getInstance(); + +System.out.println(compute.getBackend()); // METAL, OPENCL, or CPU_FALLBACK +System.out.println(compute.getBackend().isGPU()); // true if GPU backend +System.out.println(compute.isGPUAvailable()); // true if GPU ready +``` + +## Custom Kernels + +Write OpenCL kernels for operations not covered by built-ins: + +```java +String source = """ + __kernel void elementwise_multiply( + __global const float* a, + __global const float* b, + __global float* result, + const int size) + { + int i = get_global_id(0); + if (i < size) { + result[i] = a[i] * b[i]; + } + } + """; + +float[] a = {1, 2, 3, 4}; +float[] b = {2, 3, 4, 5}; + +try (var op = compute.createOperation("multiply", source, "elementwise_multiply")) { + op.setInput(0, a); // arg 0: input array a + op.setInput(1, b); // arg 1: input array b + op.setOutput(2, a.length); // arg 2: output buffer + op.setArg(3, a.length); // arg 3: size parameter + + float[] result = op.execute(a.length); // [2, 6, 12, 20] +} +``` + +### Kernel Argument Types + +```java +op.setInput(index, floatArray); // Read-only float buffer +op.setOutput(index, size); // Write-only output buffer +op.setArg(index, floatValue); // Float scalar +op.setArg(index, intValue); // Int scalar +``` + +### Work Size + +The `execute(workSize)` parameter determines how many work items run. Usually matches your array size: + +```java +op.execute(arrayLength); // 1D execution +op.execute(width * height); // 2D as flat 1D +``` + +## Low-Level API + +For more control, use the OpenCL classes directly: + +```java +import com.hellblazer.luciferase.resource.compute.opencl.*; +import static com.hellblazer.luciferase.resource.compute.opencl.OpenCLBuffer.BufferAccess.*; + +var source = KernelLoader.loadOpenCLKernel("vector_add"); + +try (var kernel = OpenCLKernel.create("vectorAdd"); + var bufA = OpenCLBuffer.createWithData(a, READ_ONLY); + var bufB = OpenCLBuffer.createWithData(b, READ_ONLY); + var bufOut = OpenCLBuffer.create(size, WRITE_ONLY)) { + + kernel.compile(source, "vectorAdd"); + kernel.setBufferArg(0, bufA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufOut, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, size); + + kernel.execute(size); + kernel.finish(); + + var result = new float[size]; + bufOut.download(result); +} +``` + +## Built-in Kernels + +Available in `kernels/opencl/`: + +| File | Functions | Description | +|------|-----------|-------------| +| `vector_add.cl` | `vectorAdd`, `vectorAddInPlace` | Element-wise addition | +| `saxpy.cl` | `saxpy`, `saxpyInPlace` | alpha*x + y | +| `reduce.cl` | `reduceSum`, `reduceMax`, `reduceMin` | Parallel reductions | +| `transform.cl` | `scale`, `scaleInPlace`, `addScalar`, `clampValues`, `absolute`, `square`, `squareRoot` | Element transforms | + +Load them: + +```java +var source = KernelLoader.loadOpenCLKernel("reduce"); +``` + +## Configuration + +### Force Backend + +Environment variable: +```bash +GPU_BACKEND=cpu ./myapp # Force CPU +GPU_BACKEND=opencl ./myapp # Force OpenCL +GPU_BACKEND=metal ./myapp # Force Metal +``` + +### Disable GPU + +```bash +GPU_DISABLE=true ./myapp +``` + +Or system property: +```bash +java -Dgpu.disable=true -jar myapp.jar +``` + +## Error Handling + +```java +// Compilation errors +try { + var op = compute.createOperation("bad", "invalid kernel", "main"); +} catch (ComputeKernel.KernelCompilationException e) { + System.err.println("Compile failed: " + e.getMessage()); +} + +// Execution errors +try { + result = op.execute(size); +} catch (ComputeKernel.KernelExecutionException e) { + System.err.println("Execution failed: " + e.getMessage()); +} + +// GPU unavailable for custom ops +if (!compute.isGPUAvailable()) { + // createOperation() will throw IllegalStateException + // Built-in ops (vectorAdd, etc.) fall back to CPU automatically +} +``` + +## Writing Kernels + +### Basic Pattern + +```c +__kernel void my_operation( + __global const float* input, // Read-only input + __global float* output, // Write-only output + const int size) // Array size +{ + int gid = get_global_id(0); // Current work item index + if (gid < size) { // Bounds check + output[gid] = input[gid] * 2.0f; + } +} +``` + +### Multiple Inputs + +```c +__kernel void blend( + __global const float* a, + __global const float* b, + __global float* result, + const float t, // Blend factor + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = a[gid] * (1.0f - t) + b[gid] * t; + } +} +``` + +### Reductions + +Parallel reductions need local memory: + +```c +__kernel void reduce_sum( + __global const float* input, + __global float* output, + __local float* scratch, // Shared within work group + const int size) +{ + int gid = get_global_id(0); + int lid = get_local_id(0); + int group_size = get_local_size(0); + int group_id = get_group_id(0); + + // Load to local memory + scratch[lid] = (gid < size) ? input[gid] : 0.0f; + barrier(CLK_LOCAL_MEM_FENCE); + + // Parallel reduction + for (int stride = group_size / 2; stride > 0; stride >>= 1) { + if (lid < stride) { + scratch[lid] += scratch[lid + stride]; + } + barrier(CLK_LOCAL_MEM_FENCE); + } + + // Write group result + if (lid == 0) { + output[group_id] = scratch[0]; + } +} +``` + +## Performance Notes + +- Small arrays (< 1000 elements): CPU may be faster due to transfer overhead +- GPU shines with large arrays (10K+ elements) and parallel operations +- Avoid frequent small transfers; batch operations when possible +- Reuse buffers in low-level API for repeated operations + +## Thread Safety + +- `ComputeService.getInstance()` is thread-safe (singleton) +- Built-in operations (`vectorAdd`, etc.) are thread-safe +- `ComputeOperation` instances are not thread-safe; create one per thread +- `OpenCLBuffer` and `OpenCLKernel` are not thread-safe + +## Testing + +```java +@Test +void testWithGPU() { + var compute = ComputeService.getInstance(); + + // Skip if no GPU + if (!compute.isGPUAvailable()) { + return; + } + + float[] a = {1, 2, 3}; + float[] b = {4, 5, 6}; + float[] result = compute.vectorAdd(a, b); + + assertArrayEquals(new float[]{5, 7, 9}, result, 0.0001f); +} +``` + +For CI environments without GPU: + +```java +@DisabledIfEnvironmentVariable(named = "CI", matches = "true") +class GPUOnlyTest { + // Tests that require actual GPU hardware +} +``` diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/BackendSelector.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/BackendSelector.java new file mode 100644 index 0000000..ed9c4d5 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/BackendSelector.java @@ -0,0 +1,219 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.slf4j.Logger; +import org.slf4j.LoggerFactory; + +import java.util.Arrays; +import java.util.Comparator; +import java.util.Optional; + +/** + * Automatic GPU backend selection. + * + *

Selects optimal backend based on platform, availability, and performance characteristics. + * + *

Priority order: + *

    + *
  1. Metal (macOS only, highest performance)
    2. + *
  2. OpenCL (cross-platform)
    3. + *
  3. CPU fallback (always available)
    4. + *
+ * + *

Environment Variables

+ *
    + *
  • {@code GPU_BACKEND} - Force a specific backend ("metal", "opencl", "cpu")
    • + *
  • {@code GPU_DISABLE} - Disable GPU and force CPU fallback ("true" or "1")
    • + *
+ * + *

Legacy ART environment variables are also supported for backwards compatibility: + *

    + *
  • {@code ART_GPU_BACKEND} - Same as GPU_BACKEND
    • + *
  • {@code ART_GPU_DISABLE} - Same as GPU_DISABLE
    • + *
+ * + * @see GPUBackend + */ +public class BackendSelector { + + private static final Logger log = LoggerFactory.getLogger(BackendSelector.class); + + private static GPUBackend selectedBackend = null; + private static boolean initialized = false; + + /** + * Get the optimal GPU backend for the current platform. + * Caches the result after first call. + * + * @return Selected backend + */ + public static GPUBackend getOptimalBackend() { + if (!initialized) { + selectedBackend = selectBackend(); + initialized = true; + + log.debug("GPU Backend Selection: {}", selectedBackend.getDisplayName()); + log.debug("Platform: {}", getPlatformDescription()); + log.debug("CI Environment: {}", isCIEnvironment()); + } + return selectedBackend; + } + + /** + * Select the best available backend. + * + * @return Selected backend + */ + private static GPUBackend selectBackend() { + // Check environment variables for forced selection + var forced = getForcedBackend(); + if (forced.isPresent()) { + var backend = forced.get(); + log.debug("Backend forced via environment: {}", backend); + return backend; + } + + // In CI, use CPU fallback + if (isCIEnvironment()) { + log.debug("CI environment detected, using CPU fallback"); + return GPUBackend.CPU_FALLBACK; + } + + // Check if GPU is disabled + if (isGPUDisabled()) { + log.debug("GPU disabled via environment, using CPU fallback"); + return GPUBackend.CPU_FALLBACK; + } + + // Select highest priority available backend + return Arrays.stream(GPUBackend.values()) + .filter(GPUBackend::isGPU) // GPU backends only + .filter(GPUBackend::isAvailable) + .max(Comparator.comparingInt(GPUBackend::getPriority)) + .orElse(GPUBackend.CPU_FALLBACK); + } + + /** + * Check if a specific backend is forced via environment variable. + * + *

Checks {@code GPU_BACKEND} first, then legacy {@code ART_GPU_BACKEND}. + * + * @return Forced backend, if any + */ + private static Optional getForcedBackend() { + // Check new generic env var first + var backend = System.getenv("GPU_BACKEND"); + + // Fall back to legacy ART env var + if (backend == null) { + backend = System.getenv("ART_GPU_BACKEND"); + if (backend != null) { + log.debug("Using legacy ART_GPU_BACKEND env var (consider using GPU_BACKEND)"); + } + } + + if (backend != null) { + return switch (backend.toLowerCase()) { + case "metal" -> Optional.of(GPUBackend.METAL); + case "opencl" -> Optional.of(GPUBackend.OPENCL); + case "cpu" -> Optional.of(GPUBackend.CPU_FALLBACK); + default -> { + log.warn("Unknown backend specified: {}. Ignoring.", backend); + yield Optional.empty(); + } + }; + } + return Optional.empty(); + } + + /** + * Check if GPU is disabled via environment variable. + * + *

Checks {@code GPU_DISABLE} first, then legacy {@code ART_GPU_DISABLE}. + * + * @return true if GPU is disabled + */ + private static boolean isGPUDisabled() { + // Check new generic env var first + var disabled = System.getenv("GPU_DISABLE"); + + // Fall back to legacy ART env var + if (disabled == null) { + disabled = System.getenv("ART_GPU_DISABLE"); + if (disabled != null) { + log.debug("Using legacy ART_GPU_DISABLE env var (consider using GPU_DISABLE)"); + } + } + + return "true".equalsIgnoreCase(disabled) || "1".equals(disabled); + } + + /** + * Check if running in a CI environment. + * + * @return true if CI environment detected + */ + public static boolean isCIEnvironment() { + return System.getenv("CI") != null || + System.getenv("GITHUB_ACTIONS") != null || + System.getenv("JENKINS_URL") != null || + System.getenv("GITLAB_CI") != null || + System.getenv("TRAVIS") != null || + System.getenv("CIRCLECI") != null; + } + + /** + * Get platform description. + * + * @return Platform description + */ + public static String getPlatformDescription() { + var os = System.getProperty("os.name"); + var arch = System.getProperty("os.arch"); + return String.format("%s %s", os, arch); + } + + /** + * Check if Metal is available on this platform. + * + * @return true if Metal is available + */ + public static boolean isMetalAvailable() { + return GPUBackend.METAL.isAvailable(); + } + + /** + * Check if OpenCL is available on this platform. + * + * @return true if OpenCL is available + */ + public static boolean isOpenCLAvailable() { + return GPUBackend.OPENCL.isAvailable(); + } + + /** + * Get environment information for debugging. + * + * @return Environment description + */ + public static String getEnvironmentInfo() { + var sb = new StringBuilder(); + sb.append("Platform: ").append(getPlatformDescription()).append("\n"); + sb.append("CI: ").append(isCIEnvironment()).append("\n"); + sb.append("Metal Available: ").append(isMetalAvailable()).append("\n"); + sb.append("OpenCL Available: ").append(isOpenCLAvailable()).append("\n"); + sb.append("Selected Backend: ").append(getOptimalBackend().getDisplayName()).append("\n"); + return sb.toString(); + } + + /** + * Reset backend selection (for testing). + * + *

WARNING: This method is for TESTING ONLY. + */ + public static void testReset() { + selectedBackend = null; + initialized = false; + // Also reset GPUBackend availability cache + GPUBackend.testResetAvailability(); + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/ComputeKernel.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/ComputeKernel.java new file mode 100644 index 0000000..fa17ba5 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/ComputeKernel.java @@ -0,0 +1,237 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.lwjgl.PointerBuffer; + +/** + * Unified interface for GPU compute kernels across Metal and OpenCL backends. + * Provides a common API for executing compute operations on the GPU. + * + *

Typical usage: + * 

{@code
+ * try (ComputeKernel kernel = context.createKernel("myKernel")) {
+ *     kernel.compile(kernelSource, "main");
+ *     kernel.setBufferArg(0, inputBuffer, BufferAccess.READ);
+ *     kernel.setBufferArg(1, outputBuffer, BufferAccess.WRITE);
+ *     kernel.setIntArg(2, dataSize);
+ *     kernel.execute(dataSize);
+ *     kernel.finish();
+ * }
+ * }
+ * + * @see GPUBuffer + * @see GPUBackend + */ +public interface ComputeKernel extends AutoCloseable { + + /** + * Compile the kernel from source code. + * + * @param source Kernel source code (Metal or OpenCL) + * @param entryPoint Kernel entry point function name + * @throws KernelCompilationException if compilation fails + */ + void compile(String source, String entryPoint) throws KernelCompilationException; + + /** + * Set a buffer argument for the kernel. + * + * @param index Argument index (0-based) + * @param buffer Buffer to bind + * @param access Access mode (READ, WRITE, READ_WRITE) + * @throws IllegalArgumentException if index is negative or buffer is null + * @throws IllegalStateException if kernel is not compiled + */ + void setBufferArg(int index, GPUBuffer buffer, BufferAccess access); + + /** + * Set a scalar float argument for the kernel. + * + * @param index Argument index (0-based) + * @param value Float value + * @throws IllegalArgumentException if index is negative + * @throws IllegalStateException if kernel is not compiled + */ + void setFloatArg(int index, float value); + + /** + * Set a scalar int argument for the kernel. + * + * @param index Argument index (0-based) + * @param value Int value + * @throws IllegalArgumentException if index is negative + * @throws IllegalStateException if kernel is not compiled + */ + void setIntArg(int index, int value); + + /** + * Execute the kernel with specified global work size (1D). + * + * @param globalWorkSize Number of work items + * @throws KernelExecutionException if execution fails + * @throws IllegalStateException if kernel is not compiled + */ + void execute(int globalWorkSize) throws KernelExecutionException; + + /** + * Execute the kernel with specified global work size (2D). + * + * @param globalWorkSizeX Number of work items in X dimension + * @param globalWorkSizeY Number of work items in Y dimension + * @throws KernelExecutionException if execution fails + * @throws IllegalStateException if kernel is not compiled + */ + void execute(int globalWorkSizeX, int globalWorkSizeY) throws KernelExecutionException; + + /** + * Execute the kernel with specified global work size (3D). + * + * @param globalWorkSizeX Number of work items in X dimension + * @param globalWorkSizeY Number of work items in Y dimension + * @param globalWorkSizeZ Number of work items in Z dimension + * @throws KernelExecutionException if execution fails + * @throws IllegalStateException if kernel is not compiled + */ + void execute(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ) + throws KernelExecutionException; + + /** + * Execute the kernel with explicit local work group sizes (3D). + * Allows tuning work group dimensions for optimal GPU occupancy. + * + *

For Apple M4 Max optimal configurations: + *

    + *
  • Total work items per group: localX × localY × localZ ≤ 1024
    • + *
  • Network dimension (X): 256-512 for best compute/memory balance
    • + *
  • Batch dimension (Y): 2-8 to reach 1024 total work items
    • + *
+ * + * @param globalWorkSizeX Number of work items in X dimension + * @param globalWorkSizeY Number of work items in Y dimension + * @param globalWorkSizeZ Number of work items in Z dimension + * @param localWorkSizeX Work group size in X dimension + * @param localWorkSizeY Work group size in Y dimension + * @param localWorkSizeZ Work group size in Z dimension + * @throws KernelExecutionException if execution fails + * @throws IllegalStateException if kernel is not compiled + */ + void execute(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ, + int localWorkSizeX, int localWorkSizeY, int localWorkSizeZ) + throws KernelExecutionException; + + /** + * Execute kernel asynchronously with event-based synchronization. + * + *

This method enables compute/transfer pipelining by passing event handles + * to OpenCL/Metal for proper dependency management. Events allow operations + * to execute in parallel while maintaining correct ordering. + * + *

Event Lifecycle

+ *
    + *
  • waitEvents: Events to wait on before execution (caller owns, we read)
    • + *
  • signalEvent: Event to signal when complete (we create, caller must release)
    • + *
+ * + *

Usage Example

+ * 
{@code
+     * try (var stack = MemoryStack.stackPush()) {
+     *     var uploadEvent = stack.mallocPointer(1);
+     *     var computeEvent = stack.mallocPointer(1);
+     *
+     *     // Upload data (creates uploadEvent)
+     *     clEnqueueWriteBuffer(..., uploadEvent);
+     *
+     *     // Compute waits on upload, signals computeEvent
+     *     kernel.executeAsync(batchSize, 1, 1, uploadEvent, computeEvent);
+     *
+     *     // Download waits on compute
+     *     clEnqueueReadBuffer(..., computeEvent, null);
+     *
+     *     // Cleanup events
+     *     clReleaseEvent(uploadEvent.get(0));
+     *     clReleaseEvent(computeEvent.get(0));
+     * }
+     * }
+ * + *

IMPORTANT: Caller is responsible for releasing signalEvent via + * {@code clReleaseEvent(signalEvent.get(0))} to prevent resource leaks. + * + *

Note: If both waitEvents and signalEvent are null, this behaves + * identically to {@link #execute(int, int, int)} (blocking execution). + * + * @param globalWorkSizeX Number of work items in X dimension + * @param globalWorkSizeY Number of work items in Y dimension + * @param globalWorkSizeZ Number of work items in Z dimension + * @param waitEvents Events to wait on before execution (null = no wait) + * @param signalEvent Event to signal on completion (null = no signal) + * @throws KernelExecutionException if execution fails + * @throws IllegalStateException if kernel is not compiled + * @see OpenCL Event Objects + */ + void executeAsync(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ, + PointerBuffer waitEvents, PointerBuffer signalEvent) + throws KernelExecutionException; + + /** + * Wait for kernel execution to complete. + * Blocks until all queued operations finish. + */ + void finish(); + + /** + * Get the backend type for this kernel. + * + * @return Backend type (METAL, OPENCL, or CPU_FALLBACK) + */ + GPUBackend getBackend(); + + /** + * Check if the kernel is compiled and ready to execute. + * + * @return true if kernel is compiled + */ + boolean isCompiled(); + + /** + * Release GPU resources. + */ + @Override + void close(); + + /** + * Buffer access modes for kernel arguments. + */ + enum BufferAccess { + /** Buffer is read-only in kernel */ + READ, + /** Buffer is write-only in kernel */ + WRITE, + /** Buffer is read-write in kernel */ + READ_WRITE + } + + /** + * Exception thrown when kernel compilation fails. + */ + class KernelCompilationException extends Exception { + public KernelCompilationException(String message) { + super(message); + } + + public KernelCompilationException(String message, Throwable cause) { + super(message, cause); + } + } + + /** + * Exception thrown when kernel execution fails. + */ + class KernelExecutionException extends Exception { + public KernelExecutionException(String message) { + super(message); + } + + public KernelExecutionException(String message, Throwable cause) { + super(message, cause); + } + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/ComputeService.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/ComputeService.java new file mode 100644 index 0000000..14fadad --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/ComputeService.java @@ -0,0 +1,448 @@ +package com.hellblazer.luciferase.resource.compute; + +import com.hellblazer.luciferase.resource.compute.opencl.OpenCLBuffer; +import com.hellblazer.luciferase.resource.compute.opencl.OpenCLContext; +import com.hellblazer.luciferase.resource.compute.opencl.OpenCLKernel; +import org.slf4j.Logger; +import org.slf4j.LoggerFactory; + +import java.util.concurrent.atomic.AtomicBoolean; + +/** + * High-level facade for GPU compute operations. + * + *

Provides a simplified API for common GPU compute tasks, hiding the complexity + * of context management, buffer allocation, and kernel execution. + * + *

Features

+ *
    + *
  • Automatic backend selection (GPU with CPU fallback)
    • + *
  • Built-in kernels for common operations (SAXPY, reduce, transform)
    • + *
  • Simplified buffer management
    • + *
  • Thread-safe singleton access
    • + *
+ * + *

Usage

+ * 
{@code
+ * var compute = ComputeService.getInstance();
+ *
+ * // Simple SAXPY: result = 2.0 * x + y
+ * float[] result = compute.saxpy(2.0f, x, y);
+ *
+ * // Vector addition
+ * float[] sum = compute.vectorAdd(a, b);
+ *
+ * // Custom kernel
+ * try (var op = compute.createOperation("myKernel", kernelSource, "entryPoint")) {
+ *     op.setInput(0, inputData);
+ *     op.setOutput(1, outputSize);
+ *     op.setArg(2, 42);
+ *     float[] result = op.execute(workSize);
+ * }
+ * }
+ * + * @see GPUBackend + * @see BackendSelector + */ +public class ComputeService { + + private static final Logger log = LoggerFactory.getLogger(ComputeService.class); + + private static volatile ComputeService INSTANCE; + private static final Object LOCK = new Object(); + + private final GPUBackend backend; + private final AtomicBoolean initialized = new AtomicBoolean(false); + + // Cached kernel sources + private String vectorAddSource; + private String saxpySource; + private String transformSource; + + private ComputeService() { + this.backend = BackendSelector.getOptimalBackend(); + log.info("ComputeService initialized with backend: {}", backend.getDisplayName()); + } + + /** + * Get the singleton instance. + * + * @return ComputeService instance + */ + public static ComputeService getInstance() { + if (INSTANCE == null) { + synchronized (LOCK) { + if (INSTANCE == null) { + INSTANCE = new ComputeService(); + } + } + } + return INSTANCE; + } + + /** + * Check if GPU compute is available. + * + * @return true if GPU backend is available + */ + public boolean isGPUAvailable() { + return backend.isGPU() && backend.isAvailable(); + } + + /** + * Get the active backend. + * + * @return Current backend + */ + public GPUBackend getBackend() { + return backend; + } + + // ========== Vector Operations ========== + + /** + * Vector addition: result[i] = a[i] + b[i] + * + * @param a First vector + * @param b Second vector (must be same length as a) + * @return Result vector + * @throws IllegalArgumentException if vectors have different lengths + */ + public float[] vectorAdd(float[] a, float[] b) { + if (a.length != b.length) { + throw new IllegalArgumentException("Vectors must have same length"); + } + + if (!isGPUAvailable()) { + return vectorAddCPU(a, b); + } + + try { + return vectorAddGPU(a, b); + } catch (Exception e) { + log.warn("GPU vectorAdd failed, falling back to CPU: {}", e.getMessage()); + return vectorAddCPU(a, b); + } + } + + /** + * SAXPY: result[i] = alpha * x[i] + y[i] + * + * @param alpha Scalar multiplier + * @param x First vector + * @param y Second vector (must be same length as x) + * @return Result vector + * @throws IllegalArgumentException if vectors have different lengths + */ + public float[] saxpy(float alpha, float[] x, float[] y) { + if (x.length != y.length) { + throw new IllegalArgumentException("Vectors must have same length"); + } + + if (!isGPUAvailable()) { + return saxpyCPU(alpha, x, y); + } + + try { + return saxpyGPU(alpha, x, y); + } catch (Exception e) { + log.warn("GPU saxpy failed, falling back to CPU: {}", e.getMessage()); + return saxpyCPU(alpha, x, y); + } + } + + /** + * Scale vector: result[i] = data[i] * scale + * + * @param data Input vector + * @param scale Scalar multiplier + * @return Scaled vector + */ + public float[] scale(float[] data, float scale) { + if (!isGPUAvailable()) { + return scaleCPU(data, scale); + } + + try { + return scaleGPU(data, scale); + } catch (Exception e) { + log.warn("GPU scale failed, falling back to CPU: {}", e.getMessage()); + return scaleCPU(data, scale); + } + } + + // ========== Reduction Operations ========== + + /** + * Compute sum of all elements. + * + * @param data Input array + * @return Sum of all elements + */ + public float sum(float[] data) { + // For now, CPU implementation (GPU reduction needs multiple passes) + float sum = 0; + for (float v : data) { + sum += v; + } + return sum; + } + + /** + * Find maximum value. + * + * @param data Input array + * @return Maximum value + */ + public float max(float[] data) { + float max = Float.NEGATIVE_INFINITY; + for (float v : data) { + if (v > max) max = v; + } + return max; + } + + /** + * Find minimum value. + * + * @param data Input array + * @return Minimum value + */ + public float min(float[] data) { + float min = Float.POSITIVE_INFINITY; + for (float v : data) { + if (v < min) min = v; + } + return min; + } + + // ========== Custom Operations ========== + + /** + * Create a custom compute operation. + * + * @param name Operation name (for logging) + * @param source Kernel source code + * @param entryPoint Kernel entry point function name + * @return Compute operation builder + * @throws ComputeKernel.KernelCompilationException if compilation fails + */ + public ComputeOperation createOperation(String name, String source, String entryPoint) + throws ComputeKernel.KernelCompilationException { + if (!isGPUAvailable()) { + throw new IllegalStateException("GPU not available for custom operations"); + } + return new ComputeOperation(name, source, entryPoint); + } + + // ========== GPU Implementations ========== + + private float[] vectorAddGPU(float[] a, float[] b) throws Exception { + ensureVectorAddKernel(); + int size = a.length; + + try (var bufferA = OpenCLBuffer.createWithData(a, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferB = OpenCLBuffer.createWithData(b, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var kernel = OpenCLKernel.create("vectorAdd")) { + + kernel.compile(vectorAddSource, "vectorAdd"); + kernel.setBufferArg(0, bufferA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, size); + + kernel.execute(size); + kernel.finish(); + + var result = new float[size]; + bufferResult.download(result); + return result; + } + } + + private float[] saxpyGPU(float alpha, float[] x, float[] y) throws Exception { + ensureSaxpyKernel(); + int size = x.length; + + try (var bufferX = OpenCLBuffer.createWithData(x, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferY = OpenCLBuffer.createWithData(y, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var kernel = OpenCLKernel.create("saxpy")) { + + kernel.compile(saxpySource, "saxpy"); + kernel.setBufferArg(0, bufferX, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferY, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setFloatArg(3, alpha); + kernel.setIntArg(4, size); + + kernel.execute(size); + kernel.finish(); + + var result = new float[size]; + bufferResult.download(result); + return result; + } + } + + private float[] scaleGPU(float[] data, float scale) throws Exception { + ensureTransformKernel(); + int size = data.length; + + try (var bufferIn = OpenCLBuffer.createWithData(data, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferOut = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var kernel = OpenCLKernel.create("scale")) { + + kernel.compile(transformSource, "scale"); + kernel.setBufferArg(0, bufferIn, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferOut, ComputeKernel.BufferAccess.WRITE); + kernel.setFloatArg(2, scale); + kernel.setIntArg(3, size); + + kernel.execute(size); + kernel.finish(); + + var result = new float[size]; + bufferOut.download(result); + return result; + } + } + + // ========== CPU Fallback Implementations ========== + + private float[] vectorAddCPU(float[] a, float[] b) { + var result = new float[a.length]; + for (int i = 0; i < a.length; i++) { + result[i] = a[i] + b[i]; + } + return result; + } + + private float[] saxpyCPU(float alpha, float[] x, float[] y) { + var result = new float[x.length]; + for (int i = 0; i < x.length; i++) { + result[i] = alpha * x[i] + y[i]; + } + return result; + } + + private float[] scaleCPU(float[] data, float scale) { + var result = new float[data.length]; + for (int i = 0; i < data.length; i++) { + result[i] = data[i] * scale; + } + return result; + } + + // ========== Kernel Loading ========== + + private void ensureVectorAddKernel() { + if (vectorAddSource == null) { + vectorAddSource = KernelLoader.loadOpenCLKernel("vector_add"); + } + } + + private void ensureSaxpyKernel() { + if (saxpySource == null) { + saxpySource = KernelLoader.loadOpenCLKernel("saxpy"); + } + } + + private void ensureTransformKernel() { + if (transformSource == null) { + transformSource = KernelLoader.loadOpenCLKernel("transform"); + } + } + + /** + * Reset for testing. + */ + public static void testReset() { + synchronized (LOCK) { + INSTANCE = null; + } + BackendSelector.testReset(); + } + + // ========== Custom Operation Builder ========== + + /** + * Builder for custom compute operations. + */ + public static class ComputeOperation implements AutoCloseable { + private final OpenCLKernel kernel; + private final java.util.Map buffers = new java.util.HashMap<>(); + private int outputIndex = -1; + private int outputSize = 0; + + ComputeOperation(String name, String source, String entryPoint) + throws ComputeKernel.KernelCompilationException { + this.kernel = OpenCLKernel.create(name); + this.kernel.compile(source, entryPoint); + } + + /** + * Set input buffer at argument index. + */ + public ComputeOperation setInput(int index, float[] data) { + var buffer = OpenCLBuffer.createWithData(data, OpenCLBuffer.BufferAccess.READ_ONLY); + buffers.put(index, buffer); + kernel.setBufferArg(index, buffer, ComputeKernel.BufferAccess.READ); + return this; + } + + /** + * Set output buffer at argument index. + */ + public ComputeOperation setOutput(int index, int size) { + var buffer = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + buffers.put(index, buffer); + kernel.setBufferArg(index, buffer, ComputeKernel.BufferAccess.WRITE); + this.outputIndex = index; + this.outputSize = size; + return this; + } + + /** + * Set float argument. + */ + public ComputeOperation setArg(int index, float value) { + kernel.setFloatArg(index, value); + return this; + } + + /** + * Set int argument. + */ + public ComputeOperation setArg(int index, int value) { + kernel.setIntArg(index, value); + return this; + } + + /** + * Execute and return output. + */ + public float[] execute(int workSize) throws ComputeKernel.KernelExecutionException { + if (outputIndex < 0) { + throw new IllegalStateException("No output buffer set"); + } + + kernel.execute(workSize); + kernel.finish(); + + var result = new float[outputSize]; + buffers.get(outputIndex).download(result); + return result; + } + + @Override + public void close() { + kernel.close(); + for (var buffer : buffers.values()) { + buffer.close(); + } + buffers.clear(); + } + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUBackend.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUBackend.java new file mode 100644 index 0000000..53735ca --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUBackend.java @@ -0,0 +1,212 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.slf4j.Logger; +import org.slf4j.LoggerFactory; + +import static org.lwjgl.opencl.CL10.*; + +/** + * Supported GPU compute backends. + * + *

Priority ordering (higher = preferred): + *

    + *
  1. METAL (100) - macOS only, highest performance
    2. + *
  2. OPENCL (90) - cross-platform
    3. + *
  3. CPU_FALLBACK (10) - always available
    4. + *
+ * + *

Use {@link BackendSelector#getOptimalBackend()} for automatic selection + * based on platform availability. + * + * @see BackendSelector + */ +public enum GPUBackend { + /** + * Metal 3 (macOS only, highest performance). + */ + METAL("Metal", 100, true), + + /** + * OpenCL 1.2+ (cross-platform). + */ + OPENCL("OpenCL", 90, true), + + /** + * CPU fallback (no GPU required). + */ + CPU_FALLBACK("CPU Fallback", 10, false); + + private final String displayName; + private final int priority; + private final boolean isGPU; + + GPUBackend(String displayName, int priority, boolean isGPU) { + this.displayName = displayName; + this.priority = priority; + this.isGPU = isGPU; + } + + /** + * Get human-readable display name. + * + * @return Display name + */ + public String getDisplayName() { + return displayName; + } + + /** + * Get priority for automatic backend selection. + * Higher values are preferred. + * + * @return Priority value + */ + public int getPriority() { + return priority; + } + + /** + * Check if this is a GPU backend (vs CPU fallback). + * + * @return true if GPU-accelerated + */ + public boolean isGPU() { + return isGPU; + } + + /** + * Check if this backend is available on the current platform. + * + *

Availability checks are cached after first call. + * + * @return true if this backend can be used + */ + public boolean isAvailable() { + return switch (this) { + case METAL -> MetalDetector.isAvailable(); + case OPENCL -> OpenCLDetector.isAvailable(); + case CPU_FALLBACK -> true; + }; + } + + /** + * Reset cached availability state for testing. + * Forces re-detection on next isAvailable() call. + * + *

WARNING: This method is for TESTING ONLY. + */ + public static void testResetAvailability() { + MetalDetector.reset(); + OpenCLDetector.reset(); + } + + /** + * Metal availability detector. + * Metal is only available on macOS. + */ + private static final class MetalDetector { + private static final Logger log = LoggerFactory.getLogger(MetalDetector.class); + private static volatile Boolean available; + + static boolean isAvailable() { + if (available == null) { + synchronized (MetalDetector.class) { + if (available == null) { + available = detectMetal(); + } + } + } + return available; + } + + static void reset() { + synchronized (MetalDetector.class) { + available = null; + } + } + + private static boolean detectMetal() { + var os = System.getProperty("os.name", "").toLowerCase(); + var isMacOS = os.contains("mac"); + if (!isMacOS) { + log.debug("Metal not available - not macOS (os.name={})", os); + return false; + } + // Note: Full Metal detection would require native calls + // For now, assume Metal available on macOS 10.14+ + log.debug("Metal potentially available on macOS"); + return true; + } + } + + /** + * OpenCL availability detector. + * Uses lightweight probe that doesn't require full context initialization. + */ + private static final class OpenCLDetector { + private static final Logger log = LoggerFactory.getLogger(OpenCLDetector.class); + private static volatile Boolean available; + + static boolean isAvailable() { + if (available == null) { + synchronized (OpenCLDetector.class) { + if (available == null) { + available = detectOpenCL(); + } + } + } + return available; + } + + static void reset() { + synchronized (OpenCLDetector.class) { + available = null; + } + } + + private static boolean detectOpenCL() { + // Check if disabled via system property + if (Boolean.getBoolean("gpu.disable") || + Boolean.getBoolean("luciferase.gpu.disable") || + Boolean.getBoolean("art.gpu.disable")) { + log.debug("OpenCL disabled via system property"); + return false; + } + + try { + // Try to enumerate platforms using MemoryStack + // NOTE: We deliberately do NOT call CL.create() here because it causes + // SIGSEGV crashes in Apple's GPU drivers when running in forked JVM processes + // (like Maven Surefire). Instead, we directly call clGetPlatformIDs which + // works correctly and detects OpenCL availability. + try (var stack = org.lwjgl.system.MemoryStack.stackPush()) { + var numPlatforms = stack.mallocInt(1); + var result = clGetPlatformIDs((org.lwjgl.PointerBuffer) null, numPlatforms); + + if (result != CL_SUCCESS) { + log.debug("OpenCL not available - clGetPlatformIDs returned {}", result); + return false; + } + + if (numPlatforms.get(0) == 0) { + log.debug("OpenCL not available - no platforms found"); + return false; + } + + log.debug("OpenCL available - found {} platform(s)", numPlatforms.get(0)); + return true; + } + + } catch (UnsatisfiedLinkError e) { + log.debug("OpenCL not available - native library not found: {}", e.getMessage()); + return false; + } catch (NoClassDefFoundError e) { + log.debug("OpenCL not available - LWJGL OpenCL classes not found: {}", e.getMessage()); + return false; + } catch (Exception e) { + log.debug("OpenCL not available - detection failed: {}", e.getMessage()); + return false; + } + } + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUBuffer.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUBuffer.java new file mode 100644 index 0000000..6a1531e --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUBuffer.java @@ -0,0 +1,90 @@ +package com.hellblazer.luciferase.resource.compute; + +import java.nio.FloatBuffer; + +/** + * GPU buffer abstraction for compute backends (OpenCL, Metal). + * Manages host-device memory transfers and buffer lifecycle. + * + *

This interface provides a backend-agnostic API for GPU buffer operations. + * Implementations exist for OpenCL ({@code OpenCLBuffer}) with Metal stubs + * for future expansion. + * + *

Typical usage: + * 

{@code
+ * try (GPUBuffer buffer = context.createBuffer(1024)) {
+ *     buffer.upload(hostData);
+ *     // ... GPU computation ...
+ *     buffer.download(results);
+ * }
+ * }
+ * + * @see com.hellblazer.luciferase.resource.compute.opencl.OpenCLBuffer + */ +public interface GPUBuffer extends AutoCloseable { + + /** + * Upload data from host to device. + * + * @param data Host data to upload + * @throws IllegalStateException if buffer is not valid + * @throws IllegalArgumentException if data size exceeds buffer capacity + */ + void upload(FloatBuffer data); + + /** + * Upload data from host to device. + * + * @param data Host data to upload + * @throws IllegalStateException if buffer is not valid + * @throws IllegalArgumentException if data size exceeds buffer capacity + */ + void upload(float[] data); + + /** + * Download data from device to host. + * + * @param data Host buffer to receive data + * @throws IllegalStateException if buffer is not valid + * @throws IllegalArgumentException if buffer capacity is insufficient + */ + void download(FloatBuffer data); + + /** + * Download data from device to host. + * + * @param data Host array to receive data + * @throws IllegalStateException if buffer is not valid + * @throws IllegalArgumentException if array size is insufficient + */ + void download(float[] data); + + /** + * Get the size of the buffer in elements (floats). + * + * @return Buffer size in float elements + */ + int size(); + + /** + * Get the size of the buffer in bytes. + * + * @return Buffer size in bytes + */ + int sizeInBytes(); + + /** + * Check if the buffer is valid and allocated. + * A buffer becomes invalid after {@link #close()} is called. + * + * @return true if buffer is valid and can be used + */ + boolean isValid(); + + /** + * Release GPU resources. + * After calling this method, {@link #isValid()} returns false. + */ + @Override + void close(); +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUErrorClassifier.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUErrorClassifier.java new file mode 100644 index 0000000..bbf6255 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/GPUErrorClassifier.java @@ -0,0 +1,262 @@ +package com.hellblazer.luciferase.resource.compute; + +import java.util.Set; +import java.util.regex.Pattern; + +/** + * Classifies GPU exceptions to determine appropriate error handling strategy. + * + *

Exception categories: + *

    + *
  • Programming errors: Bugs that should fail fast (never fallback)
    • + *
  • Recoverable errors: Transient issues that can fallback to CPU
    • + *
+ * + *

Usage: + * 

{@code
+ * try {
+ *     return processGPU(input);
+ * } catch (Exception e) {
+ *     if (GPUErrorClassifier.isProgrammingError(e)) {
+ *         throw new IllegalStateException("GPU programming error", e);
+ *     }
+ *     log.warn("Recoverable GPU error, falling back: {}",
+ *              GPUErrorClassifier.getErrorCategory(e));
+ *     return processCPU(input);
+ * }
+ * }
+ * + * @see ComputeKernel.KernelCompilationException + * @see ComputeKernel.KernelExecutionException + */ +public final class GPUErrorClassifier { + + private GPUErrorClassifier() {} // Utility class + + // OpenCL error codes that indicate programming errors + private static final Set PROGRAMMING_ERROR_CODES = Set.of( + -11, // CL_BUILD_PROGRAM_FAILURE + -44, // CL_INVALID_PROGRAM + -45, // CL_INVALID_PROGRAM_EXECUTABLE + -46, // CL_INVALID_KERNEL_NAME + -47, // CL_INVALID_KERNEL_DEFINITION + -48, // CL_INVALID_KERNEL + -49, // CL_INVALID_ARG_INDEX + -50, // CL_INVALID_ARG_VALUE + -51, // CL_INVALID_ARG_SIZE + -52, // CL_INVALID_KERNEL_ARGS + -53, // CL_INVALID_WORK_DIMENSION + -54, // CL_INVALID_WORK_GROUP_SIZE + -55, // CL_INVALID_WORK_ITEM_SIZE + -56, // CL_INVALID_GLOBAL_OFFSET + -57, // CL_INVALID_EVENT_WAIT_LIST + -58, // CL_INVALID_EVENT + -30, // CL_INVALID_VALUE + -33, // CL_INVALID_DEVICE + -34, // CL_INVALID_CONTEXT + -35, // CL_INVALID_QUEUE_PROPERTIES + -36, // CL_INVALID_COMMAND_QUEUE + -37, // CL_INVALID_HOST_PTR + -38, // CL_INVALID_MEM_OBJECT + -63 // CL_INVALID_GLOBAL_WORK_SIZE + ); + + // OpenCL error codes that indicate recoverable errors + private static final Set RECOVERABLE_ERROR_CODES = Set.of( + -4, // CL_MEM_OBJECT_ALLOCATION_FAILURE + -5, // CL_OUT_OF_RESOURCES + -6, // CL_OUT_OF_HOST_MEMORY + -59 // CL_INVALID_OPERATION (context-dependent) + ); + + // Pattern to extract error codes from exception messages + private static final Pattern ERROR_CODE_PATTERN = + Pattern.compile("error code[:\\s]+(-?\\d+)", Pattern.CASE_INSENSITIVE); + + /** + * Determine if an exception represents a programming error. + * Programming errors should fail fast and not be masked by CPU fallback. + * + * @param t the throwable to classify + * @return true if this is a programming error that should be rethrown + */ + public static boolean isProgrammingError(Throwable t) { + if (t == null) { + return false; + } + + // Check exception type first + if (t instanceof IllegalStateException || + t instanceof IllegalArgumentException || + t instanceof NullPointerException || + t instanceof AssertionError || + t instanceof ComputeKernel.KernelCompilationException) { + return true; + } + + // Check for RuntimeException wrapping programming errors + if (t instanceof RuntimeException && t.getCause() != null && t.getCause() != t) { + if (isProgrammingError(t.getCause())) { + return true; + } + } + + // Extract and check OpenCL error code from message + var errorCode = extractOpenCLErrorCode(t); + if (errorCode != 0 && PROGRAMMING_ERROR_CODES.contains(errorCode)) { + return true; + } + + // Check cause chain + if (t.getCause() != null && t.getCause() != t) { + return isProgrammingError(t.getCause()); + } + + return false; + } + + /** + * Determine if an exception represents a recoverable error. + * Recoverable errors can legitimately fall back to CPU processing. + * + * @param t the throwable to classify + * @return true if this error can be recovered by CPU fallback + */ + public static boolean isRecoverable(Throwable t) { + if (t == null) { + return false; + } + + // Programming errors are never recoverable + if (isProgrammingError(t)) { + return false; + } + + // Check for specific recoverable error codes + var errorCode = extractOpenCLErrorCode(t); + if (errorCode != 0 && RECOVERABLE_ERROR_CODES.contains(errorCode)) { + return true; + } + + // OutOfMemoryError related to GPU is recoverable + if (t instanceof OutOfMemoryError) { + var message = t.getMessage(); + if (message != null && + (message.contains("GPU") || message.contains("OpenCL") || + message.contains("buffer"))) { + return true; + } + } + + // KernelExecutionException without programming error code is recoverable + if (t instanceof ComputeKernel.KernelExecutionException) { + return true; + } + + // Check cause chain + if (t.getCause() != null && t.getCause() != t) { + return isRecoverable(t.getCause()); + } + + // Default: unknown errors are recoverable (conservative approach) + return true; + } + + /** + * Get a human-readable description of the error category. + * + * @param t the throwable to describe + * @return descriptive string for logging + */ + public static String getErrorCategory(Throwable t) { + if (t == null) { + return "null exception"; + } + + var errorCode = extractOpenCLErrorCode(t); + var codeDesc = errorCode != 0 ? " (CL error " + errorCode + ")" : ""; + + if (isProgrammingError(t)) { + return "PROGRAMMING ERROR" + codeDesc + ": " + t.getClass().getSimpleName(); + } + + if (errorCode != 0 && RECOVERABLE_ERROR_CODES.contains(errorCode)) { + return switch (errorCode) { + case -4 -> "MEMORY ALLOCATION FAILURE" + codeDesc; + case -5 -> "OUT OF GPU RESOURCES" + codeDesc; + case -6 -> "OUT OF HOST MEMORY" + codeDesc; + case -59 -> "INVALID OPERATION" + codeDesc; + default -> "RECOVERABLE ERROR" + codeDesc; + }; + } + + return "UNKNOWN ERROR" + codeDesc + ": " + t.getClass().getSimpleName(); + } + + /** + * Get the OpenCL error code name for logging. + * + * @param errorCode the OpenCL error code + * @return human-readable error name + */ + public static String getOpenCLErrorName(int errorCode) { + return switch (errorCode) { + case 0 -> "CL_SUCCESS"; + case -4 -> "CL_MEM_OBJECT_ALLOCATION_FAILURE"; + case -5 -> "CL_OUT_OF_RESOURCES"; + case -6 -> "CL_OUT_OF_HOST_MEMORY"; + case -11 -> "CL_BUILD_PROGRAM_FAILURE"; + case -30 -> "CL_INVALID_VALUE"; + case -33 -> "CL_INVALID_DEVICE"; + case -34 -> "CL_INVALID_CONTEXT"; + case -35 -> "CL_INVALID_QUEUE_PROPERTIES"; + case -36 -> "CL_INVALID_COMMAND_QUEUE"; + case -37 -> "CL_INVALID_HOST_PTR"; + case -38 -> "CL_INVALID_MEM_OBJECT"; + case -44 -> "CL_INVALID_PROGRAM"; + case -45 -> "CL_INVALID_PROGRAM_EXECUTABLE"; + case -46 -> "CL_INVALID_KERNEL_NAME"; + case -47 -> "CL_INVALID_KERNEL_DEFINITION"; + case -48 -> "CL_INVALID_KERNEL"; + case -49 -> "CL_INVALID_ARG_INDEX"; + case -50 -> "CL_INVALID_ARG_VALUE"; + case -51 -> "CL_INVALID_ARG_SIZE"; + case -52 -> "CL_INVALID_KERNEL_ARGS"; + case -53 -> "CL_INVALID_WORK_DIMENSION"; + case -54 -> "CL_INVALID_WORK_GROUP_SIZE"; + case -55 -> "CL_INVALID_WORK_ITEM_SIZE"; + case -56 -> "CL_INVALID_GLOBAL_OFFSET"; + case -57 -> "CL_INVALID_EVENT_WAIT_LIST"; + case -58 -> "CL_INVALID_EVENT"; + case -59 -> "CL_INVALID_OPERATION"; + case -63 -> "CL_INVALID_GLOBAL_WORK_SIZE"; + default -> "UNKNOWN_CL_ERROR_" + errorCode; + }; + } + + /** + * Extract OpenCL error code from exception message. + * Searches the exception chain for patterns like "error code: -49". + * + * @param t the throwable to search + * @return the error code, or 0 if not found + */ + static int extractOpenCLErrorCode(Throwable t) { + var current = t; + while (current != null) { + var message = current.getMessage(); + if (message != null) { + var matcher = ERROR_CODE_PATTERN.matcher(message); + if (matcher.find()) { + try { + return Integer.parseInt(matcher.group(1)); + } catch (NumberFormatException e) { + // Continue searching + } + } + } + current = (current.getCause() != current) ? current.getCause() : null; + } + return 0; + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/KernelLoader.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/KernelLoader.java new file mode 100644 index 0000000..0547ff8 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/KernelLoader.java @@ -0,0 +1,182 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.slf4j.Logger; +import org.slf4j.LoggerFactory; + +import java.io.IOException; +import java.io.InputStream; +import java.nio.charset.StandardCharsets; +import java.util.concurrent.ConcurrentHashMap; + +/** + * Utility for loading GPU kernel source code from classpath resources. + * + *

Provides caching to avoid repeated I/O operations and convenience methods + * for loading kernels following standard path conventions. + * + *

Path Conventions

+ *
    + *
  • OpenCL: {@code kernels/opencl/{name}.cl}
    • + *
  • Metal: {@code kernels/metal/{name}.metal}
    • + *
  • Generic: {@code kernels/{name}.cl} (for test kernels)
    • + *
+ * + *

Usage

+ * 
{@code
+ * // Load OpenCL kernel following convention
+ * String source = KernelLoader.loadOpenCLKernel("vector_add");
+ * // Loads from: kernels/opencl/vector_add.cl
+ *
+ * // Load from explicit path
+ * String source = KernelLoader.loadKernel("kernels/my_kernel.cl");
+ *
+ * // Load from test convention (flat structure)
+ * String source = KernelLoader.loadTestKernel("vector_add");
+ * // Loads from: kernels/vector_add.cl
+ * }
+ * + * @see com.hellblazer.luciferase.resource.compute.opencl.OpenCLKernel + */ +public class KernelLoader { + + private static final Logger log = LoggerFactory.getLogger(KernelLoader.class); + + private static final ConcurrentHashMap cache = new ConcurrentHashMap<>(); + + /** + * Load an OpenCL kernel from the standard location. + * + *

Follows convention: {@code kernels/opencl/{kernelName}.cl} + * + * @param kernelName Kernel name without path or extension (e.g., "vector_add") + * @return Kernel source code + * @throws KernelLoadException if kernel cannot be loaded + */ + public static String loadOpenCLKernel(String kernelName) { + return loadKernel("kernels/opencl/" + kernelName + ".cl"); + } + + /** + * Load a Metal kernel from the standard location. + * + *

Follows convention: {@code kernels/metal/{kernelName}.metal} + * + * @param kernelName Kernel name without path or extension (e.g., "vector_add") + * @return Kernel source code + * @throws KernelLoadException if kernel cannot be loaded + */ + public static String loadMetalKernel(String kernelName) { + return loadKernel("kernels/metal/" + kernelName + ".metal"); + } + + /** + * Load a test kernel from the flat test location. + * + *

Follows convention: {@code kernels/{kernelName}.cl} + * + *

This is the convention used by gpu-test-framework test kernels. + * + * @param kernelName Kernel name without path or extension (e.g., "vector_add") + * @return Kernel source code + * @throws KernelLoadException if kernel cannot be loaded + */ + public static String loadTestKernel(String kernelName) { + return loadKernel("kernels/" + kernelName + ".cl"); + } + + /** + * Load a kernel from an explicit classpath resource path. + * + *

Results are cached; subsequent calls for the same path return cached content. + * + * @param resourcePath Full resource path (e.g., "kernels/opencl/vector_add.cl") + * @return Kernel source code + * @throws KernelLoadException if kernel cannot be loaded + */ + public static String loadKernel(String resourcePath) { + return cache.computeIfAbsent(resourcePath, KernelLoader::doLoadKernel); + } + + /** + * Check if a kernel resource exists. + * + * @param resourcePath Full resource path + * @return true if the resource exists + */ + public static boolean kernelExists(String resourcePath) { + // Check cache first + if (cache.containsKey(resourcePath)) { + return true; + } + + // Check classpath + try (var is = getResourceStream(resourcePath)) { + return is != null; + } catch (Exception e) { + return false; + } + } + + /** + * Clear the kernel cache. + * + *

WARNING: This method is for TESTING ONLY. + */ + public static void testClearCache() { + cache.clear(); + log.debug("Kernel cache cleared"); + } + + /** + * Get the number of cached kernels. + * + * @return Number of cached kernel sources + */ + public static int getCacheSize() { + return cache.size(); + } + + private static String doLoadKernel(String resourcePath) { + log.debug("Loading kernel: {}", resourcePath); + + try (var is = getResourceStream(resourcePath)) { + if (is == null) { + throw new KernelLoadException("Kernel not found: " + resourcePath); + } + + var source = new String(is.readAllBytes(), StandardCharsets.UTF_8); + log.debug("Loaded kernel {} ({} bytes)", resourcePath, source.length()); + return source; + + } catch (IOException e) { + throw new KernelLoadException("Failed to read kernel: " + resourcePath, e); + } + } + + private static InputStream getResourceStream(String resourcePath) { + // Try context classloader first (works in most frameworks) + var contextLoader = Thread.currentThread().getContextClassLoader(); + if (contextLoader != null) { + var is = contextLoader.getResourceAsStream(resourcePath); + if (is != null) { + return is; + } + } + + // Fall back to class's classloader + return KernelLoader.class.getClassLoader().getResourceAsStream(resourcePath); + } + + /** + * Exception thrown when kernel loading fails. + */ + public static class KernelLoadException extends RuntimeException { + public KernelLoadException(String message) { + super(message); + } + + public KernelLoadException(String message, Throwable cause) { + super(message, cause); + } + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLBuffer.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLBuffer.java new file mode 100644 index 0000000..99adcf4 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLBuffer.java @@ -0,0 +1,315 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import com.hellblazer.luciferase.resource.compute.GPUBuffer; +import org.lwjgl.system.MemoryStack; +import org.lwjgl.system.MemoryUtil; +import org.slf4j.Logger; +import org.slf4j.LoggerFactory; + +import java.nio.ByteBuffer; +import java.nio.FloatBuffer; +import java.util.concurrent.atomic.AtomicBoolean; + +import static org.lwjgl.opencl.CL10.*; + +/** + * OpenCL implementation of GPUBuffer. + * + *

Provides RAII lifecycle management for OpenCL buffer objects. + * Uses the singleton {@link OpenCLContext} for context and command queue. + * + *

Usage: + * 

{@code
+ * try (var buffer = OpenCLBuffer.create(1024, BufferAccess.READ_WRITE)) {
+ *     float[] data = new float[256];
+ *     buffer.upload(data);
+ *     // Use buffer in kernel...
+ *     buffer.download(data);
+ * }
+ * }
+ * + * @see GPUBuffer + * @see OpenCLContext + */ +public class OpenCLBuffer implements GPUBuffer { + + private static final Logger log = LoggerFactory.getLogger(OpenCLBuffer.class); + + /** + * Buffer access mode flags. + */ + public enum BufferAccess { + READ_ONLY(CL_MEM_READ_ONLY), + WRITE_ONLY(CL_MEM_WRITE_ONLY), + READ_WRITE(CL_MEM_READ_WRITE); + + private final int flag; + + BufferAccess(int flag) { + this.flag = flag; + } + + public int getFlag() { + return flag; + } + } + + private final long buffer; + private final int sizeInFloats; + private final int sizeInBytes; + private final BufferAccess access; + private final AtomicBoolean closed = new AtomicBoolean(false); + + /** + * Create a new OpenCL buffer. + * + * @param sizeInFloats Number of floats the buffer can hold + * @param access Buffer access mode + * @return New OpenCL buffer + * @throws IllegalStateException if OpenCL is not available + */ + public static OpenCLBuffer create(int sizeInFloats, BufferAccess access) { + var ctx = OpenCLContext.getInstance(); + if (!ctx.isInitialized()) { + ctx.acquire(); + } + + try (var stack = MemoryStack.stackPush()) { + var errcode = stack.mallocInt(1); + long sizeBytes = (long) sizeInFloats * Float.BYTES; + + long buffer = clCreateBuffer( + ctx.getContext(), + access.getFlag(), + sizeBytes, + errcode + ); + + checkError(errcode.get(0), "clCreateBuffer"); + + if (buffer == 0) { + throw new IllegalStateException("Failed to create OpenCL buffer - null handle returned"); + } + + log.debug("Created OpenCL buffer: handle={}, size={} floats ({} bytes), access={}", + buffer, sizeInFloats, sizeBytes, access); + + return new OpenCLBuffer(buffer, sizeInFloats, (int) sizeBytes, access); + } + } + + /** + * Create a buffer initialized with data. + * + * @param data Initial data to upload + * @param access Buffer access mode + * @return New OpenCL buffer with data + */ + public static OpenCLBuffer createWithData(float[] data, BufferAccess access) { + var buffer = create(data.length, access); + buffer.upload(data); + return buffer; + } + + private OpenCLBuffer(long buffer, int sizeInFloats, int sizeInBytes, BufferAccess access) { + this.buffer = buffer; + this.sizeInFloats = sizeInFloats; + this.sizeInBytes = sizeInBytes; + this.access = access; + } + + @Override + public void upload(FloatBuffer data) { + checkValid(); + if (data.remaining() > sizeInFloats) { + throw new IllegalArgumentException("Data size " + data.remaining() + + " exceeds buffer size " + sizeInFloats); + } + + var ctx = OpenCLContext.getInstance(); + int bytesToWrite = data.remaining() * Float.BYTES; + + int error = clEnqueueWriteBuffer( + ctx.getCommandQueue(), + buffer, + true, // blocking + 0, + data, + null, + null + ); + checkError(error, "clEnqueueWriteBuffer"); + + log.trace("Uploaded {} floats ({} bytes) to buffer", data.remaining(), bytesToWrite); + } + + @Override + public void upload(float[] data) { + checkValid(); + if (data.length > sizeInFloats) { + throw new IllegalArgumentException("Data size " + data.length + + " exceeds buffer size " + sizeInFloats); + } + + // Allocate direct buffer for OpenCL + ByteBuffer byteBuffer = MemoryUtil.memAlloc(data.length * Float.BYTES); + try { + byteBuffer.asFloatBuffer().put(data); + byteBuffer.rewind(); + + var ctx = OpenCLContext.getInstance(); + int error = clEnqueueWriteBuffer( + ctx.getCommandQueue(), + buffer, + true, // blocking + 0, + byteBuffer, + null, + null + ); + checkError(error, "clEnqueueWriteBuffer"); + + log.trace("Uploaded {} floats ({} bytes) to buffer", data.length, data.length * Float.BYTES); + } finally { + MemoryUtil.memFree(byteBuffer); + } + } + + @Override + public void download(FloatBuffer data) { + checkValid(); + if (data.remaining() > sizeInFloats) { + throw new IllegalArgumentException("Buffer capacity " + data.remaining() + + " exceeds buffer size " + sizeInFloats); + } + + var ctx = OpenCLContext.getInstance(); + + int error = clEnqueueReadBuffer( + ctx.getCommandQueue(), + buffer, + true, // blocking + 0, + data, + null, + null + ); + checkError(error, "clEnqueueReadBuffer"); + + log.trace("Downloaded {} floats ({} bytes) from buffer", + data.remaining(), data.remaining() * Float.BYTES); + } + + @Override + public void download(float[] data) { + checkValid(); + if (data.length > sizeInFloats) { + throw new IllegalArgumentException("Array size " + data.length + + " exceeds buffer size " + sizeInFloats); + } + + // Allocate direct buffer for OpenCL + ByteBuffer byteBuffer = MemoryUtil.memAlloc(data.length * Float.BYTES); + try { + var ctx = OpenCLContext.getInstance(); + int error = clEnqueueReadBuffer( + ctx.getCommandQueue(), + buffer, + true, // blocking + 0, + byteBuffer, + null, + null + ); + checkError(error, "clEnqueueReadBuffer"); + + byteBuffer.asFloatBuffer().get(data); + + log.trace("Downloaded {} floats ({} bytes) from buffer", + data.length, data.length * Float.BYTES); + } finally { + MemoryUtil.memFree(byteBuffer); + } + } + + @Override + public int size() { + return sizeInFloats; + } + + @Override + public int sizeInBytes() { + return sizeInBytes; + } + + @Override + public boolean isValid() { + return !closed.get() && buffer != 0; + } + + /** + * Get the native OpenCL buffer handle. + * + * @return Native buffer pointer + * @throws IllegalStateException if buffer is closed + */ + public long getHandle() { + checkValid(); + return buffer; + } + + /** + * Get the buffer access mode. + * + * @return Buffer access mode + */ + public BufferAccess getAccess() { + return access; + } + + @Override + public void close() { + if (closed.compareAndSet(false, true)) { + int error = clReleaseMemObject(buffer); + if (error != CL_SUCCESS) { + log.error("Failed to release OpenCL buffer {}: error {}", buffer, error); + } else { + log.debug("Released OpenCL buffer: handle={}", buffer); + } + } + } + + private void checkValid() { + if (closed.get()) { + throw new IllegalStateException("Buffer has been closed"); + } + } + + private static void checkError(int error, String operation) { + if (error != CL_SUCCESS) { + throw new RuntimeException("OpenCL " + operation + " failed with error code: " + error + + " (" + translateError(error) + ")"); + } + } + + /** + * Translate OpenCL error code to human-readable message. + */ + public static String translateError(int error) { + return switch (error) { + case CL_SUCCESS -> "Success"; + case CL_DEVICE_NOT_FOUND -> "Device not found"; + case CL_DEVICE_NOT_AVAILABLE -> "Device not available"; + case CL_COMPILER_NOT_AVAILABLE -> "Compiler not available"; + case CL_MEM_OBJECT_ALLOCATION_FAILURE -> "Memory object allocation failure"; + case CL_OUT_OF_RESOURCES -> "Out of resources"; + case CL_OUT_OF_HOST_MEMORY -> "Out of host memory"; + case CL_INVALID_VALUE -> "Invalid value"; + case CL_INVALID_CONTEXT -> "Invalid context"; + case CL_INVALID_COMMAND_QUEUE -> "Invalid command queue"; + case CL_INVALID_MEM_OBJECT -> "Invalid memory object"; + case CL_INVALID_BUFFER_SIZE -> "Invalid buffer size"; + default -> "Unknown error (" + error + ")"; + }; + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLContext.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLContext.java new file mode 100644 index 0000000..18207e5 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLContext.java @@ -0,0 +1,339 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import org.lwjgl.PointerBuffer; +import org.slf4j.Logger; +import org.slf4j.LoggerFactory; + +import java.util.concurrent.atomic.AtomicInteger; + +import static org.lwjgl.opencl.CL10.*; +import static org.lwjgl.system.MemoryStack.stackPush; + +/** + * Singleton OpenCL context manager. + * + *

Ensures only ONE OpenCL context and command queue exist per JVM, + * preventing resource leaks and state corruption. + * + *

The command queue is created with CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE flag + * when supported, which allows GPU drivers to automatically schedule kernels across + * all GPU cores. This enables better hardware utilization without sacrificing + * correctness, as OpenCL event dependencies ensure proper kernel ordering where needed. + * + *

Usage: + * 

{@code
+ * var ctx = OpenCLContext.getInstance();
+ * ctx.acquire();
+ * try {
+ *     long context = ctx.getContext();
+ *     long commandQueue = ctx.getCommandQueue();
+ *     // Use OpenCL resources...
+ * } finally {
+ *     ctx.release();
+ * }
+ * }
+ * + *

Test Isolation

+ *

For test isolation, use {@link #testReset()} which resets internal state + * without releasing OpenCL resources (avoiding macOS SIGABRT crashes). + * For true isolation, use JVM fork (Maven Surefire forkCount). + * + * @see com.hellblazer.luciferase.resource.compute.GPUBackend + * @see com.hellblazer.luciferase.resource.compute.BackendSelector + */ +public class OpenCLContext { + + private static final Logger log = LoggerFactory.getLogger(OpenCLContext.class); + private static final Object LOCK = new Object(); + private static volatile OpenCLContext INSTANCE; + + private long context = 0L; + private long commandQueue = 0L; + private long device = 0L; + private final AtomicInteger refCount = new AtomicInteger(0); + private volatile boolean initialized = false; + private volatile boolean outOfOrderSupported = false; + + private OpenCLContext() { + // Private constructor for singleton + } + + /** + * Get the singleton instance. + * + * @return The singleton OpenCLContext instance + */ + public static OpenCLContext getInstance() { + if (INSTANCE == null) { + synchronized (LOCK) { + if (INSTANCE == null) { + INSTANCE = new OpenCLContext(); + } + } + } + return INSTANCE; + } + + /** + * Acquire the OpenCL context (increment reference count). + * Initializes OpenCL on first acquisition. + * + * @throws RuntimeException if OpenCL initialization fails + */ + public void acquire() { + synchronized (LOCK) { + if (refCount.incrementAndGet() == 1) { + // First acquire - initialize OpenCL + initialize(); + } + } + } + + /** + * Release the OpenCL context (decrement reference count). + * + *

NOTE: Does NOT cleanup on last release - context persists until JVM shutdown. + * OpenCL cannot be reliably re-initialized after cleanup in the same JVM, + * and cleanup causes SIGABRT on macOS. + */ + public void release() { + synchronized (LOCK) { + int count = refCount.decrementAndGet(); + if (count < 0) { + log.warn("OpenCL context released more times than acquired!"); + refCount.set(0); + return; + } + + // NOTE: Do NOT cleanup when count reaches 0! + // OpenCL context persists across test classes to avoid re-initialization issues. + // Cleanup happens via shutdown hook or forceCleanup() only. + if (count == 0) { + log.debug("OpenCL context reference count reached 0 - context remains active until JVM shutdown"); + } + } + } + + /** + * Get the OpenCL context handle. + * + * @return The native OpenCL context pointer + * @throws IllegalStateException if not initialized + */ + public long getContext() { + if (!initialized || context == 0L) { + throw new IllegalStateException("OpenCL context not initialized. Call acquire() first."); + } + return context; + } + + /** + * Get the OpenCL command queue handle. + * + * @return The native OpenCL command queue pointer + * @throws IllegalStateException if not initialized + */ + public long getCommandQueue() { + if (!initialized || commandQueue == 0L) { + throw new IllegalStateException("OpenCL command queue not initialized. Call acquire() first."); + } + return commandQueue; + } + + /** + * Get the OpenCL device handle. + * + * @return The native OpenCL device pointer + * @throws IllegalStateException if not initialized + */ + public long getDevice() { + if (!initialized || device == 0L) { + throw new IllegalStateException("OpenCL device not initialized. Call acquire() first."); + } + return device; + } + + /** + * Check if OpenCL is initialized. + * + * @return true if context, command queue, and device are all initialized + */ + public boolean isInitialized() { + return initialized && context != 0L && commandQueue != 0L; + } + + /** + * Get current reference count (for debugging). + * + * @return The current reference count + */ + public int getRefCount() { + return refCount.get(); + } + + /** + * Check if out-of-order execution is supported on this device. + * + * @return true if command queue was created with CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE + */ + public boolean isOutOfOrderSupported() { + return outOfOrderSupported; + } + + /** + * Reset context state for testing. DOES NOT release OpenCL resources. + * + *

WARNING: This method is for TESTING ONLY. It resets internal state + * without releasing OpenCL resources to avoid macOS SIGABRT crashes. + * For true isolation, use separate JVM forks. + * + *

After calling this method: + *

    + *
  • getInstance() returns a new instance
    • + *
  • The old context/queue leak intentionally (macOS cleanup crashes JVM)
    • + *
  • Next acquire() will initialize a fresh context
    • + *
+ */ + public static void testReset() { + synchronized (LOCK) { + if (INSTANCE != null) { + log.debug("testReset() called - resetting singleton state (resources leak intentionally)"); + INSTANCE = null; + } + } + } + + private void initialize() { + if (initialized) { + log.debug("OpenCL already initialized"); + return; + } + + // Check system property to completely disable GPU (useful for headless testing) + // Support both new and legacy property names + var gpuDisabled = Boolean.getBoolean("gpu.disable"); + if (!gpuDisabled) { + gpuDisabled = Boolean.getBoolean("luciferase.gpu.disable"); + } + if (!gpuDisabled) { + // Legacy support for ART + gpuDisabled = Boolean.getBoolean("art.gpu.disable"); + if (gpuDisabled) { + log.warn("art.gpu.disable is deprecated, use gpu.disable instead"); + } + } + + if (gpuDisabled) { + log.info("GPU disabled via system property - OpenCL not initialized"); + return; + } + + // NOTE: We deliberately do NOT call CL.create() here. + // CL.create() causes SIGSEGV crashes in Apple's GPU drivers when running + // in forked JVM processes (like Maven Surefire). The OpenCL ICD loader + // handles initialization automatically when we call clGetPlatformIDs, + // and this approach is used by the existing gpu-test-framework tests. + + try (var stack = stackPush()) { + // Get platform count first + var numPlatforms = stack.mallocInt(1); + var errcode = clGetPlatformIDs((PointerBuffer) null, numPlatforms); + + if (errcode != CL_SUCCESS || numPlatforms.get(0) == 0) { + throw new RuntimeException("No OpenCL platforms found"); + } + + // Get first platform + var platforms = stack.mallocPointer(1); + clGetPlatformIDs(platforms, (int[]) null); + var platform = platforms.get(0); + + // Get GPU device count first + var numDevices = stack.mallocInt(1); + var result = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, null, numDevices); + + boolean useGPU = (result == CL_SUCCESS && numDevices.get(0) > 0); + + if (!useGPU) { + // Fallback to CPU - check device count + result = clGetDeviceIDs(platform, CL_DEVICE_TYPE_CPU, null, numDevices); + if (result != CL_SUCCESS || numDevices.get(0) == 0) { + throw new RuntimeException("No OpenCL devices found (tried GPU and CPU)"); + } + } + + // Now get the actual device + var devices = stack.mallocPointer(1); + if (useGPU) { + clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, devices, (int[]) null); + } else { + clGetDeviceIDs(platform, CL_DEVICE_TYPE_CPU, devices, (int[]) null); + } + + device = devices.get(0); + + // Create context + var error = stack.mallocInt(1); + context = clCreateContext(null, device, null, 0L, error); + + if (error.get(0) != CL_SUCCESS) { + throw new RuntimeException("Failed to create OpenCL context: " + error.get(0)); + } + + // Create command queue with out-of-order execution if supported + // Try with out-of-order flag first (better performance on Apple GPUs) + commandQueue = clCreateCommandQueue(context, device, + CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, error); + + if (error.get(0) != CL_SUCCESS) { + // Out-of-order execution not supported, try without flags + log.debug("Out-of-order execution not supported (error {}), creating standard queue", error.get(0)); + commandQueue = clCreateCommandQueue(context, device, 0, error); + + if (error.get(0) != CL_SUCCESS) { + throw new RuntimeException("Failed to create OpenCL command queue: " + error.get(0)); + } + outOfOrderSupported = false; + log.info("OpenCL command queue created without out-of-order execution"); + } else { + outOfOrderSupported = true; + log.info("OpenCL command queue created with out-of-order execution enabled"); + } + + initialized = true; + log.info("OpenCL context initialized successfully (refCount=1)"); + + } catch (Exception e) { + log.error("OpenCL initialization exception", e); + cleanup(); + throw new RuntimeException("Failed to initialize OpenCL", e); + } + } + + /** + * Cleanup is intentionally NOT implemented to avoid macOS OpenCL driver crashes. + * The OS will reclaim OpenCL resources when the process terminates. + * + *

Attempting to call clReleaseCommandQueue() or clReleaseContext() on macOS + * causes SIGABRT (Exit Code 134) during JVM shutdown due to Apple's deprecated + * and buggy OpenCL implementation. + * + *

This is the recommended approach for LWJGL OpenCL on macOS. + */ + private void cleanup() { + log.debug("OpenCL cleanup skipped - resources will be reclaimed by OS on process exit"); + // Intentionally do nothing - let OS clean up on process termination + } + + /** + * Force cleanup is disabled to prevent macOS OpenCL driver crashes. + * This method now does nothing - OpenCL resources persist until process exit. + * + * @deprecated Cleanup causes SIGABRT on macOS - resources are reclaimed by OS + */ + @Deprecated + public void forceCleanup() { + log.debug("forceCleanup() called but ignored - OpenCL cleanup disabled to prevent macOS crashes"); + // Intentionally do nothing + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLKernel.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLKernel.java new file mode 100644 index 0000000..419f216 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLKernel.java @@ -0,0 +1,444 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import com.hellblazer.luciferase.resource.compute.ComputeKernel; +import com.hellblazer.luciferase.resource.compute.GPUBackend; +import com.hellblazer.luciferase.resource.compute.GPUBuffer; +import org.lwjgl.PointerBuffer; +import org.slf4j.Logger; +import org.slf4j.LoggerFactory; + +import java.util.HashMap; +import java.util.Map; +import java.util.concurrent.atomic.AtomicBoolean; + +import static org.lwjgl.opencl.CL10.*; +import static org.lwjgl.system.MemoryStack.stackPush; +import static org.lwjgl.system.MemoryUtil.*; + +/** + * OpenCL compute kernel implementation. + * + *

Provides OpenCL kernel compilation and execution for cross-platform GPU compute. + * Uses the singleton {@link OpenCLContext} for context and command queue. + * + *

Usage: + * 

{@code
+ * try (var kernel = OpenCLKernel.create("myKernel")) {
+ *     kernel.compile(kernelSource, "main");
+ *     kernel.setBufferArg(0, inputBuffer, BufferAccess.READ);
+ *     kernel.setBufferArg(1, outputBuffer, BufferAccess.WRITE);
+ *     kernel.setIntArg(2, dataSize);
+ *     kernel.execute(dataSize);
+ *     kernel.finish();
+ * }
+ * }
+ * + * @see ComputeKernel + * @see OpenCLContext + * @see OpenCLBuffer + */ +public class OpenCLKernel implements ComputeKernel { + + private static final Logger log = LoggerFactory.getLogger(OpenCLKernel.class); + + private final String name; + private final long context; + private final long commandQueue; + private final long device; + private long program = NULL; + long kernel = NULL; // Package-private for testing + private final AtomicBoolean compiled = new AtomicBoolean(false); + private final AtomicBoolean closed = new AtomicBoolean(false); + private final Map bufferBindings = new HashMap<>(); + + /** + * Create a new OpenCL kernel. + * + * @param name Descriptive name for the kernel (for logging) + * @return New OpenCL kernel + * @throws IllegalStateException if OpenCL is not available + */ + public static OpenCLKernel create(String name) { + var ctx = OpenCLContext.getInstance(); + if (!ctx.isInitialized()) { + ctx.acquire(); + } + + return new OpenCLKernel( + name, + ctx.getContext(), + ctx.getCommandQueue(), + ctx.getDevice() + ); + } + + private OpenCLKernel(String name, long context, long commandQueue, long device) { + this.name = name; + this.context = context; + this.commandQueue = commandQueue; + this.device = device; + log.debug("Created OpenCL kernel: {}", name); + } + + @Override + public void compile(String source, String entryPoint) throws KernelCompilationException { + checkNotClosed(); + if (compiled.get()) { + throw new KernelCompilationException("Kernel already compiled"); + } + + try (var stack = stackPush()) { + // Create program from source + var errcode = stack.mallocInt(1); + program = clCreateProgramWithSource(context, source, errcode); + checkCLError(errcode.get(0), "Failed to create OpenCL program"); + + // Build program for specific device + var devices = stack.mallocPointer(1); + devices.put(0, device); + var buildStatus = clBuildProgram(program, devices, "", null, NULL); + if (buildStatus != CL_SUCCESS) { + // Get build log + var logSize = stack.mallocPointer(1); + clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, (PointerBuffer) null, logSize); + + if (logSize.get(0) > 0) { + var logBuffer = stack.malloc((int) logSize.get(0)); + clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, logBuffer, null); + + var buildLog = memUTF8(logBuffer); + throw new KernelCompilationException( + "OpenCL kernel compilation failed:\n" + buildLog + ); + } else { + throw new KernelCompilationException( + "OpenCL kernel compilation failed (no build log available)" + ); + } + } + + // Create kernel + kernel = clCreateKernel(program, entryPoint, errcode); + checkCLError(errcode.get(0), "Failed to create OpenCL kernel: " + entryPoint); + + compiled.set(true); + log.debug("Compiled OpenCL kernel: {} (entry point: {})", name, entryPoint); + + } catch (Exception e) { + cleanup(); + if (e instanceof KernelCompilationException kce) { + throw kce; + } + throw new KernelCompilationException("OpenCL kernel compilation failed: " + name, e); + } + } + + @Override + public void setBufferArg(int index, GPUBuffer buffer, BufferAccess access) { + checkNotClosed(); + checkCompiled(); + + if (!(buffer instanceof OpenCLBuffer openCLBuffer)) { + throw new IllegalArgumentException("Buffer must be OpenCLBuffer"); + } + + var binding = new BufferBinding(openCLBuffer, access); + bufferBindings.put(index, binding); + + // Set kernel argument + try { + checkCLError( + clSetKernelArg1p(kernel, index, openCLBuffer.getHandle()), + "Failed to set buffer argument " + index + ); + } catch (KernelCompilationException e) { + throw new RuntimeException(e); // Convert to unchecked for setter + } + } + + @Override + public void setFloatArg(int index, float value) { + checkNotClosed(); + checkCompiled(); + + try (var stack = stackPush()) { + var buffer = stack.mallocFloat(1).put(0, value); + try { + checkCLError( + clSetKernelArg(kernel, index, buffer), + "Failed to set float argument " + index + ); + } catch (KernelCompilationException e) { + throw new RuntimeException(e); // Convert to unchecked for setter + } + } + } + + @Override + public void setIntArg(int index, int value) { + checkNotClosed(); + checkCompiled(); + + try (var stack = stackPush()) { + var buffer = stack.mallocInt(1).put(0, value); + try { + checkCLError( + clSetKernelArg(kernel, index, buffer), + "Failed to set int argument " + index + ); + } catch (KernelCompilationException e) { + throw new RuntimeException(e); // Convert to unchecked for setter + } + } + } + + /** + * Set local memory argument (for __local buffers). + * + * @param index Argument index + * @param sizeInBytes Size of local memory in bytes + */ + public void setLocalMemoryArg(int index, int sizeInBytes) { + checkNotClosed(); + checkCompiled(); + + try { + checkCLError( + clSetKernelArg(kernel, index, sizeInBytes), + "Failed to set local memory argument " + index + ); + } catch (KernelCompilationException e) { + throw new RuntimeException(e); // Convert to unchecked for setter + } + } + + @Override + public void execute(int globalWorkSize) throws KernelExecutionException { + execute(globalWorkSize, 1, 1); + } + + @Override + public void execute(int globalWorkSizeX, int globalWorkSizeY) throws KernelExecutionException { + execute(globalWorkSizeX, globalWorkSizeY, 1); + } + + @Override + public void execute(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ) + throws KernelExecutionException { + + checkNotClosed(); + if (!compiled.get()) { + throw new KernelExecutionException("Kernel not compiled"); + } + + try (var stack = stackPush()) { + // Set global work size + var globalWorkSize = stack.mallocPointer(3); + globalWorkSize.put(0, globalWorkSizeX); + globalWorkSize.put(1, globalWorkSizeY); + globalWorkSize.put(2, globalWorkSizeZ); + + // Enqueue kernel + var errcode = clEnqueueNDRangeKernel( + commandQueue, + kernel, + 3, // work_dim + null, // global_work_offset + globalWorkSize, + null, // local_work_size (auto-select) + null, // event_wait_list + null // event + ); + + checkCLErrorExecution(errcode, "Failed to enqueue OpenCL kernel"); + + log.trace("Executed OpenCL kernel: {} with work size [{}, {}, {}]", + name, globalWorkSizeX, globalWorkSizeY, globalWorkSizeZ); + + } catch (Exception e) { + throw new KernelExecutionException("OpenCL kernel execution failed: " + name, e); + } + } + + @Override + public void execute(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ, + int localWorkSizeX, int localWorkSizeY, int localWorkSizeZ) + throws KernelExecutionException { + + checkNotClosed(); + if (!compiled.get()) { + throw new KernelExecutionException("Kernel not compiled"); + } + + try (var stack = stackPush()) { + // Set global work size + var globalWorkSize = stack.mallocPointer(3); + globalWorkSize.put(0, globalWorkSizeX); + globalWorkSize.put(1, globalWorkSizeY); + globalWorkSize.put(2, globalWorkSizeZ); + + // Set local work size (explicit work group dimensions) + var localWorkSize = stack.mallocPointer(3); + localWorkSize.put(0, localWorkSizeX); + localWorkSize.put(1, localWorkSizeY); + localWorkSize.put(2, localWorkSizeZ); + + // Enqueue kernel with explicit local work size + var errcode = clEnqueueNDRangeKernel( + commandQueue, + kernel, + 3, // work_dim + null, // global_work_offset + globalWorkSize, + localWorkSize, // explicit local work size for optimal GPU occupancy + null, // event_wait_list + null // event + ); + + checkCLErrorExecution(errcode, "Failed to enqueue OpenCL kernel"); + + log.trace("Executed OpenCL kernel: {} with global work size [{}, {}, {}] and local work size [{}, {}, {}]", + name, globalWorkSizeX, globalWorkSizeY, globalWorkSizeZ, + localWorkSizeX, localWorkSizeY, localWorkSizeZ); + + } catch (Exception e) { + throw new KernelExecutionException("OpenCL kernel execution failed: " + name, e); + } + } + + @Override + public void executeAsync(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ, + PointerBuffer waitEvents, PointerBuffer signalEvent) + throws KernelExecutionException { + + checkNotClosed(); + if (!compiled.get()) { + throw new KernelExecutionException("Kernel not compiled"); + } + + try (var stack = stackPush()) { + // Set up global work size + var globalWorkSize = stack.mallocPointer(3); + globalWorkSize.put(0, globalWorkSizeX); + globalWorkSize.put(1, globalWorkSizeY); + globalWorkSize.put(2, globalWorkSizeZ); + + // Execute kernel with event parameters (enables async pipelining) + var errcode = clEnqueueNDRangeKernel( + commandQueue, + kernel, + 3, // work_dim + null, // global_work_offset + globalWorkSize, // global_work_size + null, // local_work_size (auto) + waitEvents, // event_wait_list (passed through) + signalEvent // event (passed through) + ); + + if (errcode != CL_SUCCESS) { + throw new KernelExecutionException( + String.format("Failed to enqueue async kernel: %s (error code: %d)", + name, errcode) + ); + } + + log.trace("Executed async OpenCL kernel: {} with work size [{}, {}, {}]", + name, globalWorkSizeX, globalWorkSizeY, globalWorkSizeZ); + + } catch (KernelExecutionException e) { + throw e; + } catch (Exception e) { + throw new KernelExecutionException( + "OpenCL async kernel execution failed: " + name, e); + } + } + + @Override + public void finish() { + checkNotClosed(); + // Wait for command queue to finish + clFinish(commandQueue); + } + + @Override + public GPUBackend getBackend() { + return GPUBackend.OPENCL; + } + + @Override + public boolean isCompiled() { + return compiled.get(); + } + + /** + * Get the kernel name. + * + * @return Kernel name + */ + public String getName() { + return name; + } + + /** + * Check if the kernel is still valid (not closed). + * + * @return true if kernel is valid + */ + public boolean isValid() { + return !closed.get(); + } + + @Override + public void close() { + if (closed.compareAndSet(false, true)) { + cleanup(); + } + } + + private void cleanup() { + if (kernel != NULL) { + clReleaseKernel(kernel); + kernel = NULL; + } + if (program != NULL) { + clReleaseProgram(program); + program = NULL; + } + compiled.set(false); + bufferBindings.clear(); + log.debug("Released OpenCL kernel: {}", name); + } + + private void checkNotClosed() { + if (closed.get()) { + throw new IllegalStateException("Kernel has been closed"); + } + } + + private void checkCompiled() { + if (!compiled.get()) { + throw new IllegalStateException("Kernel not compiled"); + } + } + + private void checkCLError(int errcode, String message) throws KernelCompilationException { + if (errcode != CL_SUCCESS) { + throw new KernelCompilationException( + String.format("%s (error code: %d - %s)", message, errcode, + OpenCLBuffer.translateError(errcode)) + ); + } + } + + private void checkCLErrorExecution(int errcode, String message) throws KernelExecutionException { + if (errcode != CL_SUCCESS) { + throw new KernelExecutionException( + String.format("%s (error code: %d - %s)", message, errcode, + OpenCLBuffer.translateError(errcode)) + ); + } + } + + private record BufferBinding(OpenCLBuffer buffer, BufferAccess access) { + } +} diff --git a/resource/src/main/java/com/hellblazer/luciferase/resource/compute/package-info.java b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/package-info.java new file mode 100644 index 0000000..4ac3a54 --- /dev/null +++ b/resource/src/main/java/com/hellblazer/luciferase/resource/compute/package-info.java @@ -0,0 +1,130 @@ +/** + * GPU compute infrastructure for cross-platform GPU acceleration. + * + *

Overview

+ *

This package provides a unified API for GPU compute operations supporting + * OpenCL and Metal backends with automatic fallback to CPU when GPU is unavailable. + * + *

Quick Start - High-Level API

+ *

For common operations, use the {@link com.hellblazer.luciferase.resource.compute.ComputeService} + * facade which provides a simple API with automatic GPU/CPU fallback: + * + * 

{@code
+ * // Get the compute service (singleton)
+ * var compute = ComputeService.getInstance();
+ *
+ * // Vector addition
+ * float[] sum = compute.vectorAdd(a, b);
+ *
+ * // SAXPY: result = 2.0 * x + y
+ * float[] result = compute.saxpy(2.0f, x, y);
+ *
+ * // Scale a vector
+ * float[] scaled = compute.scale(data, 2.5f);
+ *
+ * // Reductions
+ * float total = compute.sum(data);
+ * float maximum = compute.max(data);
+ * float minimum = compute.min(data);
+ *
+ * // Check if GPU is available
+ * if (compute.isGPUAvailable()) {
+ *     System.out.println("Using: " + compute.getBackend().getDisplayName());
+ * }
+ * }
+ * + *

Custom Kernels

+ *

For custom operations, use the {@code createOperation} method: + * + * 

{@code
+ * String source = """
+ *     __kernel void multiply(__global const float* a,
+ *                            __global const float* b,
+ *                            __global float* result,
+ *                            const int size) {
+ *         int gid = get_global_id(0);
+ *         if (gid < size) {
+ *             result[gid] = a[gid] * b[gid];
+ *         }
+ *     }
+ *     """;
+ *
+ * try (var op = compute.createOperation("multiply", source, "multiply")) {
+ *     op.setInput(0, a);
+ *     op.setInput(1, b);
+ *     op.setOutput(2, a.length);
+ *     op.setArg(3, a.length);
+ *     float[] result = op.execute(a.length);
+ * }
+ * }
+ * + *

Core Components

+ *
    + *
  • {@link com.hellblazer.luciferase.resource.compute.ComputeService} - High-level facade (recommended)
    • + *
  • {@link com.hellblazer.luciferase.resource.compute.GPUBackend} - Backend enum (METAL, OPENCL, CPU_FALLBACK)
    • + *
  • {@link com.hellblazer.luciferase.resource.compute.BackendSelector} - Automatic backend selection
    • + *
  • {@link com.hellblazer.luciferase.resource.compute.ComputeKernel} - Kernel interface
    • + *
  • {@link com.hellblazer.luciferase.resource.compute.GPUBuffer} - Buffer interface
    • + *
  • {@link com.hellblazer.luciferase.resource.compute.KernelLoader} - Kernel source loading
    • + *
+ * + *

OpenCL Implementation

+ *

The {@code opencl} subpackage provides OpenCL-specific implementations: + *

    + *
  • {@link com.hellblazer.luciferase.resource.compute.opencl.OpenCLContext} - Singleton context manager
    • + *
  • {@link com.hellblazer.luciferase.resource.compute.opencl.OpenCLKernel} - Kernel compilation and execution
    • + *
  • {@link com.hellblazer.luciferase.resource.compute.opencl.OpenCLBuffer} - GPU buffer management
    • + *
+ * + *

Built-in Kernels

+ *

The following OpenCL kernels are available as resources: + *

    + *
  • {@code vector_add.cl} - Element-wise vector addition
    • + *
  • {@code saxpy.cl} - SAXPY (alpha*x + y) operations
    • + *
  • {@code reduce.cl} - Parallel sum, min, max reductions
    • + *
  • {@code transform.cl} - Scale, clamp, abs, square, sqrt transforms
    • + *
+ * + *

Kernel Resource Conventions

+ *

Kernels should be placed in classpath resources following these conventions: + * + * + * + * + * + *
BackendPath ConventionExample
OpenCL{@code kernels/opencl/{name}.cl}{@code kernels/opencl/vector_add.cl}
Metal{@code kernels/metal/{name}.metal}{@code kernels/metal/vector_add.metal}
Test{@code kernels/{name}.cl}{@code kernels/vector_add.cl}
+ * + *

Environment Variables

+ *
    + *
  • {@code GPU_BACKEND} - Force backend: "metal", "opencl", or "cpu"
    • + *
  • {@code GPU_DISABLE} - Disable GPU: "true" or "1"
    • + *
  • {@code gpu.disable} - System property to disable GPU
    • + *
+ * + *

Low-Level Usage Example

+ *

For direct control over buffers and kernels: + * 

{@code
+ * var source = KernelLoader.loadOpenCLKernel("vector_add");
+ *
+ * try (var kernel = OpenCLKernel.create("vectorAdd");
+ *      var bufferA = OpenCLBuffer.createWithData(dataA, READ_ONLY);
+ *      var bufferB = OpenCLBuffer.createWithData(dataB, READ_ONLY);
+ *      var bufferResult = OpenCLBuffer.create(size, WRITE_ONLY)) {
+ *
+ *     kernel.compile(source, "vectorAdd");
+ *     kernel.setBufferArg(0, bufferA, READ);
+ *     kernel.setBufferArg(1, bufferB, READ);
+ *     kernel.setBufferArg(2, bufferResult, WRITE);
+ *     kernel.setIntArg(3, size);
+ *
+ *     kernel.execute(size);
+ *     kernel.finish();
+ *
+ *     bufferResult.download(result);
+ * }
+ * }
+ * + * @see com.hellblazer.luciferase.resource.compute.ComputeService + * @see com.hellblazer.luciferase.resource.compute.opencl + */ +package com.hellblazer.luciferase.resource.compute; diff --git a/resource/src/main/resources/kernels/opencl/reduce.cl b/resource/src/main/resources/kernels/opencl/reduce.cl new file mode 100644 index 0000000..8fa7098 --- /dev/null +++ b/resource/src/main/resources/kernels/opencl/reduce.cl @@ -0,0 +1,100 @@ +/** + * Parallel reduction kernels for computing sum, min, max of arrays. + * Uses local memory for efficient work-group reduction. + */ + +/** + * Sum reduction - computes partial sums per work group. + * Final reduction must be done on host or with another kernel call. + * + * @param input Input array + * @param output Partial sums (one per work group) + * @param scratch Local memory for work-group reduction + * @param size Input array size + */ +__kernel void reduceSum( + __global const float* input, + __global float* output, + __local float* scratch, + const int size) +{ + int gid = get_global_id(0); + int lid = get_local_id(0); + int groupSize = get_local_size(0); + int groupId = get_group_id(0); + + // Load into local memory + scratch[lid] = (gid < size) ? input[gid] : 0.0f; + barrier(CLK_LOCAL_MEM_FENCE); + + // Parallel reduction in local memory + for (int stride = groupSize / 2; stride > 0; stride >>= 1) { + if (lid < stride) { + scratch[lid] += scratch[lid + stride]; + } + barrier(CLK_LOCAL_MEM_FENCE); + } + + // Write result + if (lid == 0) { + output[groupId] = scratch[0]; + } +} + +/** + * Max reduction - computes partial max per work group. + */ +__kernel void reduceMax( + __global const float* input, + __global float* output, + __local float* scratch, + const int size) +{ + int gid = get_global_id(0); + int lid = get_local_id(0); + int groupSize = get_local_size(0); + int groupId = get_group_id(0); + + scratch[lid] = (gid < size) ? input[gid] : -INFINITY; + barrier(CLK_LOCAL_MEM_FENCE); + + for (int stride = groupSize / 2; stride > 0; stride >>= 1) { + if (lid < stride) { + scratch[lid] = fmax(scratch[lid], scratch[lid + stride]); + } + barrier(CLK_LOCAL_MEM_FENCE); + } + + if (lid == 0) { + output[groupId] = scratch[0]; + } +} + +/** + * Min reduction - computes partial min per work group. + */ +__kernel void reduceMin( + __global const float* input, + __global float* output, + __local float* scratch, + const int size) +{ + int gid = get_global_id(0); + int lid = get_local_id(0); + int groupSize = get_local_size(0); + int groupId = get_group_id(0); + + scratch[lid] = (gid < size) ? input[gid] : INFINITY; + barrier(CLK_LOCAL_MEM_FENCE); + + for (int stride = groupSize / 2; stride > 0; stride >>= 1) { + if (lid < stride) { + scratch[lid] = fmin(scratch[lid], scratch[lid + stride]); + } + barrier(CLK_LOCAL_MEM_FENCE); + } + + if (lid == 0) { + output[groupId] = scratch[0]; + } +} diff --git a/resource/src/main/resources/kernels/opencl/saxpy.cl b/resource/src/main/resources/kernels/opencl/saxpy.cl new file mode 100644 index 0000000..c9550dc --- /dev/null +++ b/resource/src/main/resources/kernels/opencl/saxpy.cl @@ -0,0 +1,33 @@ +/** + * SAXPY kernel: Single-precision A*X Plus Y + * result[i] = alpha * x[i] + y[i] + * + * Classic BLAS operation, useful for benchmarking. + */ +__kernel void saxpy( + __global const float* x, + __global const float* y, + __global float* result, + const float alpha, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = alpha * x[gid] + y[gid]; + } +} + +/** + * In-place SAXPY: y[i] = alpha * x[i] + y[i] + */ +__kernel void saxpyInPlace( + __global const float* x, + __global float* y, + const float alpha, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + y[gid] = alpha * x[gid] + y[gid]; + } +} diff --git a/resource/src/main/resources/kernels/opencl/transform.cl b/resource/src/main/resources/kernels/opencl/transform.cl new file mode 100644 index 0000000..d6d9c30 --- /dev/null +++ b/resource/src/main/resources/kernels/opencl/transform.cl @@ -0,0 +1,106 @@ +/** + * Element-wise transformation kernels. + */ + +/** + * Scale: result[i] = data[i] * scale + */ +__kernel void scale( + __global const float* data, + __global float* result, + const float scale, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = data[gid] * scale; + } +} + +/** + * Scale in-place: data[i] *= scale + */ +__kernel void scaleInPlace( + __global float* data, + const float scale, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + data[gid] *= scale; + } +} + +/** + * Add scalar: result[i] = data[i] + value + */ +__kernel void addScalar( + __global const float* data, + __global float* result, + const float value, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = data[gid] + value; + } +} + +/** + * Clamp: result[i] = clamp(data[i], minVal, maxVal) + */ +__kernel void clampValues( + __global const float* data, + __global float* result, + const float minVal, + const float maxVal, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = clamp(data[gid], minVal, maxVal); + } +} + +/** + * Absolute value: result[i] = |data[i]| + */ +__kernel void absolute( + __global const float* data, + __global float* result, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = fabs(data[gid]); + } +} + +/** + * Square: result[i] = data[i]^2 + */ +__kernel void square( + __global const float* data, + __global float* result, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + float v = data[gid]; + result[gid] = v * v; + } +} + +/** + * Square root: result[i] = sqrt(data[i]) + */ +__kernel void squareRoot( + __global const float* data, + __global float* result, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = sqrt(data[gid]); + } +} diff --git a/resource/src/main/resources/kernels/opencl/vector_add.cl b/resource/src/main/resources/kernels/opencl/vector_add.cl new file mode 100644 index 0000000..0ace99a --- /dev/null +++ b/resource/src/main/resources/kernels/opencl/vector_add.cl @@ -0,0 +1,15 @@ +/** + * Vector addition kernel. + * result[i] = a[i] + b[i] + */ +__kernel void vectorAdd( + __global const float* a, + __global const float* b, + __global float* result, + const int size) +{ + int gid = get_global_id(0); + if (gid < size) { + result[gid] = a[gid] + b[gid]; + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/BackendSelectorTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/BackendSelectorTest.java new file mode 100644 index 0000000..2a0e567 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/BackendSelectorTest.java @@ -0,0 +1,159 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.junit.jupiter.api.AfterEach; +import org.junit.jupiter.api.BeforeEach; +import org.junit.jupiter.api.Test; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for BackendSelector. + */ +class BackendSelectorTest { + + @BeforeEach + void setUp() { + BackendSelector.testReset(); + } + + @AfterEach + void tearDown() { + BackendSelector.testReset(); + } + + // --- Basic Selection Tests --- + + @Test + void testGetOptimalBackendReturnsNonNull() { + var backend = BackendSelector.getOptimalBackend(); + assertNotNull(backend); + } + + @Test + void testGetOptimalBackendIsCached() { + var first = BackendSelector.getOptimalBackend(); + var second = BackendSelector.getOptimalBackend(); + + assertSame(first, second, "Backend selection should be cached"); + } + + @Test + void testCpuFallbackIsAlwaysValid() { + // CPU fallback should always be available + assertTrue(GPUBackend.CPU_FALLBACK.isAvailable()); + } + + // --- CI Environment Tests --- + + @Test + void testIsCIEnvironmentDetectsCI() { + // This test depends on actual environment + // Just verify the method doesn't throw + boolean result = BackendSelector.isCIEnvironment(); + // Result depends on whether we're in CI or not + // The method should return a boolean + assertTrue(result || !result); + } + + // --- Platform Description Tests --- + + @Test + void testGetPlatformDescriptionNotNull() { + var desc = BackendSelector.getPlatformDescription(); + assertNotNull(desc); + assertFalse(desc.isEmpty()); + } + + @Test + void testGetPlatformDescriptionContainsOS() { + var desc = BackendSelector.getPlatformDescription(); + var osName = System.getProperty("os.name"); + + assertTrue(desc.contains(osName), + "Platform description should contain OS name"); + } + + // --- Availability Convenience Methods --- + + @Test + void testIsMetalAvailableMatchesGPUBackend() { + assertEquals(GPUBackend.METAL.isAvailable(), + BackendSelector.isMetalAvailable()); + } + + @Test + void testIsOpenCLAvailableMatchesGPUBackend() { + assertEquals(GPUBackend.OPENCL.isAvailable(), + BackendSelector.isOpenCLAvailable()); + } + + // --- Environment Info Tests --- + + @Test + void testGetEnvironmentInfoNotNull() { + var info = BackendSelector.getEnvironmentInfo(); + assertNotNull(info); + assertFalse(info.isEmpty()); + } + + @Test + void testGetEnvironmentInfoContainsRequiredFields() { + var info = BackendSelector.getEnvironmentInfo(); + + assertTrue(info.contains("Platform:"), "Should contain Platform"); + assertTrue(info.contains("CI:"), "Should contain CI"); + assertTrue(info.contains("Metal Available:"), "Should contain Metal status"); + assertTrue(info.contains("OpenCL Available:"), "Should contain OpenCL status"); + assertTrue(info.contains("Selected Backend:"), "Should contain selected backend"); + } + + // --- Backend Priority Tests --- + + @Test + void testMetalHasHighestPriority() { + // Metal should have highest priority among GPU backends + assertEquals(100, GPUBackend.METAL.getPriority()); + assertTrue(GPUBackend.METAL.getPriority() > GPUBackend.OPENCL.getPriority()); + assertTrue(GPUBackend.METAL.getPriority() > GPUBackend.CPU_FALLBACK.getPriority()); + } + + @Test + void testOpenCLHasMiddlePriority() { + assertEquals(90, GPUBackend.OPENCL.getPriority()); + assertTrue(GPUBackend.OPENCL.getPriority() > GPUBackend.CPU_FALLBACK.getPriority()); + } + + @Test + void testCpuFallbackHasLowestPriority() { + assertEquals(10, GPUBackend.CPU_FALLBACK.getPriority()); + } + + // --- GPU Backend Classification Tests --- + + @Test + void testMetalIsGPU() { + assertTrue(GPUBackend.METAL.isGPU()); + } + + @Test + void testOpenCLIsGPU() { + assertTrue(GPUBackend.OPENCL.isGPU()); + } + + @Test + void testCpuFallbackIsNotGPU() { + assertFalse(GPUBackend.CPU_FALLBACK.isGPU()); + } + + // --- Reset Tests --- + + @Test + void testResetAllowsReselection() { + var first = BackendSelector.getOptimalBackend(); + BackendSelector.testReset(); + var second = BackendSelector.getOptimalBackend(); + + // After reset, it should re-select (result may be same or different) + assertNotNull(second); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeKernelTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeKernelTest.java new file mode 100644 index 0000000..67a41be --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeKernelTest.java @@ -0,0 +1,233 @@ +package com.hellblazer.luciferase.resource.compute; + +import com.hellblazer.luciferase.resource.compute.ComputeKernel.BufferAccess; +import com.hellblazer.luciferase.resource.compute.ComputeKernel.KernelCompilationException; +import com.hellblazer.luciferase.resource.compute.ComputeKernel.KernelExecutionException; +import org.junit.jupiter.api.Test; +import org.lwjgl.PointerBuffer; + +import java.nio.FloatBuffer; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for ComputeKernel interface contract. + */ +class ComputeKernelTest { + + /** + * Mock implementation for testing interface contracts. + */ + static class MockComputeKernel implements ComputeKernel { + private boolean compiled = false; + private boolean closed = false; + private int executeCount = 0; + + @Override + public void compile(String source, String entryPoint) throws KernelCompilationException { + if (source == null || source.isEmpty()) { + throw new KernelCompilationException("Source cannot be null or empty"); + } + compiled = true; + } + + @Override + public void setBufferArg(int index, GPUBuffer buffer, BufferAccess access) { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + if (index < 0) throw new IllegalArgumentException("Index must be non-negative"); + if (buffer == null) throw new IllegalArgumentException("Buffer cannot be null"); + } + + @Override + public void setFloatArg(int index, float value) { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + if (index < 0) throw new IllegalArgumentException("Index must be non-negative"); + } + + @Override + public void setIntArg(int index, int value) { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + if (index < 0) throw new IllegalArgumentException("Index must be non-negative"); + } + + @Override + public void execute(int globalWorkSize) throws KernelExecutionException { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + executeCount++; + } + + @Override + public void execute(int globalWorkSizeX, int globalWorkSizeY) throws KernelExecutionException { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + executeCount++; + } + + @Override + public void execute(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ) + throws KernelExecutionException { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + executeCount++; + } + + @Override + public void execute(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ, + int localWorkSizeX, int localWorkSizeY, int localWorkSizeZ) + throws KernelExecutionException { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + executeCount++; + } + + @Override + public void executeAsync(int globalWorkSizeX, int globalWorkSizeY, int globalWorkSizeZ, + PointerBuffer waitEvents, PointerBuffer signalEvent) + throws KernelExecutionException { + if (!compiled) throw new IllegalStateException("Kernel not compiled"); + executeCount++; + } + + @Override + public void finish() { + // No-op for mock + } + + @Override + public GPUBackend getBackend() { + return GPUBackend.CPU_FALLBACK; + } + + @Override + public boolean isCompiled() { + return compiled; + } + + @Override + public void close() { + closed = true; + compiled = false; + } + + public int getExecuteCount() { + return executeCount; + } + + public boolean isClosed() { + return closed; + } + } + + @Test + void testKernelCompilation() throws KernelCompilationException { + try (var kernel = new MockComputeKernel()) { + assertFalse(kernel.isCompiled()); + + kernel.compile("__kernel void test() {}", "test"); + + assertTrue(kernel.isCompiled()); + } + } + + @Test + void testCompilationWithEmptySourceThrows() { + try (var kernel = new MockComputeKernel()) { + assertThrows(KernelCompilationException.class, () -> kernel.compile("", "test")); + } + } + + @Test + void testCompilationWithNullSourceThrows() { + try (var kernel = new MockComputeKernel()) { + assertThrows(KernelCompilationException.class, () -> kernel.compile(null, "test")); + } + } + + @Test + void testSetArgumentsBeforeCompileThrows() { + try (var kernel = new MockComputeKernel()) { + assertThrows(IllegalStateException.class, () -> kernel.setFloatArg(0, 1.0f)); + assertThrows(IllegalStateException.class, () -> kernel.setIntArg(0, 1)); + } + } + + @Test + void testExecuteBeforeCompileThrows() { + try (var kernel = new MockComputeKernel()) { + assertThrows(IllegalStateException.class, () -> kernel.execute(1024)); + } + } + + @Test + void testExecuteVariants() throws KernelCompilationException, KernelExecutionException { + try (var kernel = new MockComputeKernel()) { + kernel.compile("kernel", "main"); + + kernel.execute(1024); + kernel.execute(32, 32); + kernel.execute(8, 8, 8); + kernel.execute(8, 8, 8, 4, 4, 4); + + assertEquals(4, kernel.getExecuteCount()); + } + } + + @Test + void testBufferAccessEnum() { + assertEquals(3, BufferAccess.values().length); + assertNotNull(BufferAccess.READ); + assertNotNull(BufferAccess.WRITE); + assertNotNull(BufferAccess.READ_WRITE); + } + + @Test + void testExceptionConstruction() { + var compileEx = new KernelCompilationException("test"); + assertEquals("test", compileEx.getMessage()); + + var cause = new RuntimeException("cause"); + var compileExWithCause = new KernelCompilationException("test", cause); + assertEquals(cause, compileExWithCause.getCause()); + + var execEx = new KernelExecutionException("test"); + assertEquals("test", execEx.getMessage()); + + var execExWithCause = new KernelExecutionException("test", cause); + assertEquals(cause, execExWithCause.getCause()); + } + + @Test + void testCloseInvalidatesKernel() throws KernelCompilationException { + var kernel = new MockComputeKernel(); + kernel.compile("kernel", "main"); + assertTrue(kernel.isCompiled()); + + kernel.close(); + assertFalse(kernel.isCompiled()); + assertTrue(kernel.isClosed()); + } + + @Test + void testAutoCloseable() throws KernelCompilationException { + MockComputeKernel kernel; + try (var k = new MockComputeKernel()) { + kernel = k; + kernel.compile("kernel", "main"); + assertTrue(kernel.isCompiled()); + } + assertTrue(kernel.isClosed()); + } + + @Test + void testGetBackend() { + try (var kernel = new MockComputeKernel()) { + assertEquals(GPUBackend.CPU_FALLBACK, kernel.getBackend()); + } + } + + @Test + void testNegativeIndexThrows() throws KernelCompilationException { + try (var kernel = new MockComputeKernel()) { + kernel.compile("kernel", "main"); + assertThrows(IllegalArgumentException.class, () -> kernel.setFloatArg(-1, 1.0f)); + assertThrows(IllegalArgumentException.class, () -> kernel.setIntArg(-1, 1)); + } + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeServiceStressTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeServiceStressTest.java new file mode 100644 index 0000000..e5cfaf7 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeServiceStressTest.java @@ -0,0 +1,400 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.junit.jupiter.api.AfterEach; +import org.junit.jupiter.api.BeforeAll; +import org.junit.jupiter.api.BeforeEach; +import org.junit.jupiter.api.Test; +import org.junit.jupiter.api.condition.DisabledIfEnvironmentVariable; +import org.lwjgl.PointerBuffer; +import org.lwjgl.opencl.CL10; +import org.lwjgl.system.MemoryStack; + +import java.util.Random; +import java.util.concurrent.CountDownLatch; +import java.util.concurrent.ExecutorService; +import java.util.concurrent.Executors; +import java.util.concurrent.TimeUnit; +import java.util.concurrent.atomic.AtomicInteger; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Stress and edge case tests for ComputeService. + * + *

Tests boundary conditions, large arrays, memory pressure, + * and concurrent access patterns. + */ +@DisabledIfEnvironmentVariable(named = "CI", matches = "true", disabledReason = "OpenCL not available in CI") +class ComputeServiceStressTest { + + private static boolean openCLAvailable; + private ComputeService compute; + + @BeforeAll + static void checkOpenCL() { + try (var stack = MemoryStack.stackPush()) { + var numPlatforms = stack.mallocInt(1); + var errcode = CL10.clGetPlatformIDs((PointerBuffer) null, numPlatforms); + openCLAvailable = errcode == CL10.CL_SUCCESS && numPlatforms.get(0) > 0; + } catch (Exception e) { + openCLAvailable = false; + } + } + + @BeforeEach + void setUp() { + ComputeService.testReset(); + compute = ComputeService.getInstance(); + } + + @AfterEach + void tearDown() { + ComputeService.testReset(); + } + + // ========== Edge Case Tests ========== + + @Test + void testVectorAdd_SingleElement() { + float[] a = {42.0f}; + float[] b = {8.0f}; + + float[] result = compute.vectorAdd(a, b); + + assertEquals(1, result.length); + assertEquals(50.0f, result[0], 0.0001f); + } + + @Test + void testVectorAdd_EmptyArrays() { + float[] a = {}; + float[] b = {}; + + float[] result = compute.vectorAdd(a, b); + + assertEquals(0, result.length); + } + + @Test + void testVectorAdd_OddSize() { + // Non-power-of-2 size + float[] a = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f}; + float[] b = {7.0f, 6.0f, 5.0f, 4.0f, 3.0f, 2.0f, 1.0f}; + + float[] result = compute.vectorAdd(a, b); + + assertEquals(7, result.length); + for (float v : result) { + assertEquals(8.0f, v, 0.0001f); + } + } + + @Test + void testVectorAdd_PrimeSize() { + // Prime number size (not power of 2, odd, prime) + int size = 997; + float[] a = new float[size]; + float[] b = new float[size]; + for (int i = 0; i < size; i++) { + a[i] = i; + b[i] = 1.0f; + } + + float[] result = compute.vectorAdd(a, b); + + assertEquals(size, result.length); + for (int i = 0; i < size; i++) { + assertEquals(i + 1.0f, result[i], 0.0001f); + } + } + + @Test + void testSaxpy_ZeroAlpha() { + float[] x = {1.0f, 2.0f, 3.0f, 4.0f}; + float[] y = {10.0f, 20.0f, 30.0f, 40.0f}; + + // When alpha is 0, result should equal y + float[] result = compute.saxpy(0.0f, x, y); + + assertArrayEquals(y, result, 0.0001f); + } + + @Test + void testSaxpy_NegativeAlpha() { + float[] x = {1.0f, 2.0f, 3.0f, 4.0f}; + float[] y = {10.0f, 10.0f, 10.0f, 10.0f}; + + // result = -2 * x + y + float[] result = compute.saxpy(-2.0f, x, y); + + assertArrayEquals(new float[]{8.0f, 6.0f, 4.0f, 2.0f}, result, 0.0001f); + } + + @Test + void testScale_ZeroScale() { + float[] data = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f}; + + float[] result = compute.scale(data, 0.0f); + + for (float v : result) { + assertEquals(0.0f, v, 0.0001f); + } + } + + @Test + void testScale_NegativeScale() { + float[] data = {1.0f, -2.0f, 3.0f, -4.0f}; + + float[] result = compute.scale(data, -1.0f); + + assertArrayEquals(new float[]{-1.0f, 2.0f, -3.0f, 4.0f}, result, 0.0001f); + } + + @Test + void testSum_SingleElement() { + assertEquals(42.0f, compute.sum(new float[]{42.0f}), 0.0001f); + } + + @Test + void testSum_EmptyArray() { + assertEquals(0.0f, compute.sum(new float[]{}), 0.0001f); + } + + @Test + void testSum_AllNegatives() { + float[] data = {-1.0f, -2.0f, -3.0f, -4.0f, -5.0f}; + assertEquals(-15.0f, compute.sum(data), 0.0001f); + } + + @Test + void testMinMax_SingleElement() { + float[] data = {42.0f}; + assertEquals(42.0f, compute.min(data), 0.0001f); + assertEquals(42.0f, compute.max(data), 0.0001f); + } + + @Test + void testMinMax_AllSameValue() { + float[] data = {7.0f, 7.0f, 7.0f, 7.0f}; + assertEquals(7.0f, compute.min(data), 0.0001f); + assertEquals(7.0f, compute.max(data), 0.0001f); + } + + @Test + void testMinMax_Extremes() { + float[] data = {Float.MIN_VALUE, 0.0f, Float.MAX_VALUE}; + assertEquals(Float.MIN_VALUE, compute.min(data), 0.0001f); + assertEquals(Float.MAX_VALUE, compute.max(data), 0.0001f); + } + + // ========== Large Array Tests ========== + + @Test + void testVectorAdd_LargeArray_PowerOf2() { + if (!openCLAvailable) return; + + int size = 1 << 20; // 1 million elements + float[] a = new float[size]; + float[] b = new float[size]; + + for (int i = 0; i < size; i++) { + a[i] = 1.0f; + b[i] = 2.0f; + } + + float[] result = compute.vectorAdd(a, b); + + assertEquals(size, result.length); + // Spot check + for (int i = 0; i < size; i += 100_000) { + assertEquals(3.0f, result[i], 0.0001f); + } + } + + @Test + void testSaxpy_LargeArray() { + if (!openCLAvailable) return; + + int size = 500_000; + float[] x = new float[size]; + float[] y = new float[size]; + + for (int i = 0; i < size; i++) { + x[i] = i * 0.001f; + y[i] = 100.0f; + } + + float[] result = compute.saxpy(2.0f, x, y); + + // Spot check: result[i] = 2 * (i * 0.001) + 100 + assertEquals(100.0f, result[0], 0.0001f); + assertEquals(100.002f, result[1], 0.0001f); + assertEquals(101.0f, result[500], 0.0001f); + } + + // ========== Numerical Precision Tests ========== + + @Test + void testVectorAdd_SmallValues() { + float[] a = {1e-7f, 2e-7f, 3e-7f}; + float[] b = {4e-7f, 5e-7f, 6e-7f}; + + float[] result = compute.vectorAdd(a, b); + + assertEquals(5e-7f, result[0], 1e-12f); + assertEquals(7e-7f, result[1], 1e-12f); + assertEquals(9e-7f, result[2], 1e-12f); + } + + @Test + void testVectorAdd_LargeValues() { + float[] a = {1e30f, 2e30f}; + float[] b = {3e30f, 4e30f}; + + float[] result = compute.vectorAdd(a, b); + + assertEquals(4e30f, result[0], 1e25f); + assertEquals(6e30f, result[1], 1e25f); + } + + @Test + void testVectorAdd_MixedMagnitudes() { + // Tests precision when adding very different magnitudes + float[] a = {1e20f, 1e-20f}; + float[] b = {1.0f, 1.0f}; + + float[] result = compute.vectorAdd(a, b); + + // Large + small: the small value is lost due to float precision + assertEquals(1e20f, result[0], 1e15f); + // Small + 1: should be ~1.0 + assertEquals(1.0f, result[1], 0.0001f); + } + + // ========== Concurrent Access Tests ========== + + @Test + void testConcurrentAccess_Singleton() throws InterruptedException { + int threadCount = 10; + var latch = new CountDownLatch(threadCount); + var instances = new ComputeService[threadCount]; + var errors = new AtomicInteger(0); + + for (int i = 0; i < threadCount; i++) { + int idx = i; + new Thread(() -> { + try { + instances[idx] = ComputeService.getInstance(); + } catch (Exception e) { + errors.incrementAndGet(); + } finally { + latch.countDown(); + } + }).start(); + } + + assertTrue(latch.await(5, TimeUnit.SECONDS)); + assertEquals(0, errors.get()); + + // All threads should get the same instance + for (int i = 1; i < threadCount; i++) { + assertSame(instances[0], instances[i]); + } + } + + @Test + void testConcurrentOperations() throws InterruptedException { + if (!openCLAvailable) return; + + int threadCount = 4; + int opsPerThread = 10; + var latch = new CountDownLatch(threadCount); + var errors = new AtomicInteger(0); + + var executor = Executors.newFixedThreadPool(threadCount); + try { + for (int t = 0; t < threadCount; t++) { + executor.submit(() -> { + try { + var rand = new Random(); + for (int i = 0; i < opsPerThread; i++) { + int size = 100 + rand.nextInt(900); + float[] a = new float[size]; + float[] b = new float[size]; + for (int j = 0; j < size; j++) { + a[j] = rand.nextFloat(); + b[j] = rand.nextFloat(); + } + + // Just ensure no exceptions + float[] result = compute.vectorAdd(a, b); + assertEquals(size, result.length); + } + } catch (Exception e) { + e.printStackTrace(); + errors.incrementAndGet(); + } finally { + latch.countDown(); + } + }); + } + + assertTrue(latch.await(30, TimeUnit.SECONDS)); + assertEquals(0, errors.get(), "Some concurrent operations failed"); + } finally { + executor.shutdown(); + } + } + + // ========== Memory Pressure Tests ========== + + @Test + void testRepeatedOperations_NoMemoryLeak() { + if (!openCLAvailable) return; + + // Run many operations to check for memory leaks + int iterations = 100; + float[] a = new float[10_000]; + float[] b = new float[10_000]; + + for (int i = 0; i < a.length; i++) { + a[i] = 1.0f; + b[i] = 2.0f; + } + + for (int i = 0; i < iterations; i++) { + float[] result = compute.vectorAdd(a, b); + assertEquals(10_000, result.length); + } + // If we get here without OOM, we're good + } + + @Test + void testCustomOperation_ResourceCleanup() throws Exception { + if (!openCLAvailable) return; + + String source = """ + __kernel void add(__global const float* a, + __global float* b, + const int size) { + int gid = get_global_id(0); + if (gid < size) { + b[gid] = a[gid] + 1.0f; + } + } + """; + + // Create and close many operations to verify cleanup + for (int i = 0; i < 50; i++) { + try (var op = compute.createOperation("cleanup_test", source, "add")) { + float[] input = {1.0f, 2.0f, 3.0f}; + op.setInput(0, input); + op.setOutput(1, 3); + op.setArg(2, 3); + float[] result = op.execute(3); + assertEquals(3, result.length); + } + } + // If we get here without resource exhaustion, cleanup works + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeServiceTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeServiceTest.java new file mode 100644 index 0000000..65bc66b --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/ComputeServiceTest.java @@ -0,0 +1,328 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.junit.jupiter.api.AfterEach; +import org.junit.jupiter.api.BeforeAll; +import org.junit.jupiter.api.BeforeEach; +import org.junit.jupiter.api.Test; +import org.junit.jupiter.api.condition.DisabledIfEnvironmentVariable; +import org.lwjgl.PointerBuffer; +import org.lwjgl.opencl.CL10; +import org.lwjgl.system.MemoryStack; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Tests for ComputeService high-level API. + * + *

These tests also serve as examples of how to use the ComputeService. + */ +@DisabledIfEnvironmentVariable(named = "CI", matches = "true", disabledReason = "OpenCL not available in CI") +class ComputeServiceTest { + + private static boolean openCLAvailable; + private ComputeService compute; + + @BeforeAll + static void checkOpenCL() { + try (var stack = MemoryStack.stackPush()) { + var numPlatforms = stack.mallocInt(1); + var errcode = CL10.clGetPlatformIDs((PointerBuffer) null, numPlatforms); + openCLAvailable = errcode == CL10.CL_SUCCESS && numPlatforms.get(0) > 0; + } catch (Exception e) { + openCLAvailable = false; + } + System.out.println("OpenCL available: " + openCLAvailable); + } + + @BeforeEach + void setUp() { + ComputeService.testReset(); + compute = ComputeService.getInstance(); + } + + @AfterEach + void tearDown() { + ComputeService.testReset(); + } + + // ========== Basic API Tests ========== + + @Test + void testSingletonInstance() { + var instance1 = ComputeService.getInstance(); + var instance2 = ComputeService.getInstance(); + assertSame(instance1, instance2); + } + + @Test + void testBackendAvailable() { + assertNotNull(compute.getBackend()); + // Should always have at least CPU fallback + assertTrue(compute.getBackend() == GPUBackend.CPU_FALLBACK || + compute.getBackend().isAvailable()); + } + + // ========== Vector Add Examples ========== + + @Test + void testVectorAdd_Simple() { + if (!openCLAvailable) return; + + // Example: Add two vectors + float[] a = {1.0f, 2.0f, 3.0f, 4.0f}; + float[] b = {5.0f, 6.0f, 7.0f, 8.0f}; + + float[] result = compute.vectorAdd(a, b); + + assertArrayEquals(new float[]{6.0f, 8.0f, 10.0f, 12.0f}, result, 0.0001f); + } + + @Test + void testVectorAdd_LargeArray() { + if (!openCLAvailable) return; + + // Example: Large array addition + int size = 100_000; + float[] a = new float[size]; + float[] b = new float[size]; + + for (int i = 0; i < size; i++) { + a[i] = i; + b[i] = size - i; + } + + float[] result = compute.vectorAdd(a, b); + + // Every element should equal size + for (int i = 0; i < size; i++) { + assertEquals(size, result[i], 0.0001f); + } + } + + @Test + void testVectorAdd_DifferentLengthsThrows() { + float[] a = {1.0f, 2.0f}; + float[] b = {1.0f, 2.0f, 3.0f}; + + assertThrows(IllegalArgumentException.class, + () -> compute.vectorAdd(a, b)); + } + + // ========== SAXPY Examples ========== + + @Test + void testSaxpy_Simple() { + if (!openCLAvailable) return; + + // Example: SAXPY - Single-precision A*X Plus Y + // result = 2.0 * x + y + float alpha = 2.0f; + float[] x = {1.0f, 2.0f, 3.0f, 4.0f}; + float[] y = {10.0f, 20.0f, 30.0f, 40.0f}; + + float[] result = compute.saxpy(alpha, x, y); + + // result[i] = 2.0 * x[i] + y[i] + assertArrayEquals(new float[]{12.0f, 24.0f, 36.0f, 48.0f}, result, 0.0001f); + } + + @Test + void testSaxpy_LinearCombination() { + if (!openCLAvailable) return; + + // Example: Linear interpolation using SAXPY + // lerp(a, b, t) = (1-t)*a + t*b = a + t*(b-a) + // Using saxpy: result = t * (b-a) + a + + float[] a = {0.0f, 0.0f, 0.0f}; + float[] b = {10.0f, 20.0f, 30.0f}; + float t = 0.5f; + + // First compute b - a + float[] diff = new float[a.length]; + for (int i = 0; i < a.length; i++) { + diff[i] = b[i] - a[i]; + } + + // Then compute t * diff + a + float[] result = compute.saxpy(t, diff, a); + + // At t=0.5, result should be midpoint + assertArrayEquals(new float[]{5.0f, 10.0f, 15.0f}, result, 0.0001f); + } + + // ========== Scale Examples ========== + + @Test + void testScale_Simple() { + if (!openCLAvailable) return; + + // Example: Scale a vector + float[] data = {1.0f, 2.0f, 3.0f, 4.0f}; + + float[] result = compute.scale(data, 2.5f); + + assertArrayEquals(new float[]{2.5f, 5.0f, 7.5f, 10.0f}, result, 0.0001f); + } + + @Test + void testScale_Normalize() { + if (!openCLAvailable) return; + + // Example: Normalize by dividing by max + float[] data = {2.0f, 4.0f, 8.0f, 10.0f}; + float maxVal = compute.max(data); + + float[] normalized = compute.scale(data, 1.0f / maxVal); + + assertEquals(1.0f, compute.max(normalized), 0.0001f); + } + + // ========== Reduction Examples ========== + + @Test + void testSum() { + float[] data = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f}; + + float sum = compute.sum(data); + + assertEquals(15.0f, sum, 0.0001f); + } + + @Test + void testMinMax() { + float[] data = {3.0f, 1.0f, 4.0f, 1.0f, 5.0f, 9.0f, 2.0f, 6.0f}; + + assertEquals(1.0f, compute.min(data), 0.0001f); + assertEquals(9.0f, compute.max(data), 0.0001f); + } + + // ========== Custom Operation Examples ========== + + @Test + void testCustomOperation() throws Exception { + if (!openCLAvailable) return; + + // Example: Custom kernel for element-wise multiply + String source = """ + __kernel void multiply(__global const float* a, + __global const float* b, + __global float* result, + const int size) { + int gid = get_global_id(0); + if (gid < size) { + result[gid] = a[gid] * b[gid]; + } + } + """; + + float[] a = {1.0f, 2.0f, 3.0f, 4.0f}; + float[] b = {2.0f, 3.0f, 4.0f, 5.0f}; + + try (var op = compute.createOperation("multiply", source, "multiply")) { + op.setInput(0, a); + op.setInput(1, b); + op.setOutput(2, a.length); + op.setArg(3, a.length); + + float[] result = op.execute(a.length); + + assertArrayEquals(new float[]{2.0f, 6.0f, 12.0f, 20.0f}, result, 0.0001f); + } + } + + @Test + void testCustomOperation_WithConstants() throws Exception { + if (!openCLAvailable) return; + + // Example: Apply threshold + String source = """ + __kernel void threshold(__global const float* input, + __global float* output, + const float thresh, + const int size) { + int gid = get_global_id(0); + if (gid < size) { + output[gid] = (input[gid] > thresh) ? 1.0f : 0.0f; + } + } + """; + + float[] input = {0.1f, 0.6f, 0.3f, 0.8f, 0.4f}; + + try (var op = compute.createOperation("threshold", source, "threshold")) { + op.setInput(0, input); + op.setOutput(1, input.length); + op.setArg(2, 0.5f); // threshold value + op.setArg(3, input.length); + + float[] result = op.execute(input.length); + + assertArrayEquals(new float[]{0.0f, 1.0f, 0.0f, 1.0f, 0.0f}, result, 0.0001f); + } + } + + // ========== Performance Comparison Example ========== + + @Test + void testPerformanceComparison() { + if (!openCLAvailable) return; + + int size = 1_000_000; + float[] a = new float[size]; + float[] b = new float[size]; + + for (int i = 0; i < size; i++) { + a[i] = i * 0.001f; + b[i] = (size - i) * 0.001f; + } + + // Warm up + compute.vectorAdd(a, b); + + // Time GPU + long gpuStart = System.nanoTime(); + float[] gpuResult = compute.vectorAdd(a, b); + long gpuTime = System.nanoTime() - gpuStart; + + // Time CPU + long cpuStart = System.nanoTime(); + float[] cpuResult = new float[size]; + for (int i = 0; i < size; i++) { + cpuResult[i] = a[i] + b[i]; + } + long cpuTime = System.nanoTime() - cpuStart; + + System.out.printf("Vector add (%d elements):%n", size); + System.out.printf(" GPU: %.2f ms%n", gpuTime / 1_000_000.0); + System.out.printf(" CPU: %.2f ms%n", cpuTime / 1_000_000.0); + System.out.printf(" Speedup: %.2fx%n", (double) cpuTime / gpuTime); + + // Verify correctness + assertArrayEquals(cpuResult, gpuResult, 0.0001f); + } + + // ========== CPU Fallback Tests ========== + + @Test + void testCPUFallback_VectorAdd() { + // Force CPU by resetting and checking behavior + // This works even without GPU + float[] a = {1.0f, 2.0f, 3.0f}; + float[] b = {4.0f, 5.0f, 6.0f}; + + float[] result = compute.vectorAdd(a, b); + + assertArrayEquals(new float[]{5.0f, 7.0f, 9.0f}, result, 0.0001f); + } + + @Test + void testCPUFallback_Saxpy() { + float[] x = {1.0f, 2.0f, 3.0f}; + float[] y = {10.0f, 20.0f, 30.0f}; + + float[] result = compute.saxpy(2.0f, x, y); + + assertArrayEquals(new float[]{12.0f, 24.0f, 36.0f}, result, 0.0001f); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUBackendTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUBackendTest.java new file mode 100644 index 0000000..49311d1 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUBackendTest.java @@ -0,0 +1,121 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.junit.jupiter.api.AfterEach; +import org.junit.jupiter.api.BeforeEach; +import org.junit.jupiter.api.Test; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for GPUBackend enum. + */ +class GPUBackendTest { + + @BeforeEach + void setUp() { + GPUBackend.testResetAvailability(); + } + + @AfterEach + void tearDown() { + GPUBackend.testResetAvailability(); + } + + @Test + void testEnumValues() { + assertEquals(3, GPUBackend.values().length); + assertNotNull(GPUBackend.METAL); + assertNotNull(GPUBackend.OPENCL); + assertNotNull(GPUBackend.CPU_FALLBACK); + } + + @Test + void testDisplayNames() { + assertEquals("Metal", GPUBackend.METAL.getDisplayName()); + assertEquals("OpenCL", GPUBackend.OPENCL.getDisplayName()); + assertEquals("CPU Fallback", GPUBackend.CPU_FALLBACK.getDisplayName()); + } + + @Test + void testPriorities() { + // Metal has highest priority + assertTrue(GPUBackend.METAL.getPriority() > GPUBackend.OPENCL.getPriority()); + // OpenCL higher than CPU fallback + assertTrue(GPUBackend.OPENCL.getPriority() > GPUBackend.CPU_FALLBACK.getPriority()); + // CPU fallback has lowest priority + assertEquals(10, GPUBackend.CPU_FALLBACK.getPriority()); + } + + @Test + void testIsGPU() { + assertTrue(GPUBackend.METAL.isGPU()); + assertTrue(GPUBackend.OPENCL.isGPU()); + assertFalse(GPUBackend.CPU_FALLBACK.isGPU()); + } + + @Test + void testValueOf() { + assertEquals(GPUBackend.METAL, GPUBackend.valueOf("METAL")); + assertEquals(GPUBackend.OPENCL, GPUBackend.valueOf("OPENCL")); + assertEquals(GPUBackend.CPU_FALLBACK, GPUBackend.valueOf("CPU_FALLBACK")); + } + + // --- Availability Tests --- + + @Test + void testCPUFallbackAlwaysAvailable() { + assertTrue(GPUBackend.CPU_FALLBACK.isAvailable(), + "CPU_FALLBACK should always be available"); + } + + @Test + void testMetalAvailabilityMatchesPlatform() { + var isMacOS = System.getProperty("os.name", "").toLowerCase().contains("mac"); + assertEquals(isMacOS, GPUBackend.METAL.isAvailable(), + "Metal availability should match macOS platform"); + } + + @Test + void testOpenCLAvailabilityIsCached() { + // First call triggers detection + boolean first = GPUBackend.OPENCL.isAvailable(); + // Second call should return cached value + boolean second = GPUBackend.OPENCL.isAvailable(); + assertEquals(first, second, "Cached availability should be consistent"); + } + + @Test + void testMetalAvailabilityIsCached() { + // First call triggers detection + boolean first = GPUBackend.METAL.isAvailable(); + // Second call should return cached value + boolean second = GPUBackend.METAL.isAvailable(); + assertEquals(first, second, "Cached availability should be consistent"); + } + + @Test + void testTestResetAvailabilityClearsCache() { + // Get initial availability + GPUBackend.CPU_FALLBACK.isAvailable(); + GPUBackend.METAL.isAvailable(); + GPUBackend.OPENCL.isAvailable(); + + // Reset should not throw + GPUBackend.testResetAvailability(); + + // Should still work after reset + assertTrue(GPUBackend.CPU_FALLBACK.isAvailable()); + } + + @Test + void testAllBackendsHaveConsistentAvailability() { + // All backends should return consistent values on repeated calls + for (var backend : GPUBackend.values()) { + boolean first = backend.isAvailable(); + boolean second = backend.isAvailable(); + boolean third = backend.isAvailable(); + assertEquals(first, second); + assertEquals(second, third); + } + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUBufferTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUBufferTest.java new file mode 100644 index 0000000..4d71843 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUBufferTest.java @@ -0,0 +1,156 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.junit.jupiter.api.Test; + +import java.nio.FloatBuffer; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for GPUBuffer interface contract. + * Tests use a mock implementation to verify interface behavior. + */ +class GPUBufferTest { + + /** + * Mock implementation for testing interface contracts. + */ + static class MockGPUBuffer implements GPUBuffer { + private final int size; + private boolean valid = true; + private float[] data; + + MockGPUBuffer(int size) { + this.size = size; + this.data = new float[size]; + } + + @Override + public void upload(FloatBuffer data) { + if (!valid) throw new IllegalStateException("Buffer not valid"); + if (data.remaining() > size) throw new IllegalArgumentException("Data exceeds capacity"); + data.get(this.data, 0, Math.min(data.remaining(), size)); + } + + @Override + public void upload(float[] data) { + if (!valid) throw new IllegalStateException("Buffer not valid"); + if (data.length > size) throw new IllegalArgumentException("Data exceeds capacity"); + System.arraycopy(data, 0, this.data, 0, Math.min(data.length, size)); + } + + @Override + public void download(FloatBuffer data) { + if (!valid) throw new IllegalStateException("Buffer not valid"); + if (data.remaining() < size) throw new IllegalArgumentException("Buffer capacity insufficient"); + data.put(this.data); + } + + @Override + public void download(float[] data) { + if (!valid) throw new IllegalStateException("Buffer not valid"); + if (data.length < size) throw new IllegalArgumentException("Array size insufficient"); + System.arraycopy(this.data, 0, data, 0, size); + } + + @Override + public int size() { + return size; + } + + @Override + public int sizeInBytes() { + return size * Float.BYTES; + } + + @Override + public boolean isValid() { + return valid; + } + + @Override + public void close() { + valid = false; + data = null; + } + } + + @Test + void testBufferCreation() { + try (var buffer = new MockGPUBuffer(1024)) { + assertTrue(buffer.isValid()); + assertEquals(1024, buffer.size()); + assertEquals(1024 * Float.BYTES, buffer.sizeInBytes()); + } + } + + @Test + void testUploadDownloadArray() { + try (var buffer = new MockGPUBuffer(4)) { + var input = new float[]{1.0f, 2.0f, 3.0f, 4.0f}; + buffer.upload(input); + + var output = new float[4]; + buffer.download(output); + + assertArrayEquals(input, output); + } + } + + @Test + void testUploadDownloadFloatBuffer() { + try (var buffer = new MockGPUBuffer(4)) { + var input = FloatBuffer.wrap(new float[]{1.0f, 2.0f, 3.0f, 4.0f}); + buffer.upload(input); + + var output = FloatBuffer.allocate(4); + buffer.download(output); + output.flip(); + + assertEquals(1.0f, output.get(0)); + assertEquals(4.0f, output.get(3)); + } + } + + @Test + void testCloseInvalidatesBuffer() { + var buffer = new MockGPUBuffer(4); + assertTrue(buffer.isValid()); + + buffer.close(); + assertFalse(buffer.isValid()); + } + + @Test + void testOperationsOnClosedBufferThrow() { + var buffer = new MockGPUBuffer(4); + buffer.close(); + + assertThrows(IllegalStateException.class, () -> buffer.upload(new float[4])); + assertThrows(IllegalStateException.class, () -> buffer.download(new float[4])); + } + + @Test + void testUploadExceedingCapacityThrows() { + try (var buffer = new MockGPUBuffer(4)) { + assertThrows(IllegalArgumentException.class, () -> buffer.upload(new float[10])); + } + } + + @Test + void testDownloadInsufficientCapacityThrows() { + try (var buffer = new MockGPUBuffer(4)) { + assertThrows(IllegalArgumentException.class, () -> buffer.download(new float[2])); + } + } + + @Test + void testAutoCloseable() { + MockGPUBuffer buffer; + try (var b = new MockGPUBuffer(4)) { + buffer = b; + assertTrue(buffer.isValid()); + } + assertFalse(buffer.isValid()); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUErrorClassifierTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUErrorClassifierTest.java new file mode 100644 index 0000000..b4dcfd1 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/GPUErrorClassifierTest.java @@ -0,0 +1,253 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.junit.jupiter.api.Test; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for GPUErrorClassifier. + * Tests error classification, code extraction, and edge cases. + */ +class GPUErrorClassifierTest { + + // --- Programming Error Tests --- + + @Test + void testNullInputReturnsFalse() { + assertFalse(GPUErrorClassifier.isProgrammingError(null)); + assertFalse(GPUErrorClassifier.isRecoverable(null)); + } + + @Test + void testIllegalStateExceptionIsProgrammingError() { + var ex = new IllegalStateException("test"); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + assertFalse(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testIllegalArgumentExceptionIsProgrammingError() { + var ex = new IllegalArgumentException("test"); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + assertFalse(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testNullPointerExceptionIsProgrammingError() { + var ex = new NullPointerException("test"); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + assertFalse(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testAssertionErrorIsProgrammingError() { + var ex = new AssertionError("test"); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + assertFalse(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testKernelCompilationExceptionIsProgrammingError() { + var ex = new ComputeKernel.KernelCompilationException("compilation failed"); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + assertFalse(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testProgrammingErrorCodes() { + int[] programmingCodes = {-11, -44, -45, -46, -47, -48, -49, -50, -51, -52, + -53, -54, -55, -56, -57, -58, -30, -33, -34, -35, + -36, -37, -38, -63}; + for (int errorCode : programmingCodes) { + var ex = new RuntimeException("OpenCL error code: " + errorCode); + assertTrue(GPUErrorClassifier.isProgrammingError(ex), + "Error code " + errorCode + " should be programming error"); + } + } + + // --- Recoverable Error Tests --- + + @Test + void testKernelExecutionExceptionIsRecoverable() { + var ex = new ComputeKernel.KernelExecutionException("execution failed"); + assertFalse(GPUErrorClassifier.isProgrammingError(ex)); + assertTrue(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testRecoverableErrorCodes() { + int[] recoverableCodes = {-4, -5, -6, -59}; + for (int errorCode : recoverableCodes) { + var ex = new RuntimeException("OpenCL error code: " + errorCode); + assertFalse(GPUErrorClassifier.isProgrammingError(ex), + "Error code " + errorCode + " should not be programming error"); + assertTrue(GPUErrorClassifier.isRecoverable(ex), + "Error code " + errorCode + " should be recoverable"); + } + } + + @Test + void testGPUOutOfMemoryIsRecoverable() { + var ex = new OutOfMemoryError("GPU buffer allocation failed"); + assertFalse(GPUErrorClassifier.isProgrammingError(ex)); + assertTrue(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testOpenCLOutOfMemoryIsRecoverable() { + var ex = new OutOfMemoryError("OpenCL memory exhausted"); + assertFalse(GPUErrorClassifier.isProgrammingError(ex)); + assertTrue(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testUnknownErrorIsRecoverable() { + // Unknown errors default to recoverable (conservative approach) + var ex = new RuntimeException("some unknown error"); + assertFalse(GPUErrorClassifier.isProgrammingError(ex)); + assertTrue(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testUnknownErrorCodeIsRecoverable() { + // Unknown error code that's not in either set + var ex = new RuntimeException("OpenCL error code: -999"); + assertFalse(GPUErrorClassifier.isProgrammingError(ex)); + assertTrue(GPUErrorClassifier.isRecoverable(ex)); + } + + // --- Exception Chain Tests --- + + @Test + void testWrappedProgrammingErrorDetected() { + var cause = new IllegalArgumentException("inner"); + var ex = new RuntimeException("wrapper", cause); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + } + + @Test + void testDeepChainProgrammingErrorDetected() { + var deepCause = new NullPointerException("deep"); + var midCause = new RuntimeException("mid", deepCause); + var ex = new RuntimeException("outer", midCause); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + } + + @Test + void testChainWithErrorCodeDetected() { + var cause = new RuntimeException("OpenCL error code: -48"); + var ex = new RuntimeException("wrapper", cause); + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + } + + @Test + void testSelfReferencingCauseHandled() { + // Create exception with self-referencing cause (shouldn't infinite loop) + var ex = new RuntimeException("self") { + @Override + public synchronized Throwable getCause() { + return this; + } + }; + // Should complete without stack overflow + assertFalse(GPUErrorClassifier.isProgrammingError(ex)); + } + + // --- Error Code Extraction Tests --- + + @Test + void testExtractErrorCodeFromMessage() { + assertEquals(-48, GPUErrorClassifier.extractOpenCLErrorCode( + new RuntimeException("error code: -48"))); + assertEquals(-5, GPUErrorClassifier.extractOpenCLErrorCode( + new RuntimeException("Error Code -5"))); + assertEquals(-11, GPUErrorClassifier.extractOpenCLErrorCode( + new RuntimeException("OpenCL error code: -11 occurred"))); + } + + @Test + void testExtractErrorCodeReturnsZeroWhenNotFound() { + assertEquals(0, GPUErrorClassifier.extractOpenCLErrorCode( + new RuntimeException("no error code here"))); + assertEquals(0, GPUErrorClassifier.extractOpenCLErrorCode( + new RuntimeException((String) null))); + } + + // --- Error Category Description Tests --- + + @Test + void testNullExceptionCategory() { + assertEquals("null exception", GPUErrorClassifier.getErrorCategory(null)); + } + + @Test + void testProgrammingErrorCategory() { + var category = GPUErrorClassifier.getErrorCategory(new IllegalStateException("test")); + assertTrue(category.contains("PROGRAMMING ERROR")); + } + + @Test + void testMemoryAllocationFailureCategory() { + var category = GPUErrorClassifier.getErrorCategory( + new RuntimeException("error code: -4")); + assertTrue(category.contains("MEMORY ALLOCATION FAILURE")); + } + + @Test + void testOutOfResourcesCategory() { + var category = GPUErrorClassifier.getErrorCategory( + new RuntimeException("error code: -5")); + assertTrue(category.contains("OUT OF GPU RESOURCES")); + } + + @Test + void testOutOfHostMemoryCategory() { + var category = GPUErrorClassifier.getErrorCategory( + new RuntimeException("error code: -6")); + assertTrue(category.contains("OUT OF HOST MEMORY")); + } + + @Test + void testUnknownErrorCategory() { + var category = GPUErrorClassifier.getErrorCategory( + new RuntimeException("something unknown")); + assertTrue(category.contains("UNKNOWN ERROR")); + } + + // --- Error Name Tests --- + + @Test + void testGetOpenCLErrorName() { + assertEquals("CL_SUCCESS", GPUErrorClassifier.getOpenCLErrorName(0)); + assertEquals("CL_BUILD_PROGRAM_FAILURE", GPUErrorClassifier.getOpenCLErrorName(-11)); + assertEquals("CL_INVALID_KERNEL", GPUErrorClassifier.getOpenCLErrorName(-48)); + assertEquals("CL_OUT_OF_RESOURCES", GPUErrorClassifier.getOpenCLErrorName(-5)); + assertEquals("CL_OUT_OF_HOST_MEMORY", GPUErrorClassifier.getOpenCLErrorName(-6)); + } + + @Test + void testGetOpenCLErrorNameUnknown() { + var name = GPUErrorClassifier.getOpenCLErrorName(-999); + assertTrue(name.contains("UNKNOWN")); + assertTrue(name.contains("-999")); + } + + // --- Edge Cases --- + + @Test + void testNonGPUOutOfMemoryIsRecoverable() { + // Regular OOM without GPU keywords + var ex = new OutOfMemoryError("Java heap space"); + // Still recoverable by default conservative approach + assertTrue(GPUErrorClassifier.isRecoverable(ex)); + } + + @Test + void testProgrammingErrorCodeOverridesRecoverable() { + // A KernelExecutionException with programming error code in message + var ex = new ComputeKernel.KernelExecutionException("error code: -48"); + // The error code takes precedence + assertTrue(GPUErrorClassifier.isProgrammingError(ex)); + assertFalse(GPUErrorClassifier.isRecoverable(ex)); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/KernelLoaderTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/KernelLoaderTest.java new file mode 100644 index 0000000..8aeaf98 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/KernelLoaderTest.java @@ -0,0 +1,128 @@ +package com.hellblazer.luciferase.resource.compute; + +import org.junit.jupiter.api.AfterEach; +import org.junit.jupiter.api.BeforeEach; +import org.junit.jupiter.api.Test; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for KernelLoader. + */ +class KernelLoaderTest { + + @BeforeEach + void setUp() { + KernelLoader.testClearCache(); + } + + @AfterEach + void tearDown() { + KernelLoader.testClearCache(); + } + + // --- Basic Loading Tests --- + + @Test + void testLoadKernelFromPath() { + // Load test kernel from test resources + var source = KernelLoader.loadKernel("kernels/test_kernel.cl"); + + assertNotNull(source); + assertFalse(source.isEmpty()); + assertTrue(source.contains("__kernel")); + } + + @Test + void testLoadTestKernel() { + // Uses convention: kernels/{name}.cl + var source = KernelLoader.loadTestKernel("test_kernel"); + + assertNotNull(source); + assertTrue(source.contains("__kernel")); + } + + @Test + void testLoadNonExistentKernelThrows() { + assertThrows(KernelLoader.KernelLoadException.class, + () -> KernelLoader.loadKernel("kernels/does_not_exist.cl")); + } + + @Test + void testLoadOpenCLKernelThrowsIfNotFound() { + // No kernel at kernels/opencl/ path in test resources + assertThrows(KernelLoader.KernelLoadException.class, + () -> KernelLoader.loadOpenCLKernel("nonexistent")); + } + + // --- Caching Tests --- + + @Test + void testCaching() { + assertEquals(0, KernelLoader.getCacheSize()); + + KernelLoader.loadKernel("kernels/test_kernel.cl"); + assertEquals(1, KernelLoader.getCacheSize()); + + // Load again - should use cache + KernelLoader.loadKernel("kernels/test_kernel.cl"); + assertEquals(1, KernelLoader.getCacheSize()); + } + + @Test + void testCacheClear() { + KernelLoader.loadKernel("kernels/test_kernel.cl"); + assertEquals(1, KernelLoader.getCacheSize()); + + KernelLoader.testClearCache(); + assertEquals(0, KernelLoader.getCacheSize()); + } + + @Test + void testCachedContentSame() { + var first = KernelLoader.loadKernel("kernels/test_kernel.cl"); + var second = KernelLoader.loadKernel("kernels/test_kernel.cl"); + + assertSame(first, second, "Cached content should be same object"); + } + + // --- Kernel Exists Tests --- + + @Test + void testKernelExistsTrue() { + assertTrue(KernelLoader.kernelExists("kernels/test_kernel.cl")); + } + + @Test + void testKernelExistsFalse() { + assertFalse(KernelLoader.kernelExists("kernels/nonexistent.cl")); + } + + @Test + void testKernelExistsUsesCache() { + // Pre-load into cache + KernelLoader.loadKernel("kernels/test_kernel.cl"); + + // kernelExists should find it in cache + assertTrue(KernelLoader.kernelExists("kernels/test_kernel.cl")); + } + + // --- Content Validation Tests --- + + @Test + void testLoadedContentIsValid() { + var source = KernelLoader.loadKernel("kernels/test_kernel.cl"); + + // Should be valid OpenCL source + assertTrue(source.contains("__kernel")); + assertTrue(source.contains("void")); + } + + @Test + void testLoadedContentPreservesNewlines() { + var source = KernelLoader.loadKernel("kernels/test_kernel.cl"); + + // Multi-line kernel should have newlines + assertTrue(source.contains("\n")); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/CustomKernelExample.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/CustomKernelExample.java new file mode 100644 index 0000000..f07689a --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/CustomKernelExample.java @@ -0,0 +1,101 @@ +package com.hellblazer.luciferase.resource.compute.examples; + +import com.hellblazer.luciferase.resource.compute.ComputeService; + +/** + * Custom kernel for operations not in the built-in set. + * + *

Run with: mvn test -Dtest=CustomKernelExample + */ +public class CustomKernelExample { + + public static void main(String[] args) throws Exception { + var compute = ComputeService.getInstance(); + + if (!compute.isGPUAvailable()) { + System.out.println("GPU not available, skipping custom kernel example"); + return; + } + + System.out.println("Backend: " + compute.getBackend().getDisplayName()); + + // Element-wise multiply kernel + String multiplyKernel = """ + __kernel void multiply( + __global const float* a, + __global const float* b, + __global float* result, + const int size) + { + int i = get_global_id(0); + if (i < size) { + result[i] = a[i] * b[i]; + } + } + """; + + float[] prices = {10.50f, 25.00f, 8.99f, 42.00f, 15.75f}; + float[] quantities = {2, 1, 5, 1, 3}; + + System.out.println("\nElement-wise Multiply (prices * quantities):"); + System.out.println(" prices = " + formatArray(prices)); + System.out.println(" quantities = " + formatArray(quantities)); + + try (var op = compute.createOperation("multiply", multiplyKernel, "multiply")) { + op.setInput(0, prices); + op.setInput(1, quantities); + op.setOutput(2, prices.length); + op.setArg(3, prices.length); + + float[] totals = op.execute(prices.length); + System.out.println(" totals = " + formatArray(totals)); + + // Sum the totals + float grandTotal = compute.sum(totals); + System.out.println(" grand total = " + String.format("%.2f", grandTotal)); + } + + // Threshold kernel + String thresholdKernel = """ + __kernel void threshold( + __global const float* input, + __global float* output, + const float thresh, + const int size) + { + int i = get_global_id(0); + if (i < size) { + output[i] = (input[i] > thresh) ? 1.0f : 0.0f; + } + } + """; + + float[] values = {0.1f, 0.6f, 0.3f, 0.8f, 0.4f, 0.9f, 0.2f}; + float threshold = 0.5f; + + System.out.println("\nThreshold (values > 0.5):"); + System.out.println(" input = " + formatArray(values)); + + try (var op = compute.createOperation("threshold", thresholdKernel, "threshold")) { + op.setInput(0, values); + op.setOutput(1, values.length); + op.setArg(2, threshold); + op.setArg(3, values.length); + + float[] binary = op.execute(values.length); + System.out.println(" output = " + formatArray(binary)); + } + + // Cleanup + ComputeService.testReset(); + } + + private static String formatArray(float[] arr) { + var sb = new StringBuilder("["); + for (int i = 0; i < arr.length; i++) { + if (i > 0) sb.append(", "); + sb.append(String.format("%.2f", arr[i])); + } + return sb.append("]").toString(); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/LowLevelExample.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/LowLevelExample.java new file mode 100644 index 0000000..3ac8058 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/LowLevelExample.java @@ -0,0 +1,105 @@ +package com.hellblazer.luciferase.resource.compute.examples; + +import com.hellblazer.luciferase.resource.compute.ComputeKernel; +import com.hellblazer.luciferase.resource.compute.KernelLoader; +import com.hellblazer.luciferase.resource.compute.opencl.OpenCLBuffer; +import com.hellblazer.luciferase.resource.compute.opencl.OpenCLContext; +import com.hellblazer.luciferase.resource.compute.opencl.OpenCLKernel; + +import static com.hellblazer.luciferase.resource.compute.opencl.OpenCLBuffer.BufferAccess.*; + +/** + * Low-level API for direct buffer and kernel control. + * + *

Use when you need: + *

    + *
  • Buffer reuse across operations
    • + *
  • Fine-grained timing
    • + *
  • Multiple kernels sharing buffers
    • + *
+ * + *

Run with: mvn test -Dtest=LowLevelExample + */ +public class LowLevelExample { + + public static void main(String[] args) throws Exception { + // Check if OpenCL available + if (!OpenCLContext.getInstance().isInitialized()) { + System.out.println("OpenCL not available"); + return; + } + + System.out.println("OpenCL initialized"); + + // Load kernel source from resources + var source = KernelLoader.loadOpenCLKernel("vector_add"); + + // Test data + float[] a = {1, 2, 3, 4, 5, 6, 7, 8}; + float[] b = {8, 7, 6, 5, 4, 3, 2, 1}; + int size = a.length; + + System.out.println("\nLow-Level Vector Addition"); + System.out.println("========================="); + System.out.println("Input a: " + formatArray(a)); + System.out.println("Input b: " + formatArray(b)); + + // Create buffers and kernel + try (var bufA = OpenCLBuffer.createWithData(a, READ_ONLY); + var bufB = OpenCLBuffer.createWithData(b, READ_ONLY); + var bufResult = OpenCLBuffer.create(size, WRITE_ONLY); + var kernel = OpenCLKernel.create("vectorAdd")) { + + // Compile kernel + kernel.compile(source, "vectorAdd"); + + // Set arguments + kernel.setBufferArg(0, bufA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, size); + + // Execute + long start = System.nanoTime(); + kernel.execute(size); + kernel.finish(); + long elapsed = System.nanoTime() - start; + + // Download result + var result = new float[size]; + bufResult.download(result); + + System.out.println("Result: " + formatArray(result)); + System.out.printf("Time: %.3f ms%n", elapsed / 1_000_000.0); + + // Reuse buffers for another operation + System.out.println("\nBuffer Reuse Example"); + System.out.println("===================="); + + // Upload new data to same buffer + float[] c = {10, 20, 30, 40, 50, 60, 70, 80}; + bufA.upload(c); + + // Re-execute with same kernel setup + kernel.execute(size); + kernel.finish(); + bufResult.download(result); + + System.out.println("New a: " + formatArray(c)); + System.out.println("Same b: " + formatArray(b)); + System.out.println("Result: " + formatArray(result)); + } + + // Cleanup + OpenCLContext.testReset(); + } + + private static String formatArray(float[] arr) { + var sb = new StringBuilder("["); + for (int i = 0; i < arr.length; i++) { + if (i > 0) sb.append(", "); + sb.append(String.format("%.0f", arr[i])); + } + return sb.append("]").toString(); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/PerformanceExample.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/PerformanceExample.java new file mode 100644 index 0000000..85a50ae --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/PerformanceExample.java @@ -0,0 +1,87 @@ +package com.hellblazer.luciferase.resource.compute.examples; + +import com.hellblazer.luciferase.resource.compute.ComputeService; + +import java.util.Random; + +/** + * GPU vs CPU performance comparison. + * + *

Run with: mvn test -Dtest=PerformanceExample + */ +public class PerformanceExample { + + public static void main(String[] args) { + var compute = ComputeService.getInstance(); + System.out.println("Backend: " + compute.getBackend().getDisplayName()); + System.out.println("GPU Available: " + compute.isGPUAvailable()); + + // Test different array sizes + int[] sizes = {1_000, 10_000, 100_000, 1_000_000, 10_000_000}; + + System.out.println("\nVector Addition Performance (GPU vs CPU)"); + System.out.println("========================================="); + System.out.printf("%-12s %12s %12s %10s%n", "Size", "GPU (ms)", "CPU (ms)", "Speedup"); + System.out.println("-".repeat(48)); + + var rand = new Random(42); + + for (int size : sizes) { + // Generate test data + float[] a = new float[size]; + float[] b = new float[size]; + for (int i = 0; i < size; i++) { + a[i] = rand.nextFloat(); + b[i] = rand.nextFloat(); + } + + // Warm up + compute.vectorAdd(a, b); + + // Time GPU (via ComputeService) + long gpuStart = System.nanoTime(); + float[] gpuResult = compute.vectorAdd(a, b); + long gpuTime = System.nanoTime() - gpuStart; + + // Time CPU + long cpuStart = System.nanoTime(); + float[] cpuResult = new float[size]; + for (int i = 0; i < size; i++) { + cpuResult[i] = a[i] + b[i]; + } + long cpuTime = System.nanoTime() - cpuStart; + + // Verify + boolean correct = true; + for (int i = 0; i < size; i++) { + if (Math.abs(gpuResult[i] - cpuResult[i]) > 0.0001f) { + correct = false; + break; + } + } + + double gpuMs = gpuTime / 1_000_000.0; + double cpuMs = cpuTime / 1_000_000.0; + double speedup = cpuMs / gpuMs; + + System.out.printf("%-12s %12.2f %12.2f %9.2fx %s%n", + formatSize(size), gpuMs, cpuMs, speedup, + correct ? "" : "[MISMATCH]"); + } + + System.out.println("\nNote: GPU includes data transfer overhead."); + System.out.println(" Larger arrays amortize transfer cost better."); + + // Cleanup + ComputeService.testReset(); + } + + private static String formatSize(int size) { + if (size >= 1_000_000) { + return (size / 1_000_000) + "M"; + } else if (size >= 1_000) { + return (size / 1_000) + "K"; + } + return String.valueOf(size); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/VectorMathExample.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/VectorMathExample.java new file mode 100644 index 0000000..cd08d76 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/VectorMathExample.java @@ -0,0 +1,65 @@ +package com.hellblazer.luciferase.resource.compute.examples; + +import com.hellblazer.luciferase.resource.compute.ComputeService; + +/** + * Basic vector math operations. + * + *

Run with: mvn test -Dtest=VectorMathExample + */ +public class VectorMathExample { + + public static void main(String[] args) { + var compute = ComputeService.getInstance(); + System.out.println("Backend: " + compute.getBackend().getDisplayName()); + + // Sample data + float[] prices = {10.50f, 25.00f, 8.99f, 42.00f, 15.75f}; + float[] quantities = {2, 1, 5, 1, 3}; + + // Vector multiply (prices * quantities) using SAXPY trick + // We want: result = prices * quantities + // SAXPY gives: result = alpha * x + y + // So: set y to zeros, alpha to 1, then element-wise wouldn't work... + // Actually, let's use custom kernel for multiply + + // For now, demonstrate what's built-in: + + // Addition + float[] a = {1, 2, 3, 4, 5}; + float[] b = {10, 20, 30, 40, 50}; + float[] sum = compute.vectorAdd(a, b); + System.out.println("\nVector Add:"); + System.out.println(" a = " + formatArray(a)); + System.out.println(" b = " + formatArray(b)); + System.out.println(" a + b = " + formatArray(sum)); + + // SAXPY: result = 2*a + b + float[] saxpyResult = compute.saxpy(2.0f, a, b); + System.out.println("\nSAXPY (2*a + b):"); + System.out.println(" result = " + formatArray(saxpyResult)); + + // Scale + float[] scaled = compute.scale(a, 0.5f); + System.out.println("\nScale (a * 0.5):"); + System.out.println(" result = " + formatArray(scaled)); + + // Reductions + System.out.println("\nReductions on a:"); + System.out.println(" sum = " + compute.sum(a)); + System.out.println(" min = " + compute.min(a)); + System.out.println(" max = " + compute.max(a)); + + // Cleanup + ComputeService.testReset(); + } + + private static String formatArray(float[] arr) { + var sb = new StringBuilder("["); + for (int i = 0; i < arr.length; i++) { + if (i > 0) sb.append(", "); + sb.append(String.format("%.2f", arr[i])); + } + return sb.append("]").toString(); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/package-info.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/package-info.java new file mode 100644 index 0000000..4485e95 --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/examples/package-info.java @@ -0,0 +1,21 @@ +/** + * Runnable examples demonstrating GPU compute usage. + * + *

Examples

+ *
    + *
  • {@link VectorMathExample} - Built-in vector operations
    • + *
  • {@link CustomKernelExample} - Writing custom OpenCL kernels
    • + *
  • {@link PerformanceExample} - GPU vs CPU timing comparison
    • + *
  • {@link LowLevelExample} - Direct buffer and kernel control
    • + *
+ * + *

Running

+ * 
+ * # Run a specific example
+ * mvn exec:java -Dexec.mainClass="...examples.VectorMathExample" -pl resource
+ *
+ * # Or run as test
+ * mvn test -Dtest=VectorMathExample -pl resource
+ *
+ */ +package com.hellblazer.luciferase.resource.compute.examples; diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/ComputeIntegrationTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/ComputeIntegrationTest.java new file mode 100644 index 0000000..38e24ea --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/ComputeIntegrationTest.java @@ -0,0 +1,440 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import com.hellblazer.luciferase.resource.compute.BackendSelector; +import com.hellblazer.luciferase.resource.compute.ComputeKernel; +import com.hellblazer.luciferase.resource.compute.GPUBackend; +import org.junit.jupiter.api.BeforeAll; +import org.junit.jupiter.api.Test; +import org.junit.jupiter.api.condition.DisabledIfEnvironmentVariable; +import org.lwjgl.PointerBuffer; +import org.lwjgl.opencl.CL10; +import org.lwjgl.system.MemoryStack; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Integration tests for Phase 2 OpenCL components. + * + *

Tests the complete workflow: context → buffer → kernel → execute → read. + * Tests require OpenCL availability - skipped if not available and in CI environments. + */ +@DisabledIfEnvironmentVariable(named = "CI", matches = "true", disabledReason = "OpenCL not available in CI") +class ComputeIntegrationTest { + + private static boolean openCLAvailable; + + // Vector addition kernel + private static final String VECTOR_ADD_KERNEL = """ + __kernel void vectorAdd(__global const float* a, + __global const float* b, + __global float* result, + const int size) { + int gid = get_global_id(0); + if (gid < size) { + result[gid] = a[gid] + b[gid]; + } + } + """; + + // SAXPY kernel: result = a * x + y + private static final String SAXPY_KERNEL = """ + __kernel void saxpy(__global const float* x, + __global const float* y, + __global float* result, + const float a, + const int size) { + int gid = get_global_id(0); + if (gid < size) { + result[gid] = a * x[gid] + y[gid]; + } + } + """; + + // Matrix multiply kernel (simplified for small matrices) + private static final String DOT_PRODUCT_KERNEL = """ + __kernel void dotProduct(__global const float* a, + __global const float* b, + __global float* result, + __local float* scratch, + const int size) { + int gid = get_global_id(0); + int lid = get_local_id(0); + int groupSize = get_local_size(0); + + // Each work item computes partial product + float partial = 0.0f; + for (int i = gid; i < size; i += get_global_size(0)) { + partial += a[i] * b[i]; + } + scratch[lid] = partial; + + barrier(CLK_LOCAL_MEM_FENCE); + + // Reduction within work group + for (int stride = groupSize / 2; stride > 0; stride /= 2) { + if (lid < stride) { + scratch[lid] += scratch[lid + stride]; + } + barrier(CLK_LOCAL_MEM_FENCE); + } + + if (lid == 0) { + result[get_group_id(0)] = scratch[0]; + } + } + """; + + @BeforeAll + static void checkOpenCL() { + try (var stack = MemoryStack.stackPush()) { + var numPlatforms = stack.mallocInt(1); + var errcode = CL10.clGetPlatformIDs((PointerBuffer) null, numPlatforms); + + if (errcode == CL10.CL_SUCCESS && numPlatforms.get(0) > 0) { + var platformBuffer = stack.mallocPointer(1); + CL10.clGetPlatformIDs(platformBuffer, (int[]) null); + var platform = platformBuffer.get(0); + + var numDevices = stack.mallocInt(1); + var result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_GPU, null, numDevices); + if (result != CL10.CL_SUCCESS) { + result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_CPU, null, numDevices); + } + + openCLAvailable = result == CL10.CL_SUCCESS && numDevices.get(0) > 0; + } + System.out.println("OpenCL available for integration tests: " + openCLAvailable); + } catch (Exception e) { + openCLAvailable = false; + System.out.println("OpenCL check failed: " + e.getMessage()); + } + } + + // --- Full Workflow Tests --- + + @Test + void testFullVectorAddWorkflow() throws Exception { + if (!openCLAvailable) return; + + int size = 1024; + var a = new float[size]; + var b = new float[size]; + var result = new float[size]; + + // Initialize input data + for (int i = 0; i < size; i++) { + a[i] = i * 0.5f; + b[i] = i * 0.25f; + } + + // Full workflow: context -> buffers -> kernel -> execute -> read + try (var bufferA = OpenCLBuffer.createWithData(a, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferB = OpenCLBuffer.createWithData(b, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var kernel = OpenCLKernel.create("vectorAdd")) { + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + kernel.setBufferArg(0, bufferA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, size); + + kernel.execute(size); + kernel.finish(); + + bufferResult.download(result); + + // Verify results with CPU reference + for (int i = 0; i < size; i++) { + float expected = a[i] + b[i]; + assertEquals(expected, result[i], 0.0001f, + "Mismatch at index " + i); + } + } + } + + @Test + void testFullSAXPYWorkflow() throws Exception { + if (!openCLAvailable) return; + + int size = 2048; + float alpha = 2.5f; + var x = new float[size]; + var y = new float[size]; + var result = new float[size]; + + // Initialize input data + for (int i = 0; i < size; i++) { + x[i] = i; + y[i] = size - i; + } + + try (var bufferX = OpenCLBuffer.createWithData(x, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferY = OpenCLBuffer.createWithData(y, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var kernel = OpenCLKernel.create("saxpy")) { + + kernel.compile(SAXPY_KERNEL, "saxpy"); + + kernel.setBufferArg(0, bufferX, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferY, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setFloatArg(3, alpha); + kernel.setIntArg(4, size); + + kernel.execute(size); + kernel.finish(); + + bufferResult.download(result); + + // Verify: result = alpha * x + y + for (int i = 0; i < size; i++) { + float expected = alpha * x[i] + y[i]; + assertEquals(expected, result[i], 0.0001f, + "SAXPY mismatch at index " + i); + } + } + } + + // --- OpenCLContext Integration Tests --- + + @Test + void testOpenCLContextAcquireRelease() { + if (!openCLAvailable) return; + + var ctx = OpenCLContext.getInstance(); + + // Acquire should initialize + if (!ctx.isInitialized()) { + ctx.acquire(); + } + + assertTrue(ctx.isInitialized()); + assertTrue(ctx.getContext() != 0); + assertTrue(ctx.getCommandQueue() != 0); + assertTrue(ctx.getDevice() != 0); + + // Multiple acquires should work + ctx.acquire(); + ctx.acquire(); + assertTrue(ctx.getRefCount() >= 1); + + // Releases should work + ctx.release(); + ctx.release(); + } + + @Test + void testBackendSelectorSelectsOpenCL() { + if (!openCLAvailable) return; + + var backend = BackendSelector.getOptimalBackend(); + assertNotNull(backend); + + // Should select OpenCL or Metal (both are GPU backends) + assertTrue(backend.isGPU() || backend == GPUBackend.CPU_FALLBACK); + + if (backend == GPUBackend.OPENCL) { + assertEquals("OpenCL", backend.getDisplayName()); + } + } + + // --- Large Data Tests --- + + @Test + void testLargeVectorAdd() throws Exception { + if (!openCLAvailable) return; + + int size = 65536; // 64K elements + var a = new float[size]; + var b = new float[size]; + var result = new float[size]; + + // Initialize with pattern + for (int i = 0; i < size; i++) { + a[i] = (float) Math.sin(i * 0.01); + b[i] = (float) Math.cos(i * 0.01); + } + + try (var bufferA = OpenCLBuffer.createWithData(a, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferB = OpenCLBuffer.createWithData(b, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var kernel = OpenCLKernel.create("vectorAdd")) { + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + kernel.setBufferArg(0, bufferA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, size); + + kernel.execute(size); + kernel.finish(); + + bufferResult.download(result); + + // Spot check results + for (int i = 0; i < size; i += 1024) { + float expected = a[i] + b[i]; + assertEquals(expected, result[i], 0.0001f, + "Mismatch at index " + i); + } + } + } + + // --- Multiple Kernel Execution Tests --- + + @Test + void testMultipleKernelExecutions() throws Exception { + if (!openCLAvailable) return; + + int size = 256; + var data = new float[size]; + + // Initialize + for (int i = 0; i < size; i++) { + data[i] = i; + } + + try (var buffer = OpenCLBuffer.createWithData(data, OpenCLBuffer.BufferAccess.READ_WRITE); + var tempBuffer = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.READ_WRITE); + var kernel = OpenCLKernel.create("vectorAdd")) { + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + // Execute multiple times: data = data + data (doubling each time) + for (int iter = 0; iter < 3; iter++) { + // Copy buffer to temp + var temp = new float[size]; + buffer.download(temp); + tempBuffer.upload(temp); + + // result = buffer + temp (doubles the values) + kernel.setBufferArg(0, buffer, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, tempBuffer, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, buffer, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, size); + + kernel.execute(size); + kernel.finish(); + } + + var result = new float[size]; + buffer.download(result); + + // After 3 iterations of doubling: result = data * 2^3 = data * 8 + for (int i = 0; i < size; i++) { + float expected = data[i] * 8; + assertEquals(expected, result[i], 0.0001f, + "Mismatch at index " + i + " after multiple executions"); + } + } + } + + // 2D vector addition kernel with proper 2D indexing + private static final String VECTOR_ADD_2D_KERNEL = """ + __kernel void vectorAdd2D(__global const float* a, + __global const float* b, + __global float* result, + const int width, + const int height) { + int x = get_global_id(0); + int y = get_global_id(1); + if (x < width && y < height) { + int idx = y * width + x; + result[idx] = a[idx] + b[idx]; + } + } + """; + + // --- 2D Execution Tests --- + + @Test + void test2DExecution() throws Exception { + if (!openCLAvailable) return; + + int width = 64; + int height = 64; + int size = width * height; + + var a = new float[size]; + var b = new float[size]; + var result = new float[size]; + + for (int i = 0; i < size; i++) { + a[i] = i; + b[i] = size - i; + } + + try (var bufferA = OpenCLBuffer.createWithData(a, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferB = OpenCLBuffer.createWithData(b, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var kernel = OpenCLKernel.create("vectorAdd2D")) { + + kernel.compile(VECTOR_ADD_2D_KERNEL, "vectorAdd2D"); + + kernel.setBufferArg(0, bufferA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, width); + kernel.setIntArg(4, height); + + // Execute as 2D grid + kernel.execute(width, height); + kernel.finish(); + + bufferResult.download(result); + + // Each element should be: a[i] + b[i] = i + (size - i) = size + for (int i = 0; i < size; i++) { + assertEquals(size, result[i], 0.0001f, + "2D execution mismatch at index " + i); + } + } + } + + // --- Resource Cleanup Tests --- + + @Test + void testResourcesReleasedOnClose() throws Exception { + if (!openCLAvailable) return; + + // Create and immediately close resources + var kernel = OpenCLKernel.create("test"); + var buffer = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_WRITE); + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + assertTrue(kernel.isValid()); + assertTrue(buffer.isValid()); + + kernel.close(); + buffer.close(); + + assertFalse(kernel.isValid()); + assertFalse(buffer.isValid()); + } + + @Test + void testTryWithResourcesCleanup() throws Exception { + if (!openCLAvailable) return; + + OpenCLBuffer buffer; + OpenCLKernel kernel; + + try (var b = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_WRITE); + var k = OpenCLKernel.create("test")) { + buffer = b; + kernel = k; + k.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + assertTrue(buffer.isValid()); + assertTrue(kernel.isValid()); + } + + // After try-with-resources, resources should be released + assertFalse(buffer.isValid()); + assertFalse(kernel.isValid()); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/MinimalOpenCLTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/MinimalOpenCLTest.java new file mode 100644 index 0000000..79853cc --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/MinimalOpenCLTest.java @@ -0,0 +1,115 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import org.junit.jupiter.api.BeforeAll; +import org.junit.jupiter.api.Test; +import org.junit.jupiter.api.condition.DisabledIfEnvironmentVariable; +import org.lwjgl.PointerBuffer; +import org.lwjgl.opencl.CL10; +import org.lwjgl.system.MemoryStack; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Minimal test to verify OpenCL platform detection works. + */ +@DisabledIfEnvironmentVariable(named = "CI", matches = "true", disabledReason = "OpenCL not available in CI") +class MinimalOpenCLTest { + + private static boolean openCLAvailable; + + @BeforeAll + static void checkOpenCL() { + // Only check device availability - don't create context + // This mimics OpenCLBufferTest's @BeforeAll + try (var stack = MemoryStack.stackPush()) { + System.out.println("Step 1: Getting platform count"); + var numPlatforms = stack.mallocInt(1); + var errcode = CL10.clGetPlatformIDs((PointerBuffer) null, numPlatforms); + + if (errcode != CL10.CL_SUCCESS || numPlatforms.get(0) == 0) { + openCLAvailable = false; + System.out.println("No platforms"); + return; + } + + System.out.println("Step 2: Getting platform"); + var platformBuffer = stack.mallocPointer(1); + CL10.clGetPlatformIDs(platformBuffer, (int[]) null); + var platform = platformBuffer.get(0); + System.out.println("Got platform: " + platform); + + System.out.println("Step 3: Checking for devices"); + var numDevices = stack.mallocInt(1); + var result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_GPU, null, numDevices); + if (result != CL10.CL_SUCCESS) { + result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_CPU, null, numDevices); + } + + openCLAvailable = result == CL10.CL_SUCCESS && numDevices.get(0) > 0; + System.out.println("OpenCL available: " + openCLAvailable); + } catch (Exception e) { + openCLAvailable = false; + System.out.println("OpenCL check failed: " + e.getMessage()); + e.printStackTrace(); + } + } + + @Test + void testMinimal() { + System.out.println("Running minimal test, OpenCL=" + openCLAvailable); + assertTrue(true, "Minimal test should pass"); + } + + @Test + void testOpenCLContextBuffer() { + System.out.println("Testing OpenCLContext buffer creation"); + if (!openCLAvailable) { + System.out.println("OpenCL not available - skipping"); + return; + } + + try { + System.out.println("Acquiring OpenCLContext..."); + var ctx = OpenCLContext.getInstance(); + if (!ctx.isInitialized()) { + ctx.acquire(); + } + System.out.println("OpenCLContext acquired, context=" + ctx.getContext()); + + try (var stack = org.lwjgl.system.MemoryStack.stackPush()) { + var errcode = stack.mallocInt(1); + System.out.println("Creating buffer via OpenCLContext..."); + var buffer = CL10.clCreateBuffer(ctx.getContext(), CL10.CL_MEM_READ_WRITE, 256 * 4, errcode); + System.out.println("Buffer created: errcode=" + errcode.get(0) + ", buffer=" + buffer); + assertTrue(buffer != 0, "Buffer should be created"); + } + } catch (Exception e) { + System.out.println("Exception: " + e.getMessage()); + e.printStackTrace(); + fail("Should not throw: " + e.getMessage()); + } + } + + @Test + void testOpenCLBufferCreate() { + System.out.println("Testing OpenCLBuffer.create()"); + if (!openCLAvailable) { + System.out.println("OpenCL not available - skipping"); + return; + } + + try { + System.out.println("Calling OpenCLBuffer.create(256, READ_WRITE)..."); + try (var buffer = OpenCLBuffer.create(256, OpenCLBuffer.BufferAccess.READ_WRITE)) { + System.out.println("OpenCLBuffer created: handle=" + buffer.getHandle() + ", size=" + buffer.size()); + assertNotNull(buffer); + assertEquals(256, buffer.size()); + } + System.out.println("OpenCLBuffer test complete"); + } catch (Exception e) { + System.out.println("Exception: " + e.getMessage()); + e.printStackTrace(); + fail("Should not throw: " + e.getMessage()); + } + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLBufferTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLBufferTest.java new file mode 100644 index 0000000..7be0f5e --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLBufferTest.java @@ -0,0 +1,202 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import org.junit.jupiter.api.BeforeAll; +import org.junit.jupiter.api.Test; +import org.junit.jupiter.api.condition.DisabledIfEnvironmentVariable; +import org.lwjgl.PointerBuffer; +import org.lwjgl.opencl.CL10; +import org.lwjgl.system.MemoryStack; +import org.lwjgl.system.MemoryUtil; + +import java.nio.FloatBuffer; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for OpenCLBuffer. + * Tests require OpenCL availability - skipped if not available and in CI environments. + */ +@DisabledIfEnvironmentVariable(named = "CI", matches = "true", disabledReason = "OpenCL not available in CI") +class OpenCLBufferTest { + + private static boolean openCLAvailable; + + @BeforeAll + static void checkOpenCL() { + try (var stack = MemoryStack.stackPush()) { + var numPlatforms = stack.mallocInt(1); + var errcode = CL10.clGetPlatformIDs((PointerBuffer) null, numPlatforms); + + if (errcode == CL10.CL_SUCCESS && numPlatforms.get(0) > 0) { + var platformBuffer = stack.mallocPointer(1); + CL10.clGetPlatformIDs(platformBuffer, (int[]) null); + var platform = platformBuffer.get(0); + + var numDevices = stack.mallocInt(1); + var result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_GPU, null, numDevices); + if (result != CL10.CL_SUCCESS) { + result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_CPU, null, numDevices); + } + + openCLAvailable = result == CL10.CL_SUCCESS && numDevices.get(0) > 0; + } + System.out.println("OpenCL available: " + openCLAvailable); + } catch (Exception e) { + openCLAvailable = false; + System.out.println("OpenCL check failed: " + e.getMessage()); + } + } + + // --- Creation Tests --- + + @Test + void testCreateBuffer() { + if (!openCLAvailable) return; + + try (var buffer = OpenCLBuffer.create(256, OpenCLBuffer.BufferAccess.READ_WRITE)) { + assertNotNull(buffer); + assertEquals(256, buffer.size()); + assertEquals(256 * Float.BYTES, buffer.sizeInBytes()); + assertTrue(buffer.isValid()); + } + } + + @Test + void testCreateWithData() { + if (!openCLAvailable) return; + + var data = new float[]{1.0f, 2.0f, 3.0f, 4.0f}; + try (var buffer = OpenCLBuffer.createWithData(data, OpenCLBuffer.BufferAccess.READ_WRITE)) { + assertNotNull(buffer); + assertEquals(4, buffer.size()); + assertTrue(buffer.isValid()); + } + } + + @Test + void testBufferAccessModes() { + if (!openCLAvailable) return; + + try (var readOnly = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_ONLY); + var writeOnly = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.WRITE_ONLY); + var readWrite = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_WRITE)) { + + assertEquals(OpenCLBuffer.BufferAccess.READ_ONLY, readOnly.getAccess()); + assertEquals(OpenCLBuffer.BufferAccess.WRITE_ONLY, writeOnly.getAccess()); + assertEquals(OpenCLBuffer.BufferAccess.READ_WRITE, readWrite.getAccess()); + } + } + + // --- Upload/Download Tests --- + + @Test + void testUploadDownloadFloatArray() { + if (!openCLAvailable) return; + + try (var buffer = OpenCLBuffer.create(4, OpenCLBuffer.BufferAccess.READ_WRITE)) { + var input = new float[]{1.0f, 2.0f, 3.0f, 4.0f}; + buffer.upload(input); + + var output = new float[4]; + buffer.download(output); + + assertArrayEquals(input, output, 0.0001f); + } + } + + @Test + void testUploadDownloadFloatBuffer() { + if (!openCLAvailable) return; + + try (var buffer = OpenCLBuffer.create(4, OpenCLBuffer.BufferAccess.READ_WRITE)) { + FloatBuffer input = MemoryUtil.memAllocFloat(4); + try { + input.put(new float[]{5.0f, 6.0f, 7.0f, 8.0f}); + input.flip(); + + buffer.upload(input); + + // Reset input position for reuse check + input.rewind(); + + FloatBuffer output = MemoryUtil.memAllocFloat(4); + try { + buffer.download(output); + // Note: download doesn't flip, data is at position 0 + assertEquals(5.0f, output.get(0), 0.0001f); + assertEquals(6.0f, output.get(1), 0.0001f); + assertEquals(7.0f, output.get(2), 0.0001f); + assertEquals(8.0f, output.get(3), 0.0001f); + } finally { + MemoryUtil.memFree(output); + } + } finally { + MemoryUtil.memFree(input); + } + } + } + + // --- Lifecycle Tests --- + + @Test + void testCloseMarksInvalid() { + if (!openCLAvailable) return; + + var buffer = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_WRITE); + assertTrue(buffer.isValid()); + + buffer.close(); + assertFalse(buffer.isValid()); + } + + @Test + void testDoubleCloseIsSafe() { + if (!openCLAvailable) return; + + var buffer = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_WRITE); + buffer.close(); + buffer.close(); // Should not throw + assertFalse(buffer.isValid()); + } + + @Test + void testOperationsAfterCloseThrow() { + if (!openCLAvailable) return; + + var buffer = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_WRITE); + buffer.close(); + + assertThrows(IllegalStateException.class, () -> buffer.upload(new float[4])); + assertThrows(IllegalStateException.class, () -> buffer.download(new float[4])); + assertThrows(IllegalStateException.class, buffer::getHandle); + } + + @Test + void testGetHandle() { + if (!openCLAvailable) return; + + try (var buffer = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_WRITE)) { + long handle = buffer.getHandle(); + assertTrue(handle != 0, "Handle should be non-zero"); + } + } + + @Test + void testLargerBuffer() { + if (!openCLAvailable) return; + + int size = 16384; + try (var buffer = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.READ_WRITE)) { + var input = new float[size]; + for (int i = 0; i < size; i++) { + input[i] = (float) Math.sin(i * 0.01); + } + buffer.upload(input); + + var output = new float[size]; + buffer.download(output); + + assertArrayEquals(input, output, 0.0001f); + } + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLContextTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLContextTest.java new file mode 100644 index 0000000..2ecb18b --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLContextTest.java @@ -0,0 +1,172 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import org.junit.jupiter.api.AfterEach; +import org.junit.jupiter.api.BeforeEach; +import org.junit.jupiter.api.Test; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for OpenCLContext singleton. + * Tests the singleton pattern, reference counting, and test isolation. + * + *

Note: These tests use testReset() for isolation. Actual OpenCL + * initialization is tested in integration tests with GPU access. + */ +class OpenCLContextTest { + + @BeforeEach + void setUp() { + // Reset singleton state before each test + OpenCLContext.testReset(); + } + + @AfterEach + void tearDown() { + // Clean up after each test + OpenCLContext.testReset(); + } + + // --- Singleton Pattern Tests --- + + @Test + void testGetInstanceReturnsSameInstance() { + var instance1 = OpenCLContext.getInstance(); + var instance2 = OpenCLContext.getInstance(); + assertSame(instance1, instance2, "getInstance() should return same instance"); + } + + @Test + void testGetInstanceNotNull() { + var instance = OpenCLContext.getInstance(); + assertNotNull(instance, "getInstance() should never return null"); + } + + @Test + void testTestResetCreatesNewInstance() { + var instance1 = OpenCLContext.getInstance(); + OpenCLContext.testReset(); + var instance2 = OpenCLContext.getInstance(); + assertNotSame(instance1, instance2, "testReset() should create new instance"); + } + + // --- Reference Counting Tests --- + + @Test + void testInitialRefCountIsZero() { + var ctx = OpenCLContext.getInstance(); + assertEquals(0, ctx.getRefCount(), "Initial refCount should be 0"); + } + + @Test + void testNotInitializedBeforeAcquire() { + var ctx = OpenCLContext.getInstance(); + assertFalse(ctx.isInitialized(), "Should not be initialized before acquire()"); + } + + @Test + void testReleaseWithoutAcquireHandledGracefully() { + var ctx = OpenCLContext.getInstance(); + // Should not throw, should log warning and reset to 0 + ctx.release(); + assertEquals(0, ctx.getRefCount(), "refCount should be 0 after invalid release"); + } + + @Test + void testMultipleReleasesHandledGracefully() { + var ctx = OpenCLContext.getInstance(); + // Multiple releases without acquires should not throw + ctx.release(); + ctx.release(); + ctx.release(); + assertEquals(0, ctx.getRefCount(), "refCount should stay at 0"); + } + + // --- State Validation Tests --- + + @Test + void testGetContextThrowsWhenNotInitialized() { + var ctx = OpenCLContext.getInstance(); + assertThrows(IllegalStateException.class, ctx::getContext, + "getContext() should throw when not initialized"); + } + + @Test + void testGetCommandQueueThrowsWhenNotInitialized() { + var ctx = OpenCLContext.getInstance(); + assertThrows(IllegalStateException.class, ctx::getCommandQueue, + "getCommandQueue() should throw when not initialized"); + } + + @Test + void testGetDeviceThrowsWhenNotInitialized() { + var ctx = OpenCLContext.getInstance(); + assertThrows(IllegalStateException.class, ctx::getDevice, + "getDevice() should throw when not initialized"); + } + + // --- Thread Safety Tests (Basic) --- + + @Test + void testConcurrentGetInstance() throws InterruptedException { + var instances = new OpenCLContext[10]; + var threads = new Thread[10]; + + for (int i = 0; i < 10; i++) { + final int idx = i; + threads[i] = new Thread(() -> instances[idx] = OpenCLContext.getInstance()); + } + + for (var thread : threads) { + thread.start(); + } + + for (var thread : threads) { + thread.join(); + } + + // All threads should get the same instance + var first = instances[0]; + for (var instance : instances) { + assertSame(first, instance, "All threads should get same instance"); + } + } + + // --- Out of Order Execution Support --- + + @Test + void testOutOfOrderSupportedDefaultFalse() { + var ctx = OpenCLContext.getInstance(); + // Before initialization, out-of-order is not supported + assertFalse(ctx.isOutOfOrderSupported(), + "Out-of-order should be false before initialization"); + } + + // --- Test Reset Edge Cases --- + + @Test + void testTestResetWhenNull() { + // First reset to null + OpenCLContext.testReset(); + // Second reset should not throw + OpenCLContext.testReset(); + // Should still work + var ctx = OpenCLContext.getInstance(); + assertNotNull(ctx); + } + + @Test + void testTestResetPreservesIsolation() { + // Get instance and check state + var ctx1 = OpenCLContext.getInstance(); + assertEquals(0, ctx1.getRefCount()); + + // Reset + OpenCLContext.testReset(); + + // New instance should have fresh state + var ctx2 = OpenCLContext.getInstance(); + assertEquals(0, ctx2.getRefCount()); + assertFalse(ctx2.isInitialized()); + } +} diff --git a/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLKernelTest.java b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLKernelTest.java new file mode 100644 index 0000000..f4a880b --- /dev/null +++ b/resource/src/test/java/com/hellblazer/luciferase/resource/compute/opencl/OpenCLKernelTest.java @@ -0,0 +1,352 @@ +package com.hellblazer.luciferase.resource.compute.opencl; + +import com.hellblazer.luciferase.resource.compute.ComputeKernel; +import com.hellblazer.luciferase.resource.compute.GPUBackend; +import org.junit.jupiter.api.BeforeAll; +import org.junit.jupiter.api.Test; +import org.junit.jupiter.api.condition.DisabledIfEnvironmentVariable; +import org.lwjgl.PointerBuffer; +import org.lwjgl.opencl.CL10; +import org.lwjgl.system.MemoryStack; + +import static org.junit.jupiter.api.Assertions.*; + +/** + * Unit tests for OpenCLKernel. + * Tests require OpenCL availability - skipped if not available and in CI environments. + */ +@DisabledIfEnvironmentVariable(named = "CI", matches = "true", disabledReason = "OpenCL not available in CI") +class OpenCLKernelTest { + + private static boolean openCLAvailable; + + // Simple vector addition kernel for testing + private static final String VECTOR_ADD_KERNEL = """ + __kernel void vectorAdd(__global const float* a, + __global const float* b, + __global float* result, + const int size) { + int gid = get_global_id(0); + if (gid < size) { + result[gid] = a[gid] + b[gid]; + } + } + """; + + // Simple scale kernel for testing scalar arguments + private static final String SCALE_KERNEL = """ + __kernel void scale(__global float* data, + const float factor, + const int size) { + int gid = get_global_id(0); + if (gid < size) { + data[gid] = data[gid] * factor; + } + } + """; + + @BeforeAll + static void checkOpenCL() { + try (var stack = MemoryStack.stackPush()) { + var numPlatforms = stack.mallocInt(1); + var errcode = CL10.clGetPlatformIDs((PointerBuffer) null, numPlatforms); + + if (errcode == CL10.CL_SUCCESS && numPlatforms.get(0) > 0) { + var platformBuffer = stack.mallocPointer(1); + CL10.clGetPlatformIDs(platformBuffer, (int[]) null); + var platform = platformBuffer.get(0); + + var numDevices = stack.mallocInt(1); + var result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_GPU, null, numDevices); + if (result != CL10.CL_SUCCESS) { + result = CL10.clGetDeviceIDs(platform, CL10.CL_DEVICE_TYPE_CPU, null, numDevices); + } + + openCLAvailable = result == CL10.CL_SUCCESS && numDevices.get(0) > 0; + } + System.out.println("OpenCL available: " + openCLAvailable); + } catch (Exception e) { + openCLAvailable = false; + System.out.println("OpenCL check failed: " + e.getMessage()); + } + } + + // --- Creation Tests --- + + @Test + void testCreateKernel() { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("testKernel")) { + assertNotNull(kernel); + assertEquals("testKernel", kernel.getName()); + assertFalse(kernel.isCompiled()); + assertTrue(kernel.isValid()); + assertEquals(GPUBackend.OPENCL, kernel.getBackend()); + } + } + + // --- Compilation Tests --- + + @Test + void testCompileKernel() { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("vectorAdd")) { + assertFalse(kernel.isCompiled()); + + assertDoesNotThrow(() -> kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd")); + + assertTrue(kernel.isCompiled()); + } + } + + @Test + void testDoubleCompileThrows() { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("vectorAdd")) { + assertDoesNotThrow(() -> kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd")); + + assertThrows(ComputeKernel.KernelCompilationException.class, + () -> kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd")); + } + } + + @Test + void testCompileInvalidSourceThrows() { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("invalid")) { + var invalidSource = "__kernel void foo() { invalid_function(); }"; + + assertThrows(ComputeKernel.KernelCompilationException.class, + () -> kernel.compile(invalidSource, "foo")); + } + } + + @Test + void testCompileWrongEntryPointThrows() { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("vectorAdd")) { + // Compile with wrong entry point name + assertThrows(ComputeKernel.KernelCompilationException.class, + () -> kernel.compile(VECTOR_ADD_KERNEL, "wrongName")); + } + } + + // --- Argument Setting Tests --- + + @Test + void testSetBufferArg() throws Exception { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("vectorAdd"); + var buffer = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_ONLY)) { + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + assertDoesNotThrow(() -> + kernel.setBufferArg(0, buffer, ComputeKernel.BufferAccess.READ)); + } + } + + @Test + void testSetScalarArgs() throws Exception { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("scale")) { + kernel.compile(SCALE_KERNEL, "scale"); + + assertDoesNotThrow(() -> kernel.setFloatArg(1, 2.5f)); + assertDoesNotThrow(() -> kernel.setIntArg(2, 64)); + } + } + + @Test + void testSetArgBeforeCompileThrows() { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("test"); + var buffer = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_ONLY)) { + + assertThrows(IllegalStateException.class, + () -> kernel.setBufferArg(0, buffer, ComputeKernel.BufferAccess.READ)); + assertThrows(IllegalStateException.class, + () -> kernel.setFloatArg(0, 1.0f)); + assertThrows(IllegalStateException.class, + () -> kernel.setIntArg(0, 1)); + } + } + + // --- Execution Tests --- + + @Test + void testExecuteVectorAdd() throws Exception { + if (!openCLAvailable) return; + + int size = 64; + var a = new float[size]; + var b = new float[size]; + var result = new float[size]; + + // Initialize input data + for (int i = 0; i < size; i++) { + a[i] = i; + b[i] = i * 2; + } + + try (var kernel = OpenCLKernel.create("vectorAdd"); + var bufferA = OpenCLBuffer.createWithData(a, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferB = OpenCLBuffer.createWithData(b, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(size, OpenCLBuffer.BufferAccess.WRITE_ONLY)) { + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + kernel.setBufferArg(0, bufferA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, size); + + kernel.execute(size); + kernel.finish(); + + bufferResult.download(result); + + // Verify results + for (int i = 0; i < size; i++) { + assertEquals(a[i] + b[i], result[i], 0.0001f, + "Mismatch at index " + i); + } + } + } + + @Test + void testExecuteScale() throws Exception { + if (!openCLAvailable) return; + + int size = 64; + var data = new float[size]; + float scale = 2.5f; + + // Initialize data + for (int i = 0; i < size; i++) { + data[i] = i; + } + + try (var kernel = OpenCLKernel.create("scale"); + var buffer = OpenCLBuffer.createWithData(data, OpenCLBuffer.BufferAccess.READ_WRITE)) { + + kernel.compile(SCALE_KERNEL, "scale"); + + kernel.setBufferArg(0, buffer, ComputeKernel.BufferAccess.READ_WRITE); + kernel.setFloatArg(1, scale); + kernel.setIntArg(2, size); + + kernel.execute(size); + kernel.finish(); + + var result = new float[size]; + buffer.download(result); + + // Verify results + for (int i = 0; i < size; i++) { + assertEquals(i * scale, result[i], 0.0001f, + "Mismatch at index " + i); + } + } + } + + @Test + void testExecuteBeforeCompileThrows() { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("test")) { + assertThrows(ComputeKernel.KernelExecutionException.class, + () -> kernel.execute(64)); + } + } + + @Test + void testExecute2D() throws Exception { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("vectorAdd"); + var bufferA = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferB = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.WRITE_ONLY)) { + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + kernel.setBufferArg(0, bufferA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, 64); + + // Execute with 2D work size (8x8 = 64) + assertDoesNotThrow(() -> kernel.execute(8, 8)); + kernel.finish(); + } + } + + @Test + void testExecute3D() throws Exception { + if (!openCLAvailable) return; + + try (var kernel = OpenCLKernel.create("vectorAdd"); + var bufferA = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferB = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.READ_ONLY); + var bufferResult = OpenCLBuffer.create(64, OpenCLBuffer.BufferAccess.WRITE_ONLY)) { + + kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd"); + + kernel.setBufferArg(0, bufferA, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(1, bufferB, ComputeKernel.BufferAccess.READ); + kernel.setBufferArg(2, bufferResult, ComputeKernel.BufferAccess.WRITE); + kernel.setIntArg(3, 64); + + // Execute with 3D work size (4x4x4 = 64) + assertDoesNotThrow(() -> kernel.execute(4, 4, 4)); + kernel.finish(); + } + } + + // --- Lifecycle Tests --- + + @Test + void testCloseMarksInvalid() { + if (!openCLAvailable) return; + + var kernel = OpenCLKernel.create("test"); + assertTrue(kernel.isValid()); + + kernel.close(); + assertFalse(kernel.isValid()); + } + + @Test + void testDoubleCloseIsSafe() { + if (!openCLAvailable) return; + + var kernel = OpenCLKernel.create("test"); + kernel.close(); + kernel.close(); // Should not throw + assertFalse(kernel.isValid()); + } + + @Test + void testOperationsAfterCloseThrow() throws Exception { + if (!openCLAvailable) return; + + var kernel = OpenCLKernel.create("test"); + kernel.close(); + + assertThrows(IllegalStateException.class, + () -> kernel.compile(VECTOR_ADD_KERNEL, "vectorAdd")); + assertThrows(IllegalStateException.class, + () -> kernel.execute(64)); + assertThrows(IllegalStateException.class, + kernel::finish); + } +} diff --git a/resource/src/test/resources/kernels/test_kernel.cl b/resource/src/test/resources/kernels/test_kernel.cl new file mode 100644 index 0000000..b5fdb5d --- /dev/null +++ b/resource/src/test/resources/kernels/test_kernel.cl @@ -0,0 +1,9 @@ +/** + * Test kernel for KernelLoader unit tests. + */ +__kernel void testKernel(__global float* data, const int size) { + int gid = get_global_id(0); + if (gid < size) { + data[gid] = data[gid] * 2.0f; + } +}