cleanup code

Signed-off-by: Ceng23333 <441651826@qq.com>

Author: Ceng23333

INFLLMV2_FLASHATTN_LOCAL_WINDOW_PATCH_NOTES.md

@@ -19,7 +19,6 @@
 #include "ops/flash_attention.hpp"
 #include "ops/fmin.hpp"
 #include "ops/fmod.hpp"
-#include "ops/gla_attention.hpp"
 #include "ops/simple_gla_attention.hpp"
 #include "ops/simple_gla_decode_step.hpp"
 #include "ops/simple_gla_recurrent_state_append.hpp"

include/infinicore/ops.hpp

@@ -75,7 +75,6 @@
 from infinicore.ops.floor_divide import floor_divide
 from infinicore.ops.fmin import fmin
 from infinicore.ops.fmod import fmod
-from infinicore.ops.gla_attention import gla_attention
 from infinicore.ops.hypot import hypot
 from infinicore.ops.index_add import index_add
 from infinicore.ops.index_copy import index_copy
@@ -186,7 +185,6 @@
     "addbmm",
     "floor",
     "attention",
-    "gla_attention",
     "infllmv2_varlen",
     "infllmv2_kvcache",
     "simple_gla_attention",

include/infinicore/ops/gla_attention.hpp

@@ -36,7 +36,6 @@
 #include "ops/floor_divide.hpp"
 #include "ops/fmin.hpp"
 #include "ops/fmod.hpp"
-#include "ops/gla_attention.hpp"
 #include "ops/hardswish.hpp"
 #include "ops/hardtanh.hpp"
 #include "ops/huber_loss.hpp"
@@ -119,7 +118,6 @@ inline void bind(py::module &m) {
     bind_bilinear(m);
     bind_causal_softmax(m);
     bind_flash_attention(m);
-    bind_gla_attention(m);
     bind_infllmv2_attention(m);
     bind_simple_gla_attention(m);
     bind_simple_gla_decode_step(m);

python/infinicore/__init__.py

@@ -4,8 +4,6 @@
 #include "../../../tensor.h"
 #include "../cuda/embedding_kernel.cuh"
 #include "embedding_nvidia.cuh"
-#include <cstdio>
-#include <cstdlib>
 #include <cuda_runtime.h>
 
 template <typename T, typename IndexType>
@@ -59,14 +57,6 @@ INFINIOP_CUDA_KERNEL embeddingKernel(
                 // Fallback to scalar copy with __ldg
                 copyScalar<T, IndexType>(dst, src, embedding_dim);
             }
-        } else {
-            // Important for correctness debugging:
-            // If indices are out of range, write zeros instead of leaving output
-            // uninitialized (which can manifest as NaNs in later layers).
-            T *dst = output + idx * embedding_dim;
-            for (size_t i = 0; i < embedding_dim; ++i) {
-                dst[i] = T(0);
-            }
         }
     }
 }
@@ -225,22 +215,9 @@ infiniStatus_t Descriptor::calculate(
     // Check for kernel launch errors
     cudaError_t err = cudaGetLastError();
     if (err != cudaSuccess) {
-        std::fprintf(stderr, "infiniopEmbedding launch failed: %s\n", cudaGetErrorString(err));
-        std::fflush(stderr);
         return INFINI_STATUS_INTERNAL_ERROR;
     }
 
-    // Optional debug sync to surface runtime errors (misaligned, illegal access, etc.)
-    const char *debug_env = std::getenv("INFINIOP_DEBUG_EMBEDDING");
-    if (debug_env != nullptr && debug_env[0] != '\0' && debug_env[0] != '0') {
-        err = cudaStreamSynchronize(cuda_stream);
-        if (err != cudaSuccess) {
-            std::fprintf(stderr, "infiniopEmbedding stream sync failed: %s\n", cudaGetErrorString(err));
-            std::fflush(stderr);
-            return INFINI_STATUS_INTERNAL_ERROR;
-        }
-    }
-
     return INFINI_STATUS_SUCCESS;
 }
 

python/infinicore/ops/gla_attention.py

@@ -80,7 +80,7 @@ option("cutlass")
     set_description("Whether to compile cutlass for Nvidia GPU")
 option_end()
 
-if has_config("cutlass") then 
+if has_config("cutlass") then
     add_defines("ENABLE_CUTLASS_API")
 end
 
@@ -512,11 +512,11 @@ target("infinicore_cpp_api")
             local TORCH_DIR = outdata
 
             target:add(
-                "includedirs", 
-                path.join(TORCH_DIR, "include"), 
+                "includedirs",
+                path.join(TORCH_DIR, "include"),
                 path.join(TORCH_DIR, "include/torch/csrc/api/include"),
                 { public = true })
-            
+
             target:add(
                 "linkdirs",
                 path.join(TORCH_DIR, "lib"),
@@ -612,7 +612,7 @@ target("_infinicore")
         add_defines("BOOST_STACKTRACE_USE_NOOP")
     end
 
-    set_default(false)
+    set_default(true)
     add_rules("python.library", {soabi = true})
     add_packages("pybind11")
     set_languages("cxx17")