Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

iOS simulator shader compilation failure, subsequent crash #2352

Open
warmenhoven opened this issue Sep 26, 2024 · 0 comments
Open

iOS simulator shader compilation failure, subsequent crash #2352

warmenhoven opened this issue Sep 26, 2024 · 0 comments

Comments

@warmenhoven
Copy link
Contributor

warmenhoven commented Sep 26, 2024

As a result of c37a880, when I run in the iOS simulator (iPhone 16 Pro Max, iOS 18.0), the following shader is unable to compile; I've also attached a log:

#version 450

layout(local_size_x = 16, local_size_y = 8) in;

layout(set = 0, binding = 3, rgba8) writeonly uniform image2D uImage;
layout(set = 0, binding = 4, std430) readonly buffer SSBO
{
	uint packed_pixels[];
};

layout(set = 0, binding = 0, std140) uniform UBO
{
	uvec2 resolution;
	uint word_stride;
};

vec3 rgb565_to_rgb888(uint word)
{
	return vec3((uvec3(word) >> uvec3(11, 5, 0)) & uvec3(31, 63, 31)) / vec3(31.0, 63.0, 31.0);
}

void main()
{
	// We work on two horizonal pixels in parallel since we cannot rely on 16-bit storage.
	uvec2 first_input_pixel = gl_GlobalInvocationID.xy;
	uvec2 first_output_pixel = first_input_pixel * uvec2(2, 1);

	if (all(lessThan(first_output_pixel, resolution)))
	{
		uint word = packed_pixels[first_input_pixel.y * word_stride + first_input_pixel.x];
		uint lo_word = word & 0xffffu;
		uint hi_word = word >> 16u;
		vec3 lo = rgb565_to_rgb888(lo_word);
		vec3 hi = rgb565_to_rgb888(hi_word);

		imageStore(uImage, ivec2(first_output_pixel), vec4(lo, 1.0));
		if (first_output_pixel.x + 1u < resolution.x)
			imageStore(uImage, ivec2(first_output_pixel) + ivec2(1, 0), vec4(hi, 1.0));
	}
}

The compiler error is:

[mvk-error] VK_ERROR_INITIALIZATION_FAILED: Shader library compile failed (Error code 3):
program_source:27:71: error: cannot reserve 'buffer' resource location at index 9
kernel void main0(constant UBO& _49 [[buffer(9)]], const device SSBO& _65 [[buffer(9)]], texture2d<float, access::write> uImage [[texture(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
                                                                      ^
.
[mvk-error] VK_ERROR_INVALID_SHADER_NV: Compute shader function could not be compiled into pipeline. See previous logged error.

cross.log

Later, it leads to the following crash:

(lldb) bt
* thread #1, queue = 'com.apple.main-thread', stop reason = signal SIGSEGV
  * frame #0: 0x0000000133a606e0 MoltenVK`MVKPipeline::getDescriptorSetCount(this=0x0000000000000000) at MVKPipeline.h:146:44
    frame #1: 0x0000000133a60430 MoltenVK`MVKResourcesCommandEncoderState::encodeMetalArgumentBuffer(this=0x000000016ce012f0, stage=kMVKShaderStageCompute) at MVKCommandEncoderState.mm:669:29
    frame #2: 0x0000000133a63ef0 MoltenVK`MVKComputeResourcesCommandEncoderState::encodeImpl(this=0x000000016ce012f0, (null)=0) at MVKCommandEncoderState.mm:1163:2
    frame #3: 0x000000013395456c MoltenVK`MVKCommandEncoderState::encode(this=0x000000016ce012f0, stage=0) at MVKCommandEncoderState.h:93:9
    frame #4: 0x0000000133954828 MoltenVK`MVKCommandEncoder::finalizeDispatchState(this=0x000000016ce00450) at MVKCommandBuffer.mm:774:28
    frame #5: 0x0000000133ab4b24 MoltenVK`MVKCmdDispatch::encode(this=0x0000600000caaf10, cmdEncoder=0x000000016ce00450) at MVKCmdDispatch.mm:47:14
    frame #6: 0x00000001339521a0 MoltenVK`MVKCommandEncoder::encodeCommandsImpl(this=0x000000016ce00450, command=0x0000600000caaf10) at MVKCommandBuffer.mm:379:18
    frame #7: 0x00000001339519ac MoltenVK`MVKCommandEncoder::encodeCommands(this=0x000000016ce00450, command=0x0000600002981260) at MVKCommandBuffer.mm:372:3
    frame #8: 0x00000001339517c0 MoltenVK`MVKCommandEncoder::encode(this=0x000000016ce00450, mtlCmdBuff=0x000060000370dd40, pEncodingContext=0x000060000370de48) at MVKCommandBuffer.mm:346:5
    frame #9: 0x0000000133951a80 MoltenVK`MVKCommandBuffer::submit(this=0x0000000103db72b0, cmdBuffSubmit=0x000060000370de00, pEncodingContext=0x000060000370de48) at MVKCommandBuffer.mm:240:11
    frame #10: 0x0000000133a2b138 MoltenVK`MVKQueueFullCommandBufferSubmission<1ul>::submitCommandBuffers(this=0x000060000370de00) at MVKQueue.mm:659:59
    frame #11: 0x0000000133a28bfc MoltenVK`MVKQueueCommandBufferSubmission::execute(this=0x000060000370de00) at MVKQueue.mm:479:2
    frame #12: 0x0000000133a2590c MoltenVK`execute(qSubmit=0x000060000370de00) at MVKQueue.mm:72:98
    frame #13: 0x0000000133a25840 MoltenVK`MVKQueue::submit(this=0x0000000103e0ea30, qSubmit=0x000060000370de00) at MVKQueue.mm:99:10
    frame #14: 0x0000000133a263a0 MoltenVK`VkResult MVKQueue::submit<VkSubmitInfo>(this=0x0000000103e0ea30, submitCount=1, pSubmits=0x000000016ce030c0, fence=0x0000600003308b40, cmdUse=kMVKCommandUseQueueSubmit) at MVKQueue.mm:138:28
    frame #15: 0x00000001338d24a8 MoltenVK`vkQueueSubmit(queue=0x0000000103e0ea48, submitCount=1, pSubmits=0x000000016ce030c0, fence=0x0000600003308b40) at vulkan.mm:434:24
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

No branches or pull requests

1 participant