fix(exla): restore OutputBuffer default constructor for CUDA callback path

[blasphemetheus]

What I ran into

Building EXLA from source with XLA_TARGET=cuda12 fails to compile on current main (da1f4fb8):

c_src/exla/custom_calls/runtime_callback_cuda.cc:113:41: error: no matching constructor for initialization of 'exla::callback_bridge::OutputBuffer'
  113 |     exla::callback_bridge::OutputBuffer obuf;
      |                                         ^
c_src/exla/custom_calls/runtime_callback_bridge.h:42:3: note: candidate constructor not viable: requires single argument 'buf', but no arguments were provided
   42 |   OutputBuffer(const xla::ffi::AnyBuffer &buf);
      |   ^            ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
1 error generated.
make: *** [Makefile:126: cache/0.13.1/objs/custom_calls/runtime_callback_cuda.o] Error 1

I hit this on an RTX 5090 (sm_120), where building from source is the normal path. Every from-source CUDA build of current main is affected; CPU builds are fine, which is why CI stays green — runtime_callback_cuda.cc is only compiled when targeting CUDA (-DCUDA_ENABLED).

Cause

#1766 gave OutputBuffer a user-declared constructor taking xla::ffi::AnyBuffer. In C++, declaring any constructor suppresses the implicitly-generated default constructor. But the CUDA callback path still default-constructs an OutputBuffer and fills it field-by-field:

// runtime_callback_cuda.cc:113
exla::callback_bridge::OutputBuffer obuf;   // ← no longer compiles
obuf.data = reinterpret_cast<uint8_t *>(host_ptr);
obuf.size = size_bytes;

Fix

Re-declare the default constructor explicitly (OutputBuffer() = default;). The members already have safe initializers (nullptr/0), so the defaulted constructor is well-defined and the CPU path is untouched.

Verification

Both directions verified on Linux x86_64 (clang, CUDA 12.9, Erlang/OTP 27, Elixir 1.18.4) with XLA_TARGET=cuda12:

• This branch: compiles clean, libexla.so produced
• Same build with main's header (fix reverted): fails with the error above

No test accompanies this: it is a compile-time fix in a translation unit CI doesn't build, so the CUDA compile itself is the test.

🤖 Generated with Claude Code

[polvalente]

I think the fix is to actually use the new constructor

[blasphemetheus]

hmm okay tried that way, seemed to work

(generated text below)

Reworked as suggested — the CUDA path now uses the AnyBuffer constructor (which computes the size) and repoints data at the host staging buffer, since the callback writes to host memory before the H->D copy. The header change is reverted entirely, so OutputBuffer's API stays as it is on main, and the duplicated ByteWidth * element_count computation in runtime_callback_cuda.cc goes away.

Re-verified: XLA_TARGET=cuda12 compiles clean with this change on the same setup as before.

[blasphemetheus]

man I really gotta give more instruction not to just like reply on these threads thats annoying. This is just an error I ran into before and figured I'd upstream the short workaround on.

It's drifting into territory where I'm not sure I have the context for this code change so be aware of that, more just if this error wasn't encountered yet its documented and here's a plausible fix.