mirror of
https://github.com/Comfy-Org/comfy-kitchen.git
synced 2026-11-04 23:37:42 +00:00
[GH-ISSUE #243] INT8 GEMM not supported on this GPU for these dimensions (requires SM >= 7.5, and dimensions multiple of 4). #68
Labels
No labels
pull-request
No milestone
No project
No assignees
1 participant
Notifications
Due date
No due date set.
Dependencies
No dependencies set.
Reference
Comfy-Org/comfy-kitchen#68
Loading…
Add table
Add a link
Reference in a new issue
No description provided.
Delete branch "%!s()"
Deleting a branch is permanent. Although the deleted branch may continue to exist for a short time before it actually gets removed, it CANNOT be undone in most cases. Continue?
Originally created by @dranzerashi on GitHub (Oct 12, 2026).
Original GitHub issue: https://github.com/Comfy-Org/comfy-kitchen/issues/243
[ERROR] !!! Exception during processing !!! INT8 GEMM not supported on this GPU for these dimensions (requires SM >= 7.5, and dimensions multiple of 4).
[ERROR] Traceback (most recent call last):
File "/home/ashithr/work/ComfyUI/execution.py", line 548, in execute
output_data, output_ui, has_subgraph, has_pending_tasks = await get_output_data(prompt_id, unique_id, obj, input_data_all, execution_block_cb=execution_block_cb, pre_execute_cb=pre_execute_cb, v3_data=v3_data)
^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/execution.py", line 353, in get_output_data
return_values = await _async_map_node_over_list(prompt_id, unique_id, obj, input_data_all, obj.FUNCTION, allow_interrupt=True, execution_block_cb=execution_block_cb, pre_execute_cb=pre_execute_cb, v3_data=v3_data)
^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/execution.py", line 327, in _async_map_node_over_list
await process_inputs(input_dict, i)
File "/home/ashithr/work/ComfyUI/execution.py", line 315, in process_inputs
result = f(**inputs)
File "/home/ashithr/work/ComfyUI/comfy_api/internal/init.py", line 149, in wrapped_func
return method(locked_class, **inputs)
File "/home/ashithr/work/ComfyUI/comfy_api/latest/_io.py", line 2219, in EXECUTE_NORMALIZED
to_return = cls.execute(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/comfy_extras/nodes_custom_sampler.py", line 1055, in execute
samples = guider.sample(noise.generate_noise(latent), latent_image, sampler, sigmas, denoise_mask=noise_mask, callback=callback, disable_pbar=disable_pbar, seed=noise.seed)
File "/home/ashithr/work/ComfyUI/comfy/samplers.py", line 1335, in sample
output = executor.execute(noise, latent_image, sampler, sigmas, denoise_mask, callback, disable_pbar, seed, latent_shapes=latent_shapes)
File "/home/ashithr/work/ComfyUI/comfy/patcher_extension.py", line 113, in execute
return self.original(*args, **kwargs)
~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/comfy/samplers.py", line 1262, in outer_sample
output = self.inner_sample(noise, latent_image, device, sampler, sigmas, denoise_mask, callback, disable_pbar, seed, latent_shapes=latent_shapes)
File "/home/ashithr/work/ComfyUI/comfy/samplers.py", line 1226, in inner_sample
self.conds = process_conds(self.inner_model, noise, self.conds, device, latent_image, denoise_mask, seed, latent_shapes=latent_shapes)
~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/comfy/samplers.py", line 1048, in process_conds
conds[k] = encode_model_conds(model.extra_conds, conds[k], noise, device, k, latent_image=latent_image, denoise_mask=denoise_mask, seed=seed, latent_shapes=latent_shapes)
~~~~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/comfy/samplers.py", line 953, in encode_model_conds
out = model_function(**params)
File "/home/ashithr/work/ComfyUI/comfy/model_base.py", line 2185, in extra_conds
cross_attn = self.diffusion_model.preprocess_text_embeds(
cross_attn.to(device=kwargs["device"], dtype=self.get_dtype_inference()))
File "/home/ashithr/work/ComfyUI/comfy/ldm/minimax/model.py", line 522, in preprocess_text_embeds
return self.token_refiner(self.condition_proj(text_states[0])).unsqueeze(0)
~~~~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1783, in _wrapped_call_impl
return self._call_impl(*args, **kwargs)
~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1794, in _call_impl
return forward_call(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/comfy/ldm/minimax/model.py", line 275, in forward
x = block(x, transformer_options=transformer_options)
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1783, in _wrapped_call_impl
return self._call_impl(*args, **kwargs)
~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1794, in call_impl
return forward_call(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/comfy/ldm/minimax/model.py", line 260, in forward
x = self.attn(self.norm1(x), transformer_options=transformer_options).add(x)
~~~~~~~~~^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1783, in _wrapped_call_impl
return self._call_impl(*args, **kwargs)
~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1794, in _call_impl
return forward_call(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/comfy/ldm/minimax/model.py", line 176, in forward
q, k, v = self.qkv_proj(x).split(self.heads * self.head_dim, dim=-1)
~~~~~~~~~~~~~^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1783, in _wrapped_call_impl
return self._call_impl(*args, **kwargs)
~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/nn/modules/module.py", line 1794, in _call_impl
return forward_call(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/comfy/ops.py", line 1527, in forward
output = self.forward_comfy_cast_weights(
input,
...<2 lines>...
weight_only_quant=weight_only_quant,
)
File "/home/ashithr/work/ComfyUI/comfy/ops.py", line 1471, in forward_comfy_cast_weights
return self._forward(input, weight, bias)
~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/comfy/ops.py", line 1439, in _forward
return torch.nn.functional.linear(input, weight, bias)
~~~~~~~~~~~~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/comfy_kitchen/tensor/base.py", line 362, in torch_dispatch
return op_handlers[parent_cls](qt, args, kwargs)
~~~~~~~~~~~~~~~~~~~~~~~^^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/comfy_kitchen/tensor/int8.py", line 287, in _handle_int8_linear_tensorwise
return torch.ops.comfy_kitchen.int8_linear(
~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~^
input_tensor.contiguous(),
^^^^^^^^^^^^^^^^^^^^^^^^^^
...<5 lines>...
convrot_groupsize,
^^^^^^^^^^^^^^^^^^
)
^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/_ops.py", line 1350, in call
return self._op(*args, **kwargs)
~~~~~~~~^^^^^^^^^^^^^^^^^
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/_library/custom_ops.py", line 442, in backend_impl
result = self._backend_fns[device_type](*args, **kwargs)
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/_compile.py", line 54, in inner
return disable_fn(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/_dynamo/eval_frame.py", line 1548, in _fn
return fn(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/torch/_library/custom_ops.py", line 503, in wrapped_fn
return fn(*args, **kwargs)
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/comfy_kitchen/backends/eager/quantization.py", line 1261, in _op_int8_linear
return impl(**kwargs)
File "/home/ashithr/work/ComfyUI/.venv/lib/python3.14/site-packages/comfy_kitchen/backends/cuda/init.py", line 2270, in int8_linear
_C.cublas_gemm_int8(
~~~~~~~~~~~~~~~~~~~^
_wrap_for_dlpack(cublas_x),
^^^^^^^^^^^^^^^^^^^^^^^^^^^
...<3 lines>...
stream_ptr,
^^^^^^^^^^^
)
^
RuntimeError: INT8 GEMM not supported on this GPU for these dimensions (requires SM >= 7.5, and dimensions multiple of 4).
Additional diagnostics: INT8 GEMM failure on GTX 1660 Ti (Turing, SM 7.5)
Environment
2.14.1+cu1300.2.37The full console traceback has already been provided in the original report.
Additional findings
The failure occurs during MiniMax model execution in the
qkv_projlinear projection. The call reaches:comfy_kitchen/tensor/int8.py::_handle_int8_linear_tensorwisewhich invokes
torch.ops.comfy_kitchen.int8_linear. Execution subsequently reaches the CUDA backend'sint8_linear()implementation and fails at_C.cublas_gemm_int8().The exception reports:
INT8 GEMM not supported on this GPU for these dimensions (requires SM >= 7.5, and dimensions multiple of 4).I added a temporary diagnostic print immediately before the native GEMM call. It produced the following dimensions:
All three matrix dimensions are divisible by four, and the GPU reports the architecture level required by the error message. The padded K and N dimensions are unchanged from the original dimensions.
Findings from inspecting the installed CUDA backend
In
comfy_kitchen/backends/cuda/__init__.py, theint8_linear()implementation attempts different execution paths before reaching the cuBLAS fallback._cuda_device_should_use_turing_kernels(). The installed implementation explicitly excludes NVIDIA 16-series GPUs from that selection._cuda_device_supports_cutlass_int8_dequant(), which requires compute capability major version 8 or newer. The GTX 1660 Ti therefore does not qualify.The fallback contains Turing-specific padding logic. However, the observed dimensions already satisfy the stated multiple-of-four constraint, so simply rounding these dimensions to multiples of four would not address the demonstrated failure.