diff --git a/CMakeLists.txt b/CMakeLists.txt index ed32949873a..f67e61d4ce6 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -819,6 +819,9 @@ if (IS_HOST_PLATFORM) add_subdirectory(${TOOLS}/glslminifier) add_subdirectory(${TOOLS}/matc) add_subdirectory(${TOOLS}/matinfo) + if (NOT WIN32) # matedit not yet supported on Windows + add_subdirectory(${TOOLS}/matedit) + endif() add_subdirectory(${TOOLS}/mipgen) add_subdirectory(${TOOLS}/normal-blending) add_subdirectory(${TOOLS}/resgen) diff --git a/NEW_RELEASE_NOTES.md b/NEW_RELEASE_NOTES.md index 4a1a9c7fa7e..2594ce4088c 100644 --- a/NEW_RELEASE_NOTES.md +++ b/NEW_RELEASE_NOTES.md @@ -7,3 +7,5 @@ for next branch cut* header. appropriate header in [RELEASE_NOTES.md](./RELEASE_NOTES.md). ## Release notes for next branch cut + +- Add new matedit tool diff --git a/README.md b/README.md index b15f80e141e..c4dc01abefd 100644 --- a/README.md +++ b/README.md @@ -31,7 +31,7 @@ repositories { } dependencies { - implementation 'com.google.android.filament:filament-android:1.51.4' + implementation 'com.google.android.filament:filament-android:1.51.5' } ``` @@ -51,7 +51,7 @@ Here are all the libraries available in the group `com.google.android.filament`: iOS projects can use CocoaPods to install the latest release: ```shell -pod 'Filament', '~> 1.51.4' +pod 'Filament', '~> 1.51.5' ``` ### Snapshots diff --git a/RELEASE_NOTES.md b/RELEASE_NOTES.md index 7dae83f407e..e168a1f7af4 100644 --- a/RELEASE_NOTES.md +++ b/RELEASE_NOTES.md @@ -7,6 +7,9 @@ A new header is inserted each time a *tag* is created. Instead, if you are authoring a PR for the main branch, add your release note to [NEW_RELEASE_NOTES.md](./NEW_RELEASE_NOTES.md). +## v1.51.5 + + ## v1.51.4 diff --git a/android/gradle.properties b/android/gradle.properties index 8100edd3442..040c9248501 100644 --- a/android/gradle.properties +++ b/android/gradle.properties @@ -1,5 +1,5 @@ GROUP=com.google.android.filament -VERSION_NAME=1.51.4 +VERSION_NAME=1.51.5 POM_DESCRIPTION=Real-time physically based rendering engine for Android. diff --git a/filament/backend/include/backend/DriverEnums.h b/filament/backend/include/backend/DriverEnums.h index 6b0424ca063..ef1c655c57a 100644 --- a/filament/backend/include/backend/DriverEnums.h +++ b/filament/backend/include/backend/DriverEnums.h @@ -158,14 +158,16 @@ static constexpr const char* backendToString(Backend backend) { } /** - * Defines the shader language. Similar to the above backend enum, but the OpenGL backend can select - * between two shader languages: ESSL 1.0 and ESSL 3.0. + * Defines the shader language. Similar to the above backend enum, with some differences: + * - The OpenGL backend can select between two shader languages: ESSL 1.0 and ESSL 3.0. + * - The Metal backend can prefer precompiled Metal libraries, while falling back to MSL. */ enum class ShaderLanguage { ESSL1 = 0, ESSL3 = 1, SPIRV = 2, MSL = 3, + METAL_LIBRARY = 4, }; static constexpr const char* shaderLanguageToString(ShaderLanguage shaderLanguage) { @@ -178,6 +180,8 @@ static constexpr const char* shaderLanguageToString(ShaderLanguage shaderLanguag return "SPIR-V"; case ShaderLanguage::MSL: return "MSL"; + case ShaderLanguage::METAL_LIBRARY: + return "Metal precompiled library"; } } @@ -1237,9 +1241,6 @@ enum class Workaround : uint16_t { DISABLE_BLIT_INTO_TEXTURE_ARRAY, // Multiple workarounds needed for PowerVR GPUs POWER_VR_SHADER_WORKAROUNDS, - // The driver has some threads pinned, and we can't easily know on which core, it can hurt - // performance more if we end-up pinned on the same one. - DISABLE_THREAD_AFFINITY }; //! The type of technique for stereoscopic rendering diff --git a/filament/backend/include/backend/Program.h b/filament/backend/include/backend/Program.h index fe1c4a9b6e8..b5c1dd9babd 100644 --- a/filament/backend/include/backend/Program.h +++ b/filament/backend/include/backend/Program.h @@ -84,6 +84,9 @@ class Program { // null terminating character. Program& shader(ShaderStage shader, void const* data, size_t size); + // sets the language of the shader sources provided with shader() (defaults to ESSL3) + Program& shaderLanguage(ShaderLanguage shaderLanguage); + // Note: This is only needed for GLES3.0 backends, because the layout(binding=) syntax is // not permitted in glsl. The backend needs a way to associate a uniform block // to a binding point. @@ -136,6 +139,8 @@ class Program { utils::CString const& getName() const noexcept { return mName; } utils::CString& getName() noexcept { return mName; } + auto const& getShaderLanguage() const { return mShaderLanguage; } + utils::FixedCapacityVector const& getSpecializationConstants() const noexcept { return mSpecializationConstants; } @@ -155,6 +160,7 @@ class Program { UniformBlockInfo mUniformBlocks = {}; SamplerGroupInfo mSamplerGroups = {}; ShaderSource mShadersSource; + ShaderLanguage mShaderLanguage = ShaderLanguage::ESSL3; utils::CString mName; uint64_t mCacheId{}; utils::Invocable mLogger; diff --git a/filament/backend/src/Program.cpp b/filament/backend/src/Program.cpp index dc92e8c2a26..39a941485e8 100644 --- a/filament/backend/src/Program.cpp +++ b/filament/backend/src/Program.cpp @@ -21,7 +21,7 @@ namespace filament::backend { using namespace utils; // We want these in the .cpp file, so they're not inlined (not worth it) -Program::Program() noexcept { // NOLINT(modernize-use-equals-default) +Program::Program() noexcept { // NOLINT(modernize-use-equals-default) } Program::Program(Program&& rhs) noexcept = default; @@ -47,6 +47,11 @@ Program& Program::shader(ShaderStage shader, void const* data, size_t size) { return *this; } +Program& Program::shaderLanguage(ShaderLanguage shaderLanguage) { + mShaderLanguage = shaderLanguage; + return *this; +} + Program& Program::uniformBlockBindings( FixedCapacityVector> const& uniformBlockBindings) noexcept { for (auto const& item : uniformBlockBindings) { diff --git a/filament/backend/src/metal/MetalShaderCompiler.mm b/filament/backend/src/metal/MetalShaderCompiler.mm index 7741ae54489..1aa02f9c870 100644 --- a/filament/backend/src/metal/MetalShaderCompiler.mm +++ b/filament/backend/src/metal/MetalShaderCompiler.mm @@ -106,26 +106,41 @@ bool isReady() const noexcept { continue; } - assert_invariant(source[source.size() - 1] == '\0'); - - // the shader string is null terminated and the length includes the null character - NSString* objcSource = [[NSString alloc] initWithBytes:source.data() - length:source.size() - 1 - encoding:NSUTF8StringEncoding]; - - // By default, Metal uses the most recent language version. - MTLCompileOptions* options = [MTLCompileOptions new]; - - // Disable Fast Math optimizations. - // This ensures that operations adhere to IEEE standards for floating-point arithmetic, - // which is crucial for half precision floats in scenarios where fast math optimizations - // lead to inaccuracies, such as in handling special values like NaN or Infinity. - options.fastMathEnabled = NO; - NSError* error = nil; - id library = [device newLibraryWithSource:objcSource - options:options - error:&error]; + id library = nil; + switch (program.getShaderLanguage()) { + case ShaderLanguage::MSL: { + // By default, Metal uses the most recent language version. + MTLCompileOptions* options = [MTLCompileOptions new]; + + // Disable Fast Math optimizations. + // This ensures that operations adhere to IEEE standards for floating-point + // arithmetic, which is crucial for half precision floats in scenarios where fast + // math optimizations lead to inaccuracies, such as in handling special values like + // NaN or Infinity. + options.fastMathEnabled = NO; + + assert_invariant(source[source.size() - 1] == '\0'); + // the shader string is null terminated and the length includes the null character + NSString* objcSource = [[NSString alloc] initWithBytes:source.data() + length:source.size() - 1 + encoding:NSUTF8StringEncoding]; + library = [device newLibraryWithSource:objcSource options:options error:&error]; + break; + } + case ShaderLanguage::METAL_LIBRARY: { + dispatch_data_t data = dispatch_data_create(source.data(), source.size(), + dispatch_get_global_queue(DISPATCH_QUEUE_PRIORITY_DEFAULT, 0), + DISPATCH_DATA_DESTRUCTOR_DEFAULT); + library = [device newLibraryWithData:data error:&error]; + break; + } + case ShaderLanguage::ESSL1: + case ShaderLanguage::ESSL3: + case ShaderLanguage::SPIRV: + break; + } + if (library == nil) { NSString* errorMessage = @"unknown error"; if (error) { diff --git a/filament/backend/src/opengl/OpenGLContext.cpp b/filament/backend/src/opengl/OpenGLContext.cpp index ddf629f0f87..5355e6b1b95 100644 --- a/filament/backend/src/opengl/OpenGLContext.cpp +++ b/filament/backend/src/opengl/OpenGLContext.cpp @@ -253,6 +253,12 @@ OpenGLContext::OpenGLContext(OpenGLPlatform& platform) noexcept } OpenGLContext::~OpenGLContext() noexcept { + // note: this is called from the main thread. Can't do any GL calls. + delete mTimerQueryFactory; +} + +void OpenGLContext::terminate() noexcept { + // note: this is called from the backend thread #ifndef FILAMENT_SILENCE_NOT_SUPPORTED_BY_ES2 if (!isES2()) { for (auto& item: mSamplerMap) { @@ -262,7 +268,6 @@ OpenGLContext::~OpenGLContext() noexcept { mSamplerMap.clear(); } #endif - delete mTimerQueryFactory; } void OpenGLContext::destroyWithContext( @@ -526,8 +531,6 @@ void OpenGLContext::initBugs(Bugs* bugs, Extensions const& exts, bugs->delay_fbo_destruction = true; // PowerVR seems to have no problem with this (which is good for us) bugs->allow_read_only_ancillary_feedback_loop = true; - // PowerVR has a shader compiler thread pinned on the last core - bugs->disable_thread_affinity = true; } else if (strstr(renderer, "Apple")) { // Apple GPU } else if (strstr(renderer, "Tegra") || diff --git a/filament/backend/src/opengl/OpenGLContext.h b/filament/backend/src/opengl/OpenGLContext.h index 8f487b4781f..e6edc0c2ae9 100644 --- a/filament/backend/src/opengl/OpenGLContext.h +++ b/filament/backend/src/opengl/OpenGLContext.h @@ -92,8 +92,11 @@ class OpenGLContext final : public TimerQueryFactoryInterface { static bool queryOpenGLVersion(GLint* major, GLint* minor) noexcept; explicit OpenGLContext(OpenGLPlatform& platform) noexcept; + ~OpenGLContext() noexcept final; + void terminate() noexcept; + // TimerQueryInterface ------------------------------------------------------------------------ // note: OpenGLContext being final ensures (clang) these are not called through the vtable @@ -308,10 +311,6 @@ class OpenGLContext final : public TimerQueryFactoryInterface { // a glFinish. So we must delay the destruction until we know the GPU is finished. bool delay_fbo_destruction; - // The driver has some threads pinned, and we can't easily know on which core, it can hurt - // performance more if we end-up pinned on the same one. - bool disable_thread_affinity; - // Force feature level 0. Typically used for low end ES3 devices with significant driver // bugs or performance issues. bool force_feature_level0; @@ -552,9 +551,6 @@ class OpenGLContext final : public TimerQueryFactoryInterface { { bugs.delay_fbo_destruction, "delay_fbo_destruction", ""}, - { bugs.disable_thread_affinity, - "disable_thread_affinity", - ""}, { bugs.force_feature_level0, "force_feature_level0", ""}, diff --git a/filament/backend/src/opengl/OpenGLDriver.cpp b/filament/backend/src/opengl/OpenGLDriver.cpp index 51384e28ab0..82723bcb3e0 100644 --- a/filament/backend/src/opengl/OpenGLDriver.cpp +++ b/filament/backend/src/opengl/OpenGLDriver.cpp @@ -232,6 +232,7 @@ OpenGLDriver::OpenGLDriver(OpenGLPlatform* platform, const Platform::DriverConfi } OpenGLDriver::~OpenGLDriver() noexcept { // NOLINT(modernize-use-equals-default) + // this is called from the main thread. Can't call GL. } Dispatcher OpenGLDriver::getDispatcher() const noexcept { @@ -264,6 +265,8 @@ void OpenGLDriver::terminate() { assert_invariant(mGpuCommandCompleteOps.empty()); #endif + mContext.terminate(); + mPlatform.terminate(); } @@ -2056,8 +2059,6 @@ bool OpenGLDriver::isWorkaroundNeeded(Workaround workaround) { return mContext.bugs.disable_blit_into_texture_array; case Workaround::POWER_VR_SHADER_WORKAROUNDS: return mContext.bugs.powervr_shader_workarounds; - case Workaround::DISABLE_THREAD_AFFINITY: - return mContext.bugs.disable_thread_affinity; default: return false; } diff --git a/filament/backend/src/vulkan/VulkanCommands.cpp b/filament/backend/src/vulkan/VulkanCommands.cpp index 4f9957753d3..95c31657e13 100644 --- a/filament/backend/src/vulkan/VulkanCommands.cpp +++ b/filament/backend/src/vulkan/VulkanCommands.cpp @@ -47,7 +47,8 @@ VulkanCmdFence::VulkanCmdFence(VkFence ifence) VulkanCommandBuffer::VulkanCommandBuffer(VulkanResourceAllocator* allocator, VkDevice device, VkCommandPool pool) - : mResourceManager(allocator) { + : mResourceManager(allocator), + mPipeline(VK_NULL_HANDLE) { // Create the low-level command buffer. const VkCommandBufferAllocateInfo allocateInfo{ .sType = VK_STRUCTURE_TYPE_COMMAND_BUFFER_ALLOCATE_INFO, diff --git a/filament/backend/src/vulkan/VulkanCommands.h b/filament/backend/src/vulkan/VulkanCommands.h index c3e6697b9d4..e3c7a92f190 100644 --- a/filament/backend/src/vulkan/VulkanCommands.h +++ b/filament/backend/src/vulkan/VulkanCommands.h @@ -89,6 +89,15 @@ struct VulkanCommandBuffer { inline void reset() { fence.reset(); mResourceManager.clear(); + mPipeline = VK_NULL_HANDLE; + } + + inline void setPipeline(VkPipeline pipeline) { + mPipeline = pipeline; + } + + inline VkPipeline pipeline() const { + return mPipeline; } inline VkCommandBuffer buffer() const { @@ -103,6 +112,7 @@ struct VulkanCommandBuffer { private: VulkanAcquireOnlyResourceManager mResourceManager; VkCommandBuffer mBuffer; + VkPipeline mPipeline; }; // Allows classes to be notified after a new command buffer has been activated. diff --git a/filament/backend/src/vulkan/VulkanDriver.cpp b/filament/backend/src/vulkan/VulkanDriver.cpp index 5b8d6de9f88..e1cd3d7cf74 100644 --- a/filament/backend/src/vulkan/VulkanDriver.cpp +++ b/filament/backend/src/vulkan/VulkanDriver.cpp @@ -14,7 +14,7 @@ * limitations under the License. */ -#include "vulkan/VulkanDriver.h" +#include "VulkanDriver.h" #include "CommandStreamDispatcher.h" #include "DataReshaper.h" @@ -102,6 +102,15 @@ VulkanTexture* createEmptyTexture(VkDevice device, VkPhysicalDevice physicalDevi return emptyTexture; } +VulkanBufferObject* createEmptyBufferObject(VmaAllocator allocator, VulkanStagePool& stagePool, + VulkanCommands* commands) { + VulkanBufferObject* obj = + new VulkanBufferObject(allocator, stagePool, 1, BufferObjectBinding::UNIFORM); + uint8_t byte = 0; + obj->buffer.loadFromCpu(commands->get().buffer(), &byte, 0, 1); + return obj; +} + #if FVK_ENABLED(FVK_DEBUG_VALIDATION) VKAPI_ATTR VkBool32 VKAPI_CALL debugReportCallback(VkDebugReportFlagsEXT flags, VkDebugReportObjectTypeEXT objectType, uint64_t object, size_t location, @@ -214,12 +223,14 @@ VulkanDriver::VulkanDriver(VulkanPlatform* platform, VulkanContext const& contex mThreadSafeResourceManager(&mResourceAllocator), mCommands(mPlatform->getDevice(), mPlatform->getGraphicsQueue(), mPlatform->getGraphicsQueueFamilyIndex(), &mContext, &mResourceAllocator), - mPipelineCache(&mResourceAllocator), + mPipelineLayoutCache(mPlatform->getDevice(), &mResourceAllocator), + mPipelineCache(mPlatform->getDevice(), mAllocator), mStagePool(mAllocator, &mCommands), mFramebufferCache(mPlatform->getDevice()), mSamplerCache(mPlatform->getDevice()), mBlitter(mPlatform->getPhysicalDevice(), &mCommands), mReadPixels(mPlatform->getDevice()), + mDescriptorSetManager(mPlatform->getDevice(), &mResourceAllocator), mIsSRGBSwapChainSupported(mPlatform->getCustomization().isSRGBSwapChainSupported) { #if FVK_ENABLED(FVK_DEBUG_DEBUG_UTILS) @@ -243,13 +254,17 @@ VulkanDriver::VulkanDriver(VulkanPlatform* platform, VulkanContext const& contex #endif mTimestamps = std::make_unique(mPlatform->getDevice()); - mCommands.setObserver(&mPipelineCache); - mPipelineCache.setDevice(mPlatform->getDevice(), mAllocator); mEmptyTexture = createEmptyTexture(mPlatform->getDevice(), mPlatform->getPhysicalDevice(), mContext, mAllocator, &mCommands, mStagePool); + mEmptyBufferObject = createEmptyBufferObject(mAllocator, mStagePool, &mCommands); - mPipelineCache.setDummyTexture(mEmptyTexture->getPrimaryImageView()); + mDescriptorSetManager.setPlaceHolders(mSamplerCache.getSampler({}), mEmptyTexture, + mEmptyBufferObject); + + mGetPipelineFunction = [this](VulkanDescriptorSetLayoutList const& layouts) { + return mPipelineLayoutCache.getLayout(layouts); + }; } VulkanDriver::~VulkanDriver() noexcept = default; @@ -310,13 +325,14 @@ ShaderModel VulkanDriver::getShaderModel() const noexcept { } void VulkanDriver::terminate() { + delete mEmptyBufferObject; + delete mEmptyTexture; + // Command buffers should come first since it might have commands depending on resources that // are about to be destroyed. mCommands.terminate(); - delete mEmptyTexture; mResourceManager.clear(); - mTimestamps.reset(); mBlitter.terminate(); @@ -329,6 +345,12 @@ void VulkanDriver::terminate() { mPipelineCache.terminate(); mFramebufferCache.reset(); mSamplerCache.terminate(); + mDescriptorSetManager.terminate(); + mPipelineLayoutCache.terminate(); + +#if FVK_ENABLED(FVK_DEBUG_RESOURCE_LEAK) + mResourceAllocator.print(); +#endif vmaDestroyAllocator(mAllocator); @@ -360,6 +382,8 @@ void VulkanDriver::collectGarbage() { mCommands.gc(); mStagePool.gc(); mFramebufferCache.gc(); + mPipelineCache.gc(); + mDescriptorSetManager.gc(); #if FVK_ENABLED(FVK_DEBUG_RESOURCE_LEAK) mResourceAllocator.print(); @@ -489,9 +513,6 @@ void VulkanDriver::destroyBufferObject(Handle boh) { return; } auto bufferObject = mResourceAllocator.handle_cast(boh); - if (bufferObject->bindingType == BufferObjectBinding::UNIFORM) { - mPipelineCache.unbindUniformBuffer(bufferObject->buffer.getGpuBuffer()); - } mResourceManager.release(bufferObject); } @@ -542,6 +563,7 @@ void VulkanDriver::destroyProgram(Handle ph) { return; } auto vkprogram = mResourceAllocator.handle_cast(ph); + mDescriptorSetManager.clearProgram(vkprogram); mResourceManager.release(vkprogram); } @@ -1429,12 +1451,7 @@ void VulkanDriver::endRenderPass(int) { 0, 1, &barrier, 0, nullptr, 0, nullptr); } - if (mCurrentRenderPass.currentSubpass > 0) { - for (uint32_t i = 0; i < VulkanPipelineCache::INPUT_ATTACHMENT_COUNT; i++) { - mPipelineCache.bindInputAttachment(i, {}); - } - mCurrentRenderPass.currentSubpass = 0; - } + mDescriptorSetManager.clearState(); mCurrentRenderPass.renderTarget = nullptr; mCurrentRenderPass.renderPass = VK_NULL_HANDLE; FVK_SYSTRACE_END(); @@ -1453,15 +1470,9 @@ void VulkanDriver::nextSubpass(int) { mPipelineCache.bindRenderPass(mCurrentRenderPass.renderPass, ++mCurrentRenderPass.currentSubpass); - for (uint32_t i = 0; i < VulkanPipelineCache::INPUT_ATTACHMENT_COUNT; i++) { - if ((1 << i) & mCurrentRenderPass.params.subpassMask) { - VulkanAttachment subpassInput = renderTarget->getColor(i); - VkDescriptorImageInfo info = { - .imageView = subpassInput.getImageView(VK_IMAGE_ASPECT_COLOR_BIT), - .imageLayout = ImgUtil::getVkLayout(subpassInput.getLayout()), - }; - mPipelineCache.bindInputAttachment(i, info); - } + if (mCurrentRenderPass.params.subpassMask & 0x1) { + VulkanAttachment subpassInput = renderTarget->getColor(0); + mDescriptorSetManager.updateInputAttachment({}, subpassInput); } } @@ -1501,25 +1512,24 @@ void VulkanDriver::commit(Handle sch) { void VulkanDriver::bindUniformBuffer(uint32_t index, Handle boh) { auto* bo = mResourceAllocator.handle_cast(boh); - const VkDeviceSize offset = 0; - const VkDeviceSize size = VK_WHOLE_SIZE; - mPipelineCache.bindUniformBufferObject((uint32_t) index, bo, offset, size); + VkDeviceSize const offset = 0; + VkDeviceSize const size = VK_WHOLE_SIZE; + mDescriptorSetManager.updateBuffer({}, (uint32_t) index, bo, offset, size); } void VulkanDriver::bindBufferRange(BufferObjectBinding bindingType, uint32_t index, Handle boh, uint32_t offset, uint32_t size) { - assert_invariant(bindingType == BufferObjectBinding::SHADER_STORAGE || - bindingType == BufferObjectBinding::UNIFORM); + assert_invariant(bindingType == BufferObjectBinding::UNIFORM); // TODO: implement BufferObjectBinding::SHADER_STORAGE case auto* bo = mResourceAllocator.handle_cast(boh); - mPipelineCache.bindUniformBufferObject((uint32_t)index, bo, offset, size); + mDescriptorSetManager.updateBuffer({}, (uint32_t) index, bo, offset, size); } void VulkanDriver::unbindBuffer(BufferObjectBinding bindingType, uint32_t index) { - // TODO: implement unbindBuffer() + mDescriptorSetManager.clearBuffer((uint32_t) index); } void VulkanDriver::bindSamplers(uint32_t index, Handle sbh) { @@ -1767,21 +1777,14 @@ void VulkanDriver::bindPipeline(PipelineState pipelineState) { // where "SamplerBinding" is the integer in the GLSL, and SamplerGroupBinding is the abstract // Filament concept used to form groups of samplers. - VkDescriptorImageInfo samplerInfo[VulkanPipelineCache::SAMPLER_BINDING_COUNT] = {}; - VulkanTexture* samplerTextures[VulkanPipelineCache::SAMPLER_BINDING_COUNT] = {nullptr}; - auto const& bindingToSamplerIndex = program->getBindingToSamplerIndex(); - UsageFlags usage = program->getUsage(); - #if FVK_ENABLED_DEBUG_SAMPLER_NAME auto const& bindingToName = program->getBindingToName(); #endif for (auto binding: program->getBindings()) { uint16_t const indexPair = bindingToSamplerIndex[binding]; - if (indexPair == 0xffff) { - usage = VulkanPipelineCache::disableUsageFlags(binding, usage); continue; } @@ -1790,18 +1793,14 @@ void VulkanDriver::bindPipeline(PipelineState pipelineState) { VulkanSamplerGroup* vksb = mSamplerBindings[samplerGroupInd]; if (!vksb) { - usage = VulkanPipelineCache::disableUsageFlags(binding, usage); continue; } SamplerDescriptor const* boundSampler = ((SamplerDescriptor*) vksb->sb->data()) + samplerInd; if (UTILS_UNLIKELY(!boundSampler->t)) { - usage = VulkanPipelineCache::disableUsageFlags(binding, usage); continue; } - VulkanTexture* texture = mResourceAllocator.handle_cast(boundSampler->t); - VkImageViewType const expectedType = texture->getViewType(); // TODO: can this uninitialized check be checked in a higher layer? // This fallback path is very flaky because the dummy texture might not have @@ -1815,43 +1814,20 @@ void VulkanDriver::bindPipeline(PipelineState pipelineState) { texture = mEmptyTexture; } - SamplerParams const& samplerParams = boundSampler->s; - VkSampler const vksampler = mSamplerCache.getSampler(samplerParams); - #if FVK_ENABLED_DEBUG_SAMPLER_NAME VulkanDriver::DebugUtils::setName(VK_OBJECT_TYPE_SAMPLER, reinterpret_cast(vksampler), bindingToName[binding].c_str()); + VulkanDriver::DebugUtils::setName(VK_OBJECT_TYPE_SAMPLER, + reinterpret_cast(samplerInfo.sampler), bindingToName[binding].c_str()); #endif - VkImageView imageView = VK_NULL_HANDLE; - VkImageSubresourceRange const range = texture->getPrimaryViewRange(); - if (any(texture->usage & TextureUsage::DEPTH_ATTACHMENT) && - expectedType == VK_IMAGE_VIEW_TYPE_2D) { - // If the sampler is part of a mipmapped depth texture, where one of the level *can* be - // an attachment, then the sampler for this texture has the same view properties as a - // view for an attachment. Therefore, we can use getAttachmentView to get a - // corresponding VkImageView. - imageView = texture->getAttachmentView(range); - } else { - imageView = texture->getViewForType(range, expectedType); - } + VkSampler const vksampler = mSamplerCache.getSampler(boundSampler->s); - samplerInfo[binding] = { - .sampler = vksampler, - .imageView = imageView, - .imageLayout = ImgUtil::getVkLayout(texture->getPrimaryImageLayout()) - }; - samplerTextures[binding] = texture; + mDescriptorSetManager.updateSampler({}, binding, texture, vksampler); } - mPipelineCache.bindSamplers(samplerInfo, samplerTextures, usage); - - // Bind a new pipeline if the pipeline state changed. - // If allocation failed, skip the draw call and bail. We do not emit an error since the - // validation layer will already do so. - if (!mPipelineCache.bindPipeline(commands)) { - return; - } + mPipelineCache.bindLayout(mDescriptorSetManager.bind(commands, program, mGetPipelineFunction)); + mPipelineCache.bindPipeline(commands); FVK_SYSTRACE_END(); } @@ -1891,12 +1867,8 @@ void VulkanDriver::draw2(uint32_t indexOffset, uint32_t indexCount, uint32_t ins VulkanCommandBuffer& commands = mCommands.get(); VkCommandBuffer cmdbuffer = commands.buffer(); - // Bind new descriptor sets if they need to change. - // If descriptor set allocation failed, skip the draw call and bail. No need to emit an error - // message since the validation layers already do so. - if (!mPipelineCache.bindDescriptors(cmdbuffer)) { - return; - } + // Bind "dynamic" UBOs if they need to change. + mDescriptorSetManager.dynamicBind(&commands, {}); // Finally, make the actual draw call. TODO: support subranges const uint32_t firstIndex = indexOffset; diff --git a/filament/backend/src/vulkan/VulkanDriver.h b/filament/backend/src/vulkan/VulkanDriver.h index 98ba9efc721..fca5c45c5be 100644 --- a/filament/backend/src/vulkan/VulkanDriver.h +++ b/filament/backend/src/vulkan/VulkanDriver.h @@ -28,6 +28,8 @@ #include "VulkanSamplerCache.h" #include "VulkanStagePool.h" #include "VulkanUtility.h" +#include "caching/VulkanDescriptorSetManager.h" +#include "caching/VulkanPipelineLayoutCache.h" #include "DriverBase.h" #include "private/backend/Driver.h" @@ -70,6 +72,8 @@ class VulkanDriver final : public DriverBase { #endif // FVK_ENABLED(FVK_DEBUG_DEBUG_UTILS) private: + static constexpr uint8_t MAX_SAMPLER_BINDING_COUNT = Program::SAMPLER_BINDING_COUNT; + void debugCommandBegin(CommandStream* cmds, bool synchronous, const char* methodName) noexcept override; @@ -106,7 +110,9 @@ class VulkanDriver final : public DriverBase { VulkanPlatform* mPlatform = nullptr; std::unique_ptr mTimestamps; + // Placeholder resources VulkanTexture* mEmptyTexture; + VulkanBufferObject* mEmptyBufferObject; VulkanSwapChain* mCurrentSwapChain = nullptr; VulkanRenderTarget* mDefaultRenderTarget = nullptr; @@ -123,13 +129,17 @@ class VulkanDriver final : public DriverBase { VulkanThreadSafeResourceManager mThreadSafeResourceManager; VulkanCommands mCommands; + VulkanPipelineLayoutCache mPipelineLayoutCache; VulkanPipelineCache mPipelineCache; VulkanStagePool mStagePool; VulkanFboCache mFramebufferCache; VulkanSamplerCache mSamplerCache; VulkanBlitter mBlitter; - VulkanSamplerGroup* mSamplerBindings[VulkanPipelineCache::SAMPLER_BINDING_COUNT] = {}; + VulkanSamplerGroup* mSamplerBindings[MAX_SAMPLER_BINDING_COUNT] = {}; VulkanReadPixels mReadPixels; + VulkanDescriptorSetManager mDescriptorSetManager; + + VulkanDescriptorSetManager::GetPipelineLayoutFunction mGetPipelineFunction; bool const mIsSRGBSwapChainSupported; }; diff --git a/filament/backend/src/vulkan/VulkanHandles.cpp b/filament/backend/src/vulkan/VulkanHandles.cpp index 41f02212c88..9d615ce5151 100644 --- a/filament/backend/src/vulkan/VulkanHandles.cpp +++ b/filament/backend/src/vulkan/VulkanHandles.cpp @@ -120,8 +120,30 @@ void addDescriptors(Bitmask mask, } } +inline VkDescriptorSetLayout createDescriptorSetLayout(VkDevice device, + VkDescriptorSetLayoutCreateInfo const& info) { + VkDescriptorSetLayout layout; + vkCreateDescriptorSetLayout(device, &info, VKALLOC, &layout); + return layout; +} + } // anonymous namespace + +VulkanDescriptorSetLayout::VulkanDescriptorSetLayout(VkDevice device, VkDescriptorSetLayoutCreateInfo const& info, + Bitmask const& bitmask) + : VulkanResource(VulkanResourceType::DESCRIPTOR_SET_LAYOUT), + mDevice(device), + vklayout(createDescriptorSetLayout(device, info)), + bitmask(bitmask), + bindings(getBindings(bitmask)), + count(Count::fromLayoutBitmask(bitmask)) { +} + +VulkanDescriptorSetLayout::~VulkanDescriptorSetLayout() { + vkDestroyDescriptorSetLayout(mDevice, vklayout, VKALLOC); +} + VulkanProgram::VulkanProgram(VkDevice device, Program const& builder) noexcept : HwProgram(builder.getName()), VulkanResource(VulkanResourceType::PROGRAM), diff --git a/filament/backend/src/vulkan/VulkanHandles.h b/filament/backend/src/vulkan/VulkanHandles.h index e41112c125d..8abdccfd968 100644 --- a/filament/backend/src/vulkan/VulkanHandles.h +++ b/filament/backend/src/vulkan/VulkanHandles.h @@ -99,13 +99,12 @@ struct VulkanDescriptorSetLayout : public VulkanResource { static_assert(sizeof(Bitmask) % 8 == 0); - explicit VulkanDescriptorSetLayout(VkDescriptorSetLayout layout, Bitmask const& bitmask) - : VulkanResource(VulkanResourceType::DESCRIPTOR_SET_LAYOUT), - vklayout(layout), - bitmask(bitmask), - bindings(getBindings(bitmask)), - count(Count::fromLayoutBitmask(bitmask)) {} + explicit VulkanDescriptorSetLayout(VkDevice device, VkDescriptorSetLayoutCreateInfo const& info, + Bitmask const& bitmask); + ~VulkanDescriptorSetLayout(); + + VkDevice const mDevice; VkDescriptorSetLayout const vklayout; Bitmask const bitmask; diff --git a/filament/backend/src/vulkan/VulkanPipelineCache.cpp b/filament/backend/src/vulkan/VulkanPipelineCache.cpp index 581d6e9beef..7cfc74162e0 100644 --- a/filament/backend/src/vulkan/VulkanPipelineCache.cpp +++ b/filament/backend/src/vulkan/VulkanPipelineCache.cpp @@ -14,8 +14,9 @@ * limitations under the License. */ -#include "vulkan/VulkanMemory.h" -#include "vulkan/VulkanPipelineCache.h" +#include "VulkanPipelineCache.h" +#include "VulkanMemory.h" +#include "caching/VulkanDescriptorSetManager.h" #include #include @@ -34,46 +35,9 @@ using namespace bluevk; namespace filament::backend { -static VkShaderStageFlags getShaderStageFlags(UsageFlags key, uint16_t binding) { - // NOTE: if you modify this function, you also need to modify getUsageFlags. - assert_invariant(binding < MAX_SAMPLER_COUNT); - VkShaderStageFlags flags = 0; - if (key.test(binding)) { - flags |= VK_SHADER_STAGE_VERTEX_BIT; - } - if (key.test(MAX_SAMPLER_COUNT + binding)) { - flags |= VK_SHADER_STAGE_FRAGMENT_BIT; - } - return flags; -} - -UsageFlags VulkanPipelineCache::disableUsageFlags(uint16_t binding, UsageFlags src) { - src.unset(binding); - src.unset(MAX_SAMPLER_COUNT + binding); - return src; -} - -VulkanPipelineCache::VulkanPipelineCache(VulkanResourceAllocator* allocator) - : mResourceAllocator(allocator), - mPipelineBoundResources(allocator) { - mDummyBufferWriteInfo.sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET; - mDummyBufferWriteInfo.pNext = nullptr; - mDummyBufferWriteInfo.dstArrayElement = 0; - mDummyBufferWriteInfo.descriptorCount = 1; - mDummyBufferWriteInfo.descriptorType = VK_DESCRIPTOR_TYPE_UNIFORM_BUFFER; - mDummyBufferWriteInfo.pImageInfo = nullptr; - mDummyBufferWriteInfo.pBufferInfo = &mDummyBufferInfo; - mDummyBufferWriteInfo.pTexelBufferView = nullptr; - - mDummyTargetInfo.imageLayout = VulkanImageUtility::getVkLayout(VulkanLayout::READ_ONLY); - mDummyTargetWriteInfo.sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET; - mDummyTargetWriteInfo.pNext = nullptr; - mDummyTargetWriteInfo.dstArrayElement = 0; - mDummyTargetWriteInfo.descriptorCount = 1; - mDummyTargetWriteInfo.descriptorType = VK_DESCRIPTOR_TYPE_INPUT_ATTACHMENT; - mDummyTargetWriteInfo.pImageInfo = &mDummyTargetInfo; - mDummyTargetWriteInfo.pBufferInfo = nullptr; - mDummyTargetWriteInfo.pTexelBufferView = nullptr; +VulkanPipelineCache::VulkanPipelineCache(VkDevice device, VmaAllocator allocator) + : mDevice(device), + mAllocator(allocator) { } VulkanPipelineCache::~VulkanPipelineCache() { @@ -81,244 +45,47 @@ VulkanPipelineCache::~VulkanPipelineCache() { // be explicit about teardown order of various components. } -void VulkanPipelineCache::setDevice(VkDevice device, VmaAllocator allocator) { - assert_invariant(mDevice == VK_NULL_HANDLE); - mDevice = device; - mAllocator = allocator; - mDescriptorPool = createDescriptorPool(mDescriptorPoolSize); - - // Formulate some dummy objects and dummy descriptor info used only for clearing out unused - // bindings. This is especially crucial after a texture has been destroyed. Since core Vulkan - // does not allow specifying VK_NULL_HANDLE without the robustness2 extension, we would need to - // change the pipeline layout more frequently if we wanted to get rid of these dummy objects. - - VkBufferCreateInfo bufferInfo { - .sType = VK_STRUCTURE_TYPE_BUFFER_CREATE_INFO, - .size = 16, - .usage = VK_BUFFER_USAGE_UNIFORM_BUFFER_BIT, - }; - VmaAllocationCreateInfo allocInfo { .usage = VMA_MEMORY_USAGE_GPU_ONLY }; - vmaCreateBuffer(mAllocator, &bufferInfo, &allocInfo, &mDummyBuffer, &mDummyMemory, nullptr); - - mDummyBufferInfo.buffer = mDummyBuffer; - mDummyBufferInfo.range = bufferInfo.size; +void VulkanPipelineCache::bindLayout(VkPipelineLayout layout) noexcept { + mPipelineRequirements.layout = layout; } -bool VulkanPipelineCache::bindDescriptors(VkCommandBuffer cmdbuffer) noexcept { - DescriptorMap::iterator descriptorIter = mDescriptorSets.find(mDescriptorRequirements); - - // Check if the required descriptors are already bound. If so, there's no need to do anything. - if (DescEqual equals; UTILS_LIKELY(equals(mBoundDescriptor, mDescriptorRequirements))) { - - // If the pipeline state during an app's first draw call happens to match the default state - // vector of the cache, then the cache is uninitialized and we should not return early. - if (UTILS_LIKELY(!mDescriptorSets.empty())) { - - // Since the descriptors are already bound, they should be found in the cache. - assert_invariant(descriptorIter != mDescriptorSets.end()); - - // Update the LRU "time stamp" (really a count of cmd buf submissions) before returning. - descriptorIter.value().lastUsed = mCurrentTime; - return true; - } - } - +VulkanPipelineCache::PipelineCacheEntry* VulkanPipelineCache::getOrCreatePipeline() noexcept { // If a cached object exists, re-use it, otherwise create a new one. - DescriptorCacheEntry* cacheEntry = UTILS_LIKELY(descriptorIter != mDescriptorSets.end()) ? - &descriptorIter.value() : createDescriptorSets(); - - // If a descriptor set overflow occurred, allow higher levels to handle it gracefully. - assert_invariant(cacheEntry != nullptr); - if (UTILS_UNLIKELY(cacheEntry == nullptr)) { - return false; - } - - cacheEntry->lastUsed = mCurrentTime; - mBoundDescriptor = mDescriptorRequirements; - // This passes the currently "bound" uniform buffer objects to pipeline that will be used in the - // draw call. - auto resourceEntry = mDescriptorResources.find(cacheEntry->id); - if (resourceEntry == mDescriptorResources.end()) { - mDescriptorResources[cacheEntry->id] - = std::make_unique(mResourceAllocator); - resourceEntry = mDescriptorResources.find(cacheEntry->id); - } - resourceEntry->second->acquireAll(&mPipelineBoundResources); - - vkCmdBindDescriptorSets(cmdbuffer, VK_PIPELINE_BIND_POINT_GRAPHICS, - getOrCreatePipelineLayout()->handle, 0, VulkanPipelineCache::DESCRIPTOR_TYPE_COUNT, - cacheEntry->handles.data(), 0, nullptr); - - return true; + if (PipelineMap::iterator pipelineIter = mPipelines.find(mPipelineRequirements); + pipelineIter != mPipelines.end()) { + auto& pipeline = pipelineIter.value(); + pipeline.lastUsed = mCurrentTime; + return &pipeline; + } + auto ret = createPipeline(); + ret->lastUsed = mCurrentTime; + return ret; } -bool VulkanPipelineCache::bindPipeline(VulkanCommandBuffer* commands) noexcept { +void VulkanPipelineCache::bindPipeline(VulkanCommandBuffer* commands) { VkCommandBuffer const cmdbuffer = commands->buffer(); - PipelineMap::iterator pipelineIter = mPipelines.find(mPipelineRequirements); - + PipelineCacheEntry* cacheEntry = getOrCreatePipeline(); // Check if the required pipeline is already bound. - if (PipelineEqual equals; UTILS_LIKELY(equals(mBoundPipeline, mPipelineRequirements))) { - assert_invariant(pipelineIter != mPipelines.end()); - pipelineIter.value().lastUsed = mCurrentTime; - return true; + if (cacheEntry->handle == commands->pipeline()) { + return; } - // If a cached object exists, re-use it, otherwise create a new one. - PipelineCacheEntry* cacheEntry = UTILS_LIKELY(pipelineIter != mPipelines.end()) ? - &pipelineIter.value() : createPipeline(); - // If an error occurred, allow higher levels to handle it gracefully. - assert_invariant(cacheEntry != nullptr); - if (UTILS_UNLIKELY(cacheEntry == nullptr)) { - return false; - } - - cacheEntry->lastUsed = mCurrentTime; - getOrCreatePipelineLayout()->lastUsed = mCurrentTime; + assert_invariant(cacheEntry != nullptr && "Failed to create/find pipeline"); mBoundPipeline = mPipelineRequirements; - vkCmdBindPipeline(cmdbuffer, VK_PIPELINE_BIND_POINT_GRAPHICS, cacheEntry->handle); - return true; + commands->setPipeline(cacheEntry->handle); } void VulkanPipelineCache::bindScissor(VkCommandBuffer cmdbuffer, VkRect2D scissor) noexcept { - if (UTILS_UNLIKELY(!equivalent(mCurrentScissor, scissor))) { - mCurrentScissor = scissor; - vkCmdSetScissor(cmdbuffer, 0, 1, &scissor); - } -} - -VulkanPipelineCache::DescriptorCacheEntry* VulkanPipelineCache::createDescriptorSets() noexcept { - PipelineLayoutCacheEntry* layoutCacheEntry = getOrCreatePipelineLayout(); - - DescriptorCacheEntry descriptorCacheEntry = { - .pipelineLayout = mPipelineRequirements.layout, - .id = mDescriptorCacheEntryCount++, - }; - - // Each of the arenas for this particular layout are guaranteed to have the same size. Check - // the first arena to see if any descriptor sets are available that can be re-claimed. If not, - // create brand new ones (one for each type). They will be added to the arena later, after they - // are no longer used. This occurs during the cleanup phase during command buffer submission. - auto& descriptorSetArenas = layoutCacheEntry->descriptorSetArenas; - if (descriptorSetArenas[0].empty()) { - - // If allocating a new descriptor set from the pool would cause it to overflow, then - // recreate the pool. The number of descriptor sets that have already been allocated from - // the pool is the sum of the "active" descriptor sets (mDescriptorSets) and the "dormant" - // descriptor sets (mDescriptorArenasCount). - // - // NOTE: technically both sides of the inequality below should be multiplied by - // DESCRIPTOR_TYPE_COUNT to get the true number of descriptor sets. - if (mDescriptorSets.size() + mDescriptorArenasCount + 1 > mDescriptorPoolSize) { - growDescriptorPool(); - } - - VkDescriptorSetAllocateInfo allocInfo = {}; - allocInfo.sType = VK_STRUCTURE_TYPE_DESCRIPTOR_SET_ALLOCATE_INFO; - allocInfo.descriptorPool = mDescriptorPool; - allocInfo.descriptorSetCount = DESCRIPTOR_TYPE_COUNT; - allocInfo.pSetLayouts = layoutCacheEntry->descriptorSetLayouts.data(); - VkResult error = vkAllocateDescriptorSets(mDevice, &allocInfo, - descriptorCacheEntry.handles.data()); - assert_invariant(error == VK_SUCCESS); - if (error != VK_SUCCESS) { - return nullptr; - } - } else { - for (uint32_t i = 0; i < DESCRIPTOR_TYPE_COUNT; ++i) { - descriptorCacheEntry.handles[i] = descriptorSetArenas[i].back(); - descriptorSetArenas[i].pop_back(); - } - assert_invariant(mDescriptorArenasCount > 0); - mDescriptorArenasCount--; - } - - // Rewrite every binding in the new descriptor sets. - VkDescriptorBufferInfo descriptorBuffers[UBUFFER_BINDING_COUNT]; - VkDescriptorImageInfo descriptorSamplers[SAMPLER_BINDING_COUNT]; - VkDescriptorImageInfo descriptorInputAttachments[INPUT_ATTACHMENT_COUNT]; - VkWriteDescriptorSet descriptorWrites[UBUFFER_BINDING_COUNT + SAMPLER_BINDING_COUNT + - INPUT_ATTACHMENT_COUNT]; - uint32_t nwrites = 0; - VkWriteDescriptorSet* writes = descriptorWrites; - nwrites = 0; - for (uint32_t binding = 0; binding < UBUFFER_BINDING_COUNT; binding++) { - VkWriteDescriptorSet& writeInfo = writes[nwrites++]; - if (mDescriptorRequirements.uniformBuffers[binding]) { - VkDescriptorBufferInfo& bufferInfo = descriptorBuffers[binding]; - bufferInfo.buffer = mDescriptorRequirements.uniformBuffers[binding]; - bufferInfo.offset = mDescriptorRequirements.uniformBufferOffsets[binding]; - bufferInfo.range = mDescriptorRequirements.uniformBufferSizes[binding]; - - // We store size with 32 bits, so our "WHOLE" sentinel is different from Vk. - if (bufferInfo.range == WHOLE_SIZE) { - bufferInfo.range = VK_WHOLE_SIZE; - } - - writeInfo.sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET; - writeInfo.pNext = nullptr; - writeInfo.dstArrayElement = 0; - writeInfo.descriptorCount = 1; - writeInfo.descriptorType = VK_DESCRIPTOR_TYPE_UNIFORM_BUFFER; - writeInfo.pImageInfo = nullptr; - writeInfo.pBufferInfo = &bufferInfo; - writeInfo.pTexelBufferView = nullptr; - } else { - writeInfo = mDummyBufferWriteInfo; - assert_invariant(mDummyBufferWriteInfo.pBufferInfo->buffer); - } - assert_invariant(writeInfo.pBufferInfo->buffer); - writeInfo.dstSet = descriptorCacheEntry.handles[0]; - writeInfo.dstBinding = binding; - } - for (uint32_t binding = 0; binding < SAMPLER_BINDING_COUNT; binding++) { - if (mDescriptorRequirements.samplers[binding].sampler) { - VkWriteDescriptorSet& writeInfo = writes[nwrites++]; - VkDescriptorImageInfo& imageInfo = descriptorSamplers[binding]; - imageInfo = mDescriptorRequirements.samplers[binding]; - writeInfo.sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET; - writeInfo.pNext = nullptr; - writeInfo.dstArrayElement = 0; - writeInfo.descriptorCount = 1; - writeInfo.descriptorType = VK_DESCRIPTOR_TYPE_COMBINED_IMAGE_SAMPLER; - writeInfo.pImageInfo = &imageInfo; - writeInfo.pBufferInfo = nullptr; - writeInfo.pTexelBufferView = nullptr; - writeInfo.dstSet = descriptorCacheEntry.handles[1]; - writeInfo.dstBinding = binding; - } - } - for (uint32_t binding = 0; binding < INPUT_ATTACHMENT_COUNT; binding++) { - if (mDescriptorRequirements.inputAttachments[binding].imageView) { - VkWriteDescriptorSet& writeInfo = writes[nwrites++]; - VkDescriptorImageInfo& imageInfo = descriptorInputAttachments[binding]; - imageInfo = mDescriptorRequirements.inputAttachments[binding]; - writeInfo.sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET; - writeInfo.pNext = nullptr; - writeInfo.dstArrayElement = 0; - writeInfo.descriptorCount = 1; - writeInfo.descriptorType = VK_DESCRIPTOR_TYPE_INPUT_ATTACHMENT; - writeInfo.pImageInfo = &imageInfo; - writeInfo.pBufferInfo = nullptr; - writeInfo.pTexelBufferView = nullptr; - writeInfo.dstSet = descriptorCacheEntry.handles[2]; - writeInfo.dstBinding = binding; - } - } - - vkUpdateDescriptorSets(mDevice, nwrites, writes, 0, nullptr); - - return &mDescriptorSets.emplace(mDescriptorRequirements, descriptorCacheEntry).first.value(); + vkCmdSetScissor(cmdbuffer, 0, 1, &scissor); } VulkanPipelineCache::PipelineCacheEntry* VulkanPipelineCache::createPipeline() noexcept { assert_invariant(mPipelineRequirements.shaders[0] && "Vertex shader is not bound."); - - PipelineLayoutCacheEntry* layout = getOrCreatePipelineLayout(); - assert_invariant(layout); + assert_invariant(mPipelineRequirements.layout && "No pipeline layout specified"); VkPipelineShaderStageCreateInfo shaderStages[SHADER_MODULE_COUNT]; shaderStages[0] = VkPipelineShaderStageCreateInfo{}; @@ -387,7 +154,7 @@ VulkanPipelineCache::PipelineCacheEntry* VulkanPipelineCache::createPipeline() n VkGraphicsPipelineCreateInfo pipelineCreateInfo = {}; pipelineCreateInfo.sType = VK_STRUCTURE_TYPE_GRAPHICS_PIPELINE_CREATE_INFO; - pipelineCreateInfo.layout = layout->handle; + pipelineCreateInfo.layout = mPipelineRequirements.layout; pipelineCreateInfo.renderPass = mPipelineRequirements.renderPass; pipelineCreateInfo.subpass = mPipelineRequirements.subpassIndex; pipelineCreateInfo.stageCount = hasFragmentShader ? SHADER_MODULE_COUNT : 1; @@ -481,68 +248,6 @@ VulkanPipelineCache::PipelineCacheEntry* VulkanPipelineCache::createPipeline() n return &mPipelines.emplace(mPipelineRequirements, cacheEntry).first.value(); } -VulkanPipelineCache::PipelineLayoutCacheEntry* VulkanPipelineCache::getOrCreatePipelineLayout() noexcept { - auto iter = mPipelineLayouts.find(mPipelineRequirements.layout); - if (UTILS_LIKELY(iter != mPipelineLayouts.end())) { - return &iter.value(); - } - - PipelineLayoutCacheEntry cacheEntry = {}; - - VkDescriptorSetLayoutBinding binding = {}; - binding.descriptorCount = 1; // NOTE: We never use arrays-of-blocks. - binding.stageFlags = VK_SHADER_STAGE_ALL_GRAPHICS; // NOTE: This is potentially non-optimal. - - // First create the descriptor set layout for UBO's. - VkDescriptorSetLayoutBinding ubindings[UBUFFER_BINDING_COUNT]; - binding.descriptorType = VK_DESCRIPTOR_TYPE_UNIFORM_BUFFER; - for (uint32_t i = 0; i < UBUFFER_BINDING_COUNT; i++) { - binding.binding = i; - ubindings[i] = binding; - } - VkDescriptorSetLayoutCreateInfo dlinfo = {}; - dlinfo.sType = VK_STRUCTURE_TYPE_DESCRIPTOR_SET_LAYOUT_CREATE_INFO; - dlinfo.bindingCount = UBUFFER_BINDING_COUNT; - dlinfo.pBindings = ubindings; - vkCreateDescriptorSetLayout(mDevice, &dlinfo, VKALLOC, &cacheEntry.descriptorSetLayouts[0]); - - // Next create the descriptor set layout for samplers. - VkDescriptorSetLayoutBinding sbindings[SAMPLER_BINDING_COUNT]; - binding.descriptorType = VK_DESCRIPTOR_TYPE_COMBINED_IMAGE_SAMPLER; - for (uint32_t i = 0; i < SAMPLER_BINDING_COUNT; i++) { - binding.stageFlags = getShaderStageFlags(mPipelineRequirements.layout, i); - binding.binding = i; - sbindings[i] = binding; - } - dlinfo.bindingCount = SAMPLER_BINDING_COUNT; - dlinfo.pBindings = sbindings; - vkCreateDescriptorSetLayout(mDevice, &dlinfo, VKALLOC, &cacheEntry.descriptorSetLayouts[1]); - - // Next create the descriptor set layout for input attachments. - VkDescriptorSetLayoutBinding tbindings[INPUT_ATTACHMENT_COUNT]; - binding.descriptorType = VK_DESCRIPTOR_TYPE_INPUT_ATTACHMENT; - binding.stageFlags = VK_SHADER_STAGE_FRAGMENT_BIT; - for (uint32_t i = 0; i < INPUT_ATTACHMENT_COUNT; i++) { - binding.binding = i; - tbindings[i] = binding; - } - dlinfo.bindingCount = INPUT_ATTACHMENT_COUNT; - dlinfo.pBindings = tbindings; - vkCreateDescriptorSetLayout(mDevice, &dlinfo, VKALLOC, &cacheEntry.descriptorSetLayouts[2]); - - // Create VkPipelineLayout based on how to resources are bounded. - VkPipelineLayoutCreateInfo pPipelineLayoutCreateInfo = {}; - pPipelineLayoutCreateInfo.sType = VK_STRUCTURE_TYPE_PIPELINE_LAYOUT_CREATE_INFO; - pPipelineLayoutCreateInfo.setLayoutCount = cacheEntry.descriptorSetLayouts.size(); - pPipelineLayoutCreateInfo.pSetLayouts = cacheEntry.descriptorSetLayouts.data(); - VkResult result = vkCreatePipelineLayout(mDevice, &pPipelineLayoutCreateInfo, VKALLOC, - &cacheEntry.handle); - if (UTILS_UNLIKELY(result != VK_SUCCESS)) { - return nullptr; - } - return &mPipelineLayouts.emplace(mPipelineRequirements.layout, cacheEntry).first.value(); -} - void VulkanPipelineCache::bindProgram(VulkanProgram* program) noexcept { mPipelineRequirements.shaders[0] = program->getVertexShader(); mPipelineRequirements.shaders[1] = program->getFragmentShader(); @@ -583,97 +288,15 @@ void VulkanPipelineCache::bindVertexArray(VkVertexInputAttributeDescription cons } } -VulkanPipelineCache::UniformBufferBinding VulkanPipelineCache::getUniformBufferBinding( - uint32_t bindingIndex) const noexcept { - auto& key = mDescriptorRequirements; - return { - key.uniformBuffers[bindingIndex], - key.uniformBufferOffsets[bindingIndex], - key.uniformBufferSizes[bindingIndex], - }; -} - -void VulkanPipelineCache::unbindUniformBuffer(VkBuffer uniformBuffer) noexcept { - auto& key = mDescriptorRequirements; - for (uint32_t bindingIndex = 0u; bindingIndex < UBUFFER_BINDING_COUNT; ++bindingIndex) { - if (key.uniformBuffers[bindingIndex] == uniformBuffer) { - key.uniformBuffers[bindingIndex] = {}; - key.uniformBufferSizes[bindingIndex] = {}; - key.uniformBufferOffsets[bindingIndex] = {}; - } - } -} - -void VulkanPipelineCache::unbindImageView(VkImageView imageView) noexcept { - for (auto& sampler : mDescriptorRequirements.samplers) { - if (sampler.imageView == imageView) { - sampler = {}; - } - } - for (auto& target : mDescriptorRequirements.inputAttachments) { - if (target.imageView == imageView) { - target = {}; - } - } -} - -void VulkanPipelineCache::bindUniformBufferObject(uint32_t bindingIndex, - VulkanBufferObject* bufferObject, VkDeviceSize offset, VkDeviceSize size) noexcept { - VkBuffer buffer = bufferObject->buffer.getGpuBuffer(); - - ASSERT_POSTCONDITION(bindingIndex < UBUFFER_BINDING_COUNT, - "Uniform bindings overflow: index = %d, capacity = %d.", bindingIndex, - UBUFFER_BINDING_COUNT); - auto& key = mDescriptorRequirements; - key.uniformBuffers[bindingIndex] = buffer; - - if (size == VK_WHOLE_SIZE) { - size = WHOLE_SIZE; - } - - assert_invariant(offset <= 0xffffffffu); - assert_invariant(size <= 0xffffffffu); - - key.uniformBufferOffsets[bindingIndex] = offset; - key.uniformBufferSizes[bindingIndex] = size; - - mPipelineBoundResources.acquire(bufferObject); -} - -void VulkanPipelineCache::bindSamplers(VkDescriptorImageInfo samplers[SAMPLER_BINDING_COUNT], - VulkanTexture* textures[SAMPLER_BINDING_COUNT], UsageFlags flags) noexcept { - for (uint32_t bindingIndex = 0; bindingIndex < SAMPLER_BINDING_COUNT; bindingIndex++) { - mDescriptorRequirements.samplers[bindingIndex] = samplers[bindingIndex]; - if (textures[bindingIndex]) { - mPipelineBoundResources.acquire(textures[bindingIndex]); - } - } - mPipelineRequirements.layout = flags; -} - -void VulkanPipelineCache::bindInputAttachment(uint32_t bindingIndex, - VkDescriptorImageInfo targetInfo) noexcept { - ASSERT_POSTCONDITION(bindingIndex < INPUT_ATTACHMENT_COUNT, - "Input attachment bindings overflow: index = %d, capacity = %d.", - bindingIndex, INPUT_ATTACHMENT_COUNT); - mDescriptorRequirements.inputAttachments[bindingIndex] = targetInfo; -} - void VulkanPipelineCache::terminate() noexcept { - // Symmetric to createLayoutsAndDescriptors. - destroyLayoutsAndDescriptors(); for (auto& iter : mPipelines) { vkDestroyPipeline(mDevice, iter.second.handle, VKALLOC); } - mPipelineBoundResources.clear(); mPipelines.clear(); mBoundPipeline = {}; - vmaDestroyBuffer(mAllocator, mDummyBuffer, mDummyMemory); - mDummyBuffer = VK_NULL_HANDLE; - mDummyMemory = VK_NULL_HANDLE; } -void VulkanPipelineCache::onCommandBuffer(const VulkanCommandBuffer& commands) { +void VulkanPipelineCache::gc() noexcept { // The timestamp associated with a given cache entry represents "time" as a count of flush // events since the cache was constructed. If any cache entry was most recently used over // FVK_MAX_PIPELINE_AGE flush events in the past, then we can be sure that it is no longer @@ -683,194 +306,22 @@ void VulkanPipelineCache::onCommandBuffer(const VulkanCommandBuffer& commands) { // The Vulkan spec says: "When a command buffer begins recording, all state in that command // buffer is undefined." Therefore, we need to clear all bindings at this time. mBoundPipeline = {}; - mBoundDescriptor = {}; mCurrentScissor = {}; // NOTE: Due to robin_map restrictions, we cannot use auto or range-based loops. - // Check if any bundles in the cache are no longer in use by any command buffer. Descriptors - // from unused bundles are moved back to their respective arenas. - using ConstDescIterator = decltype(mDescriptorSets)::const_iterator; - for (ConstDescIterator iter = mDescriptorSets.begin(); iter != mDescriptorSets.end();) { - const DescriptorCacheEntry& cacheEntry = iter.value(); - if (cacheEntry.lastUsed + FVK_MAX_PIPELINE_AGE < mCurrentTime) { - auto& arenas = mPipelineLayouts[cacheEntry.pipelineLayout].descriptorSetArenas; - for (uint32_t i = 0; i < DESCRIPTOR_TYPE_COUNT; ++i) { - arenas[i].push_back(cacheEntry.handles[i]); - } - ++mDescriptorArenasCount; - mDescriptorResources.erase(cacheEntry.id); - iter = mDescriptorSets.erase(iter); - } else { - ++iter; - } - } - // Evict any pipelines that have not been used in a while. // Any pipeline older than FVK_MAX_COMMAND_BUFFERS can be safely destroyed. - using ConstPipeIterator = decltype(mPipelines)::const_iterator; - for (ConstPipeIterator iter = mPipelines.begin(); iter != mPipelines.end();) { - const PipelineCacheEntry& cacheEntry = iter.value(); - if (cacheEntry.lastUsed + FVK_MAX_PIPELINE_AGE < mCurrentTime) { - vkDestroyPipeline(mDevice, iter->second.handle, VKALLOC); - iter = mPipelines.erase(iter); - } else { - ++iter; - } - } - - // Evict any layouts that have not been used in a while. - using ConstLayoutIterator = decltype(mPipelineLayouts)::const_iterator; - for (ConstLayoutIterator iter = mPipelineLayouts.begin(); iter != mPipelineLayouts.end();) { - const PipelineLayoutCacheEntry& cacheEntry = iter.value(); - if (cacheEntry.lastUsed + FVK_MAX_PIPELINE_AGE < mCurrentTime) { - vkDestroyPipelineLayout(mDevice, iter->second.handle, VKALLOC); - for (auto setLayout : iter->second.descriptorSetLayouts) { - #if FVK_ENABLED(FVK_DEBUG_PIPELINE_CACHE) - PipelineLayoutKey key = iter.key(); - for (auto& pair : mDescriptorSets) { - assert_invariant(pair.second.pipelineLayout != key); - } - #endif - vkDestroyDescriptorSetLayout(mDevice, setLayout, VKALLOC); - } - auto& arenas = iter->second.descriptorSetArenas; - assert_invariant(mDescriptorArenasCount >= arenas[0].size()); - mDescriptorArenasCount -= arenas[0].size(); - for (auto& arena : arenas) { - vkFreeDescriptorSets(mDevice, mDescriptorPool, arena.size(), arena.data()); - } - iter = mPipelineLayouts.erase(iter); - } else { - ++iter; - } - } - - // If there are no descriptors from any extinct pool that are still in use, we can safely - // destroy the extinct pools, which implicitly frees their associated descriptor sets. - bool canPurgeExtinctPools = true; - for (auto& bundle : mExtinctDescriptorBundles) { - if (bundle.lastUsed + FVK_MAX_PIPELINE_AGE >= mCurrentTime) { - canPurgeExtinctPools = false; - break; - } - } - if (canPurgeExtinctPools) { - for (VkDescriptorPool pool : mExtinctDescriptorPools) { - vkDestroyDescriptorPool(mDevice, pool, VKALLOC); - } - mExtinctDescriptorPools.clear(); - - for (auto const& entry : mExtinctDescriptorBundles) { - mDescriptorResources.erase(entry.id); - } - mExtinctDescriptorBundles.clear(); - } -} - -VkDescriptorPool VulkanPipelineCache::createDescriptorPool(uint32_t size) const { - VkDescriptorPoolSize poolSizes[DESCRIPTOR_TYPE_COUNT] = {}; - VkDescriptorPoolCreateInfo poolInfo { - .sType = VK_STRUCTURE_TYPE_DESCRIPTOR_POOL_CREATE_INFO, - .pNext = nullptr, - .flags = VK_DESCRIPTOR_POOL_CREATE_FREE_DESCRIPTOR_SET_BIT, - .maxSets = size * DESCRIPTOR_TYPE_COUNT, - .poolSizeCount = DESCRIPTOR_TYPE_COUNT, - .pPoolSizes = poolSizes - }; - poolSizes[0].type = VK_DESCRIPTOR_TYPE_UNIFORM_BUFFER; - poolSizes[0].descriptorCount = poolInfo.maxSets * UBUFFER_BINDING_COUNT; - poolSizes[1].type = VK_DESCRIPTOR_TYPE_COMBINED_IMAGE_SAMPLER; - poolSizes[1].descriptorCount = poolInfo.maxSets * SAMPLER_BINDING_COUNT; - poolSizes[2].type = VK_DESCRIPTOR_TYPE_INPUT_ATTACHMENT; - poolSizes[2].descriptorCount = poolInfo.maxSets * INPUT_ATTACHMENT_COUNT; - - VkDescriptorPool pool; - const UTILS_UNUSED VkResult result = vkCreateDescriptorPool(mDevice, &poolInfo, VKALLOC, &pool); - assert_invariant(result == VK_SUCCESS); - return pool; -} - -void VulkanPipelineCache::destroyLayoutsAndDescriptors() noexcept { - // Our current descriptor set strategy can cause the # of descriptor sets to explode in certain - // situations, so it's interesting to report the number that get stuffed into the cache. - #if FVK_ENABLED(FVK_DEBUG_PIPELINE_CACHE) - utils::slog.d << "Destroying " << mDescriptorSets.size() << " bundles of descriptor sets." - << utils::io::endl; - #endif - - mDescriptorSets.clear(); - - // Our current layout bundle strategy can cause the # of layout bundles to explode in certain - // situations, so it's interesting to report the number that get stuffed into the cache. - #if FVK_ENABLED(FVK_DEBUG_PIPELINE_CACHE) - utils::slog.d << "Destroying " << mPipelineLayouts.size() << " pipeline layouts." - << utils::io::endl; - #endif - - for (auto& iter : mPipelineLayouts) { - vkDestroyPipelineLayout(mDevice, iter.second.handle, VKALLOC); - for (auto setLayout : iter.second.descriptorSetLayouts) { - vkDestroyDescriptorSetLayout(mDevice, setLayout, VKALLOC); - } - // There is no need to free descriptor sets individually since destroying the VkDescriptorPool - // implicitly frees them. - } - mPipelineLayouts.clear(); - vkDestroyDescriptorPool(mDevice, mDescriptorPool, VKALLOC); - mDescriptorPool = VK_NULL_HANDLE; - - for (VkDescriptorPool pool : mExtinctDescriptorPools) { - vkDestroyDescriptorPool(mDevice, pool, VKALLOC); - } - mExtinctDescriptorPools.clear(); - mExtinctDescriptorBundles.clear(); - - // Both mDescriptorSets and mExtinctDescriptorBundles have been cleared, so it's safe to call - // clear() on mDescriptorResources. - mDescriptorResources.clear(); - - mBoundDescriptor = {}; -} - -void VulkanPipelineCache::growDescriptorPool() noexcept { - // We need to destroy the old VkDescriptorPool, but we can't do so immediately because many - // of its descriptors are still in use. So, stash it in an "extinct" list. - mExtinctDescriptorPools.push_back(mDescriptorPool); - - // Create the new VkDescriptorPool, twice as big as the old one. - mDescriptorPoolSize *= 2; - mDescriptorPool = createDescriptorPool(mDescriptorPoolSize); - - // Clear out all unused descriptor sets in the arena so they don't get reclaimed. There is no - // need to free them individually since the old VkDescriptorPool will be destroyed. - for (auto iter = mPipelineLayouts.begin(); iter != mPipelineLayouts.end(); ++iter) { - for (auto& arena : iter.value().descriptorSetArenas) { - arena.clear(); - } - } - mDescriptorArenasCount = 0; - - // Move all in-use descriptors from the primary cache into an "extinct" list, so that they will - // later be destroyed rather than reclaimed. - using DescIterator = decltype(mDescriptorSets)::iterator; - for (DescIterator iter = mDescriptorSets.begin(); iter != mDescriptorSets.end(); ++iter) { - mExtinctDescriptorBundles.push_back(iter.value()); - } - mDescriptorSets.clear(); -} - -size_t VulkanPipelineCache::PipelineLayoutKeyHashFn::operator()( - const PipelineLayoutKey& key) const { - std::hash hasher; - auto h0 = hasher(key.getBitsAt(0)); - auto h1 = hasher(key.getBitsAt(1)); - return h0 ^ (h1 << 1); -} - -bool VulkanPipelineCache::PipelineLayoutKeyEqual::operator()(const PipelineLayoutKey& k1, - const PipelineLayoutKey& k2) const { - return k1 == k2; + using ConstPipeIterator = decltype(mPipelines)::const_iterator; + for (ConstPipeIterator iter = mPipelines.begin(); iter != mPipelines.end();) { + const PipelineCacheEntry& cacheEntry = iter.value(); + if (cacheEntry.lastUsed + FVK_MAX_PIPELINE_AGE < mCurrentTime) { + vkDestroyPipeline(mDevice, iter->second.handle, VKALLOC); + iter = mPipelines.erase(iter); + } else { + ++iter; + } + } } bool VulkanPipelineCache::PipelineEqual::operator()(const PipelineKey& k1, @@ -878,31 +329,6 @@ bool VulkanPipelineCache::PipelineEqual::operator()(const PipelineKey& k1, return 0 == memcmp((const void*) &k1, (const void*) &k2, sizeof(k1)); } -bool VulkanPipelineCache::DescEqual::operator()(const DescriptorKey& k1, - const DescriptorKey& k2) const { - for (uint32_t i = 0; i < UBUFFER_BINDING_COUNT; i++) { - if (k1.uniformBuffers[i] != k2.uniformBuffers[i] || - k1.uniformBufferOffsets[i] != k2.uniformBufferOffsets[i] || - k1.uniformBufferSizes[i] != k2.uniformBufferSizes[i]) { - return false; - } - } - for (uint32_t i = 0; i < SAMPLER_BINDING_COUNT; i++) { - if (k1.samplers[i].sampler != k2.samplers[i].sampler || - k1.samplers[i].imageView != k2.samplers[i].imageView || - k1.samplers[i].imageLayout != k2.samplers[i].imageLayout) { - return false; - } - } - for (uint32_t i = 0; i < INPUT_ATTACHMENT_COUNT; i++) { - if (k1.inputAttachments[i].imageView != k2.inputAttachments[i].imageView || - k1.inputAttachments[i].imageLayout != k2.inputAttachments[i].imageLayout) { - return false; - } - } - return true; -} - } // namespace filament::backend #pragma clang diagnostic pop diff --git a/filament/backend/src/vulkan/VulkanPipelineCache.h b/filament/backend/src/vulkan/VulkanPipelineCache.h index 10bc845eb3d..53eaf71287c 100644 --- a/filament/backend/src/vulkan/VulkanPipelineCache.h +++ b/filament/backend/src/vulkan/VulkanPipelineCache.h @@ -17,6 +17,11 @@ #ifndef TNT_FILAMENT_BACKEND_VULKANPIPELINECACHE_H #define TNT_FILAMENT_BACKEND_VULKANPIPELINECACHE_H +#include "VulkanCommands.h" +#include "VulkanMemory.h" +#include "VulkanResources.h" +#include "VulkanUtility.h" + #include #include @@ -34,12 +39,6 @@ #include #include -#include "VulkanCommands.h" - -VK_DEFINE_HANDLE(VmaAllocator) -VK_DEFINE_HANDLE(VmaAllocation) -VK_DEFINE_HANDLE(VmaPool) - namespace filament::backend { struct VulkanProgram; @@ -56,32 +55,14 @@ class VulkanResourceAllocator; // - Assumes that viewport and scissor should be dynamic. (not baked into VkPipeline) // - Assumes that uniform buffers should be visible across all shader stages. // -class VulkanPipelineCache : public CommandBufferObserver { +class VulkanPipelineCache { public: VulkanPipelineCache(VulkanPipelineCache const&) = delete; VulkanPipelineCache& operator=(VulkanPipelineCache const&) = delete; - static constexpr uint32_t UBUFFER_BINDING_COUNT = Program::UNIFORM_BINDING_COUNT; - static constexpr uint32_t SAMPLER_BINDING_COUNT = MAX_SAMPLER_COUNT; - - // We assume only one possible input attachment between two subpasses. See also the subpasses - // definition in VulkanFboCache. - static constexpr uint32_t INPUT_ATTACHMENT_COUNT = 1; - static constexpr uint32_t SHADER_MODULE_COUNT = 2; static constexpr uint32_t VERTEX_ATTRIBUTE_COUNT = MAX_VERTEX_ATTRIBUTE_COUNT; - // Three descriptor set layouts: uniforms, combined image samplers, and input attachments. - static constexpr uint32_t DESCRIPTOR_TYPE_COUNT = 3; - static constexpr uint32_t INITIAL_DESCRIPTOR_SET_POOL_SIZE = 512; - - // The VertexArray POD is an array of buffer targets and an array of attributes that refer to - // those targets. It does not include any references to actual buffers, so you can think of it - // as a vertex assembler configuration. For simplicity it contains fixed-size arrays and does - // not store sizes; all unused entries are simply zeroed out. - struct VertexArray { - }; - // The ProgramBundle contains weak references to the compiled vertex and fragment shaders. struct ProgramBundle { VkShaderModule vertex; @@ -89,8 +70,6 @@ class VulkanPipelineCache : public CommandBufferObserver { VkSpecializationInfo* specializationInfos = nullptr; }; - static UsageFlags disableUsageFlags(uint16_t binding, UsageFlags src); - #pragma clang diagnostic push #pragma clang diagnostic warning "-Wpadded" @@ -133,17 +112,13 @@ class VulkanPipelineCache : public CommandBufferObserver { // Upon construction, the pipeCache initializes some internal state but does not make any Vulkan // calls. On destruction it will free any cached Vulkan objects that haven't already been freed. - VulkanPipelineCache(VulkanResourceAllocator* allocator); + VulkanPipelineCache(VkDevice device, VmaAllocator allocator); ~VulkanPipelineCache(); - void setDevice(VkDevice device, VmaAllocator allocator); - // Creates new descriptor sets if necessary and binds them using vkCmdBindDescriptorSets. - // Returns false if descriptor set allocation fails. - bool bindDescriptors(VkCommandBuffer cmdbuffer) noexcept; + void bindLayout(VkPipelineLayout layout) noexcept; // Creates a new pipeline if necessary and binds it using vkCmdBindPipeline. - // Returns false if an error occurred. - bool bindPipeline(VulkanCommandBuffer* commands) noexcept; + void bindPipeline(VulkanCommandBuffer* commands); // Sets up a new scissor rectangle if it has been dirtied. void bindScissor(VkCommandBuffer cmdbuffer, VkRect2D scissor) noexcept; @@ -153,42 +128,13 @@ class VulkanPipelineCache : public CommandBufferObserver { void bindRasterState(const RasterState& rasterState) noexcept; void bindRenderPass(VkRenderPass renderPass, int subpassIndex) noexcept; void bindPrimitiveTopology(VkPrimitiveTopology topology) noexcept; - void bindUniformBufferObject(uint32_t bindingIndex, VulkanBufferObject* bufferObject, - VkDeviceSize offset = 0, VkDeviceSize size = VK_WHOLE_SIZE) noexcept; - void bindSamplers(VkDescriptorImageInfo samplers[SAMPLER_BINDING_COUNT], - VulkanTexture* textures[SAMPLER_BINDING_COUNT], UsageFlags flags) noexcept; - void bindInputAttachment(uint32_t bindingIndex, VkDescriptorImageInfo imageInfo) noexcept; + void bindVertexArray(VkVertexInputAttributeDescription const* attribDesc, VkVertexInputBindingDescription const* bufferDesc, uint8_t count); - // Gets the current UBO at the given slot, useful for push / pop. - UniformBufferBinding getUniformBufferBinding(uint32_t bindingIndex) const noexcept; - - // Checks if the given uniform is bound to any slot, and if so binds "null" to that slot. - // Also invalidates all cached descriptors that refer to the given buffer. - // This is only necessary when the client knows that the UBO is about to be destroyed. - void unbindUniformBuffer(VkBuffer uniformBuffer) noexcept; - - // Checks if an image view is bound to any sampler, and if so resets that particular slot. - // Also invalidates all cached descriptors that refer to the given image view. - // This is only necessary when the client knows that a texture is about to be destroyed. - void unbindImageView(VkImageView imageView) noexcept; - - // NOTE: In theory we should proffer "unbindSampler" but in practice we never destroy samplers. - // Destroys all managed Vulkan objects. This should be called before changing the VkDevice. void terminate() noexcept; - // vkCmdBindPipeline and vkCmdBindDescriptorSets establish bindings to a specific command - // buffer; they are not global to the device. Therefore we need to be notified when a - // new command buffer becomes active. - void onCommandBuffer(const VulkanCommandBuffer& cmdbuffer) override; - - // Injects a dummy texture that can be used to clear out old descriptor sets. - void setDummyTexture(VkImageView imageView) { - mDummyTargetInfo.imageView = imageView; - } - static VkPrimitiveTopology getPrimitiveTopology(PrimitiveType pt) noexcept { switch (pt) { case PrimitiveType::POINTS: @@ -204,22 +150,9 @@ class VulkanPipelineCache : public CommandBufferObserver { } } -private: - // PIPELINE LAYOUT CACHE KEY - // ------------------------- - - using PipelineLayoutKey = utils::bitset128; - - static_assert(PipelineLayoutKey::BIT_COUNT >= 2 * MAX_SAMPLER_COUNT); - - struct PipelineLayoutKeyHashFn { - size_t operator()(const PipelineLayoutKey& key) const; - }; - - struct PipelineLayoutKeyEqual { - bool operator()(const PipelineLayoutKey& k1, const PipelineLayoutKey& k2) const; - }; + void gc() noexcept; +private: // PIPELINE CACHE KEY // ------------------ @@ -272,10 +205,10 @@ class VulkanPipelineCache : public CommandBufferObserver { VertexInputBindingDescription vertexBuffers[VERTEX_ATTRIBUTE_COUNT]; // 128 : 156 RasterState rasterState; // 16 : 284 uint32_t padding; // 4 : 300 - PipelineLayoutKey layout; // 16 : 304 + VkPipelineLayout layout; // 8 : 304 }; - static_assert(sizeof(PipelineKey) == 320, "PipelineKey must not have implicit padding."); + static_assert(sizeof(PipelineKey) == 312, "PipelineKey must not have implicit padding."); using PipelineHashFn = utils::hash::MurmurHashFn; @@ -283,52 +216,6 @@ class VulkanPipelineCache : public CommandBufferObserver { bool operator()(const PipelineKey& k1, const PipelineKey& k2) const; }; - // DESCRIPTOR SET CACHE KEY - // ------------------------ - - // Equivalent to VkDescriptorImageInfo but with explicit padding. - struct DescriptorImageInfo { - DescriptorImageInfo& operator=(const VkDescriptorImageInfo& that) { - sampler = that.sampler; - imageView = that.imageView; - imageLayout = that.imageLayout; - padding = 0; - return *this; - } - operator VkDescriptorImageInfo() const { return { sampler, imageView, imageLayout }; } - - // TODO: replace the 64-bit sampler handle with `uint32_t samplerParams` and remove the - // padding field. This is possible if we have access to the VulkanSamplerCache. - VkSampler sampler; - - VkImageView imageView; - VkImageLayout imageLayout; - uint32_t padding; - }; - - // We store size with 32 bits, so our "WHOLE" sentinel is different from Vk. - static const uint32_t WHOLE_SIZE = 0xffffffffu; - - // Represents all the Vulkan state that comprises a bound descriptor set. - struct DescriptorKey { - VkBuffer uniformBuffers[UBUFFER_BINDING_COUNT]; // 80 0 - DescriptorImageInfo samplers[SAMPLER_BINDING_COUNT]; // 1488 80 - DescriptorImageInfo inputAttachments[INPUT_ATTACHMENT_COUNT]; // 24 1568 - uint32_t uniformBufferOffsets[UBUFFER_BINDING_COUNT]; // 40 1592 - uint32_t uniformBufferSizes[UBUFFER_BINDING_COUNT]; // 40 1632 - }; - static_assert(offsetof(DescriptorKey, samplers) == 80); - static_assert(offsetof(DescriptorKey, inputAttachments) == 1568); - static_assert(offsetof(DescriptorKey, uniformBufferOffsets) == 1592); - static_assert(offsetof(DescriptorKey, uniformBufferSizes) == 1632); - static_assert(sizeof(DescriptorKey) == 1672, "DescriptorKey must not have implicit padding."); - - using DescHashFn = utils::hash::MurmurHashFn; - - struct DescEqual { - bool operator()(const DescriptorKey& k1, const DescriptorKey& k2) const; - }; - #pragma clang diagnostic pop // CACHE ENTRY STRUCTS @@ -341,16 +228,6 @@ class VulkanPipelineCache : public CommandBufferObserver { using Timestamp = uint64_t; Timestamp mCurrentTime = 0; - // The descriptor set cache entry is a group of descriptor sets that are bound simultaneously. - struct DescriptorCacheEntry { - std::array handles; - Timestamp lastUsed; - PipelineLayoutKey pipelineLayout; - uint32_t id; - }; - uint32_t mDescriptorCacheEntryCount = 0; - - struct PipelineCacheEntry { VkPipeline handle; Timestamp lastUsed; @@ -359,98 +236,36 @@ class VulkanPipelineCache : public CommandBufferObserver { struct PipelineLayoutCacheEntry { VkPipelineLayout handle; Timestamp lastUsed; - - std::array descriptorSetLayouts; - - // Each pipeline layout has 3 arenas of unused descriptors (one for each binding type). - // - // The difference between the "arenas" and the "pool" are as follows. - // - // - The "pool" is a single, centralized factory for all descriptors (VkDescriptorPool). - // - // - Each "arena" is a set of unused (but alive) descriptors that can only be used with a - // specific pipeline layout and a specific binding type. We manually manage each arena. - // The arenas are created in an empty state, and they are gradually populated as new - // descriptors are reclaimed over time. This is quite different from the pool, which is - // given a fixed size when it is constructed. - // - std::array, DESCRIPTOR_TYPE_COUNT> descriptorSetArenas; }; // CACHE CONTAINERS // ---------------- - using PipelineLayoutMap = tsl::robin_map; using PipelineMap = tsl::robin_map; - using DescriptorMap - = tsl::robin_map; - using DescriptorResourceMap - = std::unordered_map>; - PipelineLayoutMap mPipelineLayouts; +private: + + PipelineCacheEntry* getOrCreatePipeline() noexcept; + PipelineMap mPipelines; - DescriptorMap mDescriptorSets; - DescriptorResourceMap mDescriptorResources; // These helpers all return unstable pointers that should not be stored. - DescriptorCacheEntry* createDescriptorSets() noexcept; PipelineCacheEntry* createPipeline() noexcept; PipelineLayoutCacheEntry* getOrCreatePipelineLayout() noexcept; - // Misc helper methods. - void destroyLayoutsAndDescriptors() noexcept; - VkDescriptorPool createDescriptorPool(uint32_t size) const; - void growDescriptorPool() noexcept; - // Immutable state. VkDevice mDevice = VK_NULL_HANDLE; VmaAllocator mAllocator = VK_NULL_HANDLE; // Current requirements for the pipeline layout, pipeline, and descriptor sets. PipelineKey mPipelineRequirements = {}; - DescriptorKey mDescriptorRequirements = {}; // Current bindings for the pipeline and descriptor sets. PipelineKey mBoundPipeline = {}; - DescriptorKey mBoundDescriptor = {}; // Current state for scissoring. VkRect2D mCurrentScissor = {}; - - // The descriptor set pool starts out with a decent number of descriptor sets. The cache can - // grow the pool by re-creating it with a larger size. See growDescriptorPool(). - VkDescriptorPool mDescriptorPool; - - // This describes the number of descriptor sets in mDescriptorPool. Note that this needs to be - // multiplied by DESCRIPTOR_TYPE_COUNT to get the actual number of descriptor sets. Also note - // that the number of low-level "descriptors" (not descriptor *sets*) is actually much more than - // this size. It can be computed only by factoring in UBUFFER_BINDING_COUNT etc. - uint32_t mDescriptorPoolSize = INITIAL_DESCRIPTOR_SET_POOL_SIZE; - - // To get the actual number of descriptor sets that have been allocated from the pool, - // take the sum of mDescriptorArenasCount (these are inactive descriptor sets) and the - // number of entries in the mDescriptorPool map (active descriptor sets). Multiply the result by - // DESCRIPTOR_TYPE_COUNT. - uint32_t mDescriptorArenasCount = 0; - - // After a growth event (i.e. when the VkDescriptorPool is replaced with a bigger version), all - // currently used descriptors are moved into the "extinct" sets so that they can be safely - // destroyed a few frames later. - std::list mExtinctDescriptorPools; - std::list mExtinctDescriptorBundles; - - VkDescriptorBufferInfo mDummyBufferInfo = {}; - VkWriteDescriptorSet mDummyBufferWriteInfo = {}; - VkDescriptorImageInfo mDummyTargetInfo = {}; - VkWriteDescriptorSet mDummyTargetWriteInfo = {}; - - VkBuffer mDummyBuffer; - VmaAllocation mDummyMemory; - - VulkanResourceAllocator* mResourceAllocator; - VulkanAcquireOnlyResourceManager mPipelineBoundResources; }; } // namespace filament::backend diff --git a/filament/backend/src/vulkan/VulkanResourceAllocator.h b/filament/backend/src/vulkan/VulkanResourceAllocator.h index cd63b6d76ab..3afc0d6b965 100644 --- a/filament/backend/src/vulkan/VulkanResourceAllocator.h +++ b/filament/backend/src/vulkan/VulkanResourceAllocator.h @@ -51,8 +51,8 @@ namespace filament::backend { class VulkanResourceAllocator { public: using AllocatorImpl = HandleAllocatorVK; - VulkanResourceAllocator(size_t arenaSize, bool disableUseAfterFreeCheck) - : mHandleAllocatorImpl("Handles", arenaSize, disableUseAfterFreeCheck) + VulkanResourceAllocator(size_t arenaSize, bool disableUseAfterFreeCheck) + : mHandleAllocatorImpl("Handles", arenaSize, disableUseAfterFreeCheck) #if DEBUG_RESOURCE_LEAKS , mDebugOnlyResourceCount(RESOURCE_TYPE_COUNT) { std::memset(mDebugOnlyResourceCount.data(), 0, sizeof(size_t) * RESOURCE_TYPE_COUNT); diff --git a/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.cpp b/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.cpp index 77706e92d67..cad4339a1b0 100644 --- a/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.cpp +++ b/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.cpp @@ -80,7 +80,8 @@ class DescriptorPool { mCount(count), mCapacity(capacity), mSize(0), - mUnusedCount(0) { + mUnusedCount(0), + mDisableRecycling(false) { DescriptorCount const actual = mCount * capacity; VkDescriptorPoolSize sizes[4]; uint8_t npools = 0; @@ -111,7 +112,7 @@ class DescriptorPool { VkDescriptorPoolCreateInfo info{ .sType = VK_STRUCTURE_TYPE_DESCRIPTOR_POOL_CREATE_INFO, .pNext = nullptr, - .flags = VK_DESCRIPTOR_POOL_CREATE_FREE_DESCRIPTOR_SET_BIT, + .flags = 0, .maxSets = capacity, .poolSizeCount = npools, .pPoolSizes = sizes, @@ -123,9 +124,19 @@ class DescriptorPool { DescriptorPool& operator=(DescriptorPool const&) = delete; ~DescriptorPool() { + // Note that these have to manually destroyed because they were not explicitly ref-counted. + for (auto const& [mask, sets]: mUnused) { + for (auto set: sets) { + mAllocator->destruct(set); + } + } vkDestroyDescriptorPool(mDevice, mPool, VKALLOC); } + void disableRecycling() noexcept { + mDisableRecycling = true; + } + uint16_t const& capacity() { return mCapacity; } @@ -172,6 +183,9 @@ class DescriptorPool { Handle createSet(Bitmask const& layoutMask, VkDescriptorSet vkSet) { return mAllocator->initHandle(mAllocator, vkSet, [this, layoutMask, vkSet]() { + if (mDisableRecycling) { + return; + } // We are recycling - release the set back into the pool. Note that the // vk handle has not changed, but we need to change the backend handle to allow // for proper refcounting of resources referenced in this set. @@ -200,6 +214,8 @@ class DescriptorPool { using UnusedSetMap = std::unordered_map>, BitmaskHashFn, BitmaskEqual>; UnusedSetMap mUnused; + + bool mDisableRecycling; }; // This is an ever-expanding pool of sets where it @@ -244,6 +260,12 @@ class DescriptorInfinitePool { return ret; } + void disableRecycling() noexcept { + for (auto& pool: mPools) { + pool->disableRecycling(); + } + } + private: VkDevice mDevice; VulkanResourceAllocator* mAllocator; @@ -267,21 +289,19 @@ class LayoutCache { ~LayoutCache() { for (auto [key, layout]: mLayouts) { - auto layoutPtr = mAllocator->handle_cast(layout); - vkDestroyDescriptorSetLayout(mDevice, layoutPtr->vklayout, VKALLOC); + mAllocator->destruct(layout); } mLayouts.clear(); } void destroyLayout(Handle handle) { - auto layoutPtr = mAllocator->handle_cast(handle); for (auto [key, layout]: mLayouts) { if (layout == handle) { mLayouts.erase(key); break; } } - vkDestroyDescriptorSetLayout(mDevice, layoutPtr->vklayout, VKALLOC); + mAllocator->destruct(handle); } Handle getLayout(descset::DescriptorSetLayout const& layout) { @@ -339,10 +359,8 @@ class LayoutCache { .bindingCount = count, .pBindings = toBind, }; - - VkDescriptorSetLayout outLayout; - vkCreateDescriptorSetLayout(mDevice, &dlinfo, VKALLOC, &outLayout); - return (mLayouts[key] = mAllocator->initHandle(outLayout, key)); + return (mLayouts[key] = + mAllocator->initHandle(mDevice, dlinfo, key)); } private: @@ -416,7 +434,7 @@ struct SamplerKey { ret.sampler[count] = info.sampler; ret.imageView[count] = info.imageView; ret.imageLayout[count] = info.imageLayout; - }// else keep them as VK_NULL_HANDLEs. + } // else keep them as VK_NULL_HANDLEs. count++; } return ret; @@ -431,16 +449,16 @@ struct SamplerKey { struct InputAttachmentKey { // This count should be fixed. uint8_t count; - uint8_t padding[7]; - VkImageView view = VK_NULL_HANDLE; + uint8_t padding[3]; VkImageLayout imageLayout = VK_IMAGE_LAYOUT_UNDEFINED; + VkImageView view = VK_NULL_HANDLE; static inline InputAttachmentKey key(VkDescriptorImageInfo const& info, VulkanDescriptorSetLayout* layout) { return { .count = (uint8_t) layout->count.inputAttachment, - .view = info.imageView, .imageLayout = info.imageLayout, + .view = info.imageView, }; } @@ -627,29 +645,40 @@ class DescriptorSetCache { public: DescriptorSetCache(VkDevice device, VulkanResourceAllocator* allocator) : mAllocator(allocator), - mDescriptorPool(device, allocator), - mUBOCache(allocator), - mSamplerCache(allocator), - mInputAttachmentCache(allocator) {} + mDescriptorPool(std::make_unique(device, allocator)), + mUBOCache(std::make_unique>(allocator)), + mSamplerCache(std::make_unique>(allocator)), + mInputAttachmentCache( + std::make_unique>(allocator)) {} template inline std::pair get(Key const& key, VulkanDescriptorSetLayout* layout) { if constexpr (std::is_same_v) { - return get(key, mUBOCache, layout); + return get(key, *mUBOCache, layout); } else if constexpr (std::is_same_v) { - return get(key, mSamplerCache, layout); + return get(key, *mSamplerCache, layout); } else if constexpr (std::is_same_v) { - return get(key, mInputAttachmentCache, layout); + return get(key, *mInputAttachmentCache, layout); } PANIC_POSTCONDITION("Unexpected key type"); } + ~DescriptorSetCache() { + // This will prevent the descriptor sets recycling when we destroy descriptor set caches. + mDescriptorPool->disableRecycling(); + + mInputAttachmentCache.reset(); + mSamplerCache.reset(); + mUBOCache.reset(); + mDescriptorPool.reset(); + } + // gc() should be called at the end of everyframe void gc() { - mUBOCache.gc(); - mSamplerCache.gc(); - mInputAttachmentCache.gc(); + mUBOCache->gc(); + mSamplerCache->gc(); + mInputAttachmentCache->gc(); } private: @@ -660,16 +689,18 @@ class DescriptorSetCache { return {set, true}; } auto set = mAllocator->handle_cast( - mDescriptorPool.obtainSet(layout)); + mDescriptorPool->obtainSet(layout)); cache.put(key, set); return {set, false}; } VulkanResourceAllocator* mAllocator; - DescriptorInfinitePool mDescriptorPool; - LRUDescriptorSetCache mUBOCache; - LRUDescriptorSetCache mSamplerCache; - LRUDescriptorSetCache mInputAttachmentCache; + + // We need to heap-allocate so that the destruction can be strictly ordered. + std::unique_ptr mDescriptorPool; + std::unique_ptr> mUBOCache; + std::unique_ptr> mSamplerCache; + std::unique_ptr> mInputAttachmentCache; }; } // anonymous namespace @@ -746,25 +777,38 @@ class VulkanDescriptorSetManager::Impl { mLayoutStash[program] = layouts; } + VulkanDescriptorSetLayoutList outLayouts = layouts; DescriptorSetVkHandles vkDescSets = initDescSetHandles(); VkWriteDescriptorSet descriptorWrites[MAX_BINDINGS]; uint32_t nwrites = 0; + // Use placeholders when necessary for (uint8_t i = 0; i < VulkanDescriptorSetLayout::UNIQUE_DESCRIPTOR_SET_COUNT; ++i) { - auto handle = layouts[i]; - if (!handle) { - assert_invariant(i == INPUT_ATTACHMENT_SET_ID - && "Unexpectedly absent descriptor set layout"); - continue; - } - VulkanDescriptorSetLayout* layout - = mAllocator->handle_cast(handle); - if (!((i == UBO_SET_ID && layout->bitmask.ubo) - || (i == SAMPLER_SET_ID && layout->bitmask.sampler) - || (i == INPUT_ATTACHMENT_SET_ID && layout->bitmask.inputAttachment + if (!layouts[i]) { + if (i == INPUT_ATTACHMENT_SET_ID || + (i == SAMPLER_SET_ID && !layouts[INPUT_ATTACHMENT_SET_ID])) { + continue; + } + outLayouts[i] = getPlaceHolderLayout(i); + } else { + outLayouts[i] = layouts[i]; + auto p = mAllocator->handle_cast(layouts[i]); + if (!((i == UBO_SET_ID && p->bitmask.ubo) + || (i == SAMPLER_SET_ID && p->bitmask.sampler) + || (i == INPUT_ATTACHMENT_SET_ID && p->bitmask.inputAttachment && mInputAttachment.first.texture))) { + outLayouts[i] = getPlaceHolderLayout(i); + } + } + } + + for (uint8_t i = 0; i < VulkanDescriptorSetLayout::UNIQUE_DESCRIPTOR_SET_COUNT; ++i) { + if (!outLayouts[i]) { continue; } + VulkanDescriptorSetLayout* layout + = mAllocator->handle_cast(outLayouts[i]); + bool const usePlaceholder = layouts[i] != outLayouts[i]; auto const& [set, cached] = getSet(i, layout); VkDescriptorSet const vkSet = set->vkSet; @@ -773,7 +817,8 @@ class VulkanDescriptorSetManager::Impl { // Note that we still need to bind the set, but 'cached' means that we found a set with // the exact same content already written, and we would just bind that one instead. - if (cached) { + // We also don't need to write to the placeholder set. + if (cached || usePlaceholder) { continue; } @@ -836,7 +881,7 @@ class VulkanDescriptorSetManager::Impl { vkUpdateDescriptorSets(mDevice, nwrites, descriptorWrites, 0, nullptr); } - VkPipelineLayout const pipelineLayout = getPipelineLayoutFn(layouts); + VkPipelineLayout const pipelineLayout = getPipelineLayoutFn(outLayouts); VkCommandBuffer const cmdbuffer = commands->buffer(); BoundState state{}; @@ -918,6 +963,10 @@ class VulkanDescriptorSetManager::Impl { FVK_SYSTRACE_END(); } + void clearProgram(VulkanProgram* program) noexcept { + mLayoutStash.erase(program); + } + Handle createLayout( descset::DescriptorSetLayout const& description) { return mLayoutCache.getLayout(description); @@ -1031,25 +1080,62 @@ class VulkanDescriptorSetManager::Impl { } } + inline Handle getPlaceHolderLayout(uint8_t setID) { + if (mPlaceholderLayout[setID]) { + return mPlaceholderLayout[setID]; + } + descset::DescriptorSetLayout inputLayout { + .bindings = {{}}, + }; + switch (setID) { + case UBO_SET_ID: + inputLayout.bindings[0] = { + .type = descset::DescriptorType::UNIFORM_BUFFER, + .stageFlags = descset::ShaderStageFlags2::VERTEX, + .binding = 0, + .flags = descset::DescriptorFlags::NONE, + .count = 0, + }; + break; + case SAMPLER_SET_ID: + inputLayout.bindings[0] = { + .type = descset::DescriptorType::SAMPLER, + .stageFlags = descset::ShaderStageFlags2::FRAGMENT, + .binding = 0, + .flags = descset::DescriptorFlags::NONE, + .count = 0, + }; + break; + case INPUT_ATTACHMENT_SET_ID: + inputLayout.bindings[0] = { + .type = descset::DescriptorType::INPUT_ATTACHMENT, + .stageFlags = descset::ShaderStageFlags2::FRAGMENT, + .binding = 0, + .flags = descset::DescriptorFlags::NONE, + .count = 0, + }; + break; + default: + PANIC_POSTCONDITION("Unexpected set id=%d", setID); + } + mPlaceholderLayout[setID] = mLayoutCache.getLayout(inputLayout); + return mPlaceholderLayout[setID]; + } + VkDevice mDevice; VulkanResourceAllocator* mAllocator; LayoutCache mLayoutCache; DescriptorSetCache mDescriptorSetCache; - bool mHaveDynamicUbos; - UBOMap mUboMap; SamplerMap mSamplerMap; std::pair mInputAttachment; - VulkanResourceManager mResources; - VkDescriptorBufferInfo mPlaceHolderBufferInfo; VkDescriptorImageInfo mPlaceHolderImageInfo; - std::unordered_map mLayoutStash; - BoundState mBoundState; + VulkanDescriptorSetLayoutList mPlaceholderLayout = {}; }; VulkanDescriptorSetManager::VulkanDescriptorSetManager(VkDevice device, @@ -1077,6 +1163,10 @@ void VulkanDescriptorSetManager::dynamicBind(VulkanCommandBuffer* commands, mImpl->dynamicBind(commands, uboLayout); } +void VulkanDescriptorSetManager::clearProgram(VulkanProgram* program) noexcept { + mImpl->clearProgram(program); +} + Handle VulkanDescriptorSetManager::createLayout( descset::DescriptorSetLayout const& layout) { return mImpl->createLayout(layout); diff --git a/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.h b/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.h index f8871a57aec..2fa0b020fe1 100644 --- a/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.h +++ b/filament/backend/src/vulkan/caching/VulkanDescriptorSetManager.h @@ -69,6 +69,10 @@ class VulkanDescriptorSetManager { // proper dynamic binding when Filament-side descriptor changes are completed. void dynamicBind(VulkanCommandBuffer* commands, Handle uboLayout); + // TODO: Obsolete after [GDSR]. + // Since we use program pointer as cache key, we need to clear the cache when it's freed. + void clearProgram(VulkanProgram* program) noexcept; + Handle createLayout(descset::DescriptorSetLayout const& layout); void destroyLayout(Handle layout); diff --git a/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.cpp b/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.cpp index e1d3b0be9c6..ec96418aaf1 100644 --- a/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.cpp +++ b/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.cpp @@ -24,7 +24,7 @@ VkPipelineLayout VulkanPipelineLayoutCache::getLayout( VulkanDescriptorSetLayoutList const& descriptorSetLayouts) { PipelineLayoutKey key = {VK_NULL_HANDLE}; uint8_t descSetLayoutCount = 0; - for (auto layoutHandle : descriptorSetLayouts) { + for (auto layoutHandle: descriptorSetLayouts) { if (layoutHandle) { auto layout = mAllocator->handle_cast(layoutHandle); key[descSetLayoutCount++] = layout->vklayout; @@ -55,4 +55,10 @@ VkPipelineLayout VulkanPipelineLayoutCache::getLayout( return layout; } +void VulkanPipelineLayoutCache::terminate() noexcept { + for (auto const& [key, entry]: mPipelineLayouts) { + vkDestroyPipelineLayout(mDevice, entry.handle, VKALLOC); + } +} + }// namespace filament::backend diff --git a/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.h b/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.h index fff9fd2d227..375e6124d23 100644 --- a/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.h +++ b/filament/backend/src/vulkan/caching/VulkanPipelineLayoutCache.h @@ -33,6 +33,8 @@ class VulkanPipelineLayoutCache { mAllocator(allocator), mTimestamp(0) {} + void terminate() noexcept; + using PipelineLayoutKey = std::array; diff --git a/filament/backend/test/ShaderGenerator.cpp b/filament/backend/test/ShaderGenerator.cpp index 8e329489dde..7d0ad716e1d 100644 --- a/filament/backend/test/ShaderGenerator.cpp +++ b/filament/backend/test/ShaderGenerator.cpp @@ -79,6 +79,20 @@ ShaderGenerator::ShaderGenerator(std::string vertex, std::string fragment, mVertexBlob(transpileShader(ShaderStage::VERTEX, std::move(vertex), backend, isMobile, sib)), mFragmentBlob(transpileShader(ShaderStage::FRAGMENT, std::move(fragment), backend, isMobile, sib)) { + switch (backend) { + case Backend::OPENGL: + mShaderLanguage = filament::backend::ShaderLanguage::ESSL3; + break; + case Backend::VULKAN: + mShaderLanguage = filament::backend::ShaderLanguage::SPIRV; + break; + case Backend::METAL: + mShaderLanguage = filament::backend::ShaderLanguage::MSL; + break; + case Backend::NOOP: + mShaderLanguage = filament::backend::ShaderLanguage::ESSL3; + break; + } } ShaderGenerator::Blob ShaderGenerator::transpileShader( @@ -160,6 +174,7 @@ ShaderGenerator::Blob ShaderGenerator::transpileShader( Program ShaderGenerator::getProgram(filament::backend::DriverApi&) noexcept { Program program; + program.shaderLanguage(mShaderLanguage); program.shader(ShaderStage::VERTEX, mVertexBlob.data(), mVertexBlob.size()); program.shader(ShaderStage::FRAGMENT, mFragmentBlob.data(), mFragmentBlob.size()); return program; diff --git a/filament/backend/test/ShaderGenerator.h b/filament/backend/test/ShaderGenerator.h index 19629986598..39a87a7f208 100644 --- a/filament/backend/test/ShaderGenerator.h +++ b/filament/backend/test/ShaderGenerator.h @@ -60,6 +60,7 @@ class ShaderGenerator { Blob mFragmentBlob; std::string mCompiledVertexShader; std::string mCompiledFragmentShader; + filament::backend::ShaderLanguage mShaderLanguage; }; diff --git a/filament/backend/test/test_ComputeBasic.cpp b/filament/backend/test/test_ComputeBasic.cpp index c0475908dfd..70622595592 100644 --- a/filament/backend/test/test_ComputeBasic.cpp +++ b/filament/backend/test/test_ComputeBasic.cpp @@ -58,6 +58,7 @@ kernel void main0() {} } Program program; + program.shaderLanguage(ShaderLanguage::ESSL3); program.shader(ShaderStage::COMPUTE, shader.data(), shader.size() + 1); Handle ph = driver.createProgram(std::move(program)); @@ -144,6 +145,7 @@ kernel void main0(device Output_data& output_data [[buffer(0)]], driver.updateBufferObject(input_data, { data.data(), size }, 0); Program program; + program.shaderLanguage(ShaderLanguage::ESSL3); program.shader(ShaderStage::COMPUTE, shader.data(), shader.size() + 1); Handle ph = driver.createProgram(std::move(program)); diff --git a/filament/src/MaterialParser.cpp b/filament/src/MaterialParser.cpp index 5427835a282..b5aeda47cde 100644 --- a/filament/src/MaterialParser.cpp +++ b/filament/src/MaterialParser.cpp @@ -34,6 +34,7 @@ #include #include +#include using namespace utils; using namespace filament::backend; @@ -42,32 +43,31 @@ using namespace filamat; namespace filament { -// ------------------------------------------------------------------------------------------------ - -MaterialParser::MaterialParserDetails::MaterialParserDetails(ShaderLanguage language, const void* data, size_t size) - : mManagedBuffer(data, size), - mChunkContainer(mManagedBuffer.data(), mManagedBuffer.size()), - mMaterialChunk(mChunkContainer) { +constexpr std::pair shaderLanguageToTags(ShaderLanguage language) { switch (language) { case ShaderLanguage::ESSL3: - mMaterialTag = ChunkType::MaterialGlsl; - mDictionaryTag = ChunkType::DictionaryText; - break; + return { ChunkType::MaterialGlsl, ChunkType::DictionaryText }; case ShaderLanguage::ESSL1: - mMaterialTag = ChunkType::MaterialEssl1; - mDictionaryTag = ChunkType::DictionaryText; - break; + return { ChunkType::MaterialEssl1, ChunkType::DictionaryText }; case ShaderLanguage::MSL: - mMaterialTag = ChunkType::MaterialMetal; - mDictionaryTag = ChunkType::DictionaryText; - break; + return { ChunkType::MaterialMetal, ChunkType::DictionaryText }; case ShaderLanguage::SPIRV: - mMaterialTag = ChunkType::MaterialSpirv; - mDictionaryTag = ChunkType::DictionarySpirv; - break; + return { ChunkType::MaterialSpirv, ChunkType::DictionarySpirv }; + case ShaderLanguage::METAL_LIBRARY: + return { ChunkType::MaterialMetalLibrary, ChunkType::DictionaryMetalLibrary }; } } +// ------------------------------------------------------------------------------------------------ + +MaterialParser::MaterialParserDetails::MaterialParserDetails( + const utils::FixedCapacityVector& preferredLanguages, const void* data, + size_t size) + : mManagedBuffer(data, size), + mChunkContainer(mManagedBuffer.data(), mManagedBuffer.size()), + mPreferredLanguages(preferredLanguages), + mMaterialChunk(mChunkContainer) {} + template UTILS_NOINLINE bool MaterialParser::MaterialParserDetails::getFromSimpleChunk( @@ -83,9 +83,9 @@ bool MaterialParser::MaterialParserDetails::getFromSimpleChunk( // ------------------------------------------------------------------------------------------------ -MaterialParser::MaterialParser(ShaderLanguage language, const void* data, size_t size) - : mImpl(language, data, size) { -} +MaterialParser::MaterialParser(utils::FixedCapacityVector preferredLanguages, + const void* data, size_t size) + : mImpl(preferredLanguages, data, size) {} ChunkContainer& MaterialParser::getChunkContainer() noexcept { return mImpl.mChunkContainer; @@ -100,20 +100,40 @@ MaterialParser::ParseResult MaterialParser::parse() noexcept { if (UTILS_UNLIKELY(!cc.parse())) { return ParseResult::ERROR_OTHER; } - const ChunkType matTag = mImpl.mMaterialTag; - const ChunkType dictTag = mImpl.mDictionaryTag; - if (UTILS_UNLIKELY(!cc.hasChunk(matTag) || !cc.hasChunk(dictTag))) { + + using MaybeShaderLanguageAndChunks = + std::optional>; + auto chooseLanguage = [this, &cc]() -> MaybeShaderLanguageAndChunks { + for (auto language : mImpl.mPreferredLanguages) { + const auto [matTag, dictTag] = shaderLanguageToTags(language); + if (cc.hasChunk(matTag) && cc.hasChunk(dictTag)) { + return std::make_tuple(language, matTag, dictTag); + } + } + return {}; + }; + const auto result = chooseLanguage(); + + if (!result.has_value()) { return ParseResult::ERROR_MISSING_BACKEND; } + + const auto [chosenLanguage, matTag, dictTag] = result.value(); if (UTILS_UNLIKELY(!DictionaryReader::unflatten(cc, dictTag, mImpl.mBlobDictionary))) { return ParseResult::ERROR_OTHER; } if (UTILS_UNLIKELY(!mImpl.mMaterialChunk.initialize(matTag))) { return ParseResult::ERROR_OTHER; } + + mImpl.mChosenLanguage = chosenLanguage; return ParseResult::SUCCESS; } +backend::ShaderLanguage MaterialParser::getShaderLanguage() const noexcept { + return mImpl.mChosenLanguage; +} + // Accessors bool MaterialParser::getMaterialVersion(uint32_t* value) const noexcept { return mImpl.getFromSimpleChunk(ChunkType::MaterialVersion, value); diff --git a/filament/src/MaterialParser.h b/filament/src/MaterialParser.h index 638e93d5662..2ec7f7c1cc4 100644 --- a/filament/src/MaterialParser.h +++ b/filament/src/MaterialParser.h @@ -29,10 +29,12 @@ #include #include -#include #include +#include +#include #include +#include namespace filaflat { class ChunkContainer; @@ -48,7 +50,8 @@ struct MaterialConstant; class MaterialParser { public: - MaterialParser(backend::ShaderLanguage language, const void* data, size_t size); + MaterialParser(utils::FixedCapacityVector preferredLanguages, + const void* data, size_t size); MaterialParser(MaterialParser const& rhs) noexcept = delete; MaterialParser& operator=(MaterialParser const& rhs) noexcept = delete; @@ -60,6 +63,7 @@ class MaterialParser { }; ParseResult parse() noexcept; + backend::ShaderLanguage getShaderLanguage() const noexcept; // Accessors bool getMaterialVersion(uint32_t* value) const noexcept; @@ -130,7 +134,9 @@ class MaterialParser { private: struct MaterialParserDetails { - MaterialParserDetails(backend::ShaderLanguage language, const void* data, size_t size); + MaterialParserDetails( + const utils::FixedCapacityVector& preferredLanguages, + const void* data, size_t size); template bool getFromSimpleChunk(filamat::ChunkType type, T* value) const noexcept; @@ -157,12 +163,12 @@ class MaterialParser { ManagedBuffer mManagedBuffer; filaflat::ChunkContainer mChunkContainer; + utils::FixedCapacityVector mPreferredLanguages; + backend::ShaderLanguage mChosenLanguage; // Keep MaterialChunk alive between calls to getShader to avoid reload the shader index. filaflat::MaterialChunk mMaterialChunk; filaflat::BlobDictionary mBlobDictionary; - filamat::ChunkType mMaterialTag = filamat::ChunkType::Unknown; - filamat::ChunkType mDictionaryTag = filamat::ChunkType::Unknown; }; filaflat::ChunkContainer& getChunkContainer() noexcept; diff --git a/filament/src/details/Engine.cpp b/filament/src/details/Engine.cpp index ac1b3f92b9a..fc26d745fea 100644 --- a/filament/src/details/Engine.cpp +++ b/filament/src/details/Engine.cpp @@ -682,18 +682,7 @@ int FEngine::loop() { return 0; } - // Set thread affinity for the backend thread. - // see https://developer.android.com/agi/sys-trace/threads-scheduling#cpu_core_affinity - // Certain backends already have some threads pinned, and we can't easily know on which core. - const bool disableThreadAffinity - = mDriver->isWorkaroundNeeded(Workaround::DISABLE_THREAD_AFFINITY); - - uint32_t const id = std::thread::hardware_concurrency() - 1; while (true) { - // looks like thread affinity needs to be reset regularly (on Android) - if (!disableThreadAffinity) { - JobSystem::setThreadAffinityById(id); - } if (!execute()) { break; } diff --git a/filament/src/details/Engine.h b/filament/src/details/Engine.h index b467013b995..3d02cbbd5c3 100644 --- a/filament/src/details/Engine.h +++ b/filament/src/details/Engine.h @@ -59,10 +59,11 @@ #include #include -#include #include -#include #include +#include +#include +#include #include #include @@ -231,19 +232,21 @@ class FEngine : public Engine { return mPlatform; } - backend::ShaderLanguage getShaderLanguage() const noexcept { + // Return a vector of shader languages, in order of preference. + utils::FixedCapacityVector getShaderLanguage() const noexcept { switch (mBackend) { case Backend::DEFAULT: case Backend::NOOP: default: - return backend::ShaderLanguage::ESSL3; + return { backend::ShaderLanguage::ESSL3 }; case Backend::OPENGL: - return getDriver().getFeatureLevel() == FeatureLevel::FEATURE_LEVEL_0 - ? backend::ShaderLanguage::ESSL1 : backend::ShaderLanguage::ESSL3; + return { getDriver().getFeatureLevel() == FeatureLevel::FEATURE_LEVEL_0 + ? backend::ShaderLanguage::ESSL1 + : backend::ShaderLanguage::ESSL3 }; case Backend::VULKAN: - return backend::ShaderLanguage::SPIRV; + return { backend::ShaderLanguage::SPIRV }; case Backend::METAL: - return backend::ShaderLanguage::MSL; + return { backend::ShaderLanguage::METAL_LIBRARY, backend::ShaderLanguage::MSL }; } } diff --git a/filament/src/details/Material.cpp b/filament/src/details/Material.cpp index dbb4339f465..b5597a06e7c 100644 --- a/filament/src/details/Material.cpp +++ b/filament/src/details/Material.cpp @@ -42,16 +42,16 @@ #include #include -#include -#include -#include #include #include +#include #include #include -#include #include -#include +#include +#include +#include +#include #include #include @@ -60,6 +60,7 @@ #include #include #include +#include #include #include #include @@ -74,22 +75,32 @@ using namespace backend; using namespace filaflat; using namespace utils; -static std::unique_ptr createParser( - Backend backend, ShaderLanguage language, const void* data, size_t size) { +static std::unique_ptr createParser(Backend backend, + utils::FixedCapacityVector languages, const void* data, size_t size) { // unique_ptr so we don't leak MaterialParser on failures below - auto materialParser = std::make_unique(language, data, size); + auto materialParser = std::make_unique(languages, data, size); MaterialParser::ParseResult const materialResult = materialParser->parse(); + if (UTILS_UNLIKELY(materialResult == MaterialParser::ParseResult::ERROR_MISSING_BACKEND)) { + std::stringstream languageNames; + for (auto it = languages.begin(); it != languages.end(); ++it) { + languageNames << shaderLanguageToString(*it); + if (std::next(it) != languages.end()) { + languageNames << ", "; + } + } + + ASSERT_PRECONDITION(materialResult != MaterialParser::ParseResult::ERROR_MISSING_BACKEND, + "the material was not built for any of the %s backend's supported shader " + "languages (%s)\n", + backendToString(backend), languageNames.str().c_str()); + } + if (backend == Backend::NOOP) { return materialParser; } - ASSERT_PRECONDITION(materialResult != MaterialParser::ParseResult::ERROR_MISSING_BACKEND, - "the material was not built for the %s backend and %s shader language\n", - backendToString(backend), - shaderLanguageToString(language)); - ASSERT_PRECONDITION(materialResult == MaterialParser::ParseResult::SUCCESS, "could not parse the material package"); @@ -179,7 +190,8 @@ Material* Material::Builder::build(Engine& engine) { FMaterial::FMaterial(FEngine& engine, const Material::Builder& builder, std::unique_ptr materialParser) - : mEngine(engine), + : mIsDefaultMaterial(builder->mDefaultMaterial), + mEngine(engine), mMaterialId(engine.getMaterialId()), mMaterialParser(std::move(materialParser)) { @@ -216,7 +228,7 @@ FMaterial::FMaterial(FEngine& engine, const Material::Builder& builder, success = parser->getUIB(&mUniformInterfaceBlock); assert_invariant(success); - if (UTILS_UNLIKELY(engine.getShaderLanguage() == ShaderLanguage::ESSL1)) { + if (UTILS_UNLIKELY(parser->getShaderLanguage() == ShaderLanguage::ESSL1)) { success = parser->getBindingUniformInfo(&mBindingUniformInfo); assert_invariant(success); @@ -232,111 +244,13 @@ FMaterial::FMaterial(FEngine& engine, const Material::Builder& builder, &mSamplerGroupBindingInfoList, &mSamplerBindingToNameMap); assert_invariant(success); -#if FILAMENT_ENABLE_MATDBG - // Register the material with matdbg. - matdbg::DebugServer* server = downcast(engine).debug.server; - if (UTILS_UNLIKELY(server)) { - auto details = builder.mImpl; - mDebuggerId = server->addMaterial(mName, details->mPayload, details->mSize, this); - } -#endif - // Older materials will not have a subpass chunk; this should not be an error. if (!parser->getSubpasses(&mSubpassInfo)) { mSubpassInfo.isValid = false; } - // Older materials won't have a constants chunk, but that's okay. - parser->getConstants(&mMaterialConstants); - for (size_t i = 0, c = mMaterialConstants.size(); i < c; i++) { - auto& item = mMaterialConstants[i]; - // the key can be a string_view because mMaterialConstant owns the CString - std::string_view const key{ item.name.data(), item.name.size() }; - mSpecializationConstantsNameToIndex[key] = i; - } - - // Verify that all the constant specializations exist in the material and that their types match. - // The first specialization constants are defined internally by Filament. - // The subsequent constants are user-defined in the material. - - // Feature level 0 doesn't support instancing - int const maxInstanceCount = (engine.getActiveFeatureLevel() == FeatureLevel::FEATURE_LEVEL_0) - ? 1 : CONFIG_MAX_INSTANCES; - - int const maxFroxelBufferHeight = (int)std::min( - FROXEL_BUFFER_MAX_ENTRY_COUNT / 4, - engine.getDriverApi().getMaxUniformBufferSize() / 16u); - - bool const staticTextureWorkaround = - engine.getDriverApi().isWorkaroundNeeded(Workaround::A8X_STATIC_TEXTURE_TARGET_ERROR); - - bool const powerVrShaderWorkarounds = - engine.getDriverApi().isWorkaroundNeeded(Workaround::POWER_VR_SHADER_WORKAROUNDS); - - mSpecializationConstants.reserve(mMaterialConstants.size() + CONFIG_MAX_RESERVED_SPEC_CONSTANTS); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::BACKEND_FEATURE_LEVEL, - (int)engine.getSupportedFeatureLevel() }); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_MAX_INSTANCES, - (int)maxInstanceCount }); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_FROXEL_BUFFER_HEIGHT, - (int)maxFroxelBufferHeight }); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_DEBUG_DIRECTIONAL_SHADOWMAP, - engine.debug.shadowmap.debug_directional_shadowmap }); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_DEBUG_FROXEL_VISUALIZATION, - engine.debug.lighting.debug_froxel_visualization }); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_STATIC_TEXTURE_TARGET_WORKAROUND, - staticTextureWorkaround }); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_POWER_VR_SHADER_WORKAROUNDS, - powerVrShaderWorkarounds }); - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_STEREO_EYE_COUNT, - (int)engine.getConfig().stereoscopicEyeCount }); - if (UTILS_UNLIKELY(engine.getShaderLanguage() == ShaderLanguage::ESSL1)) { - // The actual value of this spec-constant is set in the OpenGLDriver backend. - mSpecializationConstants.push_back({ - +ReservedSpecializationConstants::CONFIG_SRGB_SWAPCHAIN_EMULATION, - false}); - } - - for (auto const& [name, value] : builder->mConstantSpecializations) { - std::string_view const key{ name.data(), name.size() }; - auto pos = mSpecializationConstantsNameToIndex.find(key); - ASSERT_PRECONDITION(pos != mSpecializationConstantsNameToIndex.end(), - "The material %s does not have a constant parameter named %s.", - mName.c_str_safe(), name.c_str()); - const char* const types[3] = {"an int", "a float", "a bool"}; - const char* const errorMessage = - "The constant parameter %s on material %s is of type %s, but %s was " - "provided."; - auto& constant = mMaterialConstants[pos->second]; - switch (constant.type) { - case ConstantType::INT: - ASSERT_PRECONDITION(std::holds_alternative(value), errorMessage, - name.c_str(), mName.c_str_safe(), "int", types[value.index()]); - break; - case ConstantType::FLOAT: - ASSERT_PRECONDITION(std::holds_alternative(value), errorMessage, - name.c_str(), mName.c_str_safe(), "float", types[value.index()]); - break; - case ConstantType::BOOL: - ASSERT_PRECONDITION(std::holds_alternative(value), errorMessage, - name.c_str(), mName.c_str_safe(), "bool", types[value.index()]); - break; - } - uint32_t const index = pos->second + CONFIG_MAX_RESERVED_SPEC_CONSTANTS; - mSpecializationConstants.push_back({ index, value }); - } - parser->getShading(&mShading); parser->getMaterialProperties(&mMaterialProperties); - parser->getBlendingMode(&mBlendingMode); parser->getInterpolation(&mInterpolation); parser->getVertexDomain(&mVertexDomain); parser->getMaterialDomain(&mMaterialDomain); @@ -345,14 +259,9 @@ FMaterial::FMaterial(FEngine& engine, const Material::Builder& builder, parser->getRefractionMode(&mRefractionMode); parser->getRefractionType(&mRefractionType); parser->getReflectionMode(&mReflectionMode); - - if (mBlendingMode == BlendingMode::MASKED) { - parser->getMaskThreshold(&mMaskThreshold); - } - - if (mBlendingMode == BlendingMode::CUSTOM) { - parser->getCustomBlendFunction(&mCustomBlendFunctions); - } + parser->getTransparencyMode(&mTransparencyMode); + parser->getDoubleSided(&mDoubleSided); + parser->getCullingMode(&mCullingMode); if (mShading == Shading::UNLIT) { parser->hasShadowMultiplier(&mHasShadowMultiplier); @@ -360,74 +269,19 @@ FMaterial::FMaterial(FEngine& engine, const Material::Builder& builder, mIsVariantLit = mShading != Shading::UNLIT || mHasShadowMultiplier; - // create raster state - using BlendFunction = RasterState::BlendFunction; - using DepthFunc = RasterState::DepthFunc; - switch (mBlendingMode) { - // Do not change the MASKED behavior without checking for regressions with - // AlphaBlendModeTest and TextureLinearInterpolationTest, with and without - // View::BlendMode::TRANSLUCENT. - case BlendingMode::MASKED: - case BlendingMode::OPAQUE: - mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; - mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; - mRasterState.blendFunctionDstRGB = BlendFunction::ZERO; - mRasterState.blendFunctionDstAlpha = BlendFunction::ZERO; - mRasterState.depthWrite = true; - break; - case BlendingMode::TRANSPARENT: - case BlendingMode::FADE: - mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; - mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; - mRasterState.blendFunctionDstRGB = BlendFunction::ONE_MINUS_SRC_ALPHA; - mRasterState.blendFunctionDstAlpha = BlendFunction::ONE_MINUS_SRC_ALPHA; - mRasterState.depthWrite = false; - break; - case BlendingMode::ADD: - mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; - mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; - mRasterState.blendFunctionDstRGB = BlendFunction::ONE; - mRasterState.blendFunctionDstAlpha = BlendFunction::ONE; - mRasterState.depthWrite = false; - break; - case BlendingMode::MULTIPLY: - mRasterState.blendFunctionSrcRGB = BlendFunction::ZERO; - mRasterState.blendFunctionSrcAlpha = BlendFunction::ZERO; - mRasterState.blendFunctionDstRGB = BlendFunction::SRC_COLOR; - mRasterState.blendFunctionDstAlpha = BlendFunction::SRC_COLOR; - mRasterState.depthWrite = false; - break; - case BlendingMode::SCREEN: - mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; - mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; - mRasterState.blendFunctionDstRGB = BlendFunction::ONE_MINUS_SRC_COLOR; - mRasterState.blendFunctionDstAlpha = BlendFunction::ONE_MINUS_SRC_COLOR; - mRasterState.depthWrite = false; - break; - case BlendingMode::CUSTOM: - mRasterState.blendFunctionSrcRGB = mCustomBlendFunctions[0]; - mRasterState.blendFunctionSrcAlpha = mCustomBlendFunctions[1]; - mRasterState.blendFunctionDstRGB = mCustomBlendFunctions[2]; - mRasterState.blendFunctionDstAlpha = mCustomBlendFunctions[3]; - mRasterState.depthWrite = false; - } + // color write + bool colorWrite = false; + parser->getColorWrite(&colorWrite); + mRasterState.colorWrite = colorWrite; - bool depthWriteSet = false; - parser->getDepthWriteSet(&depthWriteSet); - if (depthWriteSet) { - bool depthWrite = false; - parser->getDepthWrite(&depthWrite); - mRasterState.depthWrite = depthWrite; - } + // depth test + bool depthTest = false; + parser->getDepthTest(&depthTest); + mRasterState.depthFunc = depthTest ? RasterState::DepthFunc::GE : RasterState::DepthFunc::A; // if doubleSided() was called we override culling() bool doubleSideSet = false; parser->getDoubleSidedSet(&doubleSideSet); - parser->getDoubleSided(&mDoubleSided); - parser->getCullingMode(&mCullingMode); - bool depthTest = false; - parser->getDepthTest(&depthTest); - if (doubleSideSet) { mDoubleSidedCapability = true; mRasterState.culling = mDoubleSided ? CullingMode::NONE : mCullingMode; @@ -435,61 +289,29 @@ FMaterial::FMaterial(FEngine& engine, const Material::Builder& builder, mRasterState.culling = mCullingMode; } - parser->getTransparencyMode(&mTransparencyMode); - parser->hasCustomDepthShader(&mHasCustomDepthShader); - mIsDefaultMaterial = builder->mDefaultMaterial; - - if (UTILS_UNLIKELY(mIsDefaultMaterial)) { - assert_invariant(mMaterialDomain == MaterialDomain::SURFACE); - filaflat::MaterialChunk const& materialChunk{ mMaterialParser->getMaterialChunk() }; - auto variants = FixedCapacityVector::with_capacity(materialChunk.getShaderCount()); - materialChunk.visitShaders([&variants]( - ShaderModel, Variant variant, ShaderStage) { - if (Variant::isValidDepthVariant(variant)) { - variants.push_back(variant); - } - }); - std::sort(variants.begin(), variants.end(), - [](Variant lhs, Variant rhs) { return lhs.key < rhs.key; }); - auto pos = std::unique(variants.begin(), variants.end()); - variants.resize(std::distance(variants.begin(), pos)); - std::swap(mDepthVariants, variants); - } - - if (mMaterialDomain == MaterialDomain::SURFACE) { - if (UTILS_UNLIKELY(!mIsDefaultMaterial && !mHasCustomDepthShader)) { - FMaterial const* const pDefaultMaterial = engine.getDefaultMaterial(); - auto& cachedPrograms = mCachedPrograms; - for (Variant const variant: pDefaultMaterial->mDepthVariants) { - pDefaultMaterial->prepareProgram(variant); - cachedPrograms[variant.key] = pDefaultMaterial->getProgram(variant); - } - } - } - - bool colorWrite = false; - parser->getColorWrite(&colorWrite); - mRasterState.colorWrite = colorWrite; - mRasterState.depthFunc = depthTest ? DepthFunc::GE : DepthFunc::A; - - bool alphaToCoverageSet = false; - parser->getAlphaToCoverageSet(&alphaToCoverageSet); - if (alphaToCoverageSet) { - bool alphaToCoverage = false; - parser->getAlphaToCoverage(&alphaToCoverage); - mRasterState.alphaToCoverage = alphaToCoverage; - } else { - mRasterState.alphaToCoverage = mBlendingMode == BlendingMode::MASKED; - } - + // specular anti-aliasing parser->hasSpecularAntiAliasing(&mSpecularAntiAliasing); if (mSpecularAntiAliasing) { parser->getSpecularAntiAliasingVariance(&mSpecularAntiAliasingVariance); parser->getSpecularAntiAliasingThreshold(&mSpecularAntiAliasingThreshold); } + processBlendingMode(parser); + processSpecializationConstants(engine, builder, parser); + processDepthVariants(engine, parser); + // we can only initialize the default instance once we're initialized ourselves mDefaultInstance.initDefaultInstance(engine, this); + + +#if FILAMENT_ENABLE_MATDBG + // Register the material with matdbg. + matdbg::DebugServer* server = downcast(engine).debug.server; + if (UTILS_UNLIKELY(server)) { + auto details = builder.mImpl; + mDebuggerId = server->addMaterial(mName, details->mPayload, details->mSize, this); + } +#endif } FMaterial::~FMaterial() noexcept = default; @@ -719,6 +541,7 @@ Program FMaterial::getProgramWithVariants( Program program; program.shader(ShaderStage::VERTEX, vsBuilder.data(), vsBuilder.size()) .shader(ShaderStage::FRAGMENT, fsBuilder.data(), fsBuilder.size()) + .shaderLanguage(mMaterialParser->getShaderLanguage()) .uniformBlockBindings(mUniformBlockBindings) .diagnostics(mName, [this, variant](io::ostream& out) -> io::ostream& { @@ -740,8 +563,7 @@ Program FMaterial::getProgramWithVariants( samplers.data(), info.count); } } - - if (UTILS_UNLIKELY(mEngine.getShaderLanguage() == ShaderLanguage::ESSL1)) { + if (UTILS_UNLIKELY(mMaterialParser->getShaderLanguage() == ShaderLanguage::ESSL1)) { assert_invariant(!mBindingUniformInfo.empty()); for (auto const& [index, uniforms] : mBindingUniformInfo) { program.uniforms(uint32_t(index), uniforms); @@ -914,6 +736,211 @@ bool FMaterial::setConstant(uint32_t id, T value) noexcept { return false; } +void FMaterial::processBlendingMode(MaterialParser const* const parser) { + parser->getBlendingMode(&mBlendingMode); + + if (mBlendingMode == BlendingMode::MASKED) { + parser->getMaskThreshold(&mMaskThreshold); + } + + if (mBlendingMode == BlendingMode::CUSTOM) { + parser->getCustomBlendFunction(&mCustomBlendFunctions); + } + + // blending mode + switch (mBlendingMode) { + // Do not change the MASKED behavior without checking for regressions with + // AlphaBlendModeTest and TextureLinearInterpolationTest, with and without + // View::BlendMode::TRANSLUCENT. + case BlendingMode::MASKED: + case BlendingMode::OPAQUE: + mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; + mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; + mRasterState.blendFunctionDstRGB = BlendFunction::ZERO; + mRasterState.blendFunctionDstAlpha = BlendFunction::ZERO; + mRasterState.depthWrite = true; + break; + case BlendingMode::TRANSPARENT: + case BlendingMode::FADE: + mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; + mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; + mRasterState.blendFunctionDstRGB = BlendFunction::ONE_MINUS_SRC_ALPHA; + mRasterState.blendFunctionDstAlpha = BlendFunction::ONE_MINUS_SRC_ALPHA; + mRasterState.depthWrite = false; + break; + case BlendingMode::ADD: + mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; + mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; + mRasterState.blendFunctionDstRGB = BlendFunction::ONE; + mRasterState.blendFunctionDstAlpha = BlendFunction::ONE; + mRasterState.depthWrite = false; + break; + case BlendingMode::MULTIPLY: + mRasterState.blendFunctionSrcRGB = BlendFunction::ZERO; + mRasterState.blendFunctionSrcAlpha = BlendFunction::ZERO; + mRasterState.blendFunctionDstRGB = BlendFunction::SRC_COLOR; + mRasterState.blendFunctionDstAlpha = BlendFunction::SRC_COLOR; + mRasterState.depthWrite = false; + break; + case BlendingMode::SCREEN: + mRasterState.blendFunctionSrcRGB = BlendFunction::ONE; + mRasterState.blendFunctionSrcAlpha = BlendFunction::ONE; + mRasterState.blendFunctionDstRGB = BlendFunction::ONE_MINUS_SRC_COLOR; + mRasterState.blendFunctionDstAlpha = BlendFunction::ONE_MINUS_SRC_COLOR; + mRasterState.depthWrite = false; + break; + case BlendingMode::CUSTOM: + mRasterState.blendFunctionSrcRGB = mCustomBlendFunctions[0]; + mRasterState.blendFunctionSrcAlpha = mCustomBlendFunctions[1]; + mRasterState.blendFunctionDstRGB = mCustomBlendFunctions[2]; + mRasterState.blendFunctionDstAlpha = mCustomBlendFunctions[3]; + mRasterState.depthWrite = false; + } + + // depth write + bool depthWriteSet = false; + parser->getDepthWriteSet(&depthWriteSet); + if (depthWriteSet) { + bool depthWrite = false; + parser->getDepthWrite(&depthWrite); + mRasterState.depthWrite = depthWrite; + } + + // alpha to coverage + bool alphaToCoverageSet = false; + parser->getAlphaToCoverageSet(&alphaToCoverageSet); + if (alphaToCoverageSet) { + bool alphaToCoverage = false; + parser->getAlphaToCoverage(&alphaToCoverage); + mRasterState.alphaToCoverage = alphaToCoverage; + } else { + mRasterState.alphaToCoverage = mBlendingMode == BlendingMode::MASKED; + } +} + +void FMaterial::processSpecializationConstants(FEngine& engine, Material::Builder const& builder, + MaterialParser const* const parser) { + // Older materials won't have a constants chunk, but that's okay. + parser->getConstants(&mMaterialConstants); + for (size_t i = 0, c = mMaterialConstants.size(); i < c; i++) { + auto& item = mMaterialConstants[i]; + // the key can be a string_view because mMaterialConstant owns the CString + std::string_view const key{ item.name.data(), item.name.size() }; + mSpecializationConstantsNameToIndex[key] = i; + } + + // Verify that all the constant specializations exist in the material and that their types match. + // The first specialization constants are defined internally by Filament. + // The subsequent constants are user-defined in the material. + + // Feature level 0 doesn't support instancing + int const maxInstanceCount = (engine.getActiveFeatureLevel() == FeatureLevel::FEATURE_LEVEL_0) + ? 1 : CONFIG_MAX_INSTANCES; + + int const maxFroxelBufferHeight = (int)std::min( + FROXEL_BUFFER_MAX_ENTRY_COUNT / 4, + engine.getDriverApi().getMaxUniformBufferSize() / 16u); + + bool const staticTextureWorkaround = + engine.getDriverApi().isWorkaroundNeeded(Workaround::A8X_STATIC_TEXTURE_TARGET_ERROR); + + bool const powerVrShaderWorkarounds = + engine.getDriverApi().isWorkaroundNeeded(Workaround::POWER_VR_SHADER_WORKAROUNDS); + + mSpecializationConstants.reserve(mMaterialConstants.size() + CONFIG_MAX_RESERVED_SPEC_CONSTANTS); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::BACKEND_FEATURE_LEVEL, + (int)engine.getSupportedFeatureLevel() }); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_MAX_INSTANCES, + (int)maxInstanceCount }); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_FROXEL_BUFFER_HEIGHT, + (int)maxFroxelBufferHeight }); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_DEBUG_DIRECTIONAL_SHADOWMAP, + engine.debug.shadowmap.debug_directional_shadowmap }); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_DEBUG_FROXEL_VISUALIZATION, + engine.debug.lighting.debug_froxel_visualization }); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_STATIC_TEXTURE_TARGET_WORKAROUND, + staticTextureWorkaround }); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_POWER_VR_SHADER_WORKAROUNDS, + powerVrShaderWorkarounds }); + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_STEREO_EYE_COUNT, + (int)engine.getConfig().stereoscopicEyeCount }); + if (UTILS_UNLIKELY(parser->getShaderLanguage() == ShaderLanguage::ESSL1)) { + // The actual value of this spec-constant is set in the OpenGLDriver backend. + mSpecializationConstants.push_back({ + +ReservedSpecializationConstants::CONFIG_SRGB_SWAPCHAIN_EMULATION, + false}); + } + + for (auto const& [name, value] : builder->mConstantSpecializations) { + std::string_view const key{ name.data(), name.size() }; + auto pos = mSpecializationConstantsNameToIndex.find(key); + ASSERT_PRECONDITION(pos != mSpecializationConstantsNameToIndex.end(), + "The material %s does not have a constant parameter named %s.", + mName.c_str_safe(), name.c_str()); + const char* const types[3] = {"an int", "a float", "a bool"}; + const char* const errorMessage = + "The constant parameter %s on material %s is of type %s, but %s was " + "provided."; + auto& constant = mMaterialConstants[pos->second]; + switch (constant.type) { + case ConstantType::INT: + ASSERT_PRECONDITION(std::holds_alternative(value), errorMessage, + name.c_str(), mName.c_str_safe(), "int", types[value.index()]); + break; + case ConstantType::FLOAT: + ASSERT_PRECONDITION(std::holds_alternative(value), errorMessage, + name.c_str(), mName.c_str_safe(), "float", types[value.index()]); + break; + case ConstantType::BOOL: + ASSERT_PRECONDITION(std::holds_alternative(value), errorMessage, + name.c_str(), mName.c_str_safe(), "bool", types[value.index()]); + break; + } + uint32_t const index = pos->second + CONFIG_MAX_RESERVED_SPEC_CONSTANTS; + mSpecializationConstants.push_back({ index, value }); + } +} + +void FMaterial::processDepthVariants(FEngine& engine, MaterialParser const* const parser) { + parser->hasCustomDepthShader(&mHasCustomDepthShader); + + if (UTILS_UNLIKELY(mIsDefaultMaterial)) { + assert_invariant(mMaterialDomain == MaterialDomain::SURFACE); + filaflat::MaterialChunk const& materialChunk{ parser->getMaterialChunk() }; + auto variants = FixedCapacityVector::with_capacity(materialChunk.getShaderCount()); + materialChunk.visitShaders([&variants]( + ShaderModel, Variant variant, ShaderStage) { + if (Variant::isValidDepthVariant(variant)) { + variants.push_back(variant); + } + }); + std::sort(variants.begin(), variants.end(), + [](Variant lhs, Variant rhs) { return lhs.key < rhs.key; }); + auto pos = std::unique(variants.begin(), variants.end()); + variants.resize(std::distance(variants.begin(), pos)); + std::swap(mDepthVariants, variants); + } + + if (mMaterialDomain == MaterialDomain::SURFACE) { + if (UTILS_UNLIKELY(!mIsDefaultMaterial && !mHasCustomDepthShader)) { + FMaterial const* const pDefaultMaterial = engine.getDefaultMaterial(); + auto& cachedPrograms = mCachedPrograms; + for (Variant const variant: pDefaultMaterial->mDepthVariants) { + pDefaultMaterial->prepareProgram(variant); + cachedPrograms[variant.key] = pDefaultMaterial->getProgram(variant); + } + } + } +} + template bool FMaterial::setConstant(uint32_t id, int32_t value) noexcept; template bool FMaterial::setConstant(uint32_t id, float value) noexcept; template bool FMaterial::setConstant(uint32_t id, bool value) noexcept; diff --git a/filament/src/details/Material.h b/filament/src/details/Material.h index 2a2819df6a7..20f1e309aad 100644 --- a/filament/src/details/Material.h +++ b/filament/src/details/Material.h @@ -238,6 +238,13 @@ class FMaterial : public Material { backend::Program getProgramWithVariants(Variant variant, Variant vertexVariant, Variant fragmentVariant) const noexcept; + void processBlendingMode(MaterialParser const* const parser); + + void processSpecializationConstants(FEngine& engine, Material::Builder const& builder, + MaterialParser const* const parser); + + void processDepthVariants(FEngine& engine, MaterialParser const* const parser); + void createAndCacheProgram(backend::Program&& p, Variant variant) const noexcept; // try to order by frequency of use diff --git a/filament/test/filament_test_material_parser.cpp b/filament/test/filament_test_material_parser.cpp index 100ac2f1cf8..853b540b300 100644 --- a/filament/test/filament_test_material_parser.cpp +++ b/filament/test/filament_test_material_parser.cpp @@ -31,7 +31,7 @@ using namespace filament; // This will re-compile the test material with the current version of matc. // To verify, rebuild and re-run test_material_parser (this test suite). TEST(MaterialParser, Parse) { - MaterialParser parser(backend::ShaderLanguage::ESSL3, + MaterialParser parser({ backend::ShaderLanguage::ESSL3 }, FILAMENT_TEST_RESOURCES_TEST_MATERIAL_DATA, FILAMENT_TEST_RESOURCES_TEST_MATERIAL_SIZE); MaterialParser::ParseResult materialOk = parser.parse(); diff --git a/ios/CocoaPods/Filament.podspec b/ios/CocoaPods/Filament.podspec index 44806433ca0..6bc053c37a9 100644 --- a/ios/CocoaPods/Filament.podspec +++ b/ios/CocoaPods/Filament.podspec @@ -1,12 +1,12 @@ Pod::Spec.new do |spec| spec.name = "Filament" - spec.version = "1.51.4" + spec.version = "1.51.5" spec.license = { :type => "Apache 2.0", :file => "LICENSE" } spec.homepage = "https://google.github.io/filament" spec.authors = "Google LLC." spec.summary = "Filament is a real-time physically based rendering engine for Android, iOS, Windows, Linux, macOS, and WASM/WebGL." spec.platform = :ios, "11.0" - spec.source = { :http => "https://github.com/google/filament/releases/download/v1.51.4/filament-v1.51.4-ios.tgz" } + spec.source = { :http => "https://github.com/google/filament/releases/download/v1.51.5/filament-v1.51.5-ios.tgz" } # Fix linking error with Xcode 12; we do not yet support the simulator on Apple silicon. spec.pod_target_xcconfig = { diff --git a/libs/filabridge/include/filament/MaterialChunkType.h b/libs/filabridge/include/filament/MaterialChunkType.h index 27d50c1683a..c80ac7d8c91 100644 --- a/libs/filabridge/include/filament/MaterialChunkType.h +++ b/libs/filabridge/include/filament/MaterialChunkType.h @@ -45,6 +45,7 @@ enum UTILS_PUBLIC ChunkType : uint64_t { MaterialEssl1 = charTo64bitNum("MAT_ESS1"), MaterialSpirv = charTo64bitNum("MAT_SPIR"), MaterialMetal = charTo64bitNum("MAT_METL"), + MaterialMetalLibrary = charTo64bitNum("MAT_MLIB"), MaterialShaderModels = charTo64bitNum("MAT_SMDL"), MaterialSamplerBindings = charTo64bitNum("MAT_SAMP"), MaterialUniformBindings = charTo64bitNum("MAT_UNIF"), @@ -93,6 +94,7 @@ enum UTILS_PUBLIC ChunkType : uint64_t { DictionaryText = charTo64bitNum("DIC_TEXT"), DictionarySpirv = charTo64bitNum("DIC_SPIR"), + DictionaryMetalLibrary = charTo64bitNum("DIC_MLIB"), }; } // namespace filamat diff --git a/libs/filaflat/include/filaflat/MaterialChunk.h b/libs/filaflat/include/filaflat/MaterialChunk.h index 2b7ceda7fa9..304819769ee 100644 --- a/libs/filaflat/include/filaflat/MaterialChunk.h +++ b/libs/filaflat/include/filaflat/MaterialChunk.h @@ -74,7 +74,7 @@ class MaterialChunk { BlobDictionary const& dictionary, ShaderContent& shaderContent, ShaderModel shaderModel, filament::Variant variant, ShaderStage shaderStage); - bool getSpirvShader( + bool getBinaryShader( BlobDictionary const& dictionary, ShaderContent& shaderContent, ShaderModel shaderModel, filament::Variant variant, ShaderStage shaderStage); }; diff --git a/libs/filaflat/src/DictionaryReader.cpp b/libs/filaflat/src/DictionaryReader.cpp index 87c8d9a2e08..6a84ca34b62 100644 --- a/libs/filaflat/src/DictionaryReader.cpp +++ b/libs/filaflat/src/DictionaryReader.cpp @@ -78,6 +78,25 @@ bool DictionaryReader::unflatten(ChunkContainer const& container, } return true; + } else if (dictionaryTag == ChunkType::DictionaryMetalLibrary) { + uint32_t blobCount; + if (!unflattener.read(&blobCount)) { + return false; + } + + dictionary.reserve(blobCount); + for (uint32_t i = 0; i < blobCount; i++) { + unflattener.skipAlignmentPadding(); + + const char* data; + size_t dataSize; + if (!unflattener.read(&data, &dataSize)) { + return false; + } + dictionary.emplace_back(dataSize); + memcpy(dictionary.back().data(), data, dictionary.back().size()); + } + return true; } else if (dictionaryTag == ChunkType::DictionaryText) { uint32_t stringCount = 0; if (!unflattener.read(&stringCount)) { diff --git a/libs/filaflat/src/MaterialChunk.cpp b/libs/filaflat/src/MaterialChunk.cpp index a7798aa6e01..b69af9abd08 100644 --- a/libs/filaflat/src/MaterialChunk.cpp +++ b/libs/filaflat/src/MaterialChunk.cpp @@ -153,7 +153,7 @@ bool MaterialChunk::getTextShader(Unflattener unflattener, return true; } -bool MaterialChunk::getSpirvShader(BlobDictionary const& dictionary, +bool MaterialChunk::getBinaryShader(BlobDictionary const& dictionary, ShaderContent& shaderContent, ShaderModel shaderModel, filament::Variant variant, ShaderStage shaderStage) { if (mBase == nullptr) { @@ -186,7 +186,8 @@ bool MaterialChunk::getShader(ShaderContent& shaderContent, BlobDictionary const case filamat::ChunkType::MaterialMetal: return getTextShader(mUnflattener, dictionary, shaderContent, shaderModel, variant, stage); case filamat::ChunkType::MaterialSpirv: - return getSpirvShader(dictionary, shaderContent, shaderModel, variant, stage); + case filamat::ChunkType::MaterialMetalLibrary: + return getBinaryShader(dictionary, shaderContent, shaderModel, variant, stage); default: return false; } diff --git a/libs/filamat/CMakeLists.txt b/libs/filamat/CMakeLists.txt index cf1e529a6e4..4e97538f726 100644 --- a/libs/filamat/CMakeLists.txt +++ b/libs/filamat/CMakeLists.txt @@ -50,7 +50,8 @@ set(PRIVATE_HDRS ${COMMON_PRIVATE_HDRS} src/eiff/BlobDictionary.h src/eiff/DictionarySpirvChunk.h - src/eiff/MaterialSpirvChunk.h + src/eiff/DictionaryMetalLibraryChunk.h + src/eiff/MaterialBinaryChunk.h src/GLSLPostProcessor.h src/MetalArgumentBuffer.h src/ShaderMinifier.h @@ -63,7 +64,8 @@ set(SRCS ${COMMON_SRCS} src/eiff/BlobDictionary.cpp src/eiff/DictionarySpirvChunk.cpp - src/eiff/MaterialSpirvChunk.cpp + src/eiff/DictionaryMetalLibraryChunk.cpp + src/eiff/MaterialBinaryChunk.cpp src/MetalArgumentBuffer.cpp src/sca/ASTHelpers.cpp src/sca/GLSLTools.cpp diff --git a/libs/filamat/src/MaterialBuilder.cpp b/libs/filamat/src/MaterialBuilder.cpp index 109d6e43da5..9acf657353a 100644 --- a/libs/filamat/src/MaterialBuilder.cpp +++ b/libs/filamat/src/MaterialBuilder.cpp @@ -33,11 +33,12 @@ #include "eiff/LineDictionary.h" #include "eiff/MaterialInterfaceBlockChunk.h" #include "eiff/MaterialTextChunk.h" -#include "eiff/MaterialSpirvChunk.h" +#include "eiff/MaterialBinaryChunk.h" #include "eiff/ChunkContainer.h" #include "eiff/SimpleFieldChunk.h" #include "eiff/DictionaryTextChunk.h" #include "eiff/DictionarySpirvChunk.h" +#include "eiff/DictionaryMetalLibraryChunk.h" #include #include @@ -821,7 +822,7 @@ bool MaterialBuilder::generateShaders(JobSystem& jobSystem, const std::vector glslEntries; std::vector essl1Entries; - std::vector spirvEntries; + std::vector spirvEntries; std::vector metalEntries; LineDictionary textDictionary; BlobDictionary spirvDictionary; @@ -872,7 +873,7 @@ bool MaterialBuilder::generateShaders(JobSystem& jobSystem, const std::vector d(reinterpret_cast(spirv.data()), + reinterpret_cast(spirv.data() + spirv.size())); spirvEntry.stage = v.stage; - spirvEntry.spirv = std::move(spirv); + spirvEntry.data = std::move(d); spirvEntries.push_back(spirvEntry); break; + } case TargetApi::METAL: assert(!spirv.empty()); assert(msl.length() > 0); @@ -1018,7 +1022,7 @@ bool MaterialBuilder::generateShaders(JobSystem& jobSystem, const std::vector spirv = std::move(s.spirv); + std::vector spirv = std::move(s.data); s.dictionaryIndex = spirvDictionary.addBlob(spirv); } for (const auto& s : metalEntries) { @@ -1041,11 +1045,11 @@ bool MaterialBuilder::generateShaders(JobSystem& jobSystem, const std::vector(std::move(spirvDictionary), stripInfo); - container.push(std::move(spirvEntries)); + container.push(std::move(spirvEntries), ChunkType::MaterialSpirv); } // Emit Metal chunk (MaterialTextChunk). diff --git a/libs/filamat/src/eiff/BlobDictionary.cpp b/libs/filamat/src/eiff/BlobDictionary.cpp index d13227a5998..692884b0d95 100644 --- a/libs/filamat/src/eiff/BlobDictionary.cpp +++ b/libs/filamat/src/eiff/BlobDictionary.cpp @@ -20,8 +20,8 @@ namespace filamat { -size_t BlobDictionary::addBlob(const std::vector& vblob) noexcept { - std::string_view blob((char*) vblob.data(), vblob.size() * 4); +size_t BlobDictionary::addBlob(const std::vector& vblob) noexcept { + std::string_view blob((char*) vblob.data(), vblob.size()); auto iter = mBlobIndices.find(blob); if (iter != mBlobIndices.end()) { return iter->second; diff --git a/libs/filamat/src/eiff/BlobDictionary.h b/libs/filamat/src/eiff/BlobDictionary.h index c204f926587..b642f188d87 100644 --- a/libs/filamat/src/eiff/BlobDictionary.h +++ b/libs/filamat/src/eiff/BlobDictionary.h @@ -36,7 +36,7 @@ class BlobDictionary { BlobDictionary(BlobDictionary&&) = default; // Adds a blob if it's not already a duplicate and returns its index. - size_t addBlob(const std::vector& blob) noexcept; + size_t addBlob(const std::vector& blob) noexcept; size_t getBlobCount() const noexcept { return mBlobs.size(); diff --git a/libs/filamat/src/eiff/DictionaryMetalLibraryChunk.cpp b/libs/filamat/src/eiff/DictionaryMetalLibraryChunk.cpp new file mode 100644 index 00000000000..fb701aab98e --- /dev/null +++ b/libs/filamat/src/eiff/DictionaryMetalLibraryChunk.cpp @@ -0,0 +1,33 @@ +/* + * Copyright (C) 2024 The Android Open Source Project + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include "DictionaryMetalLibraryChunk.h" + +namespace filamat { + +DictionaryMetalLibraryChunk::DictionaryMetalLibraryChunk(BlobDictionary&& dictionary) + : Chunk(ChunkType::DictionaryMetalLibrary), mDictionary(std::move(dictionary)) {} + +void DictionaryMetalLibraryChunk::flatten(Flattener& f) { + f.writeUint32(mDictionary.getBlobCount()); + for (size_t i = 0 ; i < mDictionary.getBlobCount() ; i++) { + std::string_view blob = mDictionary.getBlob(i); + f.writeAlignmentPadding(); + f.writeBlob((const char*) blob.data(), blob.size()); + } +} + +} // namespace filamat diff --git a/libs/filamat/src/eiff/DictionaryMetalLibraryChunk.h b/libs/filamat/src/eiff/DictionaryMetalLibraryChunk.h new file mode 100644 index 00000000000..a129974961b --- /dev/null +++ b/libs/filamat/src/eiff/DictionaryMetalLibraryChunk.h @@ -0,0 +1,43 @@ +/* + * Copyright (C) 2024 The Android Open Source Project + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#ifndef TNT_FILAMAT_DIC_METAL_LIBRARY_CHUNK_H +#define TNT_FILAMAT_DIC_METAL_LIBRARY_CHUNK_H + +#include +#include + +#include "Chunk.h" +#include "Flattener.h" +#include "BlobDictionary.h" + +namespace filamat { + +class DictionaryMetalLibraryChunk final : public Chunk { +public: + explicit DictionaryMetalLibraryChunk(BlobDictionary&& dictionary); + ~DictionaryMetalLibraryChunk() = default; + +private: + void flatten(Flattener& f) override; + + BlobDictionary mDictionary; + bool mStripDebugInfo; +}; + +} // namespace filamat + +#endif // TNT_FILAMAT_DIC_METAL_LIBRARY_CHUNK_H diff --git a/libs/filamat/src/eiff/Flattener.h b/libs/filamat/src/eiff/Flattener.h index 2e3cfa22352..e43cdb76811 100644 --- a/libs/filamat/src/eiff/Flattener.h +++ b/libs/filamat/src/eiff/Flattener.h @@ -127,6 +127,13 @@ class Flattener { mCursor += nbytes; } + void writeRaw(const char* raw, size_t nbytes) { + if (mStart != nullptr) { + memcpy(reinterpret_cast(mCursor), raw, nbytes); + } + mCursor += nbytes; + } + void writeSizePlaceholder() { mSizePlaceholders.push_back(mCursor); if (mStart != nullptr) { diff --git a/libs/filamat/src/eiff/MaterialSpirvChunk.cpp b/libs/filamat/src/eiff/MaterialBinaryChunk.cpp similarity index 75% rename from libs/filamat/src/eiff/MaterialSpirvChunk.cpp rename to libs/filamat/src/eiff/MaterialBinaryChunk.cpp index 8003ff23989..bc097e04234 100644 --- a/libs/filamat/src/eiff/MaterialSpirvChunk.cpp +++ b/libs/filamat/src/eiff/MaterialBinaryChunk.cpp @@ -1,6 +1,6 @@ /* * Copyright (C) 2018 The Android Open Source Project - *DictionaryGlsl + * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. * You may obtain a copy of the License at @@ -14,16 +14,17 @@ * limitations under the License. */ -#include "MaterialSpirvChunk.h" +#include "MaterialBinaryChunk.h" namespace filamat { -MaterialSpirvChunk::MaterialSpirvChunk(const std::vector&& entries) : - Chunk(ChunkType::MaterialSpirv), mEntries(entries) {} +MaterialBinaryChunk::MaterialBinaryChunk( + const std::vector&& entries, ChunkType chunkType) + : Chunk(chunkType), mEntries(entries) {} -void MaterialSpirvChunk::flatten(Flattener &f) { +void MaterialBinaryChunk::flatten(Flattener &f) { f.writeUint64(mEntries.size()); - for (const SpirvEntry& entry : mEntries) { + for (const BinaryEntry& entry : mEntries) { f.writeUint8(uint8_t(entry.shaderModel)); f.writeUint8(entry.variant.key); f.writeUint8(uint8_t(entry.stage)); diff --git a/libs/filamat/src/eiff/MaterialSpirvChunk.h b/libs/filamat/src/eiff/MaterialBinaryChunk.h similarity index 68% rename from libs/filamat/src/eiff/MaterialSpirvChunk.h rename to libs/filamat/src/eiff/MaterialBinaryChunk.h index 42aefdc245f..8c121947244 100644 --- a/libs/filamat/src/eiff/MaterialSpirvChunk.h +++ b/libs/filamat/src/eiff/MaterialBinaryChunk.h @@ -14,8 +14,8 @@ * limitations under the License. */ -#ifndef TNT_FILAMAT_MATERIAL_SPIRV_CHUNK_H -#define TNT_FILAMAT_MATERIAL_SPIRV_CHUNK_H +#ifndef TNT_FILAMAT_MATERIAL_BINARY_CHUNK_H +#define TNT_FILAMAT_MATERIAL_BINARY_CHUNK_H #include "Chunk.h" #include "ShaderEntry.h" @@ -24,17 +24,17 @@ namespace filamat { -class MaterialSpirvChunk final : public Chunk { +class MaterialBinaryChunk final : public Chunk { public: - explicit MaterialSpirvChunk(const std::vector&& entries); - ~MaterialSpirvChunk() = default; + explicit MaterialBinaryChunk(const std::vector&& entries, ChunkType type); + ~MaterialBinaryChunk() = default; private: void flatten(Flattener& f) override; - const std::vector mEntries; + const std::vector mEntries; }; } // namespace filamat -#endif // TNT_FILAMAT_MATERIAL_SPIRV_CHUNK_H +#endif // TNT_FILAMAT_MATERIAL_BINARY_CHUNK_H diff --git a/libs/filamat/src/eiff/ShaderEntry.h b/libs/filamat/src/eiff/ShaderEntry.h index f29b7f14a16..b8e2067609b 100644 --- a/libs/filamat/src/eiff/ShaderEntry.h +++ b/libs/filamat/src/eiff/ShaderEntry.h @@ -34,14 +34,14 @@ struct TextEntry { std::string shader; }; -struct SpirvEntry { +struct BinaryEntry { filament::backend::ShaderModel shaderModel; filament::Variant variant; filament::backend::ShaderStage stage; - size_t dictionaryIndex; + size_t dictionaryIndex; // maps to an index in the blob dictionary - // temporarily holds this entry's spirv until added to the dictionary - std::vector spirv; + // temporarily holds this entry's binary data until added to the dictionary + std::vector data; }; } // namespace filamat diff --git a/libs/iblprefilter/src/materials/iblprefilter.mat b/libs/iblprefilter/src/materials/iblprefilter.mat index aa0c25c1d67..d9fa379a8e0 100644 --- a/libs/iblprefilter/src/materials/iblprefilter.mat +++ b/libs/iblprefilter/src/materials/iblprefilter.mat @@ -122,10 +122,9 @@ void postProcess(inout PostProcessInputs postProcess) { float side = materialParams.side; // compute the view (and normal, since v = n) direction for each face - float l = inversesqrt(p.x * p.x + p.y * p.y + 1.0); - vec3 rx = vec3( side, p.y, side * -p.x) * l; - vec3 ry = vec3( p.x, side, side * -p.y) * l; - vec3 rz = vec3(side * p.x, p.y, side) * l; + vec3 rx = normalize(vec3( side, -p.y, side * -p.x)); + vec3 ry = normalize(vec3( p.x, side, side * p.y)); + vec3 rz = normalize(vec3(side * p.x, -p.y, side)); // random rotation around r mediump float a = 2.0 * PI * random(gl_FragCoord.xy); diff --git a/libs/matdbg/src/ShaderExtractor.cpp b/libs/matdbg/src/ShaderExtractor.cpp index f9dbe0ea1eb..22ba82c211f 100644 --- a/libs/matdbg/src/ShaderExtractor.cpp +++ b/libs/matdbg/src/ShaderExtractor.cpp @@ -53,6 +53,10 @@ ShaderExtractor::ShaderExtractor(backend::ShaderLanguage target, const void* dat mMaterialTag = ChunkType::MaterialMetal; mDictionaryTag = ChunkType::DictionaryText; break; + case backend::ShaderLanguage::METAL_LIBRARY: + mMaterialTag = ChunkType::MaterialMetalLibrary; + mDictionaryTag = ChunkType::DictionaryMetalLibrary; + break; case backend::ShaderLanguage::SPIRV: mMaterialTag = ChunkType::MaterialSpirv; mDictionaryTag = ChunkType::DictionarySpirv; diff --git a/libs/matdbg/src/ShaderReplacer.cpp b/libs/matdbg/src/ShaderReplacer.cpp index 8eb3497eb4c..9cd8ba4de06 100644 --- a/libs/matdbg/src/ShaderReplacer.cpp +++ b/libs/matdbg/src/ShaderReplacer.cpp @@ -35,7 +35,7 @@ #include "eiff/ChunkContainer.h" #include "eiff/DictionarySpirvChunk.h" #include "eiff/DictionaryTextChunk.h" -#include "eiff/MaterialSpirvChunk.h" +#include "eiff/MaterialBinaryChunk.h" #include "eiff/MaterialTextChunk.h" #include "eiff/LineDictionary.h" @@ -73,7 +73,7 @@ class ShaderIndex { vector mShaderRecords; }; -// Tiny database of data blobs that can import / export MaterialSpirvChunk and DictionarySpirvChunk. +// Tiny database of data blobs that can import / export MaterialBinaryChunk and DictionarySpirvChunk. // The blobs are stored *after* they have been compressed by SMOL-V. class BlobIndex { public: @@ -90,7 +90,7 @@ class BlobIndex { private: const ChunkType mDictTag; const ChunkType mMatTag; - vector mShaderRecords; + vector mShaderRecords; filaflat::BlobDictionary mDataBlobs; }; @@ -364,7 +364,7 @@ BlobIndex::BlobIndex(ChunkType dictTag, ChunkType matTag, const filaflat::ChunkC const auto& offsets = matChunk.getOffsets(); mShaderRecords.reserve(offsets.size()); for (auto [key, offset] : offsets) { - SpirvEntry info; + BinaryEntry info; filaflat::MaterialChunk::decodeKey(key, &info.shaderModel, &info.variant, &info.stage); info.dictionaryIndex = offset; mShaderRecords.emplace_back(info); @@ -378,7 +378,7 @@ void BlobIndex::writeChunks(ostream& stream) { const auto& src = mDataBlobs[record.dictionaryIndex]; assert(src.size() % 4 == 0); const uint32_t* ptr = (const uint32_t*) src.data(); - record.dictionaryIndex = blobs.addBlob(vector(ptr, ptr + src.size() / 4)); + record.dictionaryIndex = blobs.addBlob(vector(ptr, ptr + src.size())); } // Adjust start cursor of flatteners to match alignment of output stream. @@ -391,7 +391,7 @@ void BlobIndex::writeChunks(ostream& stream) { // Apply SMOL-V compression and write out the results. filamat::ChunkContainer cc; - cc.push(std::move(mShaderRecords)); + cc.push(std::move(mShaderRecords), ChunkType::MaterialSpirv); cc.push(std::move(blobs), false); Flattener prepass = Flattener::getDryRunner(); diff --git a/libs/matdbg/src/TextWriter.cpp b/libs/matdbg/src/TextWriter.cpp index b74596f6a42..fca5bb0ca19 100644 --- a/libs/matdbg/src/TextWriter.cpp +++ b/libs/matdbg/src/TextWriter.cpp @@ -421,6 +421,9 @@ static bool printShaderInfo(ostream& text, const ChunkContainer& container, Chun case ChunkType::MaterialMetal: text << "Metal shaders:" << endl; break; + case ChunkType::MaterialMetalLibrary: + text << "Metal precompiled shader libraries:" << endl; + break; default: assert(false && "Invalid shader ChunkType"); break; @@ -455,6 +458,9 @@ bool TextWriter::writeMaterialInfo(const filaflat::ChunkContainer& container) { if (!printShaderInfo(text, container, ChunkType::MaterialMetal)) { return false; } + if (!printShaderInfo(text, container, ChunkType::MaterialMetalLibrary)) { + return false; + } printChunks(text, container); diff --git a/libs/utils/CMakeLists.txt b/libs/utils/CMakeLists.txt index e385a841507..e19a943c185 100644 --- a/libs/utils/CMakeLists.txt +++ b/libs/utils/CMakeLists.txt @@ -38,6 +38,7 @@ set(DIST_HDRS ${PUBLIC_HDR_DIR}/${TARGET}/SingleInstanceComponentManager.h ${PUBLIC_HDR_DIR}/${TARGET}/Slice.h ${PUBLIC_HDR_DIR}/${TARGET}/StructureOfArrays.h + ${PUBLIC_HDR_DIR}/${TARGET}/Systrace.h ${PUBLIC_HDR_DIR}/${TARGET}/unwindows.h ) @@ -115,6 +116,11 @@ if (WIN32) target_link_libraries(${TARGET} PUBLIC Shlwapi) endif() +if (APPLE) + # Needed for NSTemporaryDirectory() + target_link_libraries(${TARGET} PRIVATE "-framework Foundation") +endif() + if (LINUX) set(THREADS_PREFER_PTHREAD_FLAG ON) find_package(Threads REQUIRED) diff --git a/libs/utils/include/utils/android/Systrace.h b/libs/utils/include/utils/android/Systrace.h index 41b64f16723..760fe5be647 100644 --- a/libs/utils/include/utils/android/Systrace.h +++ b/libs/utils/include/utils/android/Systrace.h @@ -97,7 +97,7 @@ namespace utils { namespace details { -class Systrace { +class UTILS_PUBLIC Systrace { public: enum tags { @@ -220,7 +220,7 @@ class Systrace { // ------------------------------------------------------------------------------------------------ -class ScopedTrace { +class UTILS_PUBLIC ScopedTrace { public: // we don't inline this because it's relatively heavy due to a global check ScopedTrace(uint32_t tag, const char* name) noexcept: mTrace(tag), mTag(tag) { diff --git a/tools/matedit/CMakeLists.txt b/tools/matedit/CMakeLists.txt new file mode 100644 index 00000000000..e88aad24b01 --- /dev/null +++ b/tools/matedit/CMakeLists.txt @@ -0,0 +1,28 @@ +cmake_minimum_required(VERSION 3.19) +project(matedit) + +set(TARGET matedit) + +# ================================================================================================== +# Sources and headers +# ================================================================================================== +set(SRCS + src/main.cpp + src/ExternalCompile.cpp +) + +# ================================================================================================== +# Target definitions +# ================================================================================================== +add_executable(${TARGET} ${SRCS}) + +target_include_directories(${TARGET} PRIVATE ${filamat_SOURCE_DIR}/src) + +target_link_libraries(${TARGET} matdbg getopt) + +set_target_properties(${TARGET} PROPERTIES FOLDER Tools) + +# ================================================================================================== +# Installation +# ================================================================================================== +install(TARGETS ${TARGET} RUNTIME DESTINATION bin) diff --git a/tools/matedit/src/ExternalCompile.cpp b/tools/matedit/src/ExternalCompile.cpp new file mode 100644 index 00000000000..ee0374bc2d8 --- /dev/null +++ b/tools/matedit/src/ExternalCompile.cpp @@ -0,0 +1,423 @@ +/* + * Copyright (C) 2024 The Android Open Source Project + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include "ExternalCompile.h" + +#include "backend/DriverEnums.h" +#include "eiff/BlobDictionary.h" +#include "eiff/ChunkContainer.h" +#include "eiff/DictionaryMetalLibraryChunk.h" +#include "eiff/DictionarySpirvChunk.h" +#include "eiff/DictionaryTextChunk.h" +#include "eiff/LineDictionary.h" +#include "eiff/MaterialBinaryChunk.h" +#include "eiff/MaterialTextChunk.h" +#include "eiff/ShaderEntry.h" + +#include +#include +#include + +#include +#include + +#include + +#include +#include +#include +#include + +using filamat::Flattener; +using filamat::Package; +using namespace filament; + +class PassthroughChunk final : public filamat::Chunk { +public: + explicit PassthroughChunk(const char* data, size_t size, filamat::ChunkType type) + : filamat::Chunk(type), data(data), size(size) {} + + ~PassthroughChunk() = default; + +private: + void flatten(Flattener& f) override { f.writeRaw(data, size); } + + const char* data; + size_t size; +}; + +namespace matedit { + +static std::ifstream::pos_type getFileSize(const char* filename) { + std::ifstream in(filename, std::ifstream::ate | std::ifstream::binary); + return in.tellg(); +} + +static void dumpBinary(const uint8_t* data, size_t size, utils::Path filename) { + std::ofstream out(filename, std::ofstream::binary); + out.write(reinterpret_cast(data), size); +} + +static void dumpString(const std::string& data, utils::Path filename) { + std::ofstream out(filename, std::ofstream::binary); + out << data; +} + +static bool readBinary(utils::Path filename, std::vector& buffer) { + std::ifstream in(filename, std::ifstream::binary | std::ifstream::in); + if (!in) { + return false; + } + in.seekg(0, std::ios::end); + std::ifstream::pos_type size = in.tellg(); + in.seekg(0); + buffer.resize(size); + if (!in.read((char*)buffer.data(), size)) { + return false; + } + return true; +} + +template +static std::vector getShaderRecords(const filaflat::ChunkContainer& container, + const filaflat::BlobDictionary& dictionary, filamat::ChunkType chunkType) { + if (!container.hasChunk(chunkType)) { + return {}; + } + std::vector shaderRecords; + filaflat::MaterialChunk materialChunk(container); + materialChunk.initialize(chunkType); + materialChunk.visitShaders( + [&materialChunk, &dictionary, &shaderRecords]( + backend::ShaderModel shaderModel, Variant variant, backend::ShaderStage stage) { + filaflat::ShaderContent content; + UTILS_UNUSED_IN_RELEASE bool success = + materialChunk.getShader(content, dictionary, shaderModel, variant, stage); + + std::string source { content.data(), content.data() + content.size() - 1u }; + assert_invariant(success); + + if constexpr (std::is_same_v) { + shaderRecords.push_back({ shaderModel, variant, stage, std::move(source) }); + } + if constexpr (std::is_same_v) { + filamat::BinaryEntry e {}; + e.shaderModel = shaderModel; + e.variant = variant; + e.stage = stage; + e.dictionaryIndex = 0; + e.data = std::vector(content.begin(), content.end()); + shaderRecords.push_back(std::move(e)); + } + }); + return shaderRecords; +} + +static std::string toString(backend::ShaderModel model) { + switch (model) { + case backend::ShaderModel::DESKTOP: + return "desktop"; + case backend::ShaderModel::MOBILE: + return "mobile"; + } +} + +static std::string toString(backend::ShaderStage stage) { + switch (stage) { + case backend::ShaderStage::VERTEX: + return "vertex"; + case backend::ShaderStage::FRAGMENT: + return "fragment"; + case backend::ShaderStage::COMPUTE: + return "compute"; + } +} + +static std::string toString(Variant variant) { return std::to_string(variant.key); } + +static bool invokeScript(const std::vector& userArgs, backend::ShaderStage stage, + backend::ShaderModel model, utils::Path inputPath, utils::Path outputPath) { + assert_invariant(!userArgs.empty()); + + std::vector argv; + + // The first argument is the path to the script + argv.push_back(const_cast(userArgs[0].c_str())); + + // Temporary input and output files + argv.push_back(const_cast(inputPath.c_str())); + argv.push_back(const_cast(outputPath.c_str())); + argv.push_back(const_cast(toString(stage).c_str())); + argv.push_back(const_cast(toString(model).c_str())); + + // Optional user-supplied arguments + for (int i = 1; i < userArgs.size(); i++) { + argv.push_back(const_cast(userArgs[i].c_str())); + } + + // execvp expects a null as the last element of the arguments array + argv.push_back(nullptr); + + std::cout << "Invoking script: "; + for (const char* a : argv) { + if (a) { + std::cout << a << " "; + } + } + std::cout << std::endl; + + pid_t pid = fork(); + + if (pid == -1) { + // The fork() command failed + std::cerr << "Unable to fork process." << std::endl; + return false; + } else if (pid > 0) { + // Parent process + int status; + waitpid(pid, &status, 0); // Wait for the child to finish + + if (WIFEXITED(status)) { + if (WEXITSTATUS(status) != 0) { + std::cerr << "Script exited with status: " << WEXITSTATUS(status) << std::endl; + return false; + } + } + } else { + // Child process + execvp(argv[0], argv.data()); + + // If execvp returns, it failed + std::cerr << "Unable to execute script: " << argv[0] << std::endl; + exit(1); + } + + return true; +} + +class ScopedTempFile { +public: + ScopedTempFile(utils::Path&& path) noexcept { + auto segments = path.split(); + auto ext = path.getExtension(); + segments[segments.size() - 1] = path.getNameWithoutExtension() + ".XXXXXX." + ext; + + utils::Path pathTemplate; + for (const auto& s : segments) { + pathTemplate += s; + } + + std::string pathString = pathTemplate.getPath(); + int fd = mkstemps(const_cast(pathString.c_str()), ext.size() + 1); + if (fd == -1) { + std::cerr << "Error creating temporary file: " << pathString << std::endl; + exit(1); + } + close(fd); // close the file, it's been created for us + + mPath = pathString; + } + ~ScopedTempFile() noexcept { mPath.unlinkFile(); } + + const utils::Path& getPath() const noexcept { return mPath; } + + ScopedTempFile(const ScopedTempFile& rhs) = delete; + ScopedTempFile(ScopedTempFile&& rhs) = delete; + ScopedTempFile& operator=(const ScopedTempFile& rhs) = delete; + ScopedTempFile& operator=(ScopedTempFile&& rhs) = delete; + +private: + utils::Path mPath; +}; + +bool compileMetalShaders(const std::vector& mslEntries, + std::vector& metalBinaryEntries, + const std::vector& userArgs) { + const utils::Path tempDir = utils::Path::getTemporaryDirectory(); + + for (const auto& mslEntry : mslEntries) { + const std::string fileName = toString(mslEntry.shaderModel) + "_" + + toString(mslEntry.stage) + "_" + toString(mslEntry.variant); + const std::string inputFileName = fileName + ".metal"; + const std::string outputFileName = fileName + ".metallib"; + + ScopedTempFile inputFile = tempDir + inputFileName; + ScopedTempFile outputFile = tempDir + outputFileName; + + dumpString(mslEntry.shader, inputFile.getPath()); + if (!invokeScript(userArgs, mslEntry.stage, mslEntry.shaderModel, inputFile.getPath(), + outputFile.getPath())) { + return false; + } + + std::vector buffer; + if (!readBinary(outputFile.getPath(), buffer)) { + std::cerr << "Could not read output file " << outputFile.getPath() << std::endl; + return false; + } + + if (buffer.empty()) { + std::cerr << "Output file " << outputFile.getPath() << " is empty" << std::endl; + return false; + } + + filamat::BinaryEntry metalBinaryEntry {}; + metalBinaryEntry.shaderModel = mslEntry.shaderModel; + metalBinaryEntry.variant = mslEntry.variant; + metalBinaryEntry.stage = mslEntry.stage; + metalBinaryEntry.data = std::move(buffer); + metalBinaryEntries.push_back(metalBinaryEntry); + } + + return true; +} + +int externalCompile(utils::Path input, utils::Path output, std::vector args) { + std::ifstream in(input.c_str(), std::ifstream::in | std::ios::binary); + if (!in.is_open()) { + std::cerr << "Could not open the source material " << input << std::endl; + return 1; + } + + const long fileSize = static_cast(getFileSize(input.c_str())); + + std::vector buffer(static_cast(fileSize)); + if (!in.read(buffer.data(), fileSize)) { + std::cerr << "Could not read the source material." << std::endl; + return 1; + } + + filaflat::ChunkContainer container(buffer.data(), buffer.size()); + if (!container.parse()) { + return 1; + } + + // Get all shaders from the input material. + filaflat::BlobDictionary stringBlobs; + filaflat::BlobDictionary spirvBinaryBlobs; + filaflat::DictionaryReader reader; + if (container.hasChunk(filamat::ChunkType::DictionaryText)) { + reader.unflatten(container, filamat::ChunkType::DictionaryText, stringBlobs); + } + if (container.hasChunk(filamat::ChunkType::DictionarySpirv)) { + reader.unflatten(container, filamat::ChunkType::DictionarySpirv, spirvBinaryBlobs); + } + auto mslEntries = getShaderRecords( + container, stringBlobs, filamat::ChunkType::MaterialMetal); + auto glslEntries = getShaderRecords( + container, stringBlobs, filamat::ChunkType::MaterialGlsl); + auto essl1Entries = getShaderRecords( + container, stringBlobs, filamat::ChunkType::MaterialEssl1); + auto spirvEntries = getShaderRecords( + container, spirvBinaryBlobs, filamat::ChunkType::MaterialSpirv); + + // Ask the user script to compile the MSL shaders into .metallib files. + filamat::BlobDictionary metalBinaryDictionary; + std::vector metalBinaryEntries; + if (!compileMetalShaders(mslEntries, metalBinaryEntries, args)) { + return 1; + } + + // Since we're modifying text shaders, we'll need to regenerate the text dictionary. + // We'll also need to re-emit text based shaders that rely on the dictionary. + // Here we ONLY add GLSL and ESSL 1 types, as we're removing MSL completely. + filamat::LineDictionary textDictionary; + for (const auto& s : glslEntries) { + textDictionary.addText(s.shader); + } + for (const auto& s : essl1Entries) { + textDictionary.addText(s.shader); + } + + // We'll also need to regenerate the SPIRV dictionary and SPIRV shaders. + // This is required, as the SPIRV blobs have alignment requirements. Since we're modifying other + // chunks, their alignment might have changed. + filamat::BlobDictionary spirvDictionary; + for (auto& s : spirvEntries) { + std::vector spirv = std::move(s.data); + s.dictionaryIndex = spirvDictionary.addBlob(spirv); + } + + // Generate the Metal library dictionary. + for (auto& e : metalBinaryEntries) { + std::vector data = std::move(e.data); + e.dictionaryIndex = metalBinaryDictionary.addBlob(data); + } + + // Pass through chunks that don't need to change. + filamat::ChunkContainer outputChunks; + for (int i = 0; i < container.getChunkCount(); i++) { + filaflat::ChunkContainer::Chunk c = container.getChunk(i); + if (c.type == filamat::ChunkType::MaterialMetal) { + // This chunk is being removed, skip it. + continue; + } + if (c.type == filamat::ChunkType::MaterialGlsl || + c.type == filamat::ChunkType::MaterialEssl1 || + c.type == filamat::ChunkType::MaterialSpirv || + c.type == filamat::ChunkType::DictionarySpirv || + c.type == filamat::ChunkType::DictionaryText) { + // These shader / dictionary chunks will be re-added below. + continue; + } + outputChunks.push( + reinterpret_cast(c.desc.start), c.desc.size, c.type); + } + + // Add the re-generated text dictionary chunk and text-based shaders. + if (!textDictionary.isEmpty()) { + const auto& dictionaryChunk = outputChunks.push( + std::move(textDictionary), filamat::ChunkType::DictionaryText); + + // Re-emit GLSL chunk (MaterialTextChunk). + if (!glslEntries.empty()) { + outputChunks.push(std::move(glslEntries), + dictionaryChunk.getDictionary(), filamat::ChunkType::MaterialGlsl); + } + + // Re-emit ESSL1 chunk (MaterialTextChunk). + if (!essl1Entries.empty()) { + outputChunks.push(std::move(essl1Entries), + dictionaryChunk.getDictionary(), filamat::ChunkType::MaterialEssl1); + } + } + + // Add the SPIRV chunks. + if (!spirvEntries.empty()) { + const bool stripInfo = true; + outputChunks.push(std::move(spirvDictionary), stripInfo); + outputChunks.push( + std::move(spirvEntries), filamat::ChunkType::MaterialSpirv); + } + + // Add the new Metal binary chunks. + outputChunks.push(std::move(metalBinaryDictionary)); + outputChunks.push( + std::move(metalBinaryEntries), filamat::ChunkType::MaterialMetalLibrary); + + // Flatten into a Package and write to disk. + Package package(outputChunks.getSize()); + Flattener f { package.getData() }; + outputChunks.flatten(f); + + assert_invariant(package.isValid()); + + dumpBinary(package.getData(), package.getSize(), output); + + return 0; +} + +} // namespace matedit diff --git a/tools/matedit/src/ExternalCompile.h b/tools/matedit/src/ExternalCompile.h new file mode 100644 index 00000000000..d24907c6904 --- /dev/null +++ b/tools/matedit/src/ExternalCompile.h @@ -0,0 +1,31 @@ +/* + * Copyright (C) 2024 The Android Open Source Project + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#ifndef TNT_MATEDIT_EXTERNALCOMPILE_H +#define TNT_MATEDIT_EXTERNALCOMPILE_H + +#include + +#include +#include + +namespace matedit { + +int externalCompile(utils::Path input, utils::Path output, std::vector args); + +} // namespace matedit + +#endif diff --git a/tools/matedit/src/main.cpp b/tools/matedit/src/main.cpp new file mode 100644 index 00000000000..0adc9808773 --- /dev/null +++ b/tools/matedit/src/main.cpp @@ -0,0 +1,180 @@ +/* + * Copyright (C) 2024 The Android Open Source Project + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include + +#include "ExternalCompile.h" + +#include + +#include +#include +#include + +struct Config { + utils::Path inputFile; + utils::Path outputFile; + std::vector commandArgs; +}; + +static void printUsage(const char* name) { + std::string execName(utils::Path(name).getName()); + std::string usage( + "MATEDIT allows editing material files compiled with matc\n" + "\n" + "Caution! MATEDIT was designed to operate on trusted inputs. To minimize the risk of triggering\n" + "memory corruption vulnerabilities, please make sure that the files passed to MATEDIT come from a\n" + "trusted source, or run MATEDIT in a sandboxed environment.\n" + "\n" + "Usage:\n" + " MATEDIT [options] -o -i external-compile --