Perform delayed freeing of the geometry BVHs similar to OptiX.
Previously BLAS memory was allocated in the device class as part of
device_update, but released in the BVHHIPRT destructor which gets called
when deleting geometry outside of device_update.
To avoid the GPU accessing unmapped memory, do a delayed free of this
memory in the device class as part of either device_update or device
destruction. This ensures it is in sync with other device memory changes.
Fix#148276Fix#139013Fix#138043Fix#140763
Pull Request: https://projects.blender.org/blender/blender/pulls/147247
The Cycles cpu/device.cpp `device_cpu_capabilities()` function used to
fill out a string of supported CPU capabilities separated with spaces,
with some trailing-space cleaning logic at the end of the function.
However, if no check succeeded, and especially after commit 2bf6d0fd71
which left only one check and removed the need for removing trailing
spaces, the check would run against an empty string, resulting in an
unsigned 0 - 1 operation which would then cause an out of bound access
catched by ASan.
Fixed by removing the now superflous trailing space cleaning logic and
simplifying to a direct return.
Pull Request: https://projects.blender.org/blender/blender/pulls/148227
Follow on from PR #141891. The `MTLAccelerationStructureUsagePreferFastIntersection` flag didn't exist until Xcode 26.0, so we ensure that it is defined for forward-compatibility. The runtime `if (@available(macos 26.0, *))` checks still remain.
Pull Request: https://projects.blender.org/blender/blender/pulls/147561
In some hardware configurations, it is possible that DPC++ or
Intel Drivers wrongfully report all devices twice. It is already
being worked on internally, and the fixes will be available in
the future - but for now, we need a workaround for this problem
in Blender as well, to ensure that our end-users are not impacted.
Pull Request: https://projects.blender.org/blender/blender/pulls/147731
This new version of the graphics compiler improves performance
for the majority of supported Intel devices and adds support
for upcoming Intel hardware. Such an upgrade also requires
an increase in the minimal supported driver version on Windows,
which is why these changes are combined together with
the ocloc upgrade.
Previously set minimal version 101.6557 was increased to 101.8132.
Pull Request: https://projects.blender.org/blender/blender/pulls/147460
This PR adds Vulkan/oneAPI graphics interop to Cycles. Just like for
CUDA and HIP interop, persistent memory mapping is used, as there could
potentially be some overhead of continuously mapping/unmapping buffers.
Pull Request: https://projects.blender.org/blender/blender/pulls/144442
For metal version after 3.2 it's possible to log debugging messages, it
works similar to `printf()`, except for a few differences:
- `%s` is not supported,
- `double` doesn't exist, so no casting to double for `%f`,
- no `\n` needed at the end of the format string.
To see the print in the console, environment variables `MTL_LOG_LEVEL`
should be set to `MTLLogLevelDebug`, and `MTL_LOG_TO_STDERR` should be
set to `1`. See
https://developer.apple.com/documentation/metal/logging-shader-debug-messages
Right now `printf()`, `print_float()`, `print_float2()`,
`print_float3()` and `print_float4()` are supported.
Thanks to @fclem for finding this out.
Pull Request: https://projects.blender.org/blender/blender/pulls/146585
The workaround of forcing BVH building into single thread
execution on the Blender side is not needed anymore,
because the problem was properly fixed in the upstream
since Embree upgrade in Blender 4.5
This reverts commit c0f0e2ca6f.
Pull Request: https://projects.blender.org/blender/blender/pulls/146859
MTLAccelerationStructureMotionCurveGeometryDescriptor.controlPointCount should specify the per-step control point count. Although the previous initialisation wasn't manifesting as incorrect behaviour it was technically wrong.
Pull Request: https://projects.blender.org/blender/blender/pulls/146568
HIP-RT device:
- Add missing flags from the common flags query to the final compiler options
- Switch logging utility from printf to LOG_INFO_IMPORTANT
- Remove redundant compiler options already covered by common flags
HIP device:
- Add compiler command to logging
- Update C++ standard to C++17 to resolve compiler warnings
Pull Request: https://projects.blender.org/blender/blender/pulls/145284
Experiments have shown that the OptiX denoiser performs best when
operating on images that have their origin at the top-left corner,
while Blender renders with the origin at the bottom-left corner.
Simply flipping the image vertically before and after denoising is a
relatively trivial operation, so this patch introduces this as an
additional preprocessing and postprocessing step for denoising when the
OptiX denoiser is used. Additionally, this patch also removes an unused
helper function, now that OptiX 8.0 is the minimum.
Pull Request: https://projects.blender.org/blender/blender/pulls/145358
There are several Driver versions which are constructing the wrong,
semantically, version which would force Blender to decline the Intel
device for oneAPI backend usage, based on this. Unfortunately,
the upstream fix is taking a long time to be finally delivered to
the distros and end-users, so it is better if Blender will detect
this wrong version string and parse it properly, allowing these
devices to be used - as the wrong driver version string is the only
issue here, besides this the driver functionality is fine.
Pull Request: https://projects.blender.org/blender/blender/pulls/145658
This pull request removes ROCm 5 code path and adds ROCm 7 runtime to
library search list.
ROCm 5 runtime is no longer shipped with AMD drivers, and ROCm 5 compiler
is no longer compatible with newer driver versions.
It also adds ROCm 7 runtime to the list of runtime libraries to look for.
Starting later this year, ROCm 7 runtime will be bundled with the driver
installer, and all future runtime fixes and improvements will target ROCm 7.
Once ROCm 7 runtime is rolled out, ROCm 6 compiler will continue to work
with it for about a year as a transitional measure. Beyond that, compatibility
is not guaranteed.
Pull Request: https://projects.blender.org/blender/blender/pulls/145279
This re-applies pull request #140671, but with a fix for #144713 where the
non-pointer part of IntegratorStateGPU was not initialized.
This PR is a more extensive follow on from #123551 (removal of AMD and Intel
GPU support).
All supported Apple GPUs have Metal 3 and tier 2 argument buffer support.
The invariant resource properties `gpuAddress` and `gpuResourceID` can be
written directly into GPU structs once at setup time rather than once per
dispatch. More background info can be found in this article:
https://developer.apple.com/documentation/metal/improving-cpu-performance-by-using-argument-buffers?language=objc
Code changes:
- All code relating to `MTLArgumentEncoder` is removed
- `KernelParamsMetal` updates are directly written into
`id<MTLBuffer> launch_params_buffer` which is used for the "static"
dispatch arguments
- Dynamic dispatch arguments are small enough to be encoded using the
`MTLComputeCommandEncoder.setBytes` function, eliminating the need for
cycling temporary arg buffers
Fix#144713
Co-authored-by: Brecht Van Lommel <brecht@noreply.localhost>
Pull Request: https://projects.blender.org/blender/blender/pulls/145175
Currently, it was discovered that in the case of several different
Intel dGPUs being present in the system, the experimental L0 copy
optimization does not work correctly in the Intel Driver, which is
causing crashes in the driver and Blender application. So, to avoid
this situation and restore functionality on these platforms,
a workaround was added to disable this extension from being used if
such a configuration is detected. In the future, when this problem is
fully fixed in all Intel Drivers, this workaround can be removed from
the Blender source code to restore some performance that was lost on
configurations of several dGPUs because of this workaround.
Pull Request: https://projects.blender.org/blender/blender/pulls/144262
Guide the probability to scatter in or transmit through the volume.
Only applied for primary rays.
Co-authored-by: Brecht Van Lommel <brecht@blender.org>
Add new "Linear 3D Curves" option in the Curves panel in the render
properties. This renders curves as linear segments rather than smooth
curves, for faster render time at the cost of accuracy.
On NVIDIA Blackwell GPUs, this can give a 6x speedup compared to smooth
curves, due to hardware acceleration. On NVIDIA Ada there is still
a 3x speedup, and CPU and other GPU backends will also render this
faster.
A difference with smooth curves is that these have end caps, as this
was simpler to implement and they are usually helpful anyway.
In the future this functionality will also be used to properly support
the CURVE_TYPE_POLY on the new curves object.
Pull Request: https://projects.blender.org/blender/blender/pulls/139735
This fixes issues when using Embree on mutliple GPUs.
A previous workaround used separate contexts, this one now
lets us keep a single context for all GPUs.
Pull Request: https://projects.blender.org/blender/blender/pulls/143089
Currently MetalRT renders always use extended limits, which is needed to correctly render scenes where the max primitive count can exceed 2^28 or the instance count can exceed 2^24. This patch adopts Metal best practices of only enabling this flag if it is needed.
This PR is similar to #133364, but there are some notable differences:
1) The old PR made an overly optimistic assumption that all the relevant visibility bits could be squeezed into 8 bits. This new PR adopts the same approach that Optix takes of using 8 bits as a primary HW filter, and checking the full 32 bit mask inside the SW intersection handler.
~~2) I moved the scene scanning check from Scene into MetalDevice. This avoids platform specific details leaking into platform agnostic areas.~~
~~3) In live viewport mode, we always use extended limits in case we tip over the threshold.~~
_EDIT:_
2) The limits are scanned in `Scene::update_kernel_features`, and given to the device by a new `set_bvh_limits` method which returns true if the BVH and kernels need to be reloaded.
Pull Request: https://projects.blender.org/blender/blender/pulls/142401
This improves the existing scene specialisation mechanism by replacing "kernel_data.kernel_features" with a function constant. It doesn't cause any additional compilation requests, but allows the backend compiler to eliminate more dead code. An additional compiler hint is provided for dead-stripping "volume_stack_enter_exit" which results in slightly faster rendering of non-volumetric scenes.
Pull Request: https://projects.blender.org/blender/blender/pulls/142235
These changes introduce modifications to the SYCL queue creation
in OneapiDevice::create_queue. In case several DPC++ devices are
detected by Blender and exposed through it, we are now creating
a new SYCL context for each device, which allows us to prevent
execution failures due to some known issues in the DPC++ runtime
regarding multi GPU support. As this would have some small
performance impact, few percents, it is only applied to
multi GPU configurations, while the behavior for a single
GPU configuration remains the same.
Pull Request: https://projects.blender.org/blender/blender/pulls/141834
On systems with multiple Intel GPUs with a mix of recent and old
unsupported drivers (such as 101.3302), the Level-Zero stack may have
troubles initializing, leading to a crash while enumerating devices.
Luckily this condition actually leads to an exception we can catch,
as implemented here in this commit.
Pull Request: https://projects.blender.org/blender/blender/pulls/141674
