diff --git a/base/sources/backends/direct3d12_gpu.c b/base/sources/backends/direct3d12_gpu.c
index f02b64ec..4ab733e8 100644
--- a/base/sources/backends/direct3d12_gpu.c
+++ b/base/sources/backends/direct3d12_gpu.c
@@ -63,7 +63,6 @@
 
 ID3D12Device *device = NULL;
 static ID3D12RootSignature *globalRootSignature = NULL;
-static ID3D12RootSignature *globalComputeRootSignature = NULL;
 // extern ID3D12GraphicsCommandList* commandList;
 static int framebuffer_count = 0;
 
@@ -79,8 +78,6 @@ inline struct dx_window *iron_dx_current_window() {
 	return &dx_ctx.windows[0];
 }
 
-static bool compute_pipeline_set = false;
-
 ID3D12CommandQueue *commandQueue;
 
 struct RenderEnvironment {
@@ -266,100 +263,9 @@ static void createRootSignature() {
 	                                    &globalRootSignature);
 }
 
-static void createComputeRootSignature() {
-	ID3DBlob *rootBlob;
-	ID3DBlob *errorBlob;
-
-	D3D12_ROOT_PARAMETER parameters[4] = {};
-
-	D3D12_DESCRIPTOR_RANGE range;
-	range.RangeType = D3D12_DESCRIPTOR_RANGE_TYPE_SRV;
-	range.NumDescriptors = (UINT)IRON_INTERNAL_G5_TEXTURE_COUNT;
-	range.BaseShaderRegister = 0;
-	range.RegisterSpace = 0;
-	range.OffsetInDescriptorsFromTableStart = D3D12_DESCRIPTOR_RANGE_OFFSET_APPEND;
-	parameters[0].ParameterType = D3D12_ROOT_PARAMETER_TYPE_DESCRIPTOR_TABLE;
-	parameters[0].ShaderVisibility = D3D12_SHADER_VISIBILITY_ALL;
-	parameters[0].DescriptorTable.NumDescriptorRanges = 1;
-	parameters[0].DescriptorTable.pDescriptorRanges = ⦥
-
-	D3D12_DESCRIPTOR_RANGE uav_range;
-	uav_range.RangeType = D3D12_DESCRIPTOR_RANGE_TYPE_UAV;
-	uav_range.NumDescriptors = (UINT)IRON_INTERNAL_G5_TEXTURE_COUNT;
-	uav_range.BaseShaderRegister = 0;
-	uav_range.RegisterSpace = 0;
-	uav_range.OffsetInDescriptorsFromTableStart = D3D12_DESCRIPTOR_RANGE_OFFSET_APPEND;
-	parameters[1].ParameterType = D3D12_ROOT_PARAMETER_TYPE_DESCRIPTOR_TABLE;
-	parameters[1].ShaderVisibility = D3D12_SHADER_VISIBILITY_ALL;
-	parameters[1].DescriptorTable.NumDescriptorRanges = 1;
-	parameters[1].DescriptorTable.pDescriptorRanges = &uav_range;
-
-	D3D12_DESCRIPTOR_RANGE range2;
-	range2.RangeType = D3D12_DESCRIPTOR_RANGE_TYPE_SAMPLER;
-	range2.NumDescriptors = (UINT)IRON_INTERNAL_G5_TEXTURE_COUNT;
-	range2.BaseShaderRegister = 0;
-	range2.RegisterSpace = 0;
-	range2.OffsetInDescriptorsFromTableStart = D3D12_DESCRIPTOR_RANGE_OFFSET_APPEND;
-	parameters[2].ParameterType = D3D12_ROOT_PARAMETER_TYPE_DESCRIPTOR_TABLE;
-	parameters[2].ShaderVisibility = D3D12_SHADER_VISIBILITY_ALL;
-	parameters[2].DescriptorTable.NumDescriptorRanges = 1;
-	parameters[2].DescriptorTable.pDescriptorRanges = &range2;
-
-	parameters[3].ParameterType = D3D12_ROOT_PARAMETER_TYPE_CBV;
-	parameters[3].ShaderVisibility = D3D12_SHADER_VISIBILITY_ALL;
-	parameters[3].Descriptor.ShaderRegister = 0;
-	parameters[3].Descriptor.RegisterSpace = 0;
-
-	D3D12_STATIC_SAMPLER_DESC samplers[IRON_INTERNAL_G5_TEXTURE_COUNT * 2];
-	for (int i = 0; i < IRON_INTERNAL_G5_TEXTURE_COUNT; ++i) {
-		samplers[i].ShaderRegister = i;
-		samplers[i].Filter = D3D12_FILTER_MIN_MAG_MIP_POINT;
-		samplers[i].AddressU = D3D12_TEXTURE_ADDRESS_MODE_WRAP;
-		samplers[i].AddressV = D3D12_TEXTURE_ADDRESS_MODE_WRAP;
-		samplers[i].AddressW = D3D12_TEXTURE_ADDRESS_MODE_WRAP;
-		samplers[i].MipLODBias = 0;
-		samplers[i].MaxAnisotropy = 16;
-		samplers[i].ComparisonFunc = D3D12_COMPARISON_FUNC_NEVER;
-		samplers[i].BorderColor = D3D12_STATIC_BORDER_COLOR_OPAQUE_WHITE;
-		samplers[i].MinLOD = 0.0f;
-		samplers[i].MaxLOD = D3D12_FLOAT32_MAX;
-		samplers[i].ShaderVisibility = D3D12_SHADER_VISIBILITY_ALL;
-		samplers[i].RegisterSpace = 0;
-	}
-	for (int i = IRON_INTERNAL_G5_TEXTURE_COUNT; i < IRON_INTERNAL_G5_TEXTURE_COUNT * 2; ++i) {
-		samplers[i].ShaderRegister = i;
-		samplers[i].Filter = D3D12_FILTER_MIN_MAG_LINEAR_MIP_POINT;
-		samplers[i].AddressU = D3D12_TEXTURE_ADDRESS_MODE_WRAP;
-		samplers[i].AddressV = D3D12_TEXTURE_ADDRESS_MODE_WRAP;
-		samplers[i].AddressW = D3D12_TEXTURE_ADDRESS_MODE_WRAP;
-		samplers[i].MipLODBias = 0;
-		samplers[i].MaxAnisotropy = 16;
-		samplers[i].ComparisonFunc = D3D12_COMPARISON_FUNC_NEVER;
-		samplers[i].BorderColor = D3D12_STATIC_BORDER_COLOR_OPAQUE_WHITE;
-		samplers[i].MinLOD = 0.0f;
-		samplers[i].MaxLOD = D3D12_FLOAT32_MAX;
-		samplers[i].ShaderVisibility = D3D12_SHADER_VISIBILITY_ALL;
-		samplers[i].RegisterSpace = 0;
-	}
-
-	D3D12_ROOT_SIGNATURE_DESC descRootSignature;
-	descRootSignature.NumParameters = 4;
-	descRootSignature.pParameters = parameters;
-	descRootSignature.NumStaticSamplers = 0;
-	descRootSignature.pStaticSamplers = NULL;
-	descRootSignature.Flags = D3D12_ROOT_SIGNATURE_FLAG_ALLOW_INPUT_ASSEMBLER_INPUT_LAYOUT;
-
-	iron_microsoft_affirm(D3D12SerializeRootSignature(&descRootSignature, D3D_ROOT_SIGNATURE_VERSION_1, &rootBlob, &errorBlob));
-	device->lpVtbl->CreateRootSignature(device, 0, rootBlob->lpVtbl->GetBufferPointer(rootBlob), rootBlob->lpVtbl->GetBufferSize(rootBlob), &IID_ID3D12RootSignature,
-	                                    &globalComputeRootSignature);
-
-	// createSamplersAndHeaps();
-}
-
 static void initialize(struct dx_window *window) {
 	createDeviceAndSwapChain(window);
 	createRootSignature();
-	createComputeRootSignature();
 
 	device->lpVtbl->CreateFence(device, 0, D3D12_FENCE_FLAG_NONE, &IID_ID3D12Fence, &uploadFence);
 
@@ -435,7 +341,6 @@ void iron_gpu_internal_init() {
 	iron_microsoft_affirm(D3D12CreateDevice(NULL, D3D_FEATURE_LEVEL_11_0, &IID_ID3D12Device, &device));
 
 	createRootSignature();
-	createComputeRootSignature();
 
 	D3D12_COMMAND_QUEUE_DESC queueDesc = {};
 	queueDesc.Flags = D3D12_COMMAND_QUEUE_FLAG_NONE;
@@ -595,8 +500,6 @@ void iron_gpu_internal_reset_textures(struct iron_gpu_command_list *list);
 void iron_gpu_command_list_begin(struct iron_gpu_command_list *list) {
 	assert(!list->impl.open);
 
-	compute_pipeline_set = false;
-
 	if (list->impl.fence_value > 0) {
 		waitForFence(list->impl.fence, list->impl.fence_value, list->impl.fence_event);
 		list->impl._commandAllocator->lpVtbl->Reset(list->impl._commandAllocator);
@@ -718,11 +621,6 @@ void iron_gpu_command_list_set_fragment_constant_buffer(struct iron_gpu_command_
 	list->impl._commandList->lpVtbl->SetGraphicsRootConstantBufferView(list->impl._commandList, 3, buffer->impl.constant_buffer->lpVtbl->GetGPUVirtualAddress(buffer->impl.constant_buffer) + offset);
 }
 
-void iron_gpu_command_list_set_compute_constant_buffer(struct iron_gpu_command_list *list, iron_gpu_constant_buffer_t *buffer, int offset, size_t size) {
-	assert(list->impl.open);
-	list->impl._commandList->lpVtbl->SetComputeRootConstantBufferView(list->impl._commandList, 3, buffer->impl.constant_buffer->lpVtbl->GetGPUVirtualAddress(buffer->impl.constant_buffer) + offset);
-}
-
 void iron_gpu_command_list_draw_indexed_vertices(struct iron_gpu_command_list *list) {
 	iron_gpu_command_list_draw_indexed_vertices_from_to(list, 0, list->impl._indexCount);
 }
@@ -795,7 +693,6 @@ void iron_gpu_command_list_set_pipeline(struct iron_gpu_command_list *list, iron
 
 	list->impl._currentPipeline = pipeline;
 	list->impl._commandList->lpVtbl->SetPipelineState(list->impl._commandList, pipeline->impl.pso);
-	compute_pipeline_set = false;
 
 	for (int i = 0; i < IRON_INTERNAL_G5_TEXTURE_COUNT; ++i) {
 		list->impl.currentTextures[i] = NULL;
@@ -1002,23 +899,6 @@ void iron_gpu_command_list_get_render_target_pixels(iron_gpu_command_list_t *lis
 	render_target->impl.renderTargetReadback->lpVtbl->Unmap(render_target->impl.renderTargetReadback, 0, NULL);
 }
 
-void iron_gpu_internal_set_compute_constants(iron_gpu_command_list_t *commandList);
-
-void iron_gpu_command_list_set_compute_shader(iron_gpu_command_list_t *list, iron_gpu_compute_shader *shader) {
-	list->impl._commandList->lpVtbl->SetPipelineState(list->impl._commandList, shader->impl.pso);
-	compute_pipeline_set = true;
-
-	for (int i = 0; i < IRON_INTERNAL_G5_TEXTURE_COUNT; ++i) {
-		list->impl.currentTextures[i] = NULL;
-	}
-	iron_gpu_internal_set_compute_constants(list);
-}
-
-void iron_gpu_command_list_compute(iron_gpu_command_list_t *list, int x, int y, int z) {
-	assert(list->impl.open);
-	list->impl._commandList->lpVtbl->Dispatch(list->impl._commandList, x, y, z);
-}
-
 void iron_gpu_internal_sampler_set(iron_gpu_command_list_t *list, iron_gpu_sampler_t *sampler, int unit);
 
 void iron_gpu_command_list_set_sampler(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_sampler_t *sampler) {
@@ -1038,9 +918,6 @@ void iron_gpu_command_list_set_texture(iron_gpu_command_list_t *list, iron_gpu_t
 	else if (unit.stages[IRON_GPU_SHADER_TYPE_VERTEX] >= 0) {
 		texture->impl.stage = unit.stages[IRON_GPU_SHADER_TYPE_VERTEX];
 	}
-	else if (unit.stages[IRON_GPU_SHADER_TYPE_COMPUTE] >= 0) {
-		texture->impl.stage = unit.stages[IRON_GPU_SHADER_TYPE_COMPUTE];
-	}
 	list->impl.currentTextures[texture->impl.stage] = texture;
 	iron_gpu_internal_set_textures(list);
 }
@@ -1055,166 +932,6 @@ void iron_gpu_command_list_set_texture_from_render_target_depth(iron_gpu_command
 	list->impl.currentTextures[texture->impl.stage_depth] = texture;
 }
 
-void iron_gpu_compute_shader_init(iron_gpu_compute_shader *shader, void *_data, int length) {
-	unsigned index = 0;
-	uint8_t *data = (uint8_t *)_data;
-
-	memset(&shader->impl.attributes, 0, sizeof(shader->impl.attributes));
-	int attributesCount = data[index++];
-	for (int i = 0; i < attributesCount; ++i) {
-		unsigned char name[256];
-		for (unsigned i2 = 0; i2 < 255; ++i2) {
-			name[i2] = data[index++];
-			if (name[i2] == 0)
-				break;
-		}
-		shader->impl.attributes[i].hash = iron_internal_hash_name(name);
-		shader->impl.attributes[i].index = data[index++];
-	}
-
-	memset(&shader->impl.textures, 0, sizeof(shader->impl.textures));
-	uint8_t texCount = data[index++];
-	for (unsigned i = 0; i < texCount; ++i) {
-		unsigned char name[256];
-		for (unsigned i2 = 0; i2 < 255; ++i2) {
-			name[i2] = data[index++];
-			if (name[i2] == 0)
-				break;
-		}
-		shader->impl.textures[i].hash = iron_internal_hash_name(name);
-		shader->impl.textures[i].index = data[index++];
-	}
-
-	memset(&shader->impl.constants, 0, sizeof(shader->impl.constants));
-	uint8_t constantCount = data[index++];
-	shader->impl.constantsSize = 0;
-	for (unsigned i = 0; i < constantCount; ++i) {
-		unsigned char name[256];
-		for (unsigned i2 = 0; i2 < 255; ++i2) {
-			name[i2] = data[index++];
-			if (name[i2] == 0)
-				break;
-		}
-		iron_compute_internal_shader_constant_t constant;
-		constant.hash = iron_internal_hash_name(name);
-		memcpy(&constant.offset, &data[index], sizeof(constant.offset));
-		index += 4;
-		memcpy(&constant.size, &data[index], sizeof(constant.size));
-		index += 4;
-		constant.columns = data[index];
-		index += 1;
-		constant.rows = data[index];
-		index += 1;
-
-		shader->impl.constants[i] = constant;
-		shader->impl.constantsSize = constant.offset + constant.size;
-	}
-
-	shader->impl.length = (int)(length - index);
-	shader->impl.data = (uint8_t *)malloc(shader->impl.length);
-	assert(shader->impl.data != NULL);
-	memcpy(shader->impl.data, &data[index], shader->impl.length);
-
-	D3D12_COMPUTE_PIPELINE_STATE_DESC desc = {0};
-	desc.CS.BytecodeLength = shader->impl.length;
-	desc.CS.pShaderBytecode = shader->impl.data;
-	desc.pRootSignature = globalComputeRootSignature;
-	HRESULT hr = device->lpVtbl->CreateComputePipelineState(device , &desc, &IID_ID3D12PipelineState, &shader->impl.pso);
-
-	if (hr != S_OK) {
-		iron_log("Could not initialize compute shader.");
-		return;
-	}
-
-	// iron_microsoft_affirm(device->CreateBuffer(&CD3D11_BUFFER_DESC(getMultipleOf16(shader->impl.constantsSize), D3D11_BIND_CONSTANT_BUFFER), nullptr,
-	// &shader->impl.constantBuffer));
-}
-
-void iron_gpu_compute_shader_destroy(iron_gpu_compute_shader *shader) {
-	if (shader->impl.pso != NULL) {
-		shader->impl.pso->lpVtbl->Release(shader->impl.pso);
-		shader->impl.pso = NULL;
-	}
-}
-
-static iron_compute_internal_shader_constant_t *findComputeConstant(iron_compute_internal_shader_constant_t *constants, uint32_t hash) {
-	for (int i = 0; i < 64; ++i) {
-		if (constants[i].hash == hash) {
-			return &constants[i];
-		}
-	}
-	return NULL;
-}
-
-static iron_internal_hash_index_t *findComputeTextureUnit(iron_internal_hash_index_t *units, uint32_t hash) {
-	for (int i = 0; i < 64; ++i) {
-		if (units[i].hash == hash) {
-			return &units[i];
-		}
-	}
-	return NULL;
-}
-
-iron_gpu_constant_location_t iron_gpu_compute_shader_get_constant_location(iron_gpu_compute_shader *shader, const char *name) {
-	iron_gpu_constant_location_t location = {0};
-
-	uint32_t hash = iron_internal_hash_name((unsigned char *)name);
-
-	iron_compute_internal_shader_constant_t *constant = findComputeConstant(shader->impl.constants, hash);
-	if (constant == NULL) {
-		location.impl.computeOffset = 0;
-		location.impl.computeSize = 0;
-	}
-	else {
-		location.impl.computeOffset = constant->offset;
-		location.impl.computeSize = constant->size;
-	}
-
-	if (location.impl.computeSize == 0) {
-		iron_log("Uniform %s not found.", name);
-	}
-
-	return location;
-}
-
-iron_gpu_texture_unit_t iron_gpu_compute_shader_get_texture_unit(iron_gpu_compute_shader *shader, const char *name) {
-	char unitName[64];
-	int unitOffset = 0;
-	size_t len = strlen(name);
-	if (len > 63)
-		len = 63;
-	strncpy(unitName, name, len + 1);
-	if (unitName[len - 1] == ']') {                  // Check for array - mySampler[2]
-		unitOffset = (int)(unitName[len - 2] - '0'); // Array index is unit offset
-		unitName[len - 3] = 0;                       // Strip array from name
-	}
-
-	uint32_t hash = iron_internal_hash_name((unsigned char *)unitName);
-
-	iron_gpu_texture_unit_t unit;
-	for (int i = 0; i < IRON_GPU_SHADER_TYPE_COUNT; ++i) {
-		unit.stages[i] = -1;
-	}
-	iron_internal_hash_index_t *computeUnit = findComputeTextureUnit(shader->impl.textures, hash);
-	if (computeUnit == NULL) {
-#ifndef NDEBUG
-		static int notFoundCount = 0;
-		if (notFoundCount < 10) {
-			iron_log("Sampler %s not found.", unitName);
-			++notFoundCount;
-		}
-		else if (notFoundCount == 10) {
-			iron_log("Giving up on sampler not found messages.", unitName);
-			++notFoundCount;
-		}
-#endif
-	}
-	else {
-		unit.stages[IRON_GPU_SHADER_TYPE_COMPUTE] = computeUnit->index + unitOffset;
-	}
-	return unit;
-}
-
 void iron_gpu_internal_setConstants(iron_gpu_command_list_t *commandList, iron_gpu_pipeline_t *pipeline) {
 	commandList->impl._commandList->lpVtbl->SetGraphicsRootSignature(commandList->impl._commandList, globalRootSignature);
 
@@ -1223,15 +940,6 @@ void iron_gpu_internal_setConstants(iron_gpu_command_list_t *commandList, iron_g
 	}
 }
 
-void iron_gpu_internal_set_compute_constants(iron_gpu_command_list_t *commandList) {
-
-	commandList->impl._commandList->lpVtbl->SetComputeRootSignature(commandList->impl._commandList, globalComputeRootSignature);
-
-	//if (pipeline->impl.textures > 0) {
-		iron_gpu_internal_set_textures(commandList);
-	//}
-}
-
 void iron_gpu_pipeline_init(iron_gpu_pipeline_t *pipe) {
 	iron_gpu_internal_pipeline_set_defaults(pipe);
 	iron_gpu_internal_pipeline_init(pipe);
@@ -1303,9 +1011,6 @@ iron_gpu_constant_location_t iron_gpu_pipeline_get_constant_location(struct iron
 		location.impl.fragmentSize = constant.size;
 	}
 
-	location.impl.computeOffset = 0;
-	location.impl.computeSize = 0;
-
 	return location;
 }
 
@@ -1846,15 +1551,8 @@ void iron_gpu_internal_set_textures(iron_gpu_command_list_t *list) {
 
 		ID3D12DescriptorHeap *heaps[2] = {list->impl.srvHeap, list->impl.samplerHeap};
 		list->impl._commandList->lpVtbl->SetDescriptorHeaps(list->impl._commandList, 2, heaps);
-		if (compute_pipeline_set) {
-			list->impl._commandList->lpVtbl->SetComputeRootDescriptorTable(list->impl._commandList, 0, srvGpu);
-			list->impl._commandList->lpVtbl->SetComputeRootDescriptorTable(list->impl._commandList, 1, srvGpu);
-			list->impl._commandList->lpVtbl->SetComputeRootDescriptorTable(list->impl._commandList, 2, samplerGpu);
-		}
-		else {
-			list->impl._commandList->lpVtbl->SetGraphicsRootDescriptorTable(list->impl._commandList, 0, srvGpu);
-			list->impl._commandList->lpVtbl->SetGraphicsRootDescriptorTable(list->impl._commandList, 1, samplerGpu);
-		}
+		list->impl._commandList->lpVtbl->SetGraphicsRootDescriptorTable(list->impl._commandList, 0, srvGpu);
+		list->impl._commandList->lpVtbl->SetGraphicsRootDescriptorTable(list->impl._commandList, 1, samplerGpu);
 	}
 }
 
diff --git a/base/sources/backends/direct3d12_gpu.h b/base/sources/backends/direct3d12_gpu.h
index 55d38c78..feb89dfb 100644
--- a/base/sources/backends/direct3d12_gpu.h
+++ b/base/sources/backends/direct3d12_gpu.h
@@ -72,30 +72,11 @@ typedef struct {
 	struct ID3D12DescriptorHeap *samplerHeap;
 } CommandList5Impl;
 
-typedef struct {
-	uint32_t hash;
-	uint32_t offset;
-	uint32_t size;
-	uint8_t columns;
-	uint8_t rows;
-} iron_compute_internal_shader_constant_t;
-
 typedef struct {
 	uint32_t hash;
 	uint32_t index;
 } iron_internal_hash_index_t;
 
-typedef struct iron_gpu_compute_shader_impl {
-	iron_compute_internal_shader_constant_t constants[64];
-	int constantsSize;
-	iron_internal_hash_index_t attributes[64];
-	iron_internal_hash_index_t textures[64];
-	uint8_t *data;
-	int length;
-	struct ID3D12Buffer *constantBuffer;
-	struct ID3D12PipelineState *pso;
-} iron_gpu_compute_shader_impl;
-
 struct iron_gpu_shader;
 
 struct ID3D12PipelineState;
@@ -108,18 +89,11 @@ typedef struct {
 	int textures;
 } PipelineState5Impl;
 
-typedef struct {
-	struct ID3D12PipelineState *pso;
-	int textures;
-} ComputePipelineState5Impl;
-
 typedef struct {
 	int vertexOffset;
 	uint32_t vertexSize;
 	int fragmentOffset;
 	uint32_t fragmentSize;
-	int computeOffset;
-	uint32_t computeSize;
 	int geometryOffset;
 	uint32_t geometrySize;
 	int tessEvalOffset;
diff --git a/base/sources/backends/metal_gpu.h b/base/sources/backends/metal_gpu.h
index f1e72d87..4a37b86e 100644
--- a/base/sources/backends/metal_gpu.h
+++ b/base/sources/backends/metal_gpu.h
@@ -3,21 +3,12 @@
 #include <iron_gpu.h>
 #include <iron_math.h>
 
-#pragma once
-
 struct iron_gpu_index_buffer;
 
 typedef struct {
 	struct iron_gpu_index_buffer *current_index_buffer;
 } CommandList5Impl;
 
-typedef struct iron_gpu_compute_shader_impl {
-	char name[1024];
-	void *_function;
-	void *_pipeline;
-	void *_reflection;
-} iron_gpu_compute_shader_impl;
-
 struct iron_gpu_shader;
 
 typedef struct {
@@ -30,14 +21,9 @@ typedef struct {
 	void *_depthStencilNone;
 } PipelineState5Impl;
 
-typedef struct {
-	int a;
-} ComputePipelineState5Impl;
-
 typedef struct {
 	int vertexOffset;
 	int fragmentOffset;
-	int computeOffset;
 } ConstantLocation5Impl;
 
 typedef struct iron_gpu_sampler_impl {
diff --git a/base/sources/backends/metal_gpu.m b/base/sources/backends/metal_gpu.m
index e61472b8..6f6c7d51 100644
--- a/base/sources/backends/metal_gpu.m
+++ b/base/sources/backends/metal_gpu.m
@@ -9,7 +9,6 @@
 
 static id<MTLCommandBuffer> command_buffer = nil;
 static id<MTLRenderCommandEncoder> render_command_encoder = nil;
-static id<MTLComputeCommandEncoder> compute_command_encoder = nil;
 
 static void start_render_pass(void);
 static void end_render_pass(void);
@@ -472,12 +471,6 @@ void iron_gpu_command_list_set_fragment_constant_buffer(iron_gpu_command_list_t
 	[encoder setFragmentBuffer:buf offset:offset atIndex:0];
 }
 
-void iron_gpu_command_list_set_compute_constant_buffer(iron_gpu_command_list_t *list, struct iron_gpu_constant_buffer *buffer, int offset, size_t size) {
-	assert(compute_command_encoder != nil);
-	id<MTLBuffer> buf = (__bridge id<MTLBuffer>)buffer->impl._buffer;
-	[compute_command_encoder setBuffer:buf offset:offset atIndex:1];
-}
-
 void iron_gpu_command_list_render_target_to_texture_barrier(iron_gpu_command_list_t *list, struct iron_gpu_texture *renderTarget) {
 }
 
@@ -485,16 +478,11 @@ void iron_gpu_command_list_texture_to_render_target_barrier(iron_gpu_command_lis
 
 void iron_gpu_command_list_set_texture(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_texture_t *texture) {
 	id<MTLTexture> tex = (__bridge id<MTLTexture>)texture->impl._tex;
-	if (compute_command_encoder != nil) {
-		[compute_command_encoder setTexture:tex atIndex:unit.stages[IRON_GPU_SHADER_TYPE_COMPUTE]];
+	if (unit.stages[IRON_GPU_SHADER_TYPE_VERTEX] >= 0) {
+		[render_command_encoder setVertexTexture:tex atIndex:unit.stages[IRON_GPU_SHADER_TYPE_VERTEX]];
 	}
-	else {
-		if (unit.stages[IRON_GPU_SHADER_TYPE_VERTEX] >= 0) {
-			[render_command_encoder setVertexTexture:tex atIndex:unit.stages[IRON_GPU_SHADER_TYPE_VERTEX]];
-		}
-		if (unit.stages[IRON_GPU_SHADER_TYPE_FRAGMENT] >= 0) {
-			[render_command_encoder setFragmentTexture:tex atIndex:unit.stages[IRON_GPU_SHADER_TYPE_FRAGMENT]];
-		}
+	if (unit.stages[IRON_GPU_SHADER_TYPE_FRAGMENT] >= 0) {
+		[render_command_encoder setFragmentTexture:tex atIndex:unit.stages[IRON_GPU_SHADER_TYPE_FRAGMENT]];
 	}
 }
 
@@ -521,140 +509,6 @@ void iron_gpu_command_list_set_sampler(iron_gpu_command_list_t *list, iron_gpu_t
 	}
 }
 
-void iron_gpu_command_list_set_compute_shader(iron_gpu_command_list_t *list, iron_gpu_compute_shader *shader) {
-	if (compute_command_encoder == nil) {
-		end_render_pass();
-		compute_command_encoder = [command_buffer computeCommandEncoder];
-	}
-
-	id<MTLComputePipelineState> pipeline = (__bridge id<MTLComputePipelineState>)shader->impl._pipeline;
-	[compute_command_encoder setComputePipelineState:pipeline];
-}
-
-void iron_gpu_command_list_compute(iron_gpu_command_list_t *list, int x, int y, int z) {
-	assert(compute_command_encoder != nil);
-
-	MTLSize perGrid;
-	perGrid.width = x;
-	perGrid.height = y;
-	perGrid.depth = z;
-	MTLSize perGroup;
-	perGroup.width = 16;
-	perGroup.height = 16;
-	perGroup.depth = 1;
-	[compute_command_encoder dispatchThreadgroups:perGrid threadsPerThreadgroup:perGroup];
-
-	[compute_command_encoder endEncoding];
-
-	compute_command_encoder = nil;
-
-	start_render_pass();
-}
-
-void iron_gpu_compute_shader_init(iron_gpu_compute_shader *shader, void *_data, int length) {
-	shader->impl.name[0] = 0;
-
-	{
-		uint8_t *data = (uint8_t *)_data;
-		if (length > 1 && data[0] == '>') {
-			memcpy(shader->impl.name, data + 1, length - 1);
-			shader->impl.name[length - 1] = 0;
-		}
-		else {
-			for (int i = 3; i < length; ++i) {
-				if (data[i] == '\n') {
-					shader->impl.name[i - 3] = 0;
-					break;
-				}
-				else {
-					shader->impl.name[i - 3] = data[i];
-				}
-			}
-		}
-	}
-
-	char *data = (char *)_data;
-	id<MTLLibrary> library = nil;
-	if (length > 1 && data[0] == '>') {
-		library = getMetalLibrary();
-	}
-	else {
-		id<MTLDevice> device = getMetalDevice();
-		library = [device newLibraryWithSource:[[NSString alloc] initWithBytes:data length:length encoding:NSUTF8StringEncoding] options:nil error:nil];
-	}
-	id<MTLFunction> function = [library newFunctionWithName:[NSString stringWithCString:shader->impl.name encoding:NSUTF8StringEncoding]];
-	assert(function != nil);
-	shader->impl._function = (__bridge_retained void *)function;
-
-	id<MTLDevice> device = getMetalDevice();
-	MTLComputePipelineReflection *reflection = nil;
-	NSError *error = nil;
-	shader->impl._pipeline = (__bridge_retained void *)[device newComputePipelineStateWithFunction:function
-	                                                                                       options:MTLPipelineOptionBufferTypeInfo
-	                                                                                    reflection:&reflection
-	                                                                                         error:&error];
-	if (error != nil)
-		NSLog(@"%@", [error localizedDescription]);
-	assert(shader->impl._pipeline != NULL && !error);
-	shader->impl._reflection = (__bridge_retained void *)reflection;
-}
-
-void iron_gpu_compute_shader_destroy(iron_gpu_compute_shader *shader) {
-	id<MTLFunction> function = (__bridge_transfer id<MTLFunction>)shader->impl._function;
-	function = nil;
-	shader->impl._function = NULL;
-
-	id<MTLComputePipelineState> pipeline = (__bridge_transfer id<MTLComputePipelineState>)shader->impl._pipeline;
-	pipeline = nil;
-	shader->impl._pipeline = NULL;
-
-	MTLComputePipelineReflection *reflection = (__bridge_transfer MTLComputePipelineReflection *)shader->impl._reflection;
-	reflection = nil;
-	shader->impl._reflection = NULL;
-}
-
-iron_gpu_constant_location_t iron_gpu_compute_shader_get_constant_location(iron_gpu_compute_shader *shader, const char *name) {
-	iron_gpu_constant_location_t location;
-	location.impl.vertexOffset = -1;
-	location.impl.fragmentOffset = -1;
-	location.impl.computeOffset = -1;
-
-	MTLComputePipelineReflection *reflection = (__bridge MTLComputePipelineReflection *)shader->impl._reflection;
-
-	for (MTLArgument *arg in reflection.arguments) {
-		if (arg.type == MTLArgumentTypeBuffer && [arg.name isEqualToString:@"uniforms"]) {
-			if ([arg bufferDataType] == MTLDataTypeStruct) {
-				MTLStructType *structObj = [arg bufferStructType];
-				for (MTLStructMember *member in structObj.members) {
-					if (strcmp([[member name] UTF8String], name) == 0) {
-						location.impl.computeOffset = (int)[member offset];
-						break;
-					}
-				}
-			}
-			break;
-		}
-	}
-
-	return location;
-}
-
-iron_gpu_texture_unit_t iron_gpu_compute_shader_get_texture_unit(iron_gpu_compute_shader *shader, const char *name) {
-	iron_gpu_texture_unit_t unit;
-	for (int i = 0; i < IRON_GPU_SHADER_TYPE_COUNT; ++i) {
-		unit.stages[i] = -1;
-	}
-
-	MTLComputePipelineReflection *reflection = (__bridge MTLComputePipelineReflection *)shader->impl._reflection;
-	for (MTLArgument *arg in reflection.arguments) {
-		if ([arg type] == MTLArgumentTypeTexture && strcmp([[arg name] UTF8String], name) == 0) {
-			unit.stages[IRON_GPU_SHADER_TYPE_COMPUTE] = (int)[arg index];
-		}
-	}
-
-	return unit;
-}
-
 static MTLBlendFactor convert_blending_factor(iron_gpu_blending_factor_t factor) {
 	switch (factor) {
 	case IRON_GPU_BLEND_ONE:
@@ -920,7 +774,6 @@ iron_gpu_constant_location_t iron_gpu_pipeline_get_constant_location(iron_gpu_pi
 	iron_gpu_constant_location_t location;
 	location.impl.vertexOffset = -1;
 	location.impl.fragmentOffset = -1;
-	location.impl.computeOffset = -1;
 
 	MTLRenderPipelineReflection *reflection = (__bridge MTLRenderPipelineReflection *)pipeline->impl._reflection;
 
diff --git a/base/sources/backends/vulkan_gpu.c b/base/sources/backends/vulkan_gpu.c
index 418d9c21..17e0e753 100644
--- a/base/sources/backends/vulkan_gpu.c
+++ b/base/sources/backends/vulkan_gpu.c
@@ -19,7 +19,6 @@
 static VkSemaphore framebuffer_available;
 static VkSemaphore relay_semaphore;
 static bool wait_for_relay = false;
-static VkDescriptorSetLayout compute_descriptor_layout;
 static int framebuffer_count = 0;
 
 iron_gpu_texture_t *vulkan_textures[16] = {
@@ -35,7 +34,6 @@ VkSampler vulkan_samplers[16] = {
 };
 
 VkDescriptorSet get_descriptor_set(void);
-static VkDescriptorSet get_compute_descriptor_set(void);
 
 VkRenderPassBeginInfo current_render_pass_begin_info;
 VkPipeline current_vulkan_pipeline;
@@ -45,9 +43,7 @@ iron_gpu_texture_t *current_render_targets[8] = {
 
 static uint32_t last_vertex_constant_buffer_offset = 0;
 static uint32_t last_fragment_constant_buffer_offset = 0;
-static uint32_t last_compute_constant_buffer_offset = 0;
 static iron_gpu_pipeline_t *current_pipeline = NULL;
-static iron_gpu_compute_shader *current_compute_shader = NULL;
 static int mrt_index = 0;
 static VkFramebuffer mrt_framebuffer[16];
 static VkRenderPass mrt_render_pass[16];
@@ -57,7 +53,6 @@ static bool in_render_pass = false;
 static bool wait_for_framebuffer = false;
 bool iron_gpu_transpose_mat = true;
 
-static VkDescriptorPool compute_descriptor_pool;
 VkDescriptorSetLayout desc_layout;
 static VkDescriptorPool descriptor_pool;
 
@@ -95,11 +90,6 @@ struct descriptor_set {
 static struct descriptor_set descriptor_sets[MAX_DESCRIPTOR_SETS] = {0};
 static int descriptor_sets_count = 0;
 
-static struct descriptor_set compute_descriptor_sets[MAX_DESCRIPTOR_SETS] = {0};
-static int compute_descriptor_sets_count = 0;
-
-
-
 struct vk_funs vk = {0};
 struct vk_context vk_ctx = {0};
 
@@ -117,7 +107,6 @@ void iron_gpu_internal_resize(int, int);
 	}
 
 void create_descriptor_layout(void);
-static void create_compute_descriptor_layout(void);
 
 static bool began = false;
 static VkPhysicalDeviceProperties gpu_props;
@@ -1107,7 +1096,6 @@ void iron_gpu_internal_init() {
 	err = vkCreateCommandPool(vk_ctx.device, &cmd_pool_info, NULL, &vk_ctx.cmd_pool);
 
 	create_descriptor_layout();
-	create_compute_descriptor_layout();
 	assert(!err);
 
 	VkSemaphoreCreateInfo semInfo = {0};
@@ -1256,7 +1244,6 @@ void iron_gpu_end() {
 	}
 
 	reuse_descriptor_sets();
-	reuse_compute_descriptor_sets();
 	began = false;
 }
 
@@ -1921,15 +1908,6 @@ void iron_gpu_command_list_set_fragment_constant_buffer(iron_gpu_command_list_t
 	vkCmdBindDescriptorSets(list->impl._buffer, VK_PIPELINE_BIND_POINT_GRAPHICS, current_pipeline->impl.pipeline_layout, 0, 1, &descriptor_set, 2, offsets);
 }
 
-void iron_gpu_command_list_set_compute_constant_buffer(iron_gpu_command_list_t *list, struct iron_gpu_constant_buffer *buffer, int offset, size_t size) {
-	last_compute_constant_buffer_offset = offset;
-
-	VkDescriptorSet descriptor_set = get_compute_descriptor_set();
-	uint32_t offsets[2] = {last_compute_constant_buffer_offset, last_compute_constant_buffer_offset};
-	vkCmdBindDescriptorSets(list->impl._buffer, VK_PIPELINE_BIND_POINT_COMPUTE, current_compute_shader->impl.pipeline_layout, 0, 1, &descriptor_set, 2,
-	                        offsets);
-}
-
 void iron_gpu_command_list_execute(iron_gpu_command_list_t *list) {
 	// Make sure the previous execution is done, so we can reuse the fence
 	// Not optimal of course
@@ -1999,31 +1977,6 @@ void iron_gpu_command_list_set_texture_from_render_target_depth(iron_gpu_command
 	}
 }
 
-void iron_gpu_command_list_set_compute_shader(iron_gpu_command_list_t *list, iron_gpu_compute_shader *shader) {
-	current_compute_shader = shader;
-	vkCmdBindPipeline(list->impl._buffer, VK_PIPELINE_BIND_POINT_COMPUTE, shader->impl.pipeline);
-}
-
-void iron_gpu_command_list_compute(iron_gpu_command_list_t *list, int x, int y, int z) {
-	if (in_render_pass) {
-		vkCmdEndRenderPass(list->impl._buffer);
-		in_render_pass = false;
-	}
-
-	vkCmdDispatch(list->impl._buffer, x, y, z);
-
-	int render_target_count = 0;
-	for (int i = 0; i < 8; ++i) {
-		if (current_render_targets[i] == NULL) {
-			break;
-		}
-		++render_target_count;
-	}
-	if (render_target_count > 0) {
-		iron_gpu_command_list_set_render_targets(list, current_render_targets, render_target_count);
-	}
-}
-
 static bool has_number(iron_internal_named_number *named_numbers, const char *name) {
 	for (int i = 0; i < IRON_INTERNAL_NAMED_NUMBER_COUNT; ++i) {
 		if (strcmp(named_numbers[i].name, name) == 0) {
@@ -2217,128 +2170,6 @@ static VkShaderModule create_shader_module(const void *code, size_t size) {
 	return module;
 }
 
-static void create_compute_descriptor_layout(void) {
-	VkDescriptorSetLayoutBinding layoutBindings[18];
-	memset(layoutBindings, 0, sizeof(layoutBindings));
-
-	layoutBindings[0].binding = 0;
-	layoutBindings[0].descriptorType = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER_DYNAMIC;
-	layoutBindings[0].descriptorCount = 1;
-	layoutBindings[0].stageFlags = VK_SHADER_STAGE_COMPUTE_BIT;
-	layoutBindings[0].pImmutableSamplers = NULL;
-
-	layoutBindings[1].binding = 1;
-	layoutBindings[1].descriptorType = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER_DYNAMIC;
-	layoutBindings[1].descriptorCount = 1;
-	layoutBindings[1].stageFlags = VK_SHADER_STAGE_COMPUTE_BIT;
-	layoutBindings[1].pImmutableSamplers = NULL;
-
-	for (int i = 2; i < 18; ++i) {
-		layoutBindings[i].binding = i;
-		layoutBindings[i].descriptorType = VK_DESCRIPTOR_TYPE_STORAGE_IMAGE;
-		layoutBindings[i].descriptorCount = 1;
-		layoutBindings[i].stageFlags = VK_SHADER_STAGE_COMPUTE_BIT;
-		layoutBindings[i].pImmutableSamplers = NULL;
-	}
-
-	VkDescriptorSetLayoutCreateInfo descriptor_layout = {0};
-	descriptor_layout.sType = VK_STRUCTURE_TYPE_DESCRIPTOR_SET_LAYOUT_CREATE_INFO;
-	descriptor_layout.pNext = NULL;
-	descriptor_layout.bindingCount = 18;
-	descriptor_layout.pBindings = layoutBindings;
-
-	VkResult err = vkCreateDescriptorSetLayout(vk_ctx.device, &descriptor_layout, NULL, &compute_descriptor_layout);
-	assert(!err);
-
-	VkDescriptorPoolSize typeCounts[2];
-	memset(typeCounts, 0, sizeof(typeCounts));
-
-	typeCounts[0].type = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER_DYNAMIC;
-	typeCounts[0].descriptorCount = 2 * 1024;
-
-	typeCounts[1].type = VK_DESCRIPTOR_TYPE_STORAGE_IMAGE;
-	typeCounts[1].descriptorCount = 16 * 1024;
-
-	VkDescriptorPoolCreateInfo pool_info = {0};
-	pool_info.sType = VK_STRUCTURE_TYPE_DESCRIPTOR_POOL_CREATE_INFO;
-	pool_info.pNext = NULL;
-	pool_info.maxSets = 1024;
-	pool_info.poolSizeCount = 2;
-	pool_info.pPoolSizes = typeCounts;
-
-	err = vkCreateDescriptorPool(vk_ctx.device, &pool_info, NULL, &compute_descriptor_pool);
-	assert(!err);
-}
-
-void iron_gpu_compute_shader_init(iron_gpu_compute_shader *shader, void *_data, int length) {
-	memset(shader->impl.locations, 0, sizeof(iron_internal_named_number) * IRON_INTERNAL_NAMED_NUMBER_COUNT);
-	memset(shader->impl.offsets, 0, sizeof(iron_internal_named_number) * IRON_INTERNAL_NAMED_NUMBER_COUNT);
-	memset(shader->impl.texture_bindings, 0, sizeof(iron_internal_named_number) * IRON_INTERNAL_NAMED_NUMBER_COUNT);
-	parse_shader((uint32_t *)_data, length, shader->impl.locations, shader->impl.texture_bindings, shader->impl.offsets);
-
-	VkPipelineLayoutCreateInfo pPipelineLayoutCreateInfo = {0};
-	pPipelineLayoutCreateInfo.sType = VK_STRUCTURE_TYPE_PIPELINE_LAYOUT_CREATE_INFO;
-	pPipelineLayoutCreateInfo.pNext = NULL;
-	pPipelineLayoutCreateInfo.setLayoutCount = 1;
-	pPipelineLayoutCreateInfo.pSetLayouts = &compute_descriptor_layout;
-
-	VkResult err = vkCreatePipelineLayout(vk_ctx.device, &pPipelineLayoutCreateInfo, NULL, &shader->impl.pipeline_layout);
-	assert(!err);
-
-	VkComputePipelineCreateInfo pipeline_info = {0};
-
-	memset(&pipeline_info, 0, sizeof(pipeline_info));
-	pipeline_info.sType = VK_STRUCTURE_TYPE_COMPUTE_PIPELINE_CREATE_INFO;
-	pipeline_info.layout = shader->impl.pipeline_layout;
-
-	VkPipelineShaderStageCreateInfo shader_stage;
-	memset(&shader_stage, 0, sizeof(VkPipelineShaderStageCreateInfo));
-
-	shader_stage.sType = VK_STRUCTURE_TYPE_PIPELINE_SHADER_STAGE_CREATE_INFO;
-	shader_stage.stage = VK_SHADER_STAGE_COMPUTE_BIT;
-	shader->impl.shader_module = create_shader_module(_data, (size_t)length);
-	shader_stage.module = shader->impl.shader_module;
-	shader_stage.pName = "main";
-
-	pipeline_info.stage = shader_stage;
-
-	err = vkCreateComputePipelines(vk_ctx.device, VK_NULL_HANDLE, 1, &pipeline_info, NULL, &shader->impl.pipeline);
-	assert(!err);
-
-	vkDestroyShaderModule(vk_ctx.device, shader->impl.shader_module, NULL);
-}
-
-void iron_gpu_compute_shader_destroy(iron_gpu_compute_shader *shader) {}
-
-iron_gpu_constant_location_t iron_gpu_compute_shader_get_constant_location(iron_gpu_compute_shader *shader, const char *name) {
-	iron_gpu_constant_location_t location = {0};
-	uint32_t hash = iron_hash_djb2((unsigned char *)name);
-	return location;
-}
-
-iron_gpu_texture_unit_t iron_gpu_compute_shader_get_texture_unit(iron_gpu_compute_shader *shader, const char *name) {
-	char unitName[64];
-	int unitOffset = 0;
-	size_t len = strlen(name);
-	if (len > 63)
-		len = 63;
-	strncpy(unitName, name, len + 1);
-	if (unitName[len - 1] == ']') {                  // Check for array - mySampler[2]
-		unitOffset = (int)(unitName[len - 2] - '0'); // Array index is unit offset
-		unitName[len - 3] = 0;                       // Strip array from name
-	}
-
-	uint32_t hash = iron_hash_djb2((unsigned char *)unitName);
-
-	iron_gpu_texture_unit_t unit;
-	for (int i = 0; i < IRON_GPU_SHADER_TYPE_COUNT; ++i) {
-		unit.stages[i] = -1;
-	}
-	unit.stages[IRON_GPU_SHADER_TYPE_COMPUTE] = 0;
-
-	return unit;
-}
-
 static VkShaderModule prepare_vs(VkShaderModule *vert_shader_module, iron_gpu_shader_t *vertex_shader) {
 	*vert_shader_module = create_shader_module(vertex_shader->impl.source, vertex_shader->impl.length);
 	return *vert_shader_module;
@@ -2385,7 +2216,6 @@ iron_gpu_constant_location_t iron_gpu_pipeline_get_constant_location(iron_gpu_pi
 	iron_gpu_constant_location_t location;
 	location.impl.vertexOffset = -1;
 	location.impl.fragmentOffset = -1;
-	location.impl.computeOffset = -1;
 	if (has_number(pipeline->impl.vertexOffsets, name)) {
 		location.impl.vertexOffset = find_number(pipeline->impl.vertexOffsets, name);
 	}
@@ -2840,19 +2670,6 @@ int calc_descriptor_id(void) {
 	return 1 | (texture_count << 1) | ((uniform_buffer ? 1 : 0) << 8);
 }
 
-int calc_compute_descriptor_id(void) {
-	int texture_count = 0;
-	for (int i = 0; i < 16; ++i) {
-		if (vulkan_textures[i] != NULL) {
-			texture_count++;
-		}
-	}
-
-	bool uniform_buffer = vk_ctx.compute_uniform_buffer != NULL;
-
-	return 1 | (texture_count << 1) | ((uniform_buffer ? 1 : 0) << 8);
-}
-
 static int write_tex_descs(VkDescriptorImageInfo *tex_descs) {
 	memset(tex_descs, 0, sizeof(VkDescriptorImageInfo) * 16);
 
@@ -3020,175 +2837,6 @@ VkDescriptorSet get_descriptor_set() {
 	return descriptor_set;
 }
 
-static int write_compute_tex_descs(VkDescriptorImageInfo *tex_descs) {
-	memset(tex_descs, 0, sizeof(VkDescriptorImageInfo) * 16);
-
-	int texture_count = 0;
-	for (int i = 0; i < 16; ++i) {
-		if (vulkan_textures[i] != NULL) {
-			tex_descs[i].sampler = vulkan_samplers[i];
-			if (vulkan_textures[i]->impl.stage_depth == i) {
-				tex_descs[i].imageView = vulkan_textures[i]->impl.depthView;
-				vulkan_textures[i]->impl.stage_depth = -1;
-			}
-			else {
-				tex_descs[i].imageView = vulkan_textures[i]->impl.view;
-			}
-			texture_count++;
-		}
-		tex_descs[i].imageLayout = VK_IMAGE_LAYOUT_GENERAL;
-	}
-	return texture_count;
-}
-
-static bool compute_textures_changed(struct descriptor_set *set) {
-	VkDescriptorImageInfo tex_desc[16];
-
-	write_compute_tex_descs(tex_desc);
-
-	return memcmp(&tex_desc, &set->tex_desc, sizeof(tex_desc)) != 0;
-}
-
-static void update_compute_textures(struct descriptor_set *set) {
-	memset(&set->tex_desc, 0, sizeof(set->tex_desc));
-
-	int texture_count = write_compute_tex_descs(set->tex_desc);
-
-	VkWriteDescriptorSet writes[16];
-	memset(&writes, 0, sizeof(writes));
-
-	for (int i = 0; i < 16; ++i) {
-		writes[i].sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET;
-		writes[i].dstSet = set->set;
-		writes[i].dstBinding = i + 2;
-		writes[i].descriptorCount = 1;
-		writes[i].descriptorType = VK_DESCRIPTOR_TYPE_STORAGE_IMAGE;
-		writes[i].pImageInfo = &set->tex_desc[i];
-	}
-
-	if (vulkan_textures[0] != NULL) {
-		vkUpdateDescriptorSets(vk_ctx.device, texture_count, writes, 0, NULL);
-	}
-}
-
-void reuse_compute_descriptor_sets(void) {
-	for (int i = 0; i < compute_descriptor_sets_count; ++i) {
-		compute_descriptor_sets[i].in_use = false;
-	}
-}
-
-static VkDescriptorSet get_compute_descriptor_set() {
-	int id = calc_compute_descriptor_id();
-	for (int i = 0; i < compute_descriptor_sets_count; ++i) {
-		if (compute_descriptor_sets[i].id == id) {
-			if (!compute_descriptor_sets[i].in_use) {
-				compute_descriptor_sets[i].in_use = true;
-				update_compute_textures(&compute_descriptor_sets[i]);
-				return compute_descriptor_sets[i].set;
-			}
-			else {
-				if (!compute_textures_changed(&compute_descriptor_sets[i])) {
-					return compute_descriptor_sets[i].set;
-				}
-			}
-		}
-	}
-
-	VkDescriptorSetAllocateInfo alloc_info = {0};
-	alloc_info.sType = VK_STRUCTURE_TYPE_DESCRIPTOR_SET_ALLOCATE_INFO;
-	alloc_info.pNext = NULL;
-	alloc_info.descriptorPool = descriptor_pool;
-	alloc_info.descriptorSetCount = 1;
-	alloc_info.pSetLayouts = &compute_descriptor_layout;
-	VkDescriptorSet descriptor_set;
-	VkResult err = vkAllocateDescriptorSets(vk_ctx.device, &alloc_info, &descriptor_set);
-	assert(!err);
-
-	VkDescriptorBufferInfo buffer_descs[2];
-
-	memset(&buffer_descs, 0, sizeof(buffer_descs));
-
-	if (vk_ctx.compute_uniform_buffer != NULL) {
-		buffer_descs[0].buffer = *vk_ctx.compute_uniform_buffer;
-	}
-	buffer_descs[0].offset = 0;
-	buffer_descs[0].range = 256 * sizeof(float);
-
-	if (vk_ctx.compute_uniform_buffer != NULL) {
-		buffer_descs[1].buffer = *vk_ctx.compute_uniform_buffer;
-	}
-	buffer_descs[1].offset = 0;
-	buffer_descs[1].range = 256 * sizeof(float);
-
-	VkDescriptorImageInfo tex_desc[16];
-	memset(&tex_desc, 0, sizeof(tex_desc));
-
-	int texture_count = 0;
-	for (int i = 0; i < 16; ++i) {
-		if (vulkan_textures[i] != NULL) {
-			tex_desc[i].sampler = vulkan_samplers[i];
-			if (vulkan_textures[i]->impl.stage_depth == i) {
-				tex_desc[i].imageView = vulkan_textures[i]->impl.depthView;
-				vulkan_textures[i]->impl.stage_depth = -1;
-			}
-			else {
-				tex_desc[i].imageView = vulkan_textures[i]->impl.view;
-			}
-			texture_count++;
-		}
-		tex_desc[i].imageLayout = VK_IMAGE_LAYOUT_GENERAL;
-	}
-
-	VkWriteDescriptorSet writes[18];
-	memset(&writes, 0, sizeof(writes));
-
-	writes[0].sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET;
-	writes[0].dstSet = descriptor_set;
-	writes[0].dstBinding = 0;
-	writes[0].descriptorCount = 1;
-	writes[0].descriptorType = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER_DYNAMIC;
-	writes[0].pBufferInfo = &buffer_descs[0];
-
-	writes[1].sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET;
-	writes[1].dstSet = descriptor_set;
-	writes[1].dstBinding = 1;
-	writes[1].descriptorCount = 1;
-	writes[1].descriptorType = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER_DYNAMIC;
-	writes[1].pBufferInfo = &buffer_descs[1];
-
-	for (int i = 2; i < 18; ++i) {
-		writes[i].sType = VK_STRUCTURE_TYPE_WRITE_DESCRIPTOR_SET;
-		writes[i].dstSet = descriptor_set;
-		writes[i].dstBinding = i;
-		writes[i].descriptorCount = 1;
-		writes[i].descriptorType = VK_DESCRIPTOR_TYPE_STORAGE_IMAGE;
-		writes[i].pImageInfo = &tex_desc[i - 2];
-	}
-
-	if (vulkan_textures[0] != NULL) {
-		if (vk_ctx.compute_uniform_buffer != NULL) {
-			vkUpdateDescriptorSets(vk_ctx.device, 2 + texture_count, writes, 0, NULL);
-		}
-		else {
-			vkUpdateDescriptorSets(vk_ctx.device, texture_count, writes + 2, 0, NULL);
-		}
-	}
-	else {
-		if (vk_ctx.compute_uniform_buffer != NULL) {
-			vkUpdateDescriptorSets(vk_ctx.device, 2, writes, 0, NULL);
-		}
-	}
-
-	assert(compute_descriptor_sets_count + 1 < MAX_DESCRIPTOR_SETS);
-	compute_descriptor_sets[compute_descriptor_sets_count].id = id;
-	compute_descriptor_sets[compute_descriptor_sets_count].in_use = true;
-	compute_descriptor_sets[compute_descriptor_sets_count].set = descriptor_set;
-	write_tex_descs(compute_descriptor_sets[compute_descriptor_sets_count].tex_desc);
-	compute_descriptor_sets_count += 1;
-
-	return descriptor_set;
-}
-
 void iron_gpu_shader_init(iron_gpu_shader_t *shader, const void *source, size_t length, iron_gpu_shader_type_t type) {
 	shader->impl.length = (int)length;
 	shader->impl.id = 0;
@@ -4075,9 +3723,6 @@ void iron_gpu_constant_buffer_init(iron_gpu_constant_buffer_t *buffer, int size)
 	else if (vk_ctx.fragment_uniform_buffer == NULL) {
 		vk_ctx.fragment_uniform_buffer = &buffer->impl.buf;
 	}
-	else if (vk_ctx.compute_uniform_buffer == NULL) {
-		vk_ctx.compute_uniform_buffer = &buffer->impl.buf;
-	}
 
 	void *p;
 	VkResult err = vkMapMemory(vk_ctx.device, buffer->impl.mem, 0, buffer->impl.mem_alloc.allocationSize, 0, (void **)&p);
diff --git a/base/sources/backends/vulkan_gpu.h b/base/sources/backends/vulkan_gpu.h
index 6003a6b7..c53b82c7 100644
--- a/base/sources/backends/vulkan_gpu.h
+++ b/base/sources/backends/vulkan_gpu.h
@@ -89,7 +89,6 @@ struct vk_context {
 	// buffer hack
 	VkBuffer *vertex_uniform_buffer;
 	VkBuffer *fragment_uniform_buffer;
-	VkBuffer *compute_uniform_buffer;
 
 #ifdef VALIDATE
 	bool validation_found;
@@ -102,7 +101,6 @@ extern struct vk_context vk_ctx;
 
 extern void flush_init_cmd(void);
 extern void reuse_descriptor_sets(void);
-extern void reuse_compute_descriptor_sets(void);
 
 typedef struct {
 	int _indexCount;
@@ -110,16 +108,6 @@ typedef struct {
 	VkFence fence;
 } CommandList5Impl;
 
-typedef struct iron_gpu_compute_shader_impl {
-	iron_internal_named_number locations[IRON_INTERNAL_NAMED_NUMBER_COUNT];
-	iron_internal_named_number texture_bindings[IRON_INTERNAL_NAMED_NUMBER_COUNT];
-	iron_internal_named_number offsets[IRON_INTERNAL_NAMED_NUMBER_COUNT];
-
-	VkPipelineLayout pipeline_layout;
-	VkPipeline pipeline;
-	VkShaderModule shader_module;
-} iron_gpu_compute_shader_impl;
-
 typedef struct PipelineState5Impl_s {
 	const char **textures;
 	int *textureValues;
@@ -140,14 +128,9 @@ typedef struct PipelineState5Impl_s {
 	VkPipelineLayout pipeline_layout;
 } PipelineState5Impl;
 
-typedef struct ComputePipelineState5Impl_t {
-	int a;
-} ComputePipelineState5Impl;
-
 typedef struct {
 	int vertexOffset;
 	int fragmentOffset;
-	int computeOffset;
 } ConstantLocation5Impl;
 
 typedef struct {
diff --git a/base/sources/backends/webgpu_gpu.c b/base/sources/backends/webgpu_gpu.c
index 3e06a8eb..f6651a81 100644
--- a/base/sources/backends/webgpu_gpu.c
+++ b/base/sources/backends/webgpu_gpu.c
@@ -480,10 +480,6 @@ void iron_gpu_command_list_set_fragment_constant_buffer(iron_gpu_command_list_t
 
 }
 
-void iron_gpu_command_list_set_compute_constant_buffer(iron_gpu_command_list_t *list, struct iron_gpu_constant_buffer *buffer, int offset, size_t size) {
-
-}
-
 void iron_gpu_command_list_set_texture(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_texture_t *texture) {
 
 }
@@ -491,21 +487,3 @@ void iron_gpu_command_list_set_texture(iron_gpu_command_list_t *list, iron_gpu_t
 void iron_gpu_command_list_set_sampler(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_sampler_t *sampler) {}
 
 void iron_gpu_command_list_set_texture_from_render_target_depth(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_texture_t *renderTarget) {}
-
-void iron_gpu_command_list_set_compute_shader(iron_gpu_command_list_t *list, iron_gpu_compute_shader *shader) {}
-
-void iron_gpu_command_list_compute(iron_gpu_command_list_t *list, int x, int y, int z) {}
-
-void iron_gpu_compute_shader_init(iron_gpu_compute_shader *shader, void *source, int length) {}
-
-void iron_gpu_compute_shader_destroy(iron_gpu_compute_shader *shader) {}
-
-iron_gpu_constant_location_t iron_gpu_compute_shader_get_constant_location(iron_gpu_compute_shader *shader, const char *name) {
-	iron_gpu_constant_location_t location = {0};
-	return location;
-}
-
-iron_gpu_texture_unit_t iron_gpu_compute_shader_get_texture_unit(iron_gpu_compute_shader *shader, const char *name) {
-	iron_gpu_texture_unit_t unit = {0};
-	return unit;
-}
diff --git a/base/sources/backends/webgpu_gpu.h b/base/sources/backends/webgpu_gpu.h
index 20c0a085..076c1b7a 100644
--- a/base/sources/backends/webgpu_gpu.h
+++ b/base/sources/backends/webgpu_gpu.h
@@ -27,14 +27,9 @@ typedef struct {
 	WGPURenderPipeline pipeline;
 } PipelineState5Impl;
 
-typedef struct {
-	WGPUComputePipeline pipeline;
-} ComputePipelineState5Impl;
-
 typedef struct {
 	int vertexOffset;
 	int fragmentOffset;
-	int computeOffset;
 } ConstantLocation5Impl;
 
 typedef struct iron_gpu_sampler_impl {
@@ -52,7 +47,3 @@ typedef struct {
 	WGPURenderPassEncoder pass;
 	int indexCount;
 } CommandList5Impl;
-
-typedef struct iron_gpu_compute_shader_impl {
-	int nothing;
-} iron_gpu_compute_shader_impl;
diff --git a/base/sources/iron_gpu.c b/base/sources/iron_gpu.c
index 96fa7c41..d631feea 100644
--- a/base/sources/iron_gpu.c
+++ b/base/sources/iron_gpu.c
@@ -31,7 +31,6 @@ bool waitAfterNextDraw = false;
 
 static iron_gpu_constant_buffer_t vertexConstantBuffer;
 static iron_gpu_constant_buffer_t fragmentConstantBuffer;
-static iron_gpu_constant_buffer_t computeConstantBuffer;
 static int constantBufferIndex = 0;
 
 static struct {
@@ -44,7 +43,6 @@ static struct {
 
 typedef struct render_state {
 	iron_gpu_pipeline_t *pipeline;
-	iron_gpu_compute_shader *compute_shader;
 
 	iron_gpu_index_buffer_t *index_buffer;
 	iron_gpu_vertex_buffer_t *vertex_buffer;
@@ -71,7 +69,6 @@ typedef struct render_state {
 
 	uint8_t vertex_constant_data[CONSTANT_BUFFER_SIZE];
 	uint8_t fragment_constant_data[CONSTANT_BUFFER_SIZE];
-	uint8_t compute_constant_data[CONSTANT_BUFFER_SIZE];
 } render_state;
 
 static render_state current_state;
@@ -107,7 +104,6 @@ void gpu_internal_init_window(int depthBufferBits, bool vsync) {
 	}
 	iron_gpu_constant_buffer_init(&vertexConstantBuffer, CONSTANT_BUFFER_SIZE * CONSTANT_BUFFER_MULTIPLY);
 	iron_gpu_constant_buffer_init(&fragmentConstantBuffer, CONSTANT_BUFFER_SIZE * CONSTANT_BUFFER_MULTIPLY);
-	iron_gpu_constant_buffer_init(&computeConstantBuffer, CONSTANT_BUFFER_SIZE * CONSTANT_BUFFER_MULTIPLY);
 
 	// to support doing work after gpu_end and before gpu_begin
 	iron_gpu_command_list_begin(&commandList);
@@ -133,22 +129,15 @@ static void iron_internal_start_draw(bool compute) {
 	if ((constantBufferIndex + 1) >= CONSTANT_BUFFER_MULTIPLY || waitAfterNextDraw) {
 		memcpy(current_state.vertex_constant_data, vertexConstantBuffer.data, CONSTANT_BUFFER_SIZE);
 		memcpy(current_state.fragment_constant_data, fragmentConstantBuffer.data, CONSTANT_BUFFER_SIZE);
-		memcpy(current_state.compute_constant_data, computeConstantBuffer.data, CONSTANT_BUFFER_SIZE);
 	}
 	iron_gpu_constant_buffer_unlock(&vertexConstantBuffer);
 	iron_gpu_constant_buffer_unlock(&fragmentConstantBuffer);
-	iron_gpu_constant_buffer_unlock(&computeConstantBuffer);
 
 	iron_gpu_internal_set_samplers(current_state.texture_count, current_state.texture_units);
 	iron_gpu_internal_set_samplers(current_state.depth_render_target_count, current_state.depth_render_target_units);
 
-	if (compute) {
-		iron_gpu_command_list_set_compute_constant_buffer(&commandList, &computeConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
-	}
-	else {
-		iron_gpu_command_list_set_vertex_constant_buffer(&commandList, &vertexConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
-		iron_gpu_command_list_set_fragment_constant_buffer(&commandList, &fragmentConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
-	}
+	iron_gpu_command_list_set_vertex_constant_buffer(&commandList, &vertexConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
+	iron_gpu_command_list_set_fragment_constant_buffer(&commandList, &fragmentConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
 }
 
 static void iron_internal_end_draw(bool compute) {
@@ -173,9 +162,6 @@ static void iron_internal_end_draw(bool compute) {
 		if (current_state.pipeline != NULL) {
 			iron_gpu_command_list_set_pipeline(&commandList, current_state.pipeline);
 		}
-		if (current_state.compute_shader != NULL) {
-			iron_gpu_command_list_set_compute_shader(&commandList, current_state.compute_shader);
-		}
 		if (current_state.index_buffer != NULL) {
 			iron_gpu_command_list_set_index_buffer(&commandList, current_state.index_buffer);
 		}
@@ -202,16 +188,13 @@ static void iron_internal_end_draw(bool compute) {
 
 		iron_gpu_constant_buffer_lock(&vertexConstantBuffer, 0, CONSTANT_BUFFER_SIZE);
 		iron_gpu_constant_buffer_lock(&fragmentConstantBuffer, 0, CONSTANT_BUFFER_SIZE);
-		iron_gpu_constant_buffer_lock(&computeConstantBuffer, 0, CONSTANT_BUFFER_SIZE);
 
 		memcpy(vertexConstantBuffer.data, current_state.vertex_constant_data, CONSTANT_BUFFER_SIZE);
 		memcpy(fragmentConstantBuffer.data, current_state.fragment_constant_data, CONSTANT_BUFFER_SIZE);
-		memcpy(computeConstantBuffer.data, current_state.compute_constant_data, CONSTANT_BUFFER_SIZE);
 	}
 	else {
 		iron_gpu_constant_buffer_lock(&vertexConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
 		iron_gpu_constant_buffer_lock(&fragmentConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
-		iron_gpu_constant_buffer_lock(&computeConstantBuffer, constantBufferIndex * CONSTANT_BUFFER_SIZE, CONSTANT_BUFFER_SIZE);
 	}
 }
 
@@ -271,7 +254,6 @@ void gpu_begin() {
 	windows[0].current_render_target_count = 1;
 
 	current_state.pipeline = NULL;
-	current_state.compute_shader = NULL;
 	current_state.index_buffer = NULL;
 	current_state.vertex_buffer = NULL;
 	current_state.viewport_set = false;
@@ -316,7 +298,6 @@ void gpu_disable_scissor(void) {
 void gpu_end() {
 	iron_gpu_constant_buffer_unlock(&vertexConstantBuffer);
 	iron_gpu_constant_buffer_unlock(&fragmentConstantBuffer);
-	iron_gpu_constant_buffer_unlock(&computeConstantBuffer);
 
 	iron_gpu_command_list_render_target_to_framebuffer_barrier(&commandList, &windows[0].framebuffers[windows[0].currentBuffer]);
 	iron_gpu_command_list_end(&commandList);
@@ -333,8 +314,6 @@ void gpu_set_int(iron_gpu_constant_location_t location, int value) {
 		iron_gpu_constant_buffer_set_int(&vertexConstantBuffer, location.impl.vertexOffset, value);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_int(&fragmentConstantBuffer, location.impl.fragmentOffset, value);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_int(&computeConstantBuffer, location.impl.computeOffset, value);
 }
 
 void gpu_set_int2(iron_gpu_constant_location_t location, int value1, int value2) {
@@ -342,8 +321,6 @@ void gpu_set_int2(iron_gpu_constant_location_t location, int value1, int value2)
 		iron_gpu_constant_buffer_set_int2(&vertexConstantBuffer, location.impl.vertexOffset, value1, value2);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_int2(&fragmentConstantBuffer, location.impl.fragmentOffset, value1, value2);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_int2(&computeConstantBuffer, location.impl.computeOffset, value1, value2);
 }
 
 void gpu_set_int3(iron_gpu_constant_location_t location, int value1, int value2, int value3) {
@@ -351,8 +328,6 @@ void gpu_set_int3(iron_gpu_constant_location_t location, int value1, int value2,
 		iron_gpu_constant_buffer_set_int3(&vertexConstantBuffer, location.impl.vertexOffset, value1, value2, value3);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_int3(&fragmentConstantBuffer, location.impl.fragmentOffset, value1, value2, value3);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_int3(&computeConstantBuffer, location.impl.computeOffset, value1, value2, value3);
 }
 
 void gpu_set_int4(iron_gpu_constant_location_t location, int value1, int value2, int value3, int value4) {
@@ -360,8 +335,6 @@ void gpu_set_int4(iron_gpu_constant_location_t location, int value1, int value2,
 		iron_gpu_constant_buffer_set_int4(&vertexConstantBuffer, location.impl.vertexOffset, value1, value2, value3, value4);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_int4(&fragmentConstantBuffer, location.impl.fragmentOffset, value1, value2, value3, value4);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_int4(&computeConstantBuffer, location.impl.computeOffset, value1, value2, value3, value4);
 }
 
 void gpu_set_ints(iron_gpu_constant_location_t location, int *values, int count) {
@@ -369,8 +342,6 @@ void gpu_set_ints(iron_gpu_constant_location_t location, int *values, int count)
 		iron_gpu_constant_buffer_set_ints(&vertexConstantBuffer, location.impl.vertexOffset, values, count);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_ints(&fragmentConstantBuffer, location.impl.fragmentOffset, values, count);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_ints(&computeConstantBuffer, location.impl.computeOffset, values, count);
 }
 
 void gpu_set_float(iron_gpu_constant_location_t location, float value) {
@@ -378,8 +349,6 @@ void gpu_set_float(iron_gpu_constant_location_t location, float value) {
 		iron_gpu_constant_buffer_set_float(&vertexConstantBuffer, location.impl.vertexOffset, value);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_float(&fragmentConstantBuffer, location.impl.fragmentOffset, value);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_float(&computeConstantBuffer, location.impl.computeOffset, value);
 }
 
 void gpu_set_float2(iron_gpu_constant_location_t location, float value1, float value2) {
@@ -387,8 +356,6 @@ void gpu_set_float2(iron_gpu_constant_location_t location, float value1, float v
 		iron_gpu_constant_buffer_set_float2(&vertexConstantBuffer, location.impl.vertexOffset, value1, value2);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_float2(&fragmentConstantBuffer, location.impl.fragmentOffset, value1, value2);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_float2(&computeConstantBuffer, location.impl.computeOffset, value1, value2);
 }
 
 void gpu_set_float3(iron_gpu_constant_location_t location, float value1, float value2, float value3) {
@@ -396,8 +363,6 @@ void gpu_set_float3(iron_gpu_constant_location_t location, float value1, float v
 		iron_gpu_constant_buffer_set_float3(&vertexConstantBuffer, location.impl.vertexOffset, value1, value2, value3);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_float3(&fragmentConstantBuffer, location.impl.fragmentOffset, value1, value2, value3);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_float3(&computeConstantBuffer, location.impl.computeOffset, value1, value2, value3);
 }
 
 void gpu_set_float4(iron_gpu_constant_location_t location, float value1, float value2, float value3, float value4) {
@@ -405,8 +370,6 @@ void gpu_set_float4(iron_gpu_constant_location_t location, float value1, float v
 		iron_gpu_constant_buffer_set_float4(&vertexConstantBuffer, location.impl.vertexOffset, value1, value2, value3, value4);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_float4(&fragmentConstantBuffer, location.impl.fragmentOffset, value1, value2, value3, value4);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_float4(&computeConstantBuffer, location.impl.computeOffset, value1, value2, value3, value4);
 }
 
 void gpu_set_floats(iron_gpu_constant_location_t location, float *values, int count) {
@@ -414,8 +377,6 @@ void gpu_set_floats(iron_gpu_constant_location_t location, float *values, int co
 		iron_gpu_constant_buffer_set_floats(&vertexConstantBuffer, location.impl.vertexOffset, values, count);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_floats(&fragmentConstantBuffer, location.impl.fragmentOffset, values, count);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_floats(&computeConstantBuffer, location.impl.computeOffset, values, count);
 }
 
 void gpu_set_bool(iron_gpu_constant_location_t location, bool value) {
@@ -423,8 +384,6 @@ void gpu_set_bool(iron_gpu_constant_location_t location, bool value) {
 		iron_gpu_constant_buffer_set_bool(&vertexConstantBuffer, location.impl.vertexOffset, value);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_bool(&fragmentConstantBuffer, location.impl.fragmentOffset, value);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_bool(&computeConstantBuffer, location.impl.computeOffset, value);
 }
 
 void gpu_set_matrix4(iron_gpu_constant_location_t location, iron_matrix4x4_t *value) {
@@ -432,8 +391,6 @@ void gpu_set_matrix4(iron_gpu_constant_location_t location, iron_matrix4x4_t *va
 		iron_gpu_constant_buffer_set_matrix4(&vertexConstantBuffer, location.impl.vertexOffset, value);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_matrix4(&fragmentConstantBuffer, location.impl.fragmentOffset, value);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_matrix4(&computeConstantBuffer, location.impl.computeOffset, value);
 }
 
 void gpu_set_matrix3(iron_gpu_constant_location_t location, iron_matrix3x3_t *value) {
@@ -441,8 +398,6 @@ void gpu_set_matrix3(iron_gpu_constant_location_t location, iron_matrix3x3_t *va
 		iron_gpu_constant_buffer_set_matrix3(&vertexConstantBuffer, location.impl.vertexOffset, value);
 	if (location.impl.fragmentOffset >= 0)
 		iron_gpu_constant_buffer_set_matrix3(&fragmentConstantBuffer, location.impl.fragmentOffset, value);
-	if (location.impl.computeOffset >= 0)
-		iron_gpu_constant_buffer_set_matrix3(&computeConstantBuffer, location.impl.computeOffset, value);
 }
 
 void gpu_set_texture_addressing(iron_gpu_texture_unit_t unit, gpu_texture_direction_t dir, gpu_texture_addressing_t addressing) {
@@ -584,17 +539,6 @@ void gpu_render_target_use_depth_as_texture(iron_gpu_texture_t *render_target, i
 	iron_gpu_command_list_set_texture_from_render_target_depth(&commandList, g5_unit, render_target);
 }
 
-void iron_gpu_set_compute_shader(iron_gpu_compute_shader *shader) {
-	current_state.compute_shader = shader;
-	iron_gpu_command_list_set_compute_shader(&commandList, shader);
-}
-
-void gpu_compute(int x, int y, int z) {
-	iron_internal_start_draw(true);
-	iron_gpu_command_list_compute(&commandList, x, y, z);
-	iron_internal_end_draw(true);
-}
-
 void iron_gpu_render_target_get_pixels(iron_gpu_texture_t *render_target, uint8_t *data) {
 	iron_gpu_command_list_get_render_target_pixels(&commandList, render_target, data);
 }
diff --git a/base/sources/iron_gpu.h b/base/sources/iron_gpu.h
index a2947078..24e5dd64 100644
--- a/base/sources/iron_gpu.h
+++ b/base/sources/iron_gpu.h
@@ -127,7 +127,6 @@ void gpu_set_texture_mipmap_filter(struct iron_gpu_texture_unit unit, gpu_mipmap
 void gpu_restore_render_target(void);
 void gpu_set_render_targets(struct iron_gpu_texture **targets, int count);
 void gpu_set_texture(struct iron_gpu_texture_unit unit, struct iron_gpu_texture *texture);
-void gpu_compute(int x, int y, int z);
 void gpu_set_index_buffer(struct iron_gpu_index_buffer *buffer);
 
 void gpu_internal_init_window(int depth_buffer_bits, bool vsync);
@@ -281,7 +280,6 @@ int iron_gpu_index_buffer_count(iron_gpu_index_buffer_t *buffer);
 typedef enum iron_gpu_shader_type {
 	IRON_GPU_SHADER_TYPE_FRAGMENT,
 	IRON_GPU_SHADER_TYPE_VERTEX,
-	IRON_GPU_SHADER_TYPE_COMPUTE,
 	IRON_GPU_SHADER_TYPE_COUNT
 } iron_gpu_shader_type_t;
 
@@ -415,7 +413,6 @@ void iron_gpu_sampler_destroy(iron_gpu_sampler_t *sampler);
 
 void iron_gpu_internal_pipeline_set_defaults(iron_gpu_pipeline_t *pipeline);
 
-struct iron_gpu_compute_shader;
 struct iron_gpu_constant_buffer;
 struct iron_gpu_index_buffer;
 struct iron_gpu_vertex_buffer;
@@ -442,7 +439,6 @@ void iron_gpu_command_list_viewport(iron_gpu_command_list_t *list, int x, int y,
 void iron_gpu_command_list_scissor(iron_gpu_command_list_t *list, int x, int y, int width, int height);
 void iron_gpu_command_list_disable_scissor(iron_gpu_command_list_t *list);
 void iron_gpu_command_list_set_pipeline(iron_gpu_command_list_t *list, struct iron_gpu_pipeline *pipeline);
-void iron_gpu_command_list_set_compute_shader(iron_gpu_command_list_t *list, struct iron_gpu_compute_shader *shader);
 void iron_gpu_command_list_set_vertex_buffer(iron_gpu_command_list_t *list, struct iron_gpu_vertex_buffer *buffer);
 void iron_gpu_command_list_set_index_buffer(iron_gpu_command_list_t *list, struct iron_gpu_index_buffer *buffer);
 void iron_gpu_command_list_set_render_targets(iron_gpu_command_list_t *list, struct iron_gpu_texture **targets, int count);
@@ -451,26 +447,15 @@ void iron_gpu_command_list_upload_vertex_buffer(iron_gpu_command_list_t *list, s
 void iron_gpu_command_list_upload_texture(iron_gpu_command_list_t *list, struct iron_gpu_texture *texture);
 void iron_gpu_command_list_set_vertex_constant_buffer(iron_gpu_command_list_t *list, struct iron_gpu_constant_buffer *buffer, int offset, size_t size);
 void iron_gpu_command_list_set_fragment_constant_buffer(iron_gpu_command_list_t *list, struct iron_gpu_constant_buffer *buffer, int offset, size_t size);
-void iron_gpu_command_list_set_compute_constant_buffer(iron_gpu_command_list_t *list, struct iron_gpu_constant_buffer *buffer, int offset, size_t size);
 void iron_gpu_command_list_execute(iron_gpu_command_list_t *list);
 void iron_gpu_command_list_wait_for_execution_to_finish(iron_gpu_command_list_t *list);
 void iron_gpu_command_list_get_render_target_pixels(iron_gpu_command_list_t *list, struct iron_gpu_texture *render_target, uint8_t *data);
-void iron_gpu_command_list_compute(iron_gpu_command_list_t *list, int x, int y, int z);
 void iron_gpu_command_list_set_texture(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_texture_t *texture);
 void iron_gpu_command_list_set_texture_from_render_target_depth(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_texture_t *target);
 void iron_gpu_command_list_set_sampler(iron_gpu_command_list_t *list, iron_gpu_texture_unit_t unit, iron_gpu_sampler_t *sampler);
 
 void iron_gpu_render_target_get_pixels(iron_gpu_texture_t *render_target, uint8_t *data);
 
-typedef struct iron_gpu_compute_shader {
-	iron_gpu_compute_shader_impl impl;
-} iron_gpu_compute_shader;
-
-void iron_gpu_compute_shader_init(iron_gpu_compute_shader *shader, void *source, int length);
-void iron_gpu_compute_shader_destroy(iron_gpu_compute_shader *shader);
-iron_gpu_constant_location_t iron_gpu_compute_shader_get_constant_location(iron_gpu_compute_shader *shader, const char *name);
-iron_gpu_texture_unit_t iron_gpu_compute_shader_get_texture_unit(iron_gpu_compute_shader *shader, const char *name);
-
 struct iron_gpu_command_list;
 struct iron_gpu_constant_buffer;
 struct iron_gpu_index_buffer;