diff --git a/unsloth/kernels/cross_entropy_loss.py b/unsloth/kernels/cross_entropy_loss.py
index 0c07035097..cc3dbb1d87 100644
--- a/unsloth/kernels/cross_entropy_loss.py
+++ b/unsloth/kernels/cross_entropy_loss.py
@@ -84,12 +84,12 @@ def _cross_entropy_forward(
     logsumexp = c + tl.log(tl.sum(tl.exp(logits - c), 0))
 
     if label_idx != -100:
-        x = tl.load(logits_ptr + label_idx)
+        x = tl.load(logits_ptr + label_idx).to(tl.float32)
         # Go logit scaling for Cohere: t * x
         if DO_LOGIT_SCALING: x = LOGIT_SCALE * x
         # Do logit softcapping for Gemma 2: t * tanh(1/t * x)
         if DO_SOFTCAPPING:   x = SOFTCAP * triton_tanh(x / SOFTCAP)
-        loss = logsumexp - x.to(tl.float32)
+        loss = logsumexp - x
     else:
         loss = 0.0
     tl.store(logsumexp_ptr, logsumexp)
@@ -170,7 +170,7 @@ def _chunked_cross_entropy_forward(
             if DO_LOGIT_SCALING: x = LOGIT_SCALE * x
             # Do logit softcapping for Gemma 2: t * tanh(1/t * x)
             if DO_SOFTCAPPING:   x = SOFTCAP * triton_tanh(x / SOFTCAP)
-            loss = -1.0 * x.to(tl.float32)
+            loss = -1.0 * x
         else:
             loss = 0.0
         tl.store(loss_ptr, loss)
diff --git a/unsloth/kernels/flex_attention.py b/unsloth/kernels/flex_attention.py
index 08426b69e0..887ffca1b7 100644
--- a/unsloth/kernels/flex_attention.py
+++ b/unsloth/kernels/flex_attention.py
@@ -15,12 +15,13 @@
 import torch
 from functools import lru_cache
 from transformers.models.llama.modeling_llama import logger
+import os
 
 torch_compile_options = {
     "epilogue_fusion"   : True,
     "max_autotune"      : True,
     "shape_padding"     : True,
-    "trace.enabled"     : False, # Output Triton kernel outputs!
+    "trace.enabled"     : os.environ.get("UNSLOTH_COMPILE_DEBUG", "0") == "1",
     "triton.cudagraphs" : False,
 }
 
diff --git a/unsloth/models/_utils.py b/unsloth/models/_utils.py
index 903093e60f..a6cd13d251 100644
--- a/unsloth/models/_utils.py
+++ b/unsloth/models/_utils.py
@@ -12,7 +12,7 @@
 # See the License for the specific language governing permissions and
 # limitations under the License.
 
-__version__ = "2024.11.3"
+__version__ = "2024.11.4"
 
 __all__ = [
     "prepare_model_for_kbit_training",
diff --git a/unsloth/models/llama.py b/unsloth/models/llama.py
index 3c4d8f3b38..7f07bea4c5 100644
--- a/unsloth/models/llama.py
+++ b/unsloth/models/llama.py
@@ -1376,6 +1376,15 @@ def _wrap_fast_inference(generate, device_type, dtype, model):
     @torch.inference_mode
     def _fast_generate(*args, **kwargs):
 
+        if hasattr(model, "config") and hasattr(model.config, "max_position_embeddings"):
+            if "input_ids" in kwargs and kwargs["input_ids"] is not None and "max_new_tokens" in kwargs:
+                if kwargs["input_ids"].shape[-1] + kwargs["max_new_tokens"] > model.config.max_position_embeddings:
+                    raise ValueError(
+                        f'Unsloth: input length {kwargs["input_ids"].shape[-1]} + max_new_tokens {kwargs["max_new_tokens"]} exceeds the maximum sequence length of {model.config.max_position_embeddings}!\n'\
+                        'You will need to do long context extension by increasing the `max_seq_length` in `FastLanguageModel.from_pretrained`.'
+                    )
+        pass
+
         # Set a flag for generation!
         internal_model = model
         while hasattr(internal_model, "model"):
diff --git a/unsloth/models/loader.py b/unsloth/models/loader.py
index db7259b1d9..4566302ed0 100644
--- a/unsloth/models/loader.py
+++ b/unsloth/models/loader.py
@@ -43,6 +43,24 @@ if SUPPORTS_GEMMA:
 if SUPPORTS_GEMMA2:
     from .gemma2 import FastGemma2Model
 pass
+import torch
+
+def _get_dtype(dtype):
+    __DTYPE_MAP = {
+        "float32": torch.float32,
+        torch.float32: torch.float32,
+        "float16": torch.float16,
+        torch.float16: torch.float16,
+        "bfloat16": torch.bfloat16,
+        torch.bfloat16: torch.bfloat16,
+    }
+    if dtype in __DTYPE_MAP:
+        return __DTYPE_MAP[dtype]
+    else:
+        print(f"Unsloth: {dtype} is not recognized, so we'll default to torch.float16")
+        return torch.float16
+    pass
+pass
 
 
 def __get_model_name(
@@ -332,7 +350,7 @@ class FastLanguageModel(FastLlamaModel):
         model, tokenizer = dispatch_model.from_pretrained(
             model_name        = model_name,
             max_seq_length    = max_seq_length,
-            dtype             = dtype,
+            dtype             = _get_dtype(dtype),
             load_in_4bit      = load_in_4bit,
             token             = token,
             device_map        = device_map,